Intel® oneAPI DPC++/C++ Compiler Developer Guide and Reference
A newer version of this document is available. Customers should click here to go to the newest version.
Visible to Intel only — GUID: GUID-61F09D3E-612A-4DDC-83F6-423BA5A99536
Visible to Intel only — GUID: GUID-61F09D3E-612A-4DDC-83F6-423BA5A99536
Intel® oneAPI Level Zero Backend Specification
The Intel® oneAPI Level Zero (Level Zero) extension introduces a Level Zero backend for SYCL. It is built on top of Level Zero runtime enabled with the oneAPI Level Zero Specification. The Level Zero backend aims to provide the best possible performance of SYCL application on a variety of targets supported. The currently supported targets are all Intel GPUs starting with Gen9.
This extension provides a feature-test macro as described in the SYCL spec's section, Feature Test Macros. Any implementation supporting this extension must predefine the macro SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO to one of the values defined in the table below. Applications can test for the existence of this macro to see if the implementation supports this feature, or they can test the macro's value to see the extension APIs the implementation supports:
| Value | Description |
|---|---|
1 |
Initial extension version. |
2 |
Added support for the make_buffer() API. |
3 |
Added device member to backend_input_t<backend::ext_oneapi_level_zero, queue>. |
Prerequisites
The Level Zero loader and drivers must be installed on your system for the SYCL runtime to recognize and enable the Level Zero backend. Visit Intel® oneAPI DPC++/C++ Compiler System Requirements for specific instructions.User-visible Level Zero Backend Selection and Default Backend
The Level Zero backend is added to the sycl::backend enumeration with:
enum class backend { // ... ext_oneapi_level_zero, // ... };
The sections below explain the different ways the Level Zero backend can be selected.
Through an Environment Variable
The SYCL_DEVICE_FILTER environment variable limits the SYCL runtime to use only a subset of the system's devices. By using level_zero for the backend in SYCL_DEVICE_FILTER, you can select the use of Level Zero as a SYCL backend. For more information, see the Environment Variables.
Through a Programming API
The Filter Selector extension is described in SYCL Proposals: Filter Selector. Similar to how the SYCL_DEVICE_FILTER applies filtering to the entire process, this device selector can be used to select the Level Zero backend.
When neither the environment variable nor the filtering device selector is used, the implementation chooses the Level Zero backend for GPU devices supported by the installed Level Zero runtime. The serving backend for a SYCL platform can be queried with the get_backend() member function sycl::platform.
Interoperability with the Level Zero API
The sections below describe the various interoperabilities that are possible between SYCL and Level Zero. The application must include the following headers to use any of the inter-operation APIs described in this section. These headers must be included in the order shown:
#include "level_zero/ze_api.h" #include "sycl/ext/oneapi/backend/level_zero.hpp"
Mapping of SYCL Objects to Level Zero Handles
These SYCL objects encapsulate the corresponding Level Zero handles:
SYCL Type |
backend_return_t <backend::ext_oneapi_level_zero, SyclType> |
backend_input_t<backend::ext_oneapi_level_zero, SyclType> |
|---|---|---|
| platform | ze_driver_handle_t | ze_driver_handle_t |
| device | ze_device_handle_t | ze_device_handle_t |
| context | ze_context_handle_t |
|
| queue | ze_command_queue_handle_t |
Deprecated in Version 3 of the Level Zero Backend Specification. |
Supported since Version 3 of the Level Zero Backend Specification. |
||
| event | ze_event_handle_t |
|
| kernel_bundle | std::vector<ze_module_handle_t> |
|
| kernel | ze_kernel_handle_t |
|
| buffer | void * |
|
Obtaining Built-in Level Zero Handles from SYCL Objects
The sycl::get_native<backend::ext_oneapi_level_zero> free-function is how you can use a raw built-in Level Zero handle to obtain a specific SYCL object. The function is supported for the SYCL platform, device, context, queue, event and program classes. You can use a free-function defined in the sycl:: namespace instead of the member function with:
template <backend BackendName, class SyclObjectT> auto get_native(const SyclObjectT &Obj) -> backend_return_t<BackendName, SyclObjectT>
This function is supported for SYCL platform, device, context, queue, event, kernel_bundle, and kernel classes.
The sycl::get_native<backend::ext_oneapi_level_zero> free-function is not supported for the SYCL buffer class. The built-in backend object associated with the buffer can be obtained using the interop_hande class as described in the SYCL spec's section, Class interop_handle. The pointer is returned by get_native_mem<backend::ext_oneapi_level_zero> method of the interop_handle class, which is the value returned from a call to zeMemAllocShared(), zeMemAllocDevice(), or zeMemAllocHost() and not directly accessible from the host. You may need to copy your data to the host to access the data. You can get information on the type of the allocation using the type data member of the ze_memory_allocation_properties_t struct that is returned by zeMemGetAllocProperties.
Queue.submit([&](handler &CGH) { auto BufferAcc = Buffer.get_access<access::mode::write>(CGH); CGH.host_task([=](const interop_handle &IH) { void *DevicePtr = IH.get_native_mem<backend::ext_oneapi_level_zero>(BufferAcc); ze_memory_allocation_properties_t MemAllocProperties{}; ze_result_t Res = zeMemGetAllocProperties( ZeContext, DevicePtr, &MemAllocProperties, nullptr); ze_memory_type_t ZeMemType = MemAllocProperties.type; }); }).wait();
Construct a SYCL Object from a Level Zero Handle
The following free functions, defined in the sycl namespace are specialized for the Level Zero backend to allow an application to create a SYCL object that encapsulates a corresponding Level Zero object, see the table below for specific functions.
| Level Zero Interoperability Function | Description |
|---|---|
|
Constructs a SYCL platform instance from a Level Zero ze_driver_handle_t. The SYCL execution environment contains a fixed number of platforms that are counted with sycl::platform::get_platforms(). Calling this function does not create a platform, it creates a sycl::platform object that is a copy of one of the platforms from that enumeration. |
|
Constructs a SYCL device instance from a Level Zero ze_device_handle_t. The SYCL execution environment for the Level Zero backend contains a fixed number of devices that are counted with sycl::device::get_devices() and a fixed number of sub-devices that are counted with sycl::device::create_sub_devices(...). Calling this function does not create a device, it creates a sycl::device object that is a copy of one of the devices from those enumerations. |
|
Constructs a SYCL context instance from a Level Zero ze_context_handle_t. The context is created against the devices passed in a DeviceList structure member. There must be at least one device given and all the devices must be from the same SYCL platform and from the same Level Zero driver. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed built-in handle. The default behavior is to transfer the ownership to the SYCL runtime. See section Level Zero Handle Ownership and Thread-safety for details. |
|
Constructs a SYCL queue instance from a Level Zero ze_command_queue_handle_t. The Context argument must be a valid SYCL context encapsulating a Level Zero context. The Device input structure member specifies the device to create the queue against and must be in Context. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed built-in handle. The default behavior is to transfer the ownership to the SYCL runtime. See Level Zero Handle Ownership and Thread-safety for details. If the deprecated variant of backend_input_t<backend::ext_oneapi_level_zero, queue> is passed to make_queue, the queue is attached to the first device in Context. |
|
Constructs a SYCL event instance from a Level Zero ze_event_handle_t. The Context argument must be a valid SYCL context encapsulating a Level Zero context. The Level Zero event should be allocated from an event pool created in the same context. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed built-in handle. The default behavior is to transfer the ownership to the SYCL runtime. See Level Zero Handle Ownership and Thread-safety for details. |
|
Constructs a SYCL kernel_bundle instance from a Level Zero ze_module_handle_t. The Context argument must be a valid SYCL context encapsulating a Level Zero context, and the Level Zero module must be created on the same context. The Level Zero module must be fully linked (it cannot require further linking through zeModuleDynamicLink). The SYCL kernel_bundle is created in the executable state. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed built-in handle. The default behavior is to transfer the ownership to the SYCL runtime. See Level Zero Handle Ownership and Thread-safety for details. If the behavior is transfer, then the runtime is going to destroy the input Level Zero module, and the application must not have any outstanding ze_kernel_handle_t handles to the underlying ze_module_handle_t by the time this interoperability kernel_bundle destructor is called. |
|
Constructs a SYCL kernel instance from a Level Zero ze_kernel_handle_t. The KernelBundle input structure specifies the kernel_bundle corresponding to the Level Zero module from which the kernel is created. There must be exactly one Level Zero module in the KernelBundle. The Context argument must be a valid SYCL context encapsulating a Level Zero context, and the Level Zero module must be created on the same context. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed built-in handle. The default behavior is to transfer the ownership to the SYCL runtime. See Level Zero Handle Ownership and Thread-safety for details. If the behavior is transfer, then the runtime is going to destroy the input Level Zero kernel. |
|
This API is available starting with revision 2 of the Level Zero Backend Specification. Construct a SYCL buffer instance from a pointer to a Level Zero memory allocation. The pointer must be the value returned from a previous call to zeMemAllocShared(), zeMemAllocDevice(), or zeMemAllocHost(). The input SYCL context Context must be associated with a single device, matching the device used at the prior allocation. The Context argument must be a valid SYCL context encapsulating a Level Zero context, and the Level Zero memory must be allocated on the same context. Created SYCL buffer can be accessed in another contexts, not only in the provided input context. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed built-in handle. The default behavior is to transfer the ownership to the SYCL runtime. See Level Zero Handle Ownership and Thread-safety for details. If the behavior is transfer, then the runtime is going to free the input Level Zero memory allocation. Synchronization rules for a buffer that is created with this API are described in Interoperability Buffer Synchronization Rules. |
|
This API is available starting with revision 2 of the Level Zero Backend Specification. Construct a SYCL buffer instance from a pointer to a Level Zero memory allocation. Refer to make_buffer description above for semantics and restrictions. The additional AvailableEvent argument must be a valid SYCL event. The instance of the SYCL buffer class template being constructed must wait for the SYCL event parameter to signal that the memory built-in handle is ready to be used. |
Level Zero Handle Ownership and Thread-safety
The Level Zero runtime does not do reference-counting of its objects, so it is crucial to adhere to these practices of how Level Zero handles are managed. By default, the ownership is transferred to the SYCL runtime, but some interoperability API supports overriding this behavior and keeps the ownership in the application. Use this enumeration for explicit specification of the ownership:
namespace sycl { namespace ext { namespace oneapi { namespace level_zero { enum class ownership { transfer, keep }; } // namespace level_zero } // namespace oneapi } // namespace ext } // namespace sycl
- SYCL Runtime Takes Ownership (default): Whenever the application creates a SYCL object from the corresponding Level Zero handle, with one of the make_* functions, the SYCL runtime takes ownership of the Level Zero handle if no explicit ownership::keep was specified. The application must not use the Level Zero handle after the last host copy of the SYCL object is destroyed. The application must not destroy the Level Zero handle. For more information, see the SYCL Common Reference Semantics section.
- Application Keeps Ownership (explicit): If a SYCL object is created with an interoperability API explicitly asking to keep the built-in handle ownership in the application with ownership::keep, then the SYCL runtime does not take the ownership and will not destroy the Level Zero handle at the destruction of the SYCL object. The application is responsible for destroying the built-in handle when it no longer needs it, but it must not destroy the handle before the last host copy of the SYCL object is destroyed (as described in the core SYCL specification under SYCL Common Reference Semantics.
- Obtaining Built-in Handle Does Not Change Ownership: The application may call the get_native<backend::ext_oneapi_level_zero> free function on a SYCL object to retrieve the underlying Level Zero handle. Doing so does not change the ownership of the Level Zero handle. The application may not use this handle after the last host copy of the SYCL object is destroyed (as described in the core SYCL specification under SYCL Common Reference Semantics unless the SYCL object was created by the application with ownership::keep.
- Considerations for Multi-threaded Environment: The Level Zero API is not thread-safe, refer to Multithreading and Concurrency for more information. Applications must make sure that the Level Zero handles are not used simultaneously from different threads. The SYCL runtime takes ownership of the Level Zero handles and should not attempt further direct use of those handles.
Interoperability Buffer Synchronization Rules
A SYCL buffer that is constructed with this interop API uses the Level Zero memory allocation for its full lifetime. The contents of the Level Zero memory allocation are unspecified for the lifetime of the SYCL buffer. If the application modifies the contents of that Level Zero memory allocation during the lifetime of the SYCL buffer, the behavior is undefined. The initial contents of the SYCL buffer will be the initial contents of the Level Zero memory allocation at the time of the SYCL buffer's construction.
The behavior of the SYCL buffer destructor depends on the Ownership flag. As with other SYCL buffers, this behavior is triggered only when the last reference count to the buffer is dropped, as described in the SYCL spec's section, Buffer Synchronization Rules.
- If the ownership is keep (the application retains ownership of the Level Zero memory allocation), then the SYCL buffer destructor blocks until all work in queues on the buffer have completed. The contents of the buffer is not copied back to the Level Zero memory allocation.
- If the ownership is transfer (the SYCL runtime has ownership of the Level Zero memory allocation), then the SYCL buffer destructor does not need to block, even if work on the buffer has not completed. The SYCL runtime frees the Level Zero memory allocation asynchronously when it is no longer in use in queues.
Level Zero Additional Functionality
Device Information Descriptors
The Level Zero backend provides the following device information descriptors that an application can use to query information about a Level Zero device. Applications use these queries with the device::get_backend_info<>() member function as shown in the example below, which illustrates the free_memory query:
sycl::queue Queue; auto Device = Queue.get_device(); size_t freeMemory = Device.get_backend_info<sycl::ext::oneapi::level_zero::info::device::free_memory>();
New descriptors have been added as part of this specification, and are described in the table and example below.
| Descriptor | Description |
|---|---|
| sycl::ext::oneapi::level_zero::info::device::free_memory | Returns the number of bytes of free memory for the device. |
namespace sycl{ namespace ext { namespace oneapi { namespace level_zero { namespace info { namespace device { struct free_memory { using return_type = size_t; }; } // namespace device; } // namespace info } // namespace level_zero } // namespace oneapi } // namespace ext } // namespace sycl