refactor

Author: igchor

sycl/source/detail/queue_impl.cpp

@@ -292,13 +292,16 @@ queue_impl::getLastEvent(const std::shared_ptr<queue_impl> &Self) {
   std::lock_guard<std::mutex> Lock{MMutex};
   if (MEmpty)
     return std::nullopt;
-  if (MGraph.expired() && !MDefaultGraphDeps.LastEventPtr) {
-    assert(!MHostTaskMode);
+  if (MNoEventMode) {
+    assert(MGraph.expired());
+    assert(!MDefaultGraphDeps.LastEventPtr);
+
     // We insert a marker to represent an event at end.
     return detail::createSyclObjFromImpl<event>(insertMarkerEvent(Self));
   }
   if (!MGraph.expired() && MExtGraphDeps.LastEventPtr)
     return detail::createSyclObjFromImpl<event>(MExtGraphDeps.LastEventPtr);
+  assert(MDefaultGraphDeps.LastEventPtr);
   return detail::createSyclObjFromImpl<event>(MDefaultGraphDeps.LastEventPtr);
 }
 
@@ -342,16 +345,21 @@ event queue_impl::submit_impl(const detail::type_erased_cgfo_ty &CGF,
 
   HandlerImpl->MEventMode = SubmitInfo.EventMode();
 
-  std::unique_lock<std::mutex> Lock(MMutex, std::defer_lock);
-  auto Event = finalizeHandler(Handler, SubmitInfo.PostProcessorFunc(), Lock);
+  auto isInNoEventsMode = isInOrder() && MNoEventMode;
+  auto needToStoreEvent = Type == CGType::CodeplayHostTask ||
+                          Streams.size() > 0 || SubmitInfo.PostProcessorFunc();
 
-  if (isInOrder() && !shouldRecordLastEvent() && Streams.empty()) {
-    // NOP
-  } else {
-    addEvent(Event);
+  if (isInNoEventsMode && !needToStoreEvent) {
+    std::unique_lock<std::mutex> Lock(MMutex);
+    return finalizeHandlerInOrderNoEventsUnlocked(Handler);
   }
 
-  Lock.unlock();
+  auto Event = finalizeHandler(Handler, SubmitInfo.PostProcessorFunc());
+  addEvent(Event);
+
+  // If we taken this path, it's because we need to store the last event.
+  if (isInOrder())
+    assert(!MNoEventMode);
 
   const auto &EventImpl = detail::getSyclObjImpl(Event);
   for (auto &Stream : Streams) {
@@ -405,13 +413,10 @@ event queue_impl::submit_impl(const detail::type_erased_cgfo_ty &CGF,
 
   HandlerImpl->MEventMode = SubmitInfo.EventMode();
 
-  std::unique_lock<std::mutex> Lock(MMutex, std::defer_lock);
+  auto Event = finalizeHandler(Handler, SubmitInfo.PostProcessorFunc());
 
-  auto Event = finalizeHandler(Handler, SubmitInfo.PostProcessorFunc(), Lock);
   addEvent(Event);
 
-  Lock.unlock();
-
   const auto &EventImpl = detail::getSyclObjImpl(Event);
   for (auto &Stream : Streams) {
     // We don't want stream flushing to be blocking operation that is why submit
@@ -473,8 +478,7 @@ event queue_impl::submitMemOpHelper(const std::shared_ptr<queue_impl> &Self,
     // handler rather than by-passing the scheduler.
     if (MGraph.expired() && Scheduler::areEventsSafeForSchedulerBypass(
                                 ExpandedDepEvents, MContext)) {
-      if (!CallerNeedsEvent && supportsDiscardingPiEvents() &&
-          !shouldRecordLastEvent()) {
+      if (!CallerNeedsEvent && MNoEventMode) {
         NestedCallsTracker tracker;
         MemOpFunc(MemOpArgs..., getUrEvents(ExpandedDepEvents),
                   /*PiEvent*/ nullptr, /*EventImplPtr*/ nullptr);
@@ -505,7 +509,7 @@ event queue_impl::submitMemOpHelper(const std::shared_ptr<queue_impl> &Self,
         }
       }
 
-      if (shouldRecordLastEvent()) {
+      if (!MNoEventMode) {
         auto &EventToStoreIn = MGraph.expired() ? MDefaultGraphDeps.LastEventPtr
                                                 : MExtGraphDeps.LastEventPtr;
         EventToStoreIn = EventImpl;
@@ -750,7 +754,7 @@ bool queue_impl::ext_oneapi_empty() const {
   // the status of the last event.
   if (isInOrder() && !MDiscardEvents) {
     std::lock_guard<std::mutex> Lock(MMutex);
-    assert((MDefaultGraphDeps.LastEventPtr != nullptr) == MHostTaskMode);
+    assert((MDefaultGraphDeps.LastEventPtr != nullptr) == MNoEventMode);
     // Note that we fall back to the backend query if the event was discarded,
     // which may happend despite the queue not being a discard event queue.
     if (MDefaultGraphDeps.LastEventPtr &&

sycl/source/detail/queue_impl.hpp

@@ -631,6 +631,7 @@ class queue_impl {
     std::lock_guard<std::mutex> Lock(MMutex);
     MGraph = Graph;
     MExtGraphDeps.reset();
+    MNoEventMode = false;
   }
 
   std::shared_ptr<ext::oneapi::experimental::detail::graph_impl>
@@ -721,56 +722,93 @@ class queue_impl {
   }
 
   template <typename HandlerType = handler>
-  event finalizeHandlerInOrder(HandlerType &Handler,
-                               std::unique_lock<std::mutex> &Lock) {
-    Lock.lock();
+  void synchronizeWithExternalEvent(HandlerType &Handler) {
+    // If there is an external event set, add it as a dependency and clear it.
+    // We do not need to hold the lock as MLastEventMtx will ensure the last
+    // event reflects the corresponding external event dependence as well.
+    std::optional<event> ExternalEvent = popExternalEvent();
+    if (ExternalEvent)
+      Handler.depends_on(*ExternalEvent);
+  }
+
+  template <typename HandlerType = handler>
+  event finalizeHandlerInOrderNoEventsUnlocked(HandlerType &Handler) {
+    assert(isInOrder());
+    assert(MGraph.expired());
+    assert(MNoEventMode);
 
+    MEmpty = false;
+
+    synchronizeWithExternalEvent(Handler);
+
+    return Handler.finalize();
+  }
+
+  template <typename HandlerType = handler>
+  event finalizeHandlerInOrder(HandlerType &Handler) {
+    // Accessing and changing of an event isn't atomic operation.
+    // Hence, here is the lock for thread-safety.
+    std::lock_guard<std::mutex> Lock{MMutex};
+
+    MEmpty = false;
     auto &EventToBuildDeps = MGraph.expired() ? MDefaultGraphDeps.LastEventPtr
                                               : MExtGraphDeps.LastEventPtr;
 
-    if (Handler.getType() == CGType::CodeplayHostTask) {
-      if (!MHostTaskMode && MGraph.expired() && !MEmpty) {
-        assert(EventToBuildDeps == nullptr);
-        // since we don't store any events, insert a barrier to ensure proper
-        // ordering with device execution
-        auto barrierEvent = insertHelperBarrier(Handler);
-        Handler.depends_on(barrierEvent);
-      }
-
-      MHostTaskMode = true;
+    if (MNoEventMode && Handler.getType() == CGType::CodeplayHostTask) {
+      assert(MGraph.expired());
+      assert(MDefaultGraphDeps.LastEventPtr == nullptr);
+      // There might be some operations submitted to the queue
+      // but the LastEventPtr is not set. If we are to run a host_task,
+      // we need to insert a barrier to ensure proper synchronization.
+      Handler.depends_on(insertHelperBarrier(Handler));
     }
 
-    if (EventToBuildDeps && Handler.getType() != CGType::AsyncAlloc) {
+    // This dependency is needed for the following purposes:
+    //    - host tasks are handled by the runtime and cannot be implicitly
+    //    synchronized by the backend.
+    //    - to prevent the 2nd kernel enqueue when the 1st kernel is blocked
+    //    by a host task. This dependency allows to build the enqueue order in
+    //    the RT but will not be passed to the backend. See getPIEvents in
+    //    Command.
+    if (EventToBuildDeps) {
+      // If we have last event, this means we are no longer in no-event mode.
+      assert(!MNoEventMode);
+
+      // In the case where the last event was discarded and we are to run a
+      // host_task, we insert a barrier into the queue and use the resulting
+      // event as the dependency for the host_task.
+      // Note that host_task events can never be discarded, so this will not
+      // insert barriers between host_task enqueues.
+      if (EventToBuildDeps->isDiscarded() &&
+          Handler.getType() == CGType::CodeplayHostTask)
+        EventToBuildDeps = insertHelperBarrier(Handler);
+
       // depends_on after an async alloc is explicitly disallowed. Async alloc
       // handles in order queue dependencies preemptively, so we skip them.
       // Note: This could be improved by moving the handling of dependencies
       // to before calling the CGF.
-      Handler.depends_on(EventToBuildDeps);
+      if (!EventToBuildDeps->isDiscarded() &&
+          !(Handler.getType() == CGType::AsyncAlloc))
+        Handler.depends_on(EventToBuildDeps);
     }
 
-    MEmpty = false;
+    MNoEventMode = false;
 
-    // If there is an external event set, add it as a dependency and clear it.
-    // We do not need to hold the lock as MLastEventMtx will ensure the last
-    // event reflects the corresponding external event dependence as well.
-    std::optional<event> ExternalEvent = popExternalEvent();
-    if (ExternalEvent)
-      Handler.depends_on(*ExternalEvent);
+    synchronizeWithExternalEvent(Handler);
 
     auto EventRet = Handler.finalize();
-
-    if (shouldRecordLastEvent()) {
-      EventToBuildDeps = getSyclObjImpl(EventRet);
-    }
+    EventToBuildDeps = getSyclObjImpl(EventRet);
 
     return EventRet;
   }
 
   template <typename HandlerType = handler>
-  event finalizeHandlerOutOfOrder(HandlerType &Handler,
-                                  std::unique_lock<std::mutex> &Lock) {
+  event finalizeHandlerOutOfOrder(HandlerType &Handler) {
+    // Accessing and changing of an event isn't atomic operation.
+    // Hence, here is the lock for thread-safety.
+    std::lock_guard<std::mutex> Lock{MMutex};
+
     const CGType Type = getSyclObjImpl(Handler)->MCGType;
-    Lock.lock();
 
     MEmpty = false;
 
@@ -810,8 +848,7 @@ class queue_impl {
   template <typename HandlerType = handler>
   event finalizeHandlerPostProcess(
       HandlerType &Handler,
-      const optional<SubmitPostProcessF> &PostProcessorFunc,
-      std::unique_lock<std::mutex> &Lock) {
+      const optional<SubmitPostProcessF> &PostProcessorFunc) {
     bool IsKernel = Handler.getType() == CGType::Kernel;
     bool KernelUsesAssert = false;
 
@@ -822,8 +859,8 @@ class queue_impl {
           ProgramManager::getInstance().kernelUsesAssert(
               Handler.MKernelName.data());
 
-    auto Event = MIsInorder ? finalizeHandlerInOrder(Handler, Lock)
-                            : finalizeHandlerOutOfOrder(Handler, Lock);
+    auto Event = MIsInorder ? finalizeHandlerInOrder(Handler)
+                            : finalizeHandlerOutOfOrder(Handler);
 
     auto &PostProcess = *PostProcessorFunc;
 
@@ -835,13 +872,12 @@ class queue_impl {
   // template is needed for proper unit testing
   template <typename HandlerType = handler>
   event finalizeHandler(HandlerType &Handler,
-                        const optional<SubmitPostProcessF> &PostProcessorFunc,
-                        std::unique_lock<std::mutex> &Lock) {
+                        const optional<SubmitPostProcessF> &PostProcessorFunc) {
     if (PostProcessorFunc) {
-      return finalizeHandlerPostProcess(Handler, PostProcessorFunc, Lock);
+      return finalizeHandlerPostProcess(Handler, PostProcessorFunc);
     } else {
-      return MIsInorder ? finalizeHandlerInOrder(Handler, Lock)
-                        : finalizeHandlerOutOfOrder(Handler, Lock);
+      return MIsInorder ? finalizeHandlerInOrder(Handler)
+                        : finalizeHandlerOutOfOrder(Handler);
     }
   }
 
@@ -1011,18 +1047,11 @@ class queue_impl {
 
   const bool MIsInorder;
 
-  // Specifies whether this queue uses host tasks. If yes, then event
-  // from all operations need to be recorded for proper synchronization.
-  bool MHostTaskMode = false;
-
-  bool shouldRecordLastEvent() const {
-    // For in-order queues we rely on UR queue ordering.
-    // We only need to keep the event if host task are used
-    // (to ensure proper ordering).
-
-    // TODO: do not record last event for graphs as well
-    return MIsInorder && (MHostTaskMode || !MGraph.expired());
-  }
+  // Specifies whether this queue records last event. This can only
+  // be true if the queue is in-order, the command graph is not
+  // associated with the queue and there has never been any host
+  // tasks submitted to the queue.
+  bool MNoEventMode = true;
 
   bool MEmpty = true;
 

sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp

@@ -86,18 +86,12 @@ TEST_F(SchedulerTest, InOrderQueueSyncCheck) {
     LimitedHandlerSimulation MockCGH{detail::CGType::CodeplayHostTask, Queue};
     EXPECT_CALL(MockCGH, depends_on(An<const sycl::detail::EventImplPtr &>()))
         .Times(0);
-    std::mutex mtx;
-    std::unique_lock<std::mutex> Lock(mtx, std::defer_lock);
-    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, std::nullopt,
-                                                     Lock);
+    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, std::nullopt);
   }
   {
     LimitedHandlerSimulation MockCGH{detail::CGType::CodeplayHostTask, Queue};
     EXPECT_CALL(MockCGH, depends_on(An<const sycl::detail::EventImplPtr &>()))
         .Times(1);
-    std::mutex mtx;
-    std::unique_lock<std::mutex> Lock(mtx, std::defer_lock);
-    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, std::nullopt,
-                                                     Lock);
+    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, std::nullopt);
   }
 }