{% import "utils.j2" as u with context -%}
{{ preamble }}
{% if fs.spec_name == "wgl" %}
#include <gloam/gl.h>
{%- endif %}
#include <gloam/{{ stem }}.h>
{% if fs.spec_name == "glx" %}
#ifdef GLOAM_PLATFORM_LINUX
{%- elif fs.spec_name == "wgl" %}
#ifdef GLOAM_PLATFORM_WINDOWS
{%- endif %}
{%- if loader and fs.spec_name != "wgl" %}
#if defined(__CYGWIN__) || defined(_WIN32)
#  ifndef WIN32_LEAN_AND_MEAN
#    define WIN32_LEAN_AND_MEAN
#  endif
#  undef APIENTRY       /* fix macro redefinition warning */
#  include <windows.h>  /* LoadLibrary, GetProcAddress */
#else
#  include <dlfcn.h> /* dlopen, dlsym, dlclose */
#endif
{%- endif %}

#include <stdlib.h>  /* calloc, free    */
#include <stddef.h>
#include <stdio.h>   /* sscanf          */
#include <string.h>  /* strlen, strncmp */

{% if fs.extensions | length > 0 -%}
#if defined(__x86_64__) || defined(__i386__) || defined(_M_IX86) || defined(_M_X64)
#  ifndef XXH_VECTOR
#    define XXH_VECTOR XXH_SSE2
#  endif
#  include <immintrin.h>
#elif defined(__aarch64__) || defined(__arm__) || defined(_M_ARM) || defined(_M_ARM64)
#  ifndef XXH_VECTOR
#    define XXH_VECTOR XXH_NEON
#  endif
#  include <arm_neon.h>
#endif
#ifndef GLOAM_EXTERNAL_XXHASH
#  define XXH_INLINE_ALL
#  define XXH_NO_STREAM
#endif
#include "xxhash.h"
{% endif %}
{% include "impl_util.j2" %}
{% include "hash_search.j2" %}
{% if fs.is_vulkan -%}
/*
 * Vulkan command scope — determines which vkGet*ProcAddr function and handle
 * is used when loading each command's function pointer.
 */
typedef enum {
    GloamCommandScopeUnknown  = 0,
    GloamCommandScopeGlobal   = 1,
    GloamCommandScopeInstance = 2,
    GloamCommandScopeDevice   = 3
} GloamCommandScope;

{% if loader %}
{% include "library.j2" %}

static const char * const gloam_vk_lib_names[] = {
#if defined(__APPLE__)
    "libvulkan.dylib", "libvulkan.1.dylib", "libMoltenVK.dylib",
#elif defined(GLOAM_PLATFORM_WINDOWS)
    "vulkan-1.dll",
#else
    "libvulkan.so.1", "libvulkan.so",
#endif
};
{% endif -%}
{% endif -%}
/* ---- Global context (zero-initialised at program startup) ---------------- */
#ifdef __cplusplus
{{ fs.context_name }} gloam_{{ fs.spec_name }}_context = {};
#else
{{ fs.context_name }} gloam_{{ fs.spec_name }}_context = { 0 };
#endif

/* ---- Function name table -------------------------------------------------
 * Command names stored as a single NUL-terminated string blob with a parallel
 * offset table for O(1) indexing. This avoids one pointer (8 bytes on 64-bit)
 * plus one relocation entry (~24 bytes in PIC builds) per command compared to
 * the traditional const char * const [] approach.
 */
static const uint32_t kFnCount_{{ fs.spec_name | spec_display }} = {{ fs.commands | length }};

static const char kFnNameData_{{ fs.spec_name | spec_display }}[] =
{%- for cmd in fs.commands %}
    /* {{ fn_name_offsets[cmd.index] | rjust(5) }} */ "{{ cmd.name }}\0"
{%- endfor %}
;
static const {{ fn_name_offset_type }} kFnNameOffsets_{{ fs.spec_name | spec_display }}[] = {
{%- for cmd in fs.commands %}
    /* {{ cmd.index | rjust(4) }} */ {{ fn_name_offsets[cmd.index] | rjust(5) }}{% if not loop.last %},{% endif %} /* {{ cmd.name }} */
{%- endfor %}
};
{% if fs.is_vulkan -%}
/* ---- Command scope table -------------------------------------------------
 * Indexed in lockstep with kFnNameOffsets_{{ fs.spec_name | spec_display }}[].
 * Each entry is the GloamCommandScope value for that slot — stored as uint8_t so
 * the whole table is one byte per command (compares well against the
 * alternatives: a parallel pointer array or a switch inside the loop).
 */
static const uint8_t kCommandScopes_{{ fs.spec_name | spec_display }}[] = {
{%- for cmd in fs.commands %}
    /* {{ cmd.index | rjust(4) }} */ {{ cmd.scope | ljust(25) }}, /* {{ cmd.name }} */
{%- endfor %}
};

{% endif -%}
{%- if fs.extensions | length > 0 -%}
/* ---- Extension hash table ------------------------------------------------
   One XXH3-64 hash per extension, in extArray index order.
   Pre-baked at generator time with the same algorithm used at load time. */
static const uint64_t kExtHashes_{{ fs.spec_name | spec_display }}[] = {
{%- for ext in fs.extensions %}
    /* {{ ext.index | rjust(4) }} */ {{ ext.hash }}ULL{% if not loop.last %},{% else %} {% endif %} /* {{ ext.name }} */
{%- endfor %}
};
{% endif %}
{% if fs.feature_pfn_ranges | length > 0 -%}
/* ---- Feature PFN range table ---------------------------------------------
 * Each entry maps one feature (by featArray index) to a contiguous run of
 * pfnArray slots. The loader iterates this table and bulk-loads the run
 * when featArray[entry.extension] is set.
 */
static const GloamPfnRange_t kFeatPfnRanges_{{ fs.spec_name | spec_display }}[] = {
{%- for r in fs.feature_pfn_ranges %}
    { {{ r.extension | rjust(4) }}, {{ r.start | rjust(4) }}, {{ r.count | rjust(4) }} }, /* {{ fs.features[r.extension].full_name }} */
{%- endfor %}
};
{% endif %}
{% if not fs.is_vulkan -%}
/* ---- PFN range helper (GL / EGL / GLX / WGL) ----------------------------
 * Walks a contiguous run of pfnArray slots and calls the plain load callback
 * (name only) for each one.
 */
static void gloam_load_pfn_range_{{ fs.spec_name }}({{ u.ctx_arg(', ') }}GloamLoadFunc getProcAddr, uint16_t start, uint16_t count)
{
    uint16_t i;
    for (i = start; i < (uint16_t)(start + count); ++i) {
        const char *pfnName = &kFnNameData_{{ fs.spec_name | spec_display }}[kFnNameOffsets_{{ fs.spec_name | spec_display }}[i]];
        context->pfnArray[i] = (void *)getProcAddr(pfnName);
    }
}
{% else -%}
/* ---- Vulkan scope-aware PFN range helper ---------------------------------
 * Loads a contiguous range of pfnArray slots, consulting kCommandScopes to
 * pick the right Vulkan proc-addr for each command:
 *   Global   → vkGetInstanceProcAddr(NULL, name)
 *   Instance → vkGetInstanceProcAddr(instance, name)  [skipped if instance is NULL]
 *   Device   → vkGetDeviceProcAddr(device, name)      [skipped if device is NULL]
 * Pass NULL for instance or device to skip commands of that scope.
 */
static void gloam_load_pfn_range_{{ fs.spec_name }}({{ u.ctx_arg(', ') }}VkInstance instance, VkDevice device, uint16_t start, uint16_t count)
{
    PFN_vkGetInstanceProcAddr gipa = context->GetInstanceProcAddr;
    PFN_vkGetDeviceProcAddr gdpa = context->GetDeviceProcAddr;
    uint16_t i;
    for (i = start; i < (uint16_t)(start + count); ++i) {
        const char *pfnName = &kFnNameData_{{ fs.spec_name | spec_display }}[kFnNameOffsets_{{ fs.spec_name | spec_display }}[i]];
        const GloamCommandScope cmdScope = (GloamCommandScope)kCommandScopes_{{ fs.spec_name | spec_display }}[i];
        GloamAPIProc pfn = NULL;
        switch (cmdScope) {
        case GloamCommandScopeGlobal:
            pfn = (GloamAPIProc)gipa(NULL, pfnName);
            break;
        case GloamCommandScopeInstance:
            if (instance)
                pfn = (GloamAPIProc)gipa(instance, pfnName);
            break;
        case GloamCommandScopeDevice:
            if (device && gdpa)
                pfn = (GloamAPIProc)gdpa(device, pfnName);
            break;
        default:
            break;
        }
        if (pfn)
            context->pfnArray[i] = (void *)pfn;
    }
}
{% endif %}
{% if alias and fs.alias_pairs | length > 0 -%}
/* ---- Alias resolution table ----------------------------------------------
 * Pairs are sorted by canonical index so same-canonical entries are
 * consecutive; the resolver processes them in a single forward pass.
 */
static const GloamAliasPair_t kAliases_{{ fs.spec_name | spec_display }}[] = {
{%- for p in fs.alias_pairs %}
    { {{ p.canonical | rjust(4) }}, {{ p.secondary | rjust(4) }} }, /* {{ fs.commands[p.canonical].name }} and {{ fs.commands[p.secondary].name }} */
{%- endfor %}
};

GLOAM_NO_INLINE static void gloam_resolve_aliases_{{ fs.spec_name }}({{ u.ctx_arg() }})
{
    uint32_t i;
    for (i = 0; i < GLOAM_ARRAYSIZE(kAliases_{{ fs.spec_name | spec_display }}); ++i) {
        uint16_t ci = kAliases_{{ fs.spec_name | spec_display }}[i].first;
        uint16_t si = kAliases_{{ fs.spec_name | spec_display }}[i].second;
        if (!context->pfnArray[ci] && context->pfnArray[si])
            context->pfnArray[ci] = context->pfnArray[si];
        else if (context->pfnArray[ci] && !context->pfnArray[si])
            context->pfnArray[si] = context->pfnArray[ci];
    }
}
{% endif %}
{% if fs.is_vulkan -%}
/* ==========================================================================
 * Vulkan enabled-path helpers (shared across per-API sections)
 * ==========================================================================
 */

/* Set featArray bits from a packed Vulkan API version (VK_MAKE_API_VERSION or
 * VK_API_VERSION_x_y). Extracts major.minor, packs as (major << 8 | minor),
 * and compares against the threshold for each feature.
 */
static void gloam_vk_apply_version({{ u.ctx_arg(', ') }}uint32_t api_version)
{
    const uint16_t version_value = (uint16_t)(
        (((api_version >> 22) & 0x7fu) << 8) | ((api_version >> 12) & 0x3ffu));
{% for feat in fs.features %}
    context->{{ feat.short_name }} = (unsigned char)(version_value >= {{ feat.packed | hex4 }});
{%- endfor %}
}

/* Load Global-scope PFNs via vkGetInstanceProcAddr(NULL, name).
 * context->GetInstanceProcAddr must already be set before calling.
 */
static void gloam_vk_load_global_pfns({{ u.ctx_arg(', ') }}PFN_vkGetInstanceProcAddr gipa)
{
    uint32_t i;
    for (i = 0; i < kFnCount_{{ fs.spec_name | spec_display }}; ++i) {
        if ((GloamCommandScope)kCommandScopes_{{ fs.spec_name | spec_display }}[i] == GloamCommandScopeGlobal) {
            const char *pfnName = &kFnNameData_{{ fs.spec_name | spec_display }}[kFnNameOffsets_{{ fs.spec_name | spec_display }}[i]];
            context->pfnArray[i] = (void *)gipa(NULL, pfnName);
        }
    }
}

{% endif -%}
/* ==========================================================================
 * Per-API sections
 * ==========================================================================
 */
{% for api in fs.apis %}
/* --------------------------------------------------------------------------
 * API: {{ api }}
 * --------------------------------------------------------------------------
 */
{%- set subset   = fs.ext_subset_indices[api] | default([]) %}
{% set extranges = fs.ext_pfn_ranges[api] | default([]) %}

{%- if fs.extensions | length > 0 and subset | length > 0 %}
/* Extension index subset for {{ api }}: extArray indices this API supports. */
static const uint16_t kExtIdx_{{ api }}[] = {
{%- for idx in subset %}
    {{ idx | rjust(4) }}, /* {{ fs.extensions[idx].name }} */
{%- endfor %}
};

/* Extension PFN range table for {{ api }}. */
static const GloamPfnRange_t kExtPfnRanges_{{ api }}[] = {
{%- for r in extranges %}
    { {{ r.extension | rjust(4) }}, {{ r.start | rjust(4) }}, {{ r.count | rjust(4) }} }, /* {{ fs.extensions[r.extension].name }} */
{%- endfor %}
{%- if extranges | length == 0 %}
    { 0, 0, 0 } /* sentinel — no extension commands for this API */
{%- endif %}
};
{% endif -%}
{#- ---- Extension enumeration and detection --------------------------------- -#}
{% if fs.extensions | length > 0 and subset | length > 0 %}
{%- if fs.spec_name in ["gl", "gles2", "glcore"] %}
/* Query driver-reported extension names, hash them, and fill *out_exts.
 * Returns 1 on success, 0 on failure. Caller must free(*out_exts).
 */
static int gloam_{{ fs.spec_name }}_get_extensions_{{ api }}({{ u.ctx_arg(', ') }}uint64_t **out_exts, uint32_t *out_num_exts)
{
#if defined(GL_ES_VERSION_3_0) || defined(GL_VERSION_3_0)
    /* Modern path: glGetIntegerv(GL_NUM_EXTENSIONS) + glGetStringi. */
    if (context->GetStringi != NULL && context->GetIntegerv != NULL) {
        GLint n = 0;
        uint32_t num_exts, i;
        uint64_t *exts;
        context->GetIntegerv(GL_NUM_EXTENSIONS, &n);
        num_exts = (uint32_t)n;
        if (num_exts == 0) {
            *out_exts = NULL;
            *out_num_exts = 0;
            return 1;
        }
        exts = (uint64_t *)calloc(num_exts, sizeof(uint64_t));
        if (!exts)
            return 0;
        for (i = 0; i < num_exts; ++i) {
            const char *name = (const char *)context->GetStringi(GL_EXTENSIONS, i);
            if (name)
                exts[i] = gloam_hash_string(name, strlen(name));
        }
        gloam_sort_hashes(exts, num_exts);
        *out_exts     = exts;
        *out_num_exts = num_exts;
        return 1;
    }
#endif
    /* Legacy path: glGetString(GL_EXTENSIONS) — space-separated string. */
    {
        const char *ext_str;
        if (!context->GetString)
            return 0;
        ext_str = (const char *)context->GetString(GL_EXTENSIONS);
        if (!ext_str)
            return 0;
        return gloam_hash_ext_string(ext_str, out_exts, out_num_exts);
    }
}
{% elif fs.spec_name == "gles1" %}
/* GLES 1.x: only glGetString(GL_EXTENSIONS), space-separated. */
static int gloam_{{ fs.spec_name }}_get_extensions_{{ api }}({{ u.ctx_arg(', ') }}uint64_t **out_exts, uint32_t *out_num_exts)
{
    const char *ext_str;

    if (!context->GetString)
        return 0;

    ext_str = (const char *)context->GetString(GL_EXTENSIONS);
    if (!ext_str)
        return 0;

    return gloam_hash_ext_string(ext_str, out_exts, out_num_exts);
}
{% elif fs.spec_name == "egl" %}
/* EGL: concatenate client extensions (EGL_NO_DISPLAY) and display
 * extensions, then hash the combined space-separated list.
 */
static int gloam_{{ fs.spec_name }}_get_extensions_{{ api }}({{ u.ctx_arg(', ') }}EGLDisplay display, uint64_t **out_exts, uint32_t *out_num_exts)
{
    const char *client_str, *display_str;
    char *concat = NULL;
    size_t client_len, display_len;
    int result;

    if (!context->QueryString)
        return 0;

    /* Client extensions live at EGL_NO_DISPLAY. */
    client_str  = (const char *)context->QueryString(EGL_NO_DISPLAY, EGL_EXTENSIONS);
    display_str = (display == EGL_NO_DISPLAY) ? "" :
                  (const char *)context->QueryString(display, EGL_EXTENSIONS);

    if (!client_str)
        return 0;
    if (!display_str)
        return 0;

    client_len  = strlen(client_str);
    display_len = strlen(display_str);

    /* Concatenate with a space separator. */
    concat = (char *)malloc(client_len + display_len + 2);
    if (!concat)
        return 0;
    memcpy(concat, client_str, client_len);
    size_t pos = client_len;
    if (display_len) {
        if (client_len && client_str[client_len - 1] != ' ')
            concat[pos++] = ' ';
        memcpy(concat + pos, display_str, display_len);
        pos += display_len;
    }
    concat[pos] = '\0';

    result = gloam_hash_ext_string(concat, out_exts, out_num_exts);
    free(concat);
    return result;
}
{% elif fs.spec_name == "glx" %}
/* GLX: glXQueryExtensionsString(display, screen) — space-separated. */
static int gloam_{{ fs.spec_name }}_get_extensions_{{ api }}({{ u.ctx_arg(', ') }}Display *display, int screen, uint64_t **out_exts, uint32_t *out_num_exts)
{
    const char *ext_str;

    if (!context->QueryExtensionsString)
        return 0;

    ext_str = (const char *)context->QueryExtensionsString(display, screen);
    if (!ext_str)
        return 0;

    return gloam_hash_ext_string(ext_str, out_exts, out_num_exts);
}
{% elif fs.spec_name == "wgl" %}
/* WGL: wglGetExtensionsStringARB / wglGetExtensionsStringEXT. */
static int gloam_{{ fs.spec_name }}_get_extensions_{{ api }}({{ u.ctx_arg(', ') }}HDC hdc, uint64_t **out_exts, uint32_t *out_num_exts)
{
    const char *ext_str = NULL;

    if (context->GetExtensionsStringARB)
        ext_str = (const char *)context->GetExtensionsStringARB(hdc);

    if (!ext_str && context->GetExtensionsStringEXT)
        ext_str = (const char *)context->GetExtensionsStringEXT();

    if (!ext_str)
        return 0;

    return gloam_hash_ext_string(ext_str, out_exts, out_num_exts);
}
{% elif fs.spec_name == "vk" %}
/* Enumerate Vulkan extensions and return a heap-allocated, sorted array of
 * XXH3-64 hashes. Only enumerates scopes not yet cached on the context:
 * instance extensions are skipped if vk_found_instance_exts is set, device
 * extensions are skipped if vk_found_device_exts is set. The caller
 * (find_extensions) uses |= when merging results so bits only accumulate.
 */
static int gloam_{{ fs.spec_name }}_get_extensions_{{ api }}({{ u.ctx_arg(', ') }} VkPhysicalDevice physical_device, uint64_t **out_exts, uint32_t *out_num_exts)
{
    uint32_t inst_count = 0, dev_count = 0, total, i;
    VkExtensionProperties *props = NULL;
    uint64_t *exts = NULL;

    if (!context->vk_found_instance_exts) {
        if (!context->EnumerateInstanceExtensionProperties)
            return 0;
        context->EnumerateInstanceExtensionProperties(NULL, &inst_count, NULL);
    }
    if (physical_device != NULL && !context->vk_found_device_exts &&
            context->EnumerateDeviceExtensionProperties != NULL)
        context->EnumerateDeviceExtensionProperties(physical_device, NULL, &dev_count, NULL);

    total = inst_count + dev_count;
    if (total == 0) {
        *out_exts = NULL;
        *out_num_exts = 0;
        return 1;
    }

    /* Single allocation covers the larger of inst / dev for the props buffer,
     * and the exact total for the hash array.
     */
    props = (VkExtensionProperties *)calloc(
        (inst_count > dev_count ? inst_count : dev_count), sizeof(*props));
    exts  = (uint64_t *)calloc(total, sizeof(uint64_t));
    if (!props || !exts) {
        free(props);
        free(exts);
        return 0;
    }

    if (inst_count) {
        context->EnumerateInstanceExtensionProperties(NULL, &inst_count, props);
        for (i = 0; i < inst_count; ++i)
            exts[i] = gloam_hash_string(props[i].extensionName,
                                        strlen(props[i].extensionName));
        context->vk_found_instance_exts = 1;
    }

    if (dev_count) {
        context->EnumerateDeviceExtensionProperties(physical_device, NULL, &dev_count, props);
        for (i = 0; i < dev_count; ++i)
            exts[inst_count + i] = gloam_hash_string(props[i].extensionName,
                                                     strlen(props[i].extensionName));
        context->vk_found_device_exts = 1;
    }

    free(props);
    gloam_sort_hashes(exts, total);
    *out_exts     = exts;
    *out_num_exts = total;
    return 1;
}
{% endif %}
/* Search pre-baked kExtHashes_{{ fs.spec_name | spec_display }} against the sorted driver hash list and set
 * extArray flags for every matching extension.
 */
{%- if fs.spec_name == "vk" %}
static int gloam_{{ fs.spec_name }}_find_extensions_{{ api }}({{ u.ctx_arg(', ') }}VkPhysicalDevice physical_device)
{
    uint64_t *exts = NULL;
    uint32_t  num_exts = 0, i;

    /* Skip if there is nothing new to enumerate. */
    if (context->vk_found_instance_exts &&
        (physical_device == NULL || context->vk_found_device_exts))
        return 1;

    if (!gloam_{{ fs.spec_name }}_get_extensions_{{ api }}(context, physical_device, &exts, &num_exts))
        return 0;

    /* |= so that bits from a previous call (instance-only) are preserved
       when this call adds device extensions. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kExtIdx_{{ api }}); ++i) {
        const uint16_t extIdx = kExtIdx_{{ api }}[i];
        context->extArray[extIdx] |= (unsigned char)gloam_hash_search(exts, num_exts, kExtHashes_{{ fs.spec_name | spec_display }}[extIdx]);
    }

    free(exts);
    return 1;
}
{% else %}
static int gloam_{{ fs.spec_name }}_find_extensions_{{ api }}({{ u.ctx_arg() }}{{ u.extra_load_params(fs.spec_name) }})
{
    uint64_t *exts = NULL;
    uint32_t  num_exts = 0, i;

    if (!gloam_{{ fs.spec_name }}_get_extensions_{{ api }}(context{% if fs.spec_name not in ["gl", "gles1", "gles2", "glcore", "vk"] %}{{ u.extra_load_args(fs.spec_name) }}{% endif %}, &exts, &num_exts))
        return 0;

    for (i = 0; i < GLOAM_ARRAYSIZE(kExtIdx_{{ api }}); ++i) {
        const uint16_t extIdx = kExtIdx_{{ api }}[i];
        context->extArray[extIdx] = (unsigned char)gloam_hash_search(exts, num_exts, kExtHashes_{{ fs.spec_name | spec_display }}[extIdx]);
    }

    free(exts);
    return 1;
}
{% endif %}
{% endif %}
{#- ---- find_core_* (GL / GLES only) --------------------------------------- -#}
{%- if fs.spec_name in ["gl", "gles1", "gles2", "glcore"] -%}
/* Parse the GL_VERSION string and set featArray entries for this API. */
static int gloam_{{ fs.spec_name }}_find_core_{{ api }}({{ u.ctx_arg() }})
{
    int i, major = 0, minor = 0;
    unsigned short version_value;
    static const char * const kPrefixes[] = {
        "OpenGL ES-CM ", "OpenGL ES-CL ", "OpenGL ES ",
        "OpenGL SC ",    "OpenGL ",       NULL
    };
    const char *version = (const char *)context->GetString(GL_VERSION);
    if (!version)
        return 0;
    for (i = 0; kPrefixes[i]; ++i) {
        const size_t len = strlen(kPrefixes[i]);
        if (strncmp(version, kPrefixes[i], len) == 0) {
            version += len;
            break;
        }
    }
    GLOAM_IMPL_UTIL_SSCANF(version, "%d.%d", &major, &minor);
    version_value = (unsigned short)((major << 8) | minor);
{% for feat in fs.features | selectattr("api", "equalto", api) %}
    context->{{ feat.short_name }} = (unsigned char)(version_value >= {{ feat.packed | hex4 }});
{%- endfor %}

    return (int)version_value;
}

{%- elif fs.spec_name == "glx" -%}
/* Query the GLX version via glXQueryVersion and set featArray entries for GLX.
 * Also returns the packed version (major << 8 | minor) for loader use.
 */
static int gloam_{{ fs.spec_name }}_find_core_{{ api }}({{ u.ctx_arg(', ') }}Display **display, int *screen)
{
    int major = 0, minor = 0;
    unsigned short version_value;
    if(*display == NULL) {
        *display = XOpenDisplay(0);
        if (*display == NULL) {
            return 0;
        }
        *screen = XScreenNumberOfScreen(XDefaultScreenOfDisplay(*display));
    }
    context->QueryVersion(*display, &major, &minor);
    version_value = (major << 8U) | minor;
    context->VERSION_1_0 = (unsigned char)(version_value >= 0x0100);
    context->VERSION_1_1 = (unsigned char)(version_value >= 0x0101);
    context->VERSION_1_2 = (unsigned char)(version_value >= 0x0102);
    context->VERSION_1_3 = (unsigned char)(version_value >= 0x0103);
    context->VERSION_1_4 = (unsigned char)(version_value >= 0x0104);
    return version_value;
}
{%- elif fs.spec_name == "vk" -%}
/* Determine the Vulkan API version and set featArray bits accordingly.
 *
 * Called on every gloamLoaderLoadVulkanContext call so that version fields can be
 * filled in incrementally:
 *   - Instance version: queried once via EnumerateInstanceVersion (VK 1.1+)
 *     or assumed 1.0. Cached in context->vk_instance_version.
 *   - Device version: queried via GetPhysicalDeviceProperties when
 *     physical_device is non-null. Cached in context->vk_device_version.
 *     Device version takes precedence over instance version when set.
 *
 * Returns the packed version (major << 8 | minor), or 0 on hard failure.
 */
static int gloam_vk_find_core({{ u.ctx_arg(', ') }}VkPhysicalDevice physical_device)
{
    /* The top 3 bits of apiVersion encode the variant — mask them off so
     * Vulkan SC (variant 1) and plain Vulkan (variant 0) compare the same.
     */
    const uint32_t kVariantMask = 0xe0000000u;
    int major = 1, minor = 0;
    uint16_t version_value;

#ifdef VK_VERSION_1_1
    /* EnumerateInstanceVersion is Global-scope and was bootstrapped in the
     * load function before we are called.
     */
    if (!context->vk_instance_version && context->EnumerateInstanceVersion != NULL) {
        VkResult r = context->EnumerateInstanceVersion(&context->vk_instance_version);
        if (r != VK_SUCCESS)
            context->vk_instance_version = 0;
        else
            context->vk_instance_version &= ~kVariantMask;
    }
    if (context->vk_instance_version) {
        major = (int)VK_VERSION_MAJOR(context->vk_instance_version);
        minor = (int)VK_VERSION_MINOR(context->vk_instance_version);
    }
#endif

    /* If a physical device is provided and we haven't cached its version yet,
     * query it. GetPhysicalDeviceProperties is Instance-scope and will have
     * been loaded from the feature ranges on the previous call.
     */
    if (!context->vk_device_version && physical_device != NULL &&
            context->GetPhysicalDeviceProperties != NULL) {
        VkPhysicalDeviceProperties props;
        context->GetPhysicalDeviceProperties(physical_device, &props);
        context->vk_device_version = props.apiVersion & ~kVariantMask;
    }
    /* Device version is the authoritative cap; prefer it over instance version. */
    if (context->vk_device_version) {
        major = (int)VK_VERSION_MAJOR(context->vk_device_version);
        minor = (int)VK_VERSION_MINOR(context->vk_device_version);
    }

    version_value = (uint16_t)((major << 8) | minor);
{% for feat in fs.features %}
    context->{{ feat.short_name }} = (unsigned char)(version_value >= {{ feat.packed | hex4 }});
{%- endfor %}

    return (int)version_value;
}

{%- elif fs.spec_name == "wgl" -%}
/* Faux version detection on WGL */
static int gloam_{{ fs.spec_name }}_find_core_{{ api }}({{ u.ctx_arg() }})
{
    int major = 1, minor = 0;
    unsigned short version_value;
    version_value = (unsigned short)((major << 8) | minor);
{% for feat in fs.features | selectattr("api", "equalto", api) %}
    context->{{ feat.short_name }} = (unsigned char)(version_value >= {{ feat.packed | hex4 }});
{%- endfor %}

    return (int)version_value;
}

{%- elif fs.spec_name == "egl" -%}
/* Parse the EGL version string from eglQueryString(display, EGL_VERSION). */
static int gloam_{{ fs.spec_name }}_find_core_{{ api }}({{ u.ctx_arg(', ') }}EGLDisplay display)
{
    int major = 0, minor = 0;
    unsigned short version_value;
    const char *version;
    if (!context->QueryString)
        return 0;
    version = (const char *)context->QueryString(display, EGL_VERSION);
    if (!version)
        return 0;
    GLOAM_IMPL_UTIL_SSCANF(version, "%d.%d", &major, &minor);
    version_value = (unsigned short)((major << 8) | minor);
{% for feat in fs.features | selectattr("api", "equalto", api) %}
    context->{{ feat.short_name }} = (unsigned char)(version_value >= {{ feat.packed | hex4 }});
{%- endfor %}

    return (int)version_value;
}
{% endif -%}
{# ---- Load function ------------------------------------------------------- #}
{%- if fs.spec_name in ["gl", "gles1", "gles2", "glcore"] %}

int gloamLoad{{ api | api_display }}Context({{ u.ctx_arg(', ') }}GloamLoadFunc getProcAddr)
{
    int version;
    uint32_t i;
    GLOAM_UNUSED(kFnCount_{{ fs.spec_name | spec_display }});

    memset(context, 0, sizeof(*context));

    /* Bootstrap: glGetString must be loaded before find_core can run. */
{%- for cmd in fs.commands %}{% if cmd.name == "glGetString" %}
    context->{{ cmd.short_name }} = ({{ cmd.pfn_type }})getProcAddr("glGetString");
    if (!context->{{ cmd.short_name }})
        return 0;
{% endif %}{%- endfor %}
    version = gloam_{{ fs.spec_name }}_find_core_{{ api }}(context);
    if (!version)
        return 0;

    /* Load PFNs for each enabled feature via the range table. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        if (context->featArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, getProcAddr, r->start, r->count);
    }
{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_{{ fs.spec_name }}_find_extensions_{{ api }}(context))
        return 0;

    /* Load PFNs for each detected extension via the range table. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, getProcAddr, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    return version;
}

int gloamLoad{{ api | api_display }}(GloamLoadFunc getProcAddr)
{
    return gloamLoad{{ api | api_display }}Context(&gloam_{{ fs.spec_name }}_context, getProcAddr);
}
{% elif fs.spec_name == "egl" %}
int gloamLoad{{ api | api_display }}Context({{ u.ctx_arg(', ') }}EGLDisplay display, GloamLoadFunc getProcAddr)
{
    int version;
    uint32_t i;
    GLOAM_UNUSED(kFnCount_{{ fs.spec_name | spec_display }});

    memset(context, 0, sizeof(*context));

    /* Bootstrap: QueryString must be loaded before version detection. */
{%- for cmd in fs.commands %}{% if cmd.name == "eglGetString" or cmd.name == "eglQueryString" %}
    context->{{ cmd.short_name }} = ({{ cmd.pfn_type }})getProcAddr("{{ cmd.name }}");
{% endif %}{%- endfor %}
    version = gloam_{{ fs.spec_name }}_find_core_{{ api }}(context, display);
    if (!version)
        return 0;

    /* Load PFNs for each enabled feature via the range table. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        if (context->featArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, getProcAddr, r->start, r->count);
    }
{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_{{ fs.spec_name }}_find_extensions_{{ api }}(context, display))
        return 0;

    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, getProcAddr, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    return version;
}

int gloamLoad{{ api | api_display }}(EGLDisplay display, GloamLoadFunc getProcAddr)
{
    return gloamLoad{{ api | api_display }}Context(&gloam_{{ fs.spec_name }}_context, display, getProcAddr);
}
{% elif fs.spec_name == "glx" %}
int gloamLoad{{ api | api_display }}Context({{ u.ctx_arg(', ') }}Display *display, int screen, GloamLoadFunc getProcAddr)
{
    int version;
    uint32_t i;
    memset(context, 0, sizeof(*context));
{% for cmd in fs.commands %}{% if cmd.name == "glXQueryVersion" %}
    context->{{ cmd.short_name }} = ({{ cmd.pfn_type }})getProcAddr("{{ cmd.name }}");
{% endif %}{%- endfor %}
    version = gloam_{{ fs.spec_name }}_find_core_{{ api }}(context, &display, &screen);
    if (!version)
        return 0;

    /* Load all PFNs upfront. */
    for (i = 0; i < kFnCount_{{ fs.spec_name | spec_display }}; ++i)
        context->pfnArray[i] = (void *)getProcAddr((kFnNameData_{{ fs.spec_name | spec_display }} + kFnNameOffsets_{{ fs.spec_name | spec_display }}[i]));

    /* Mark features based on PFN availability. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        uint16_t j; int ok = 1;
        for (j = r->start; j < (uint16_t)(r->start + r->count); ++j)
            ok &= (context->pfnArray[j] != NULL);
        if (ok)
            context->featArray[r->extension] = 1;
    }
{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_{{ fs.spec_name }}_find_extensions_{{ api }}(context, display, screen))
        return 0;

    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, getProcAddr, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    return 1;
}

int gloamLoad{{ api | api_display }}(Display *display, int screen, GloamLoadFunc getProcAddr)
{
    return gloamLoad{{ api | api_display }}Context(&gloam_{{ fs.spec_name }}_context, display, screen, getProcAddr);
}
{% elif fs.spec_name == "wgl" -%}
int gloamLoad{{ api | api_display }}Context({{ u.ctx_arg(', ') }}HDC hdc, GloamLoadFunc getProcAddr)
{
    int version;
    uint32_t i;
    GLOAM_UNUSED(kFnCount_{{ fs.spec_name | spec_display }});

    memset(context, 0, sizeof(*context));

    /* WGL mandatory extensions must be loaded first for extension detection. */
{%- for cmd in fs.commands %}
{%- if cmd.name in ["wglGetExtensionsStringARB", "wglGetExtensionsStringEXT"] %}
    context->{{ cmd.short_name }} = ({{ cmd.pfn_type }})getProcAddr("{{ cmd.name }}");
{%- endif %}
{%- endfor %}

    version = gloam_{{ fs.spec_name }}_find_core_{{ api }}(context);
    if (!version)
        return 0;

    /* Load all PFNs upfront. */
    for (i = 0; i < kFnCount_{{ fs.spec_name | spec_display }}; ++i)
        context->pfnArray[i] = (void *)getProcAddr((kFnNameData_{{ fs.spec_name | spec_display }} + kFnNameOffsets_{{ fs.spec_name | spec_display }}[i]));

    /* Mark features based on PFN availability. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        uint16_t j; int ok = 1;
        for (j = r->start; j < (uint16_t)(r->start + r->count); ++j)
            ok &= (context->pfnArray[j] != NULL);
        if (ok)
            context->featArray[r->extension] = 1;
    }

{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_{{ fs.spec_name }}_find_extensions_{{ api }}(context, hdc))
        return 0;

    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, getProcAddr, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    return 1;
}

int gloamLoad{{ api | api_display }}(HDC hdc, GloamLoadFunc getProcAddr)
{
    return gloamLoad{{ api | api_display }}Context(&gloam_{{ fs.spec_name }}_context, hdc, getProcAddr);
}
{% elif fs.spec_name == "vk" %}
{% if fs.extensions | length > 0 and subset | length > 0 %}
/* Hash user-provided extension name arrays, search against kExtHashes, and
 * set the corresponding extArray bits. Uses |= so bits from a previous
 * call (e.g. instance extensions) are preserved when device extensions are
 * added.
 */
static int gloam_vk_apply_extensions_{{ api }}({{ u.ctx_arg(', ') }}uint32_t num_exts, const char *const *ext_names)
{
    uint32_t i;
    uint64_t *exts;

    if (num_exts == 0)
        return 1;

    exts = (uint64_t *)calloc(num_exts, sizeof(uint64_t));
    if (!exts)
        return 0;

    for (i = 0; i < num_exts; ++i)
        exts[i] = gloam_hash_string(ext_names[i], strlen(ext_names[i]));
    gloam_sort_hashes(exts, num_exts);

    for (i = 0; i < GLOAM_ARRAYSIZE(kExtIdx_{{ api }}); ++i) {
        const uint16_t extIdx = kExtIdx_{{ api }}[i];
        context->extArray[extIdx] |= (unsigned char)gloam_hash_search(exts, num_exts, kExtHashes_{{ fs.spec_name | spec_display }}[extIdx]);
    }

    free(exts);
    return 1;
}
{% endif %}
/* gloamVulkanDiscoverContext — canonical Vulkan discovery loader.
 *
 * May be called multiple times on the same context as the application
 * progresses through Vulkan initialisation:
 *   1. (NULL, NULL, NULL) — loads Global-scope functions; detects instance
 *      extensions so the caller can choose which to enable.
 *   2. (instance, NULL, NULL) — loads Global + Instance-scope functions;
 *      detects device extensions given the now-live instance.
 *   3. (instance, physical_device, device) — loads all scopes; Device-scope
 *      commands get the fast vkGetDeviceProcAddr path.
 *
 * Each call is additive: context state from previous calls is preserved and
 * only new or better-scoped slots are updated.
 *
 * Requires context->GetInstanceProcAddr to be set before the first call.
 * context->GetDeviceProcAddr is resolved automatically when an instance is
 * provided.
 */
static int gloamVulkanDiscoverContext({{ u.ctx_arg(', ') }}VkInstance instance, VkPhysicalDevice physical_device, VkDevice device)
{
    int version;
    uint32_t i;
    GLOAM_UNUSED(kFnCount_{{ fs.spec_name | spec_display }});
    GLOAM_UNUSED(gloam_hash_ext_string);

    if (!context->GetInstanceProcAddr)
        return 0;

    /* Resolve vkGetDeviceProcAddr through the instance when available. */
    if (instance && !context->GetDeviceProcAddr)
        context->GetDeviceProcAddr =
            (PFN_vkGetDeviceProcAddr)context->GetInstanceProcAddr(
                instance, "vkGetDeviceProcAddr");

    /* Bootstrap: EnumerateInstanceVersion is Global-scope — it can be loaded
     * before any instance exists — and must be available before find_core.
     */
#ifdef VK_VERSION_1_1
    if (!context->EnumerateInstanceVersion)
        context->EnumerateInstanceVersion = (PFN_vkEnumerateInstanceVersion)
            context->GetInstanceProcAddr(NULL, "vkEnumerateInstanceVersion");
#endif

    version = gloam_vk_find_core(context, physical_device);
    if (!version)
        return 0;

    /* Load PFNs for every enabled feature via the range table. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        if (context->featArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, device, r->start, r->count);
    }
{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_{{ fs.spec_name }}_find_extensions_{{ api }}(context, physical_device))
        return 0;

    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, device, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    return version;
}

/* ==========================================================================
 * Vulkan enabled API — phased loading
 * ==========================================================================
 */
{% if loader %}
/* gloamVulkanInitializeContext — Phase 0: open libvulkan, load Global-scope
 * PFNs (vkCreateInstance, vkEnumerateInstance*, vkGetInstanceProcAddr).
 * If library_handle is non-NULL, use it without taking ownership.
 * If NULL, dlopen the platform default and take ownership.
 */
int gloamVulkanInitializeContext({{ u.ctx_arg(', ') }}void *library_handle)
{
    PFN_vkGetInstanceProcAddr gipa;

    /* In case there's an open handle in the context that we own, we should
     * tear that down now before zeroing the context. Use Finalize to do that.
     */
    gloamVulkanFinalizeContext(context);

    if (library_handle) {
        context->gloam_loader_handle = library_handle;
        context->gloam_loader_owns_handle = 0;
    } else {
        context->gloam_loader_handle = gloam_open_library(
            gloam_vk_lib_names, GLOAM_ARRAYSIZE(gloam_vk_lib_names));
        if (!context->gloam_loader_handle)
            return 0;
        context->gloam_loader_owns_handle = 1;
    }

    gipa = (PFN_vkGetInstanceProcAddr)gloam_dlsym(
        context->gloam_loader_handle, "vkGetInstanceProcAddr");
    if (!gipa) {
        if (context->gloam_loader_owns_handle)
            gloam_dlclose(context->gloam_loader_handle);
        memset(context, 0, sizeof(*context));
        return 0;
    }

    /* Store GIPA in the context, then load Global-scope PFNs. */
    context->GetInstanceProcAddr = gipa;
    gloam_vk_load_global_pfns(context, gipa);
    return 1;
}

int gloamVulkanInitialize(void *library_handle)
{
    return gloamVulkanInitializeContext(&gloam_{{ fs.spec_name }}_context, library_handle);
}
{% endif %}
/* gloamVulkanInitializeCustomContext — Phase 0 variant: caller provides
 * vkGetInstanceProcAddr directly. No library handle management.
 */
void gloamVulkanInitializeCustomContext({{ u.ctx_arg(', ') }}PFN_vkGetInstanceProcAddr getInstanceProcAddr)
{
    gloamVulkanFinalizeContext(context);
    context->GetInstanceProcAddr = getInstanceProcAddr;
    gloam_vk_load_global_pfns(context, getInstanceProcAddr);
}

void gloamVulkanInitializeCustom(PFN_vkGetInstanceProcAddr getInstanceProcAddr)
{
    gloamVulkanInitializeCustomContext(&gloam_{{ fs.spec_name }}_context, getInstanceProcAddr);
}

/* gloamVulkanLoadInstanceContext — Phase 1: load PFNs for core features and
 * instance extensions. Instance and Global scope commands are loaded via
 * vkGetInstanceProcAddr. Device-scope commands in instance extensions (e.g.
 * VK_EXT_debug_utils) are skipped here and picked up by LoadDevice.
 *
 * Sets featArray from api_version. Sets extArray for enabled instance
 * extensions. Runs alias resolution. Returns 1 on success.
 */
int gloamVulkanLoadInstanceContext({{ u.ctx_arg(', ') }}VkInstance instance, uint32_t api_version,
                                   uint32_t num_instance_extensions, const char *const *instance_extensions)
{
    uint32_t i;

    if (!context->GetInstanceProcAddr)
        return 0;

    gloam_vk_apply_version(context, api_version);
{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_vk_apply_extensions_{{ api }}(context,
            num_instance_extensions, instance_extensions))
        return 0;
{% endif %}
    /* Load PFNs for every enabled feature (instance + global scope). */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        if (context->featArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, NULL, r->start, r->count);
    }
{% if fs.extensions | length > 0 and subset | length > 0 %}
    /* Load PFNs for every enabled extension (instance + global scope). */
    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, NULL, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    context->vk_loaded_instance = instance;
    GLOAM_UNUSED(num_instance_extensions);
    GLOAM_UNUSED(instance_extensions);
    return 1;
}

int gloamVulkanLoadInstance(VkInstance instance, uint32_t api_version, uint32_t num_instance_extensions, const char *const *instance_extensions)
{
    return gloamVulkanLoadInstanceContext(&gloam_{{ fs.spec_name }}_context, instance,
        api_version, num_instance_extensions, instance_extensions);
}

/* gloamVulkanLoadPhysicalDeviceExtensionsContext — Phase 1.5 (optional):
 * pre-load PFNs for device extensions before the VkDevice exists.
 *
 * Device extensions may provide Instance-scope query functions (e.g.
 * vkGetPhysicalDeviceFragmentShadingRatesKHR) that applications need to call
 * while selecting which device extensions to enable. This function loads
 * Global and Instance-scope commands for the listed extensions via
 * vkGetInstanceProcAddr. Device-scope commands are skipped because no
 * VkDevice exists yet.
 *
 * Does NOT set extArray — the GLOAM_VK_* support macros remain unchanged.
 * Call gloamVulkanLoadDevice after device creation for full extension
 * loading.
 */
void gloamVulkanLoadPhysicalDeviceExtensionsContext({{ u.ctx_arg(', ') }} uint32_t num_device_extensions, const char *const *device_extensions)
{
{% if fs.extensions | length > 0 and subset | length > 0 %}
    VkInstance instance = context->vk_loaded_instance;
    uint32_t j;

    if (!context->GetInstanceProcAddr || !instance)
        return;

    for (j = 0; j < num_device_extensions; ++j) {
        const uint64_t hash = gloam_hash_string(device_extensions[j],
                                                 strlen(device_extensions[j]));
        uint32_t i;
        for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
            const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
            if (kExtHashes_{{ fs.spec_name | spec_display }}[r->extension] == hash)
                gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, NULL,
                    r->start, r->count);
        }
    }
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
{% endif %}
    GLOAM_UNUSED(num_device_extensions);
    GLOAM_UNUSED(device_extensions);
}

void gloamVulkanLoadPhysicalDeviceExtensions(uint32_t num_device_extensions, const char *const *device_extensions)
{
    gloamVulkanLoadPhysicalDeviceExtensionsContext(&gloam_{{ fs.spec_name }}_context,
        num_device_extensions, device_extensions);
}

void gloamVulkanLoadPhysicalDeviceExtensionContext({{ u.ctx_arg(', ') }} const char *device_extension)
{
    gloamVulkanLoadPhysicalDeviceExtensionsContext(context, 1, &device_extension);
}

void gloamVulkanLoadPhysicalDeviceExtension(const char *device_extension)
{
    gloamVulkanLoadPhysicalDeviceExtensionsContext(&gloam_{{ fs.spec_name }}_context,
        1, &device_extension);
}

/* gloamVulkanLoadDeviceContext — Phase 2: load PFNs for device extensions and
 * reload Device-scope core PFNs via vkGetDeviceProcAddr (bypassing the loader
 * trampoline).
 *
 * Extensions may contain commands of any scope. The unified loader dispatches
 * each command to the correct proc-addr function based on scope: Instance-scope
 * via vkGetInstanceProcAddr, Device-scope via vkGetDeviceProcAddr.
 *
 * Updates featArray from the device's api_version. Sets extArray for enabled
 * device extensions. Runs alias resolution. Returns 1 on success.
 */
int gloamVulkanLoadDeviceContext({{ u.ctx_arg(', ') }}VkDevice device, VkPhysicalDevice physical_device,
                                 uint32_t num_device_extensions, const char *const *device_extensions)
{
    uint32_t i;
    VkPhysicalDeviceProperties props;
    VkInstance instance = context->vk_loaded_instance;

    if (instance && !context->GetDeviceProcAddr)
        context->GetDeviceProcAddr =
            (PFN_vkGetDeviceProcAddr)context->GetInstanceProcAddr(
                instance, "vkGetDeviceProcAddr");

    if (!context->GetDeviceProcAddr || !device)
        return 0;

    context->GetPhysicalDeviceProperties(physical_device, &props);
    gloam_vk_apply_version(context, props.apiVersion);
{% if fs.extensions | length > 0 and subset | length > 0 %}
    if (!gloam_vk_apply_extensions_{{ api }}(context,
            num_device_extensions, device_extensions))
        return 0;
{% endif %}
    /* Reload PFNs for enabled features — Device-scope gets the fast path. */
    for (i = 0; i < GLOAM_ARRAYSIZE(kFeatPfnRanges_{{ fs.spec_name | spec_display }}); ++i) {
        const GloamPfnRange_t *r = &kFeatPfnRanges_{{ fs.spec_name | spec_display }}[i];
        if (context->featArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, device, r->start, r->count);
    }
{% if fs.extensions | length > 0 and subset | length > 0 %}
    /* Load PFNs for every enabled extension (all scopes). */
    for (i = 0; i < GLOAM_ARRAYSIZE(kExtPfnRanges_{{ api }}); ++i) {
        const GloamPfnRange_t *r = &kExtPfnRanges_{{ api }}[i];
        if (context->extArray[r->extension])
            gloam_load_pfn_range_{{ fs.spec_name }}(context, instance, device, r->start, r->count);
    }
{% endif -%}
{% if alias and fs.alias_pairs | length > 0 %}
    gloam_resolve_aliases_{{ fs.spec_name }}(context);
{% endif %}
    context->vk_loaded_device = device;
    GLOAM_UNUSED(num_device_extensions);
    GLOAM_UNUSED(device_extensions);
    return 1;
}

int gloamVulkanLoadDevice(VkDevice device, VkPhysicalDevice physical_device, uint32_t num_device_extensions, const char *const *device_extensions)
{
    return gloamVulkanLoadDeviceContext(&gloam_{{ fs.spec_name }}_context, device,
        physical_device, num_device_extensions, device_extensions);
}

uint32_t gloamVulkanGetInstanceVersionContext({{ u.ctx_arg() }})
{
#if defined(VK_VERSION_1_1)
    uint32_t apiVersion = 0;
    if (context->EnumerateInstanceVersion && context->EnumerateInstanceVersion(&apiVersion) == VK_SUCCESS)
        return apiVersion;
#endif

    if (context->CreateInstance)
        return VK_API_VERSION_1_0;

    return 0;
}

uint32_t gloamVulkanGetInstanceVersion(void)
{
    return gloamVulkanGetInstanceVersionContext(&gloam_{{ fs.spec_name }}_context);
}

VkDevice gloamVulkanGetLoadedDeviceContext({{ u.ctx_arg() }})
{
    return context->vk_loaded_device;
}

VkDevice gloamVulkanGetLoadedDevice(void)
{
    return gloamVulkanGetLoadedDeviceContext(&gloam_{{ fs.spec_name }}_context);
}

VkInstance gloamVulkanGetLoadedInstanceContext({{ u.ctx_arg() }})
{
    return context->vk_loaded_instance;
}

VkInstance gloamVulkanGetLoadedInstance(void)
{
    return gloamVulkanGetLoadedInstanceContext(&gloam_{{ fs.spec_name }}_context);
}

/* gloamVulkanFinalizeContext — close library handle if gloam owns it, zero
 * the context.
 */
void gloamVulkanFinalizeContext({{ u.ctx_arg() }})
{
{%- if loader %}
    if (context->gloam_loader_owns_handle && context->gloam_loader_handle)
        gloam_dlclose(context->gloam_loader_handle);
{%- endif %}
    memset(context, 0, sizeof(*context));
}

void gloamVulkanFinalize(void)
{
    gloamVulkanFinalizeContext(&gloam_{{ fs.spec_name }}_context);
}
{% endif %}
{% endfor %}{# end for api in fs.apis #}

{%- if loader -%}
/* ==========================================================================
 * Built-in loader (--loader)
 * ==========================================================================
 */
{% include "loader.j2" -%}
{% endif -%}
{% if fs.spec_name == "glx" -%}
#endif /* GLOAM_PLATFORM_LINUX */
{% elif fs.spec_name == "wgl" -%}
#endif /* GLOAM_PLATFORM_WINDOWS */
{% endif %}
