Intel® oneAPI DPC++/C++ Compiler Developer Guide and Reference

Development Reference Guides

Contents

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>
.
This extension is following SYCL 2020 backend specification. Prior APIs for interoperability with Level Zero are marked as deprecated and will be removed in the next release.

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
struct {
  ze_context_handle_t NativeHandle;
  std::vector<device> DeviceList;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
queue
ze_command_queue_handle_t
struct {
  ze_command_queue_handle_t NativeHandle;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
Deprecated in Version 3 of the Level Zero Backend Specification. 
struct {
  ze_command_queue_handle_t NativeHandle;
  device Device;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
Supported since Version 3 of the Level Zero Backend Specification.
event
ze_event_handle_t
struct {
  ze_event_handle_t NativeHandle;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
kernel_bundle
std::vector<ze_module_handle_t>
struct {
  ze_module_handle_t NativeHandle;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
kernel
ze_kernel_handle_t
struct {
  kernel_bundle<bundle_state::executable> KernelBundle;
  ze_kernel_handle_t NativeHandle;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
buffer
void *
struct {
  void *NativeHandle;
  ext::oneapi::level_zero::ownership Ownership{
      ext::oneapi::level_zero::ownership::transfer};
}
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
cl::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 
make_platform<backend::ext_oneapi_level_zero>(
    const backend_input_t<
        backend::ext_oneapi_level_zero, platform> &)
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. 
make_device<backend::ext_oneapi_level_zero>(
    const backend_input_t<
        backend::ext_oneapi_level_zero, device> &)
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. 
make_context<backend::ext_oneapi_level_zero>(
    const backend_input_t<
        backend::ext_oneapi_level_zero, context> &)
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. 
make_queue<backend::ext_oneapi_level_zero>(
    const backend_input_t<
        backend::ext_oneapi_level_zero, queue> &,
    const context &Context)
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
. 
make_event<backend::ext_oneapi_level_zero>(
    const backend_input_t<
        backend::ext_oneapi_level_zero, event> &,
    const context &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. 
make_kernel_bundle<backend::ext_oneapi_level_zero,
                   bundle_state::executable>(
    const backend_input_t<
        backend::ext_oneapi_level_zero,
        kernel_bundle<bundle_state::executable>> &,
    const context &Context)
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. 
make_kernel<backend::ext_oneapi_level_zero>(
    const backend_input_t<
        backend::ext_oneapi_level_zero, kernel> &,
    const context &Context)
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. 
make_buffer(
    const backend_input_t<backend::ext_oneapi_level_zero,
                          buffer<T, Dimensions, AllocatorT>> &,
    const context &Context)
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. 
make_buffer(
    const backend_input_t<backend::ext_oneapi_level_zero,
                          buffer<T, Dimensions, AllocatorT>> &,
    const context &Context, event AvailableEvent)
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

Product and Performance Information

1

Performance varies by use, configuration and other factors. Learn more at www.Intel.com/PerformanceIndex.