From 59d3f0356106c5fc5b24a9e2cc63b46647123ea7 Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Mon, 13 Sep 2021 16:14:30 -0700 Subject: [PATCH 1/8] [SYCL] Updated according to SYCL 2020 API Signed-off-by: Sergey V Maslov --- .../LevelZeroBackend/LevelZeroBackend.md | 196 +++++++++++++++--- 1 file changed, 164 insertions(+), 32 deletions(-) mode change 100644 => 100755 sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md old mode 100644 new mode 100755 index 78bb5392e0cc6..acddd841458d9 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -1,4 +1,4 @@ -# Level-Zero backend specification +# Level-Zero backend specification ## 1. Introduction @@ -7,8 +7,8 @@ It is built on top of Level-Zero runtime enabled with [Level-Zero API](https://s The Level-Zero backend is aimed 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. -NOTE: This specification is a draft. While describing the currently implemented behaviors it is known to be not complete nor exhaustive. - We shall continue to add more information, e.g. explain general mapping of SYCL programming model to Level-Zero API. +NOTE: This extension is following SYCL 2020 backend specification. Prior API for interoperability with Level-Zero is marked + as deprecated and will be removed in the next release. ## 2. Prerequisites @@ -17,7 +17,7 @@ For further details see + +SyclType + + +``` C++ +backend_return_t +``` + + + +``` C++ +backend_input_t +``` + + +platform +
ze_driver_handle_t
+
ze_driver_handle_t
+ +device +
ze_device_handle_t
+
ze_device_handle_t
+ +context +
ze_context_handle_t
+ + +``` C++ +struct { + ze_context_handle_t NativeHandle; + std::vector DeviceList; + ext::oneapi::level_zero::ownership Ownership; +} +``` + + +queue +
ze_command_queue_handle_t
+ + +``` C++ +struct { + ze_command_queue_handle_t NativeHandle; + ext::oneapi::level_zero::ownership Ownership; +} +``` + + +event +
ze_event_handle_t
+ + +``` C++ +struct { + ze_event_handle_t NativeHandle; + ext::oneapi::level_zero::ownership Ownership; +} +``` + + +kernel_bundle + + +``` C++ +std::vector +``` + +
ze_module_handle_t
+ + ### 4.2 Obtaining of native Level-Zero handles from SYCL objects -The ```get_native()``` member function is how a raw native Level-Zero handle can be obtained -for a specific SYCL object. It is currently supported for SYCL ```platform```, ```device```, ```context```, ```queue```, ```event``` -and ```program``` classes. There is also a free-function defined in ```cl::sycl``` namespace that can be used instead of the member function: +The ```sycl::get_native``` free-function is how a raw native Level-Zero handle can be obtained +for a specific SYCL object. ``` C++ template -auto get_native(const SyclObjectT &Obj) -> - typename interop::type; +auto get_native(const SyclObjectT &Obj) + -> backend_return_t ``` +It is currently supported for SYCL ```platform```, ```device```, ```context```, ```queue```, ```event``` +and ```kernel_bundle``` classes. + ### 4.3 Construct a SYCL object from a Level-Zero handle -The following free functions defined in the ```cl::sycl::ext::oneapi::level_zero``` namespace allow an application to create -a SYCL object that encapsulates a corresponding Level-Zero object: - -| Level-Zero interoperability function |Description| -|-------------|:------------| -|``` make(ze_driver_handle_t);```|Constructs a SYCL platform instance from a Level-Zero ```ze_driver_handle_t```.| -|``` make(const platform &, ze_device_handle_t);```|Constructs a SYCL device instance from a Level-Zero ```ze_device_handle_t```. The platform argument gives a SYCL platform, encapsulating a Level-Zero driver supporting the passed Level-Zero device.| -|``` make(const std::vector &, ze_context_handle_t, ownership = transfer);```| Constructs a SYCL context instance from a Level-Zero ```ze_context_handle_t```. The context is created against the devices passed in. There must be at least one device given and all the devices must be from the same SYCL platform and thus from the same Level-Zero driver. The ```ownership``` argument specifies if the SYCL runtime should take ownership of the passed native handle. The default behavior is to transfer the ownership to the SYCL runtime. See section 4.4 for details.| -|``` make(const context &, ze_command_queue_handle_t, ownership = transfer);```| 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 queue is attached to the first device in the passed SYCL context. The ```ownership``` argument specifies if the SYCL runtime should take ownership of the passed native handle. The default behavior is to transfer the ownership to the SYCL runtime. See section 4.4 for details.| -|``` make(const context &, ze_event_handle_t, ownership = transfer);```| 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``` argument specifies if the SYCL runtime should take ownership of the passed native handle. The default behavior is to transfer the ownership to the SYCL runtime. See section 4.4 for details.| -|``` make(const context &, ze_module_handle_t);```| Constructs a SYCL program instance from a Level-Zero ```ze_module_handle_t```. The context argument must be a valid SYCL context encapsulating a Level-Zero context. The Level-Zero module must be fully linked (i.e. not require further linking through [```zeModuleDynamicLink```](https://spec.oneapi.com/level-zero/latest/core/api.html?highlight=zemoduledynamiclink#_CPPv419zeModuleDynamicLink8uint32_tP18ze_module_handle_tP28ze_module_build_log_handle_t)), and thus the SYCL program is created in the "linked" state.| +The following free functions defined in the ```sycl``` namespace are specialized for Level-Zero backend to allow +an application to create a SYCL object that encapsulates a corresponding Level-Zero object: + + + + + + + + + + + + + + + + + + + + + + + + +
Level-Zero interoperability function Description
+ +``` C++ +make_platform( + const backend_input_t< + backend::ext_oneapi_level_zero, platform> &) +``` +Constructs a SYCL platform instance from a Level-Zero ze_driver_handle_t
+ +``` C++ +make_device( + const backend_input_t< + backend::ext_oneapi_level_zero, device> &) +``` +Constructs a SYCL device instance from a Level-Zero ze_device_handle_t.
+ +``` C++ +make_context( + 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 DeviceList structure member. There must be at least one device given and all the devices must be from the same SYCL platform and thus from the same Level-Zero driver. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed native handle. The default behavior is to transfer the ownership to the SYCL runtime. See section 4.4 for details.
+ +``` C++ +make_queue( + 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 queue is attached to the first device in the passed SYCL context. The Ownership input structure member specifies if the SYCL runtime should take ownership of the passed native handle. The default behavior is to transfer the ownership to the SYCL runtime. See section 4.4 for details.
+ +``` C++ +make_event( + 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 native handle. The default behavior is to transfer the ownership to the SYCL runtime. See section 4.4 for details.
+ +``` C++ +make_kernel_bundle( + const backend_input_t< + backend::ext_oneapi_level_zero, + kernel_bundle> &, + 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. The Level-Zero module must be fully linked (i.e. not require further linking through [zeModuleDynamicLink](https://spec.oneapi.com/level-zero/latest/core/api.html?highlight=zemoduledynamiclink#_CPPv419zeModuleDynamicLink8uint32_tP18ze_module_handle_tP28ze_module_build_log_handle_t)), and thus the SYCL kernel_bundle is created in the "executable" state.
NOTE: We shall consider adding other interoperability as needed, if possible. @@ -116,7 +247,7 @@ enum class ownership { transfer, keep }; #### 4.4.1 SYCL runtime takes ownership (default) -Whenever the application creates a SYCL object from the corresponding Level-Zero handle via one of the ```make()``` functions, +Whenever the application creates a SYCL object from the corresponding Level-Zero handle via 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 ( as described in the core SYCL specification under "Common reference semantics"), and the application must not @@ -132,7 +263,7 @@ handle before the last host copy of the SYCL object is destroyed (as described i #### 4.4.3 Obtaining native handle does not change ownership -The application may call the ```get_native()``` member function of a SYCL object to retrieve the underlying Level-Zero handle. +The application may call the ```get_native``` free function on a SYCL object to retrieve the underlying Level-Zero handle. Doing so does not change the ownership of the the Level-Zero handle. Therefore, 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 "Common reference semantics") unless the SYCL object was created by the application with ```ownership::keep```. @@ -197,3 +328,4 @@ struct free_memory { |4|2021-07-06|Rehana Begam|Introduced explicit ownership for queue |5|2021-07-25|Sergey Maslov|Introduced SYCL interop for events |6|2021-08-30|Dmitry Vodopyanov|Updated according to SYCL 2020 reqs for extensions +|7|2021-09-13|Sergey Maslov|Updated according to SYCL 2020 standard From b1172030372d519522da23679071027c85a2247b Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Tue, 14 Sep 2021 12:35:46 -0700 Subject: [PATCH 2/8] address review comments Signed-off-by: Sergey V Maslov --- .../extensions/LevelZeroBackend/LevelZeroBackend.md | 10 +++++++--- 1 file changed, 7 insertions(+), 3 deletions(-) diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index acddd841458d9..da9bb1fb429b1 100755 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -7,6 +7,10 @@ It is built on top of Level-Zero runtime enabled with [Level-Zero API](https://s The Level-Zero backend is aimed 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 core SYCL specification section 6.3.3 "Feature test macros". +Therefore, an implementation supporting this extension must predefine the macro `SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO`. +Applications can test for the existence of this macro to determine if the implementation supports this feature. + NOTE: This extension is following SYCL 2020 backend specification. Prior API for interoperability with Level-Zero is marked as deprecated and will be removed in the next release. @@ -164,7 +168,7 @@ make_platform( backend::ext_oneapi_level_zero, platform> &) ``` -Constructs a SYCL platform instance from a Level-Zero ze_driver_handle_t +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 enumerated via sycl::platform::get_platforms(). Calling this function does not create a new platform. Rather it merely creates a sycl::platform object that is a copy of one of the platforms from that enumeration. @@ -174,7 +178,7 @@ make_device( backend::ext_oneapi_level_zero, device> &) ``` -Constructs a SYCL device instance from a Level-Zero ze_device_handle_t. +Constructs a SYCL device instance from a Level-Zero ze_device_handle_t. The SYCL execution environment contains a fixed number of devices that are enumerated via sycl::device::get_devices() and sycl::device::get_sub_devices(...). Calling this function does not create a new device. Rather it merely creates a sycl::device object that is a copy of one of the devices from those enumerations. @@ -219,7 +223,7 @@ make_kernel_bundle -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. The Level-Zero module must be fully linked (i.e. not require further linking through [zeModuleDynamicLink](https://spec.oneapi.com/level-zero/latest/core/api.html?highlight=zemoduledynamiclink#_CPPv419zeModuleDynamicLink8uint32_tP18ze_module_handle_tP28ze_module_build_log_handle_t)), and thus the SYCL kernel_bundle is created in the "executable" state. +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. The Level-Zero module must be fully linked (i.e. not require further linking through zeModuleDynamicLink), and thus the SYCL kernel_bundle is created in the "executable" state. From 9cf7e95bcbab1d810d34e1d767a9b7153819f44a Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Wed, 15 Sep 2021 09:55:58 -0700 Subject: [PATCH 3/8] address comments Signed-off-by: Sergey V Maslov --- .../extensions/LevelZeroBackend/LevelZeroBackend.md | 12 ++++++++---- 1 file changed, 8 insertions(+), 4 deletions(-) diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index da9bb1fb429b1..272f3bb198cb4 100755 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -7,9 +7,13 @@ It is built on top of Level-Zero runtime enabled with [Level-Zero API](https://s The Level-Zero backend is aimed 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 core SYCL specification section 6.3.3 "Feature test macros". -Therefore, an implementation supporting this extension must predefine the macro `SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO`. -Applications can test for the existence of this macro to determine if the implementation supports this feature. +This extension provides a feature-test macro as described in the core SYCL specification section 6.3.3 "Feature test macros". Therefore, an implementation supporting this extension must predefine the macro SYCL_EXT_ONEAPI_DEVICE_IF to one of the values defined in the table below. Applications can test for the existence of this macro to determine if the implementation supports this feature, or applications can test the macro’s value to determine which of the extension’s APIs the implementation supports. + +[%header,cols="1,5"] +|=== +|Value |Description +|1 |Initial extension version. +|=== NOTE: This extension is following SYCL 2020 backend specification. Prior API for interoperability with Level-Zero is marked as deprecated and will be removed in the next release. @@ -178,7 +182,7 @@ make_device( backend::ext_oneapi_level_zero, device> &) ``` -Constructs a SYCL device instance from a Level-Zero ze_device_handle_t. The SYCL execution environment contains a fixed number of devices that are enumerated via sycl::device::get_devices() and sycl::device::get_sub_devices(...). Calling this function does not create a new device. Rather it merely creates a sycl::device object that is a copy of one of the devices from those enumerations. +Constructs a SYCL device instance from a Level-Zero ze_device_handle_t. The SYCL execution environment contains a fixed number of devices that are enumerated via sycl::device::get_devices() and sycl::device::create_sub_devices(...). Calling this function does not create a new device. Rather it merely creates a sycl::device object that is a copy of one of the devices from those enumerations. From 65901230577532d85e36210b08109aa7935978b9 Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Wed, 15 Sep 2021 09:59:06 -0700 Subject: [PATCH 4/8] address comments Signed-off-by: Sergey V Maslov --- sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md | 8 +++----- 1 file changed, 3 insertions(+), 5 deletions(-) diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index 272f3bb198cb4..8b8aa2531d130 100755 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -9,11 +9,9 @@ The currently supported targets are all Intel GPUs starting with Gen9. This extension provides a feature-test macro as described in the core SYCL specification section 6.3.3 "Feature test macros". Therefore, an implementation supporting this extension must predefine the macro SYCL_EXT_ONEAPI_DEVICE_IF to one of the values defined in the table below. Applications can test for the existence of this macro to determine if the implementation supports this feature, or applications can test the macro’s value to determine which of the extension’s APIs the implementation supports. -[%header,cols="1,5"] -|=== -|Value |Description -|1 |Initial extension version. -|=== +|Value|Description +|===|=== +|1|Initial extension version. NOTE: This extension is following SYCL 2020 backend specification. Prior API for interoperability with Level-Zero is marked as deprecated and will be removed in the next release. From 5bfc92917bb1eca974129ce818a67d06fd003295 Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Wed, 15 Sep 2021 10:00:33 -0700 Subject: [PATCH 5/8] address comments Signed-off-by: Sergey V Maslov --- sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index 8b8aa2531d130..de3b394c8c761 100755 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -9,8 +9,8 @@ The currently supported targets are all Intel GPUs starting with Gen9. This extension provides a feature-test macro as described in the core SYCL specification section 6.3.3 "Feature test macros". Therefore, an implementation supporting this extension must predefine the macro SYCL_EXT_ONEAPI_DEVICE_IF to one of the values defined in the table below. Applications can test for the existence of this macro to determine if the implementation supports this feature, or applications can test the macro’s value to determine which of the extension’s APIs the implementation supports. -|Value|Description -|===|=== +|Value|Description| +|===|:===| |1|Initial extension version. NOTE: This extension is following SYCL 2020 backend specification. Prior API for interoperability with Level-Zero is marked From 2a6648431ca0220f7cc38c834e1960a0fa28c32d Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Wed, 15 Sep 2021 10:05:00 -0700 Subject: [PATCH 6/8] address comments Signed-off-by: Sergey V Maslov --- sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index de3b394c8c761..1480349f5813a 100755 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -10,7 +10,7 @@ The currently supported targets are all Intel GPUs starting with Gen9. This extension provides a feature-test macro as described in the core SYCL specification section 6.3.3 "Feature test macros". Therefore, an implementation supporting this extension must predefine the macro SYCL_EXT_ONEAPI_DEVICE_IF to one of the values defined in the table below. Applications can test for the existence of this macro to determine if the implementation supports this feature, or applications can test the macro’s value to determine which of the extension’s APIs the implementation supports. |Value|Description| -|===|:===| +|---|:---| |1|Initial extension version. NOTE: This extension is following SYCL 2020 backend specification. Prior API for interoperability with Level-Zero is marked From cbc54ed62ae98592854e967c40cf51809ab825b9 Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Wed, 15 Sep 2021 15:25:02 -0700 Subject: [PATCH 7/8] address comments Signed-off-by: Sergey V Maslov --- sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) mode change 100755 => 100644 sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md old mode 100755 new mode 100644 index 1480349f5813a..be10d5e64815c --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -180,7 +180,7 @@ make_device( backend::ext_oneapi_level_zero, device> &) ``` -Constructs a SYCL device instance from a Level-Zero ze_device_handle_t. The SYCL execution environment contains a fixed number of devices that are enumerated via sycl::device::get_devices() and sycl::device::create_sub_devices(...). Calling this function does not create a new device. Rather it merely creates a sycl::device object that is a copy of one of the devices from those enumerations. +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 enumerated via sycl::device::get_devices() and a fixed number of sub-devices that are enumerated via sycl::device::create_sub_devices(...). Calling this function does not create a new device. Rather it merely creates a sycl::device object that is a copy of one of the devices from those enumerations. From 3545eea59ae24c086920358e10294f7977e985b9 Mon Sep 17 00:00:00 2001 From: smaslov-intel Date: Thu, 16 Sep 2021 08:00:08 -0700 Subject: [PATCH 8/8] Update sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md Co-authored-by: vladimirlaz --- sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index be10d5e64815c..b30bc7bbc1978 100644 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -7,7 +7,8 @@ It is built on top of Level-Zero runtime enabled with [Level-Zero API](https://s The Level-Zero backend is aimed 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 core SYCL specification section 6.3.3 "Feature test macros". Therefore, an implementation supporting this extension must predefine the macro SYCL_EXT_ONEAPI_DEVICE_IF to one of the values defined in the table below. Applications can test for the existence of this macro to determine if the implementation supports this feature, or applications can test the macro’s value to determine which of the extension’s APIs the implementation supports. +This extension provides a feature-test macro as described in the core SYCL specification section 6.3.3 "Feature test macros". Therefore, an 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 determine if the implementation supports this feature, or applications can test the macro’s value to determine which of the extension’s APIs the implementation supports. + |Value|Description| |---|:---|