Skip to content

[SYCL][UR][L0 v2] Add a note about make_queue support #19644

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 4 commits into from
Aug 5, 2025
Merged

[SYCL][UR][L0 v2] Add a note about make_queue support #19644

Show file tree
Hide file tree
Changes from 1 commit
Commits
Show all changes
4 commits
Select commit Hold shift + click to select a range
841c33a
[SYCL][UR][L0 v2] Add a note about make_queue support
igchor Jul 30, 2025
303b37e
Update sycl/doc/extensions/supported/sycl_ext_oneapi_backend_level_ze…
igchor Jul 31, 2025
5249789
Add in-order requirement
igchor Jul 31, 2025
99a8426
Add note about UB
igchor Aug 5, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 2 additions & 0 deletions sycl/doc/extensions/supported/sycl_ext_oneapi_backend_level_zero.md
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -388,6 +388,8 @@ If the deprecated variant of <code>backend_input_t<backend::ext_oneapi_level_zer

Starting in version 4 of this specification, ```make_queue()``` can be called by passing either a Level Zero ```ze_command_queue_handle_t``` or a Level Zero ```ze_command_list_handle_t```. Queues created from a Level Zero immediate command list (```ze_command_list_handle_t```) generally perform better than queues created from a standard Level Zero ```ze_command_queue_handle_t```. See the Level Zero documentation of these native handles for more details. Also starting in version 4 the ```make_queue()``` function accepts a ```Properties``` member variable. This can contain any of the SYCL properties that are accepted by the SYCL queue constructor, except
the ```compute_index``` property which is built into the command queue or command list.

**Warning:** <span style="color:red"> When L0 v2 adapter is used (when running on platforms with GPUs based on the Xe2 architecture or later, such as Battlemage, Lunar Lake, and Arrow Lake or when SYCL_UR_USE_LEVEL_ZERO_V2=1 is set) ```make_queue()``` accepts only ```ze_command_list_handle_t```: a handle to an **immediate**, **in-order** command list. If non-immediate or out-of-order command list support is needed, the legacy adapter should be used (by setting SYCL_UR_USE_LEVEL_ZERO_V2=0)</span>
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I found this quite hard to read. Something like the below might be clearer:

Suggested change
**Warning:** <span style="color:red"> When L0 v2 adapter is used (when running on platforms with GPUs based on the Xe2 architecture or later, such as Battlemage, Lunar Lake, and Arrow Lake or when SYCL_UR_USE_LEVEL_ZERO_V2=1 is set) ```make_queue()``` accepts only ```ze_command_list_handle_t```: a handle to an **immediate**, **in-order** command list. If non-immediate or out-of-order command list support is needed, the legacy adapter should be used (by setting SYCL_UR_USE_LEVEL_ZERO_V2=0)</span>
**Warning:** <span style="color:red"> When using the L0 v2 adapter, ```make_queue()``` only accepts ```ze_command_list_handle_t```. The L0 v2 adapter is always used when running on platforms with GPUs based on the Xe2 architecture or later, such as Battlemage, Lunar Lake, and Arrow Lake. Applications may also opt-in to the L0 v2 adapter using the `SYCL_UR_USE_LEVEL_ZERO_V2` environment variable.</span>

I think you also need to specify what happens if somebody passes a ze_command_queue_handle_t to this function and the V2 adapter is not used. Is it undefined behavior, will the program crash, will we throw an exception...?

More generally... The description of this function's behavior is now really long, with reference to different specification versions and different adapter versions. I think it's too confusing and would benefit from a rewrite, but that might be out of scope for this PR.

@gmlueck - This might be an interesting case to consider in the context of KhronosGroup/SYCL-Docs#777. If the Level Zero backend didn't have to jump through hoops to satisfy the backend_input_t interface, we would be free to define separate overloads like:

sycl::level_zero::make_queue(ze_command_queue_handle_t& handle, ...);
sycl::level_zero::make_queue(ze_command_list_handle_t& handle, ...);

I think this would make it much easier to describe what these functions actually do, which would also make it easier for developers to understand and use them.

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Is it undefined behavior, will the program crash, will we throw an exception...?

I think we should throw an exception, so applications know they did something that is not supported.

Copy link
Contributor Author

@igchor igchor Jul 31, 2025 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Thanks, I applied your suggestion @Pennycook. Right now, the behavior is that when ze_command_queue_handle_t is passed we will throw an exception (UR returns UR_RESULT_ERROR_UNSUPPORTED_FEATURE). However, when the command list is not in-order we might crash. Unfortunately, there is currently no API to check whether a command list is in-order or not. Once this API is added we can also throw an exception in that case.

Should we document this behavior as is right now?

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Yes, I think it would be valuable to capture what the current state is, and to update it later.

If you say that passing an out-of-order command list is undefined behavior, that will cover the current state (crash) and the desired future state (exception). Once we can be sure that we always throw an exception, we can tighten the specification to say that's required.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Done.

</td>
</tr><tr>
<td>
Expand Down
20 changes: 20 additions & 0 deletions unified-runtime/source/adapters/level_zero/v2/queue_create.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -94,6 +94,7 @@ ur_result_t urQueueCreateWithNativeHandle(

bool ownNativeHandle = pProperties ? pProperties->isNativeHandleOwned : false;
ur_queue_flags_t flags = 0;
bool isNativeHandleImmediate = true;

if (pProperties) {
void *pNext = pProperties->pNext;
Expand All @@ -104,11 +105,30 @@ ur_result_t urQueueCreateWithNativeHandle(
const ur_queue_properties_t *pUrProperties =
reinterpret_cast<const ur_queue_properties_t *>(extendedProperties);
flags = pUrProperties->flags;
} else if (extendedProperties->stype ==
UR_STRUCTURE_TYPE_QUEUE_NATIVE_DESC) {
const ur_queue_native_desc_t *pUrNativeDesc =
reinterpret_cast<const ur_queue_native_desc_t *>(
extendedProperties);
if (pUrNativeDesc->pNativeData) {
// The pNativeData has value if if the native handle is an immediate
// command list.
isNativeHandleImmediate =
*(reinterpret_cast<int32_t *>((pUrNativeDesc->pNativeData))) == 1;
}
}
pNext = extendedProperties->pNext;
}
}

if (!isNativeHandleImmediate) {
UR_LOG(ERR, "urQueueCreateWithNativeHandle: "
"Native handle is not an immediate command "
"list; only immediate command lists are "
"supported.");
return UR_RESULT_ERROR_UNSUPPORTED_FEATURE;
}

ze_bool_t isImmediate = false;
ZE2UR_CALL(
zeCommandListIsImmediate,
Expand Down