Skip to content

[SYCL][Graph] Avoid removing in-order event dependencies across the native recording capture boundary#22604

Open
mmichel11 wants to merge 2 commits into
intel:syclfrom
adamfidel:matt/native_graph_event_tracking_pr
Open

[SYCL][Graph] Avoid removing in-order event dependencies across the native recording capture boundary#22604
mmichel11 wants to merge 2 commits into
intel:syclfrom
adamfidel:matt/native_graph_event_tracking_pr

Conversation

@mmichel11

@mmichel11 mmichel11 commented Jul 10, 2026

Copy link
Copy Markdown
Contributor

Wait event dependencies across the native graph capture boundary cannot be removed and must be processed by the native backend. Due to transitive queue recording and interoperability with native backend calls (e.g. capturing direct L0 submissions), SYCL alone cannot track if a queue is recording. We can only determine if a queue is truly recording through an adapter call which is too expensive to perform on each command submission and muddles tracing tool output.

We solve this issue when a native recording is in progress while keeping the SYCL optimization for the general case with the following approach:

  • Track on the context if any SYCL graph is native recording. Mark any events on any queue owned by the context during this period as PotentiallyNativeRecorded.
  • For any PotentiallyNativeRecorded wait event or queue state, disable the in-order queue optimization removing the dependency.
  • Acquire the primary queue lock around begin and end capture to ensure the native recording transition and atomic update are not split by a submission to the primary queue from a separate thread.

@mmichel11 mmichel11 force-pushed the matt/native_graph_event_tracking_pr branch from 4eae1bd to 29853c5 Compare July 10, 2026 21:40
@mmichel11 mmichel11 requested a review from Copilot July 10, 2026 21:43

Copilot AI left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

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

Pull request overview

This PR fixes incorrect pruning of redundant in-order event wait-list dependencies when SYCL graphs use native backend recording, by conservatively tracking native recording at the context level and propagating that state onto events so dependencies that may cross the native capture boundary are preserved.

Changes:

  • Add a context-level “native recording active” counter and use it to disable same-queue in-order dependency pruning during native recording.
  • Tag events created while native recording is active as PotentiallyNativeRecorded and prevent pruning of those dependencies even after recording ends.
  • Refactor UR begin/end capture into queue_impl::{beginNativeRecording,endNativeRecording} and add a new end-to-end test validating the wait-list behavior via SYCL_UR_TRACE.

Reviewed changes

Copilot reviewed 7 out of 7 changed files in this pull request and generated 3 comments.

Show a summary per file
File Description
sycl/test-e2e/Graph/RecordReplay/NativeRecording/inorder_event_dependencies.cpp New E2E test asserting wait-list dependency preservation across native capture boundaries.
sycl/source/detail/scheduler/commands.cpp Gate redundant in-order dependency pruning on context native-recording activity and event tagging.
sycl/source/detail/queue_impl.hpp Expose queue-level wrappers for UR native capture begin/end.
sycl/source/detail/queue_impl.cpp Implement begin/end native capture wrappers and mark scheduler-bypass events as potentially native recorded.
sycl/source/detail/graph/graph_impl.cpp Use the new queue_impl wrappers for native capture begin/end and handle queue-destruction cleanup.
sycl/source/detail/event_impl.hpp Add PotentiallyNativeRecorded flag to events to preserve cross-boundary dependencies.
sycl/source/detail/context_impl.hpp Add atomic native-recording counter and query helpers on context.

Comment on lines +629 to +639
void queue_impl::beginNativeRecording(ur_exp_graph_handle_t Graph) {
std::lock_guard<std::mutex> Lock(MMutex);
ur_result_t Result =
getAdapter().call_nocheck<UrApiKind::urQueueBeginCaptureIntoGraphExp>(
MQueue, Graph);
if (Result != UR_RESULT_SUCCESS) {
throw sycl::exception(sycl::make_error_code(errc::runtime),
"Failed to begin native UR graph capture");
}
getContextImpl().nativeRecordingBegan();
}

Copy link
Copy Markdown
Contributor Author

Choose a reason for hiding this comment

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

This scenario occurs when the user has an application race between begin_recording(q) and a command submission to the same queue. It cannot occur between different queues because the initial submission must be done on the primary queue and forking the graph to a second queue requires some form of synchronization. Since we acquire the queue lock now in this call, I believe there is no risk.

Comment thread sycl/test-e2e/Graph/RecordReplay/NativeRecording/inorder_event_dependencies.cpp Outdated
Comment thread sycl/source/detail/event_impl.hpp Outdated
@mmichel11 mmichel11 marked this pull request as ready for review July 12, 2026 20:02
@mmichel11 mmichel11 requested review from a team as code owners July 12, 2026 20:02
@mmichel11 mmichel11 requested review from adamfidel and againull July 12, 2026 20:02
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

2 participants