From 7b85fb2475dc2680ac6898c2d7b8e3ed726ac72d Mon Sep 17 00:00:00 2001
From: Ewan Crawford <ewan@codeplay.com>
Date: Tue, 18 Jun 2024 14:14:52 +0100
Subject: [PATCH] [SYCL][Graph] Fix queue recording barrier to different graphs

Recording barrier submissions to from the same queue to a different
graph current produces the following error with added regression test:

```
Terminate called after throwing an instance of 'sycl::_V1::exception'
  what():  Graph nodes cannot depend on events from another graph.
```

This is because the queue implementation doesn't clear all the
state around what the last queue submission was between graph
recordings.

Fixed by clearing all members of the barrier book keeping struct
in the queue.
---
 sycl/source/detail/queue_impl.hpp             |  8 +++++-
 .../Extensions/CommandGraph/Regressions.cpp   | 27 +++++++++++++++++++
 2 files changed, 34 insertions(+), 1 deletion(-)

diff --git a/sycl/source/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp
index 82334e6467dfd..52d01dc2923a6 100644
--- a/sycl/source/detail/queue_impl.hpp
+++ b/sycl/source/detail/queue_impl.hpp
@@ -732,7 +732,7 @@ class queue_impl {
       std::shared_ptr<ext::oneapi::experimental::detail::graph_impl> Graph) {
     std::lock_guard<std::mutex> Lock(MMutex);
     MGraph = Graph;
-    MExtGraphDeps.LastEventPtr = nullptr;
+    MExtGraphDeps.reset();
   }
 
   std::shared_ptr<ext::oneapi::experimental::detail::graph_impl>
@@ -938,6 +938,12 @@ class queue_impl {
     // ordering
     std::vector<EventImplPtr> UnenqueuedCmdEvents;
     EventImplPtr LastBarrier;
+
+    void reset() {
+      LastEventPtr = nullptr;
+      UnenqueuedCmdEvents.clear();
+      LastBarrier = nullptr;
+    }
   } MDefaultGraphDeps, MExtGraphDeps;
 
   const bool MIsInorder;
diff --git a/sycl/unittests/Extensions/CommandGraph/Regressions.cpp b/sycl/unittests/Extensions/CommandGraph/Regressions.cpp
index 17b58f542d760..94b8549ed7c04 100644
--- a/sycl/unittests/Extensions/CommandGraph/Regressions.cpp
+++ b/sycl/unittests/Extensions/CommandGraph/Regressions.cpp
@@ -58,3 +58,30 @@ TEST_F(CommandGraphTest, AccessorModeRegression) {
   EXPECT_EQ(NodeC.get_predecessors().size(), 0ul);
   EXPECT_EQ(NodeC.get_successors().size(), 0ul);
 }
+
+TEST_F(CommandGraphTest, QueueRecordBarrierMultipleGraph) {
+  // Test that using barriers recorded from the same queue to
+  // different graphs.
+
+  Graph.begin_recording(Queue);
+  auto NodeKernel = Queue.submit(
+      [&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
+  Queue.ext_oneapi_submit_barrier({NodeKernel});
+  Graph.end_recording(Queue);
+
+  experimental::command_graph<experimental::graph_state::modifiable> GraphB{
+      Queue};
+  GraphB.begin_recording(Queue);
+  auto NodeKernelB = Queue.submit(
+      [&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
+  Queue.ext_oneapi_submit_barrier({NodeKernelB});
+  GraphB.end_recording(Queue);
+
+  experimental::command_graph<experimental::graph_state::modifiable> GraphC{
+      Queue};
+  GraphC.begin_recording(Queue);
+  auto NodeKernelC = Queue.submit(
+      [&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
+  Queue.ext_oneapi_submit_barrier();
+  GraphC.end_recording(Queue);
+}