[SYCL] Use raw pointers in enqueue flow #18306

vinser52 · 2025-05-03T11:30:23Z

For kernel_impl parameter - it's an optional input only parameter that is owned by the caller for the duration of the call and we don't extend its lifetime/create copies as can be seen by the absence of changes in the callee (all we need is to access its data).

For event_impl "out" parameter the semantics is such that the caller creates an object, and the modified interfaces update that object with the UR handles once enqueue/launch is done. The ownership remains controlled by the caller.

vinser52 · 2025-05-05T12:48:19Z

@intel/llvm-reviewers-runtime Looks like CI failures are not related to changes in this PR.

aelovikov-intel

Please split into two (or three, if kernel_impl * needs to go separately). Also, PR's detailed description should explain if this is NFC (i.e., only local temporary variable are updated) or if the whole ownership model is being changed somehow.

aelovikov-intel · 2025-05-05T14:57:29Z

sycl/source/detail/program_manager/program_manager.cpp

Suggested change

const device_impl *DeviceImpl) {

const device_impl &DeviceImpl) {

aelovikov-intel · 2025-05-05T14:57:41Z

sycl/source/detail/program_manager/program_manager.cpp

Suggested change

RTDeviceBinaryImage *BinImage, const device_impl *DeviceImpl) {

RTDeviceBinaryImage *BinImage, const device_impl &DeviceImpl) {

Same below.

aelovikov-intel · 2025-05-05T14:59:37Z

sycl/source/detail/program_manager/program_manager.cpp

IMO, the function should be updated to accept device_impl &, same as above. Also, I think that change would be a good NFC PR on its own - no helpers under source/detail should need sycl::device.

Moved the changes into a separate PR #18320

aelovikov-intel · 2025-05-05T15:00:43Z

sycl/source/detail/scheduler/commands.cpp

This isn't implied by the PR's title ;)

Ok, I changed the titile. But if you want I can split this PR into two separate PRs.

aelovikov-intel · 2025-05-05T21:12:13Z

Also, PR's detailed description should explain if this is NFC (i.e., only local temporary variable are updated) or if the whole ownership model is being changed somehow.

This piece is still missing.

vinser52 · 2025-05-06T10:21:34Z

Also, PR's detailed description should explain if this is NFC (i.e., only local temporary variable are updated) or if the whole ownership model is being changed somehow.

This piece is still missing.

Yeah, sorry. Added the PR description.

Signed-off-by: Sergei Vinogradov <[email protected]>

aelovikov-intel · 2025-05-06T14:29:10Z

sycl/source/detail/scheduler/commands.cpp

                                     void *ptr, size_t size,
                                     std::vector<ur_event_handle_t> &RawEvents,
-                                     const detail::EventImplPtr &OutEventImpl,
+                                     detail::event_impl *OutEventImpl,


I can somewhat believe that in-parameters don't change ownership model, but modifying out-parameter and saying "it's unchanged, trust me" sounds very suspicious. Can you explain more?

Or maybe in other words, is that out-parameter even necessary?

If OutEventImpl is not nullptr then this function calls OutEventImpl->setHandle(UREvent); on L2805.

The same is true for the SetKernelParamsAndLaunch function.

I've updated PR description summarizing my understanding. Can you please check if that's correct and you agree with the change?

Looks good. Thank you!
I just fixed one misprint: "cand the modified interfaces update" -> "and the modified interfaces updat"

Yeah, sorry to be a pain, but mistakes can result in leaks/use-after-free and other flaky failures, so I want to be extra careful with this stuff.

Once Sergey replies about const this is LGTM.

Blocking objections have been addressed.

aelovikov-intel · 2025-05-06T15:03:34Z

sycl/source/detail/kernel_impl.hpp

  const KernelBundleImplPtr MKernelBundleImpl;
  bool MIsInterop = false;
-  std::mutex MNoncacheableEnqueueMutex;
+  mutable std::mutex MNoncacheableEnqueueMutex;


I'm not sure what's our approach for const, we probably didn't care before because shared pointers could always give us non-const ptr even if they themselves were const.

@sergey-semenov , any preferences? I think the pointer we get below is non-const, so we can avoid this bit by simply changing the signature to accept non-const.

My preference is that if the object itself is not modified, then a const pointer should be used and use mutable std::mutex since it might be required for read-only concurrent access synchronization.

My issue is that we (most likely) don't do that consistently. I'd rather have "wrong" behavior everywhere than the "right" one somewhere only. But this isn't big deal for me.

we probably didn't care before because shared pointers could always give us non-const ptr even if they themselves were const.

That's probably true in some (most?) places. I don't have a strong opinion on this, but without knowing exactly how widespread the problem is, I'd rather not contribute to it and start fixing the pieces that we hit, so I'm leaning to sticking with mutable here.

aelovikov-intel · 2025-05-06T15:58:04Z

See #17486 (comment) for the pre-commit failure, merging.

vinser52 requested a review from a team as a code owner May 3, 2025 11:30

vinser52 requested a review from againull May 3, 2025 11:30

vinser52 temporarily deployed to WindowsCILock May 3, 2025 11:30 — with GitHub Actions Inactive

vinser52 temporarily deployed to WindowsCILock May 3, 2025 12:07 — with GitHub Actions Inactive

vinser52 temporarily deployed to WindowsCILock May 3, 2025 12:18 — with GitHub Actions Inactive

vinser52 requested a review from sergey-semenov May 5, 2025 12:34

aelovikov-intel previously requested changes May 5, 2025

View reviewed changes

vinser52 force-pushed the raw_pointers_in_program_manager branch from d295540 to 8e80ef8 Compare May 5, 2025 16:30

vinser52 temporarily deployed to WindowsCILock May 5, 2025 16:30 — with GitHub Actions Inactive

vinser52 temporarily deployed to WindowsCILock May 5, 2025 17:26 — with GitHub Actions Inactive

vinser52 force-pushed the raw_pointers_in_program_manager branch from 8e80ef8 to cd5c535 Compare May 5, 2025 19:39

vinser52 temporarily deployed to WindowsCILock May 5, 2025 19:39 — with GitHub Actions Inactive

vinser52 changed the title ~~[SYCL] Use raw pointers to event_impl~~ [SYCL] Use raw pointers in enqueue flow May 5, 2025

vinser52 requested a review from aelovikov-intel May 5, 2025 19:41

vinser52 temporarily deployed to WindowsCILock May 5, 2025 20:04 — with GitHub Actions Inactive

vinser52 added 2 commits May 6, 2025 12:22

Use raw pointers to event_impl in enqueue flow

47f9aa2

Signed-off-by: Sergei Vinogradov <[email protected]>

Use raw pointer to kernel_impl in enqueueImpKernel

58afdde

Signed-off-by: Sergei Vinogradov <[email protected]>

vinser52 force-pushed the raw_pointers_in_program_manager branch from cd5c535 to 58afdde Compare May 6, 2025 10:37

vinser52 temporarily deployed to WindowsCILock May 6, 2025 10:37 — with GitHub Actions Inactive

vinser52 temporarily deployed to WindowsCILock May 6, 2025 11:12 — with GitHub Actions Inactive

aelovikov-intel reviewed May 6, 2025

View reviewed changes

aelovikov-intel approved these changes May 6, 2025

View reviewed changes

aelovikov-intel merged commit fc7429e into intel:sycl May 6, 2025
47 of 50 checks passed

	const device_impl *DeviceImpl) {
	const device_impl &DeviceImpl) {

	RTDeviceBinaryImage BinImage, const device_impl DeviceImpl) {
	RTDeviceBinaryImage *BinImage, const device_impl &DeviceImpl) {

[SYCL] Use raw pointers in enqueue flow #18306

[SYCL] Use raw pointers in enqueue flow #18306

Uh oh!

Conversation

vinser52 commented May 3, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

vinser52 commented May 5, 2025

Uh oh!

aelovikov-intel left a comment

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

aelovikov-intel commented May 5, 2025

Uh oh!

vinser52 commented May 6, 2025

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

aelovikov-intel commented May 6, 2025

Uh oh!

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

3 participants

vinser52 commented May 3, 2025 •

edited

Loading