Kernel

Contents

Kernel#

oneAPI Level Zero Specification - Version 1.18.31

Kernel Functions#

zeKernelCreate#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelCreate(ze_module_handle_t hModule, const ze_kernel_desc_t *desc, ze_kernel_handle_t *phKernel)#

Create a kernel from the module.

  • Modules that have unresolved imports need to be dynamically linked before a kernel can be created from them. (See zeModuleDynamicLink)

  • The application may call this function from simultaneous threads.

  • The implementation of this function must be thread-safe.

Parameters:

  • hModule – [in] handle of the module

  • desc – [in] pointer to kernel descriptor

  • phKernel – [out] handle of the Function object

Returns:

zeKernelDestroy#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelDestroy(ze_kernel_handle_t hKernel)#

Destroys a kernel object.

  • The application must ensure the device is not currently referencing the kernel before it is deleted.

  • The implementation of this function may immediately free all Host and Device allocations associated with this kernel.

  • The application must not call this function from simultaneous threads with the same kernel handle.

  • The implementation of this function must be thread-safe.

Parameters:

hKernel – [in][release] handle of the kernel object

Returns:

zeKernelSetCacheConfig#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelSetCacheConfig(ze_kernel_handle_t hKernel, ze_cache_config_flags_t flags)#

Sets the preferred cache configuration.

  • The cache configuration will be used when a zeCommandListAppendLaunchKernel variant is called.

  • The application must not call this function from simultaneous threads with the same kernel handle.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • flags – [in] cache configuration. must be 0 (default configuration) or a valid combination of ze_cache_config_flag_t.

Returns:

zeKernelSetGroupSize#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelSetGroupSize(ze_kernel_handle_t hKernel, uint32_t groupSizeX, uint32_t groupSizeY, uint32_t groupSizeZ)#

Set group size for a kernel.

  • The group size will be used when a zeCommandListAppendLaunchKernel variant is called.

  • The application must not call this function from simultaneous threads with the same kernel handle.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • groupSizeX – [in] group size for X dimension to use for this kernel

  • groupSizeY – [in] group size for Y dimension to use for this kernel

  • groupSizeZ – [in] group size for Z dimension to use for this kernel

Returns:

zeKernelSuggestGroupSize#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelSuggestGroupSize(ze_kernel_handle_t hKernel, uint32_t globalSizeX, uint32_t globalSizeY, uint32_t globalSizeZ, uint32_t *groupSizeX, uint32_t *groupSizeY, uint32_t *groupSizeZ)#

Query a suggested group size for a kernel given a global size for each dimension.

  • This function ignores the group size that is set using zeKernelSetGroupSize.

  • The application may call this function from simultaneous threads.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • globalSizeX – [in] global width for X dimension

  • globalSizeY – [in] global width for Y dimension

  • globalSizeZ – [in] global width for Z dimension

  • groupSizeX – [out] recommended size of group for X dimension

  • groupSizeY – [out] recommended size of group for Y dimension

  • groupSizeZ – [out] recommended size of group for Z dimension

Returns:

zeKernelSuggestMaxCooperativeGroupCount#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelSuggestMaxCooperativeGroupCount(ze_kernel_handle_t hKernel, uint32_t *totalGroupCount)#

Query a suggested max group count for a cooperative kernel.

Parameters:

  • hKernel – [in] handle of the kernel object

  • totalGroupCount – [out] recommended total group count.

Returns:

zeKernelSetArgumentValue#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelSetArgumentValue(ze_kernel_handle_t hKernel, uint32_t argIndex, size_t argSize, const void *pArgValue)#

Set kernel argument for a kernel.

  • The argument values will be used when a zeCommandListAppendLaunchKernel variant is called.

  • The application must not call this function from simultaneous threads with the same kernel handle.

  • The implementation of this function should be lock-free.

  • If argument is SLM (size), then SLM size in bytes for this resource is provided as argument size and argument value is null

Parameters:

  • hKernel – [in] handle of the kernel object

  • argIndex – [in] argument index in range [0, num args - 1]

  • argSize – [in] size of argument type

  • pArgValue – [in][optional] argument value represented as matching arg type. If null then argument value is considered null.

Returns:

zeKernelSetIndirectAccess#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelSetIndirectAccess(ze_kernel_handle_t hKernel, ze_kernel_indirect_access_flags_t flags)#

Sets kernel indirect access flags.

  • The application should specify which allocations will be indirectly accessed by the kernel to allow driver to optimize which allocations are made resident

  • This function may not be called from simultaneous threads with the same Kernel handle.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • flags – [in] kernel indirect access flags

Returns:

zeKernelGetIndirectAccess#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelGetIndirectAccess(ze_kernel_handle_t hKernel, ze_kernel_indirect_access_flags_t *pFlags)#

Retrieve kernel indirect access flags.

  • This function may be called from simultaneous threads with the same Kernel handle.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • pFlags – [out] query result for kernel indirect access flags.

Returns:

zeKernelGetSourceAttributes#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelGetSourceAttributes(ze_kernel_handle_t hKernel, uint32_t *pSize, char **pString)#

Retrieve all declared kernel attributes (i.e. can be specified with attribute in runtime language).

  • This function may be called from simultaneous threads with the same Kernel handle.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • pSize – [in,out] pointer to size of string in bytes, including null-terminating character.

  • pString – [in,out][optional] pointer to application-managed character array (string data). If NULL, the string length of the kernel source attributes, including a null-terminating character, is returned in pSize. Otherwise, pString must point to valid application memory that is greater than or equal to *pSize bytes in length, and on return the pointed-to string will contain a space-separated list of kernel source attributes. Note: This API was originally intended to ship with a char *pString, however this typo was introduced. Thus the API has to stay this way for backwards compatible reasons. It can be corrected in v2.0. Suggestion is to create your own char *pString and then pass to this API with &pString.

Returns:

zeKernelGetProperties#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelGetProperties(ze_kernel_handle_t hKernel, ze_kernel_properties_t *pKernelProperties)#

Retrieve kernel properties.

  • The application may call this function from simultaneous threads.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • pKernelProperties – [in,out] query result for kernel properties.

Returns:

zeKernelGetName#

Added in version 1.0

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelGetName(ze_kernel_handle_t hKernel, size_t *pSize, char *pName)#

Retrieve kernel name from Kernel.

  • The caller can pass nullptr for pName when querying only for size.

  • The implementation will copy the kernel name into a buffer supplied by the caller.

  • The application may call this function from simultaneous threads.

  • The implementation of this function should be lock-free.

Parameters:

  • hKernel – [in] handle of the kernel object

  • pSize – [in,out] size of kernel name string, including null terminator, in bytes.

  • pName – [in,out][optional] char pointer to kernel name.

Returns:

zeKernelGetModuleHandle#

Added in version 1.18

ZE_APIEXPORT ze_result_t ZE_APICALL zeKernelGetModuleHandle(ze_kernel_handle_t hKernel, ze_module_handle_t *phModule)#

Retrieve the handle of the module from which the kernel was created.

  • The application may call this function from simultaneous threads.

Parameters:

  • hKernel – [in] handle of the kernel

  • phModule – [out] handle of the module the kernel was created from

Returns:

Kernel Enums#

ze_kernel_flags_t#

Added in version 1.0

enum ze_kernel_flag_t#

Values:

enumerator ZE_KERNEL_FLAG_FORCE_RESIDENCY#

force all device allocations to be resident during execution

enumerator ZE_KERNEL_FLAG_EXPLICIT_RESIDENCY#

application is responsible for all residency of device allocations. driver may disable implicit residency management.

enumerator ZE_KERNEL_FLAG_FORCE_UINT32#

Value marking end of ZE_KERNEL_FLAG_* ENUMs.

ze_kernel_indirect_access_flags_t#

Added in version 1.0

enum ze_kernel_indirect_access_flag_t#

Values:

enumerator ZE_KERNEL_INDIRECT_ACCESS_FLAG_HOST#

Indicates that the kernel accesses host allocations indirectly.

enumerator ZE_KERNEL_INDIRECT_ACCESS_FLAG_DEVICE#

Indicates that the kernel accesses device allocations indirectly.

enumerator ZE_KERNEL_INDIRECT_ACCESS_FLAG_SHARED#

Indicates that the kernel accesses shared allocations indirectly.

enumerator ZE_KERNEL_INDIRECT_ACCESS_FLAG_FORCE_UINT32#

Value marking end of ZE_KERNEL_INDIRECT_ACCESS_FLAG_* ENUMs.

ze_cache_config_flags_t#

Added in version 1.0

enum ze_cache_config_flag_t#

Values:

enumerator ZE_CACHE_CONFIG_FLAG_LARGE_SLM#

Large SLM size.

enumerator ZE_CACHE_CONFIG_FLAG_LARGE_DATA#

Large General Data size.

enumerator ZE_CACHE_CONFIG_FLAG_FORCE_UINT32#

Value marking end of ZE_CACHE_CONFIG_FLAG_* ENUMs.

Kernel Structures#

ze_kernel_desc_t#

Added in version 1.0

struct ze_kernel_desc_t#

Kernel descriptor.

Public Members

ze_structure_type_t stype#

[in] type of this structure

const void *pNext#

[in][optional] must be null or a pointer to an extension-specific structure (i.e. contains stype and pNext).

ze_kernel_flags_t flags#

[in] creation flags. must be 0 (default) or a valid combination of ze_kernel_flag_t; default behavior may use driver-based residency.

const char *pKernelName#

[in] null-terminated name of kernel in module

ze_kernel_properties_t#

Added in version 1.0

struct ze_kernel_properties_t#

Kernel properties.

Public Members

ze_structure_type_t stype#

[in] type of this structure

void *pNext#

[in,out][optional] must be null or a pointer to an extension-specific structure (i.e. contains stype and pNext).

uint32_t numKernelArgs#

[out] number of kernel arguments.

uint32_t requiredGroupSizeX#

[out] required group size in the X dimension, or zero if there is no required group size

uint32_t requiredGroupSizeY#

[out] required group size in the Y dimension, or zero if there is no required group size

uint32_t requiredGroupSizeZ#

[out] required group size in the Z dimension, or zero if there is no required group size

uint32_t requiredNumSubGroups#

[out] required number of subgroups per thread group, or zero if there is no required number of subgroups

uint32_t requiredSubgroupSize#

[out] required subgroup size, or zero if there is no required subgroup size

uint32_t maxSubgroupSize#

[out] maximum subgroup size

uint32_t maxNumSubgroups#

[out] maximum number of subgroups per thread group

uint32_t localMemSize#

[out] local memory size used by each thread group

uint32_t privateMemSize#

[out] private memory size allocated by compiler used by each thread

uint32_t spillMemSize#

[out] spill memory size allocated by compiler

ze_kernel_uuid_t uuid#

[out] universal unique identifier.

ze_kernel_preferred_group_size_properties_t#

Added in version 1.2

struct ze_kernel_preferred_group_size_properties_t#

Additional kernel preferred group size properties.

Public Members

ze_structure_type_t stype#

[in] type of this structure

void *pNext#

[in,out][optional] must be null or a pointer to an extension-specific structure (i.e. contains stype and pNext).

uint32_t preferredMultiple#

[out] preferred group size multiple