[SYCL] Avoid multiple event_impl allocations on kernel enqueue (#17312)

Reduced the number of event_impl allocations on the kernel enqueue path
to one, for the scheduler bypass fast path.

Author: slawekptak

sycl/source/detail/event_impl.cpp

@@ -178,6 +178,17 @@ event_impl::event_impl(const QueueImplPtr &Queue)
   MState.store(HES_Complete);
 }
 
+void event_impl::setQueue(const QueueImplPtr &Queue) {
+  MQueue = Queue;
+  MIsProfilingEnabled = Queue->MIsProfilingEnabled;
+  MFallbackProfiling = MIsProfilingEnabled && Queue->isProfilingFallback();
+
+  // TODO After setting the queue, the event is no longer default
+  // constructed. Consider a design change which would allow
+  // for such a change regardless of the construction method.
+  MIsDefaultConstructed = false;
+}
+
 void *event_impl::instrumentationProlog(std::string &Name, int32_t StreamID,
                                         uint64_t &IId) const {
   void *TraceEvent = nullptr;

sycl/source/detail/event_impl.hpp

@@ -68,6 +68,14 @@ class event_impl {
   event_impl(ur_event_handle_t Event, const context &SyclContext);
   event_impl(const QueueImplPtr &Queue);
 
+  /// Sets a queue associated with the event
+  ///
+  /// Please note that this function changes the event state
+  /// as it was constructed with the queue based constructor.
+  ///
+  /// \param Queue is a queue to be associated with the event
+  void setQueue(const QueueImplPtr &Queue);
+
   /// Waits for the event.
   ///
   /// Self is needed in order to pass shared_ptr to Scheduler.

sycl/source/detail/queue_impl.cpp

@@ -376,30 +376,13 @@ event queue_impl::submit_impl(const detail::type_erased_cgfo_ty &CGF,
   // Host and interop tasks, however, are not submitted to low-level runtimes
   // and require separate dependency management.
   const CGType Type = HandlerImpl->MCGType;
-  event Event = detail::createSyclObjFromImpl<event>(
-      std::make_shared<detail::event_impl>());
   std::vector<StreamImplPtr> Streams;
   if (Type == CGType::Kernel)
     Streams = std::move(Handler.MStreamStorage);
 
   HandlerImpl->MEventMode = SubmitInfo.EventMode();
 
-  if (SubmitInfo.PostProcessorFunc()) {
-    auto &PostProcess = *SubmitInfo.PostProcessorFunc();
-
-    bool IsKernel = Type == CGType::Kernel;
-    bool KernelUsesAssert = false;
-
-    if (IsKernel)
-      // Kernel only uses assert if it's non interop one
-      KernelUsesAssert = !(Handler.MKernel && Handler.MKernel->isInterop()) &&
-                         ProgramManager::getInstance().kernelUsesAssert(
-                             Handler.MKernelName.c_str());
-    finalizeHandler(Handler, Event);
-
-    PostProcess(IsKernel, KernelUsesAssert, Event);
-  } else
-    finalizeHandler(Handler, Event);
+  auto Event = finalizeHandler(Handler, SubmitInfo.PostProcessorFunc());
 
   addEvent(Event);
 

sycl/source/detail/queue_impl.hpp

@@ -781,79 +781,121 @@ class queue_impl {
     return ResEvent;
   }
 
-  // template is needed for proper unit testing
   template <typename HandlerType = handler>
-  void finalizeHandler(HandlerType &Handler, event &EventRet) {
-    if (MIsInorder) {
-      // 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};
-
-      auto &EventToBuildDeps = MGraph.expired() ? MDefaultGraphDeps.LastEventPtr
-                                                : MExtGraphDeps.LastEventPtr;
-
-      // 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) {
-        // 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() &&
-            getSyclObjImpl(Handler)->MCGType == CGType::CodeplayHostTask)
-          EventToBuildDeps = insertHelperBarrier(Handler);
-
-        if (!EventToBuildDeps->isDiscarded())
-          Handler.depends_on(EventToBuildDeps);
-      }
+  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};
+
+    auto &EventToBuildDeps = MGraph.expired() ? MDefaultGraphDeps.LastEventPtr
+                                              : MExtGraphDeps.LastEventPtr;
+
+    // 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) {
+      // 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() &&
+          getSyclObjImpl(Handler)->MCGType == CGType::CodeplayHostTask)
+        EventToBuildDeps = insertHelperBarrier(Handler);
+
+      if (!EventToBuildDeps->isDiscarded())
+        Handler.depends_on(EventToBuildDeps);
+    }
+
+    // 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);
+
+    auto EventRet = Handler.finalize();
+    EventToBuildDeps = getSyclObjImpl(EventRet);
+
+    return EventRet;
+  }
+
+  template <typename HandlerType = handler>
+  event finalizeHandlerOutOfOrder(HandlerType &Handler) {
+    const CGType Type = getSyclObjImpl(Handler)->MCGType;
+    std::lock_guard<std::mutex> Lock{MMutex};
+    // The following code supports barrier synchronization if host task is
+    // involved in the scenario. Native barriers cannot handle host task
+    // dependency so in the case where some commands were not enqueued
+    // (blocked), we track them to prevent barrier from being enqueued
+    // earlier.
+    {
+      std::lock_guard<std::mutex> RequestLock(MMissedCleanupRequestsMtx);
+      for (auto &UpdatedGraph : MMissedCleanupRequests)
+        doUnenqueuedCommandCleanup(UpdatedGraph);
+      MMissedCleanupRequests.clear();
+    }
+    auto &Deps = MGraph.expired() ? MDefaultGraphDeps : MExtGraphDeps;
+    if (Type == CGType::Barrier && !Deps.UnenqueuedCmdEvents.empty()) {
+      Handler.depends_on(Deps.UnenqueuedCmdEvents);
+    }
+    if (Deps.LastBarrier &&
+        (Type == CGType::CodeplayHostTask || (!Deps.LastBarrier->isEnqueued())))
+      Handler.depends_on(Deps.LastBarrier);
+
+    auto EventRet = Handler.finalize();
+    EventImplPtr EventRetImpl = getSyclObjImpl(EventRet);
+    if (Type == CGType::CodeplayHostTask)
+      Deps.UnenqueuedCmdEvents.push_back(EventRetImpl);
+    else if (Type == CGType::Barrier || Type == CGType::BarrierWaitlist) {
+      Deps.LastBarrier = EventRetImpl;
+      Deps.UnenqueuedCmdEvents.clear();
+    } else if (!EventRetImpl->isEnqueued()) {
+      Deps.UnenqueuedCmdEvents.push_back(EventRetImpl);
+    }
+
+    return EventRet;
+  }
+
+  template <typename HandlerType = handler>
+  event finalizeHandlerPostProcess(
+      HandlerType &Handler,
+      const optional<SubmitPostProcessF> &PostProcessorFunc) {
+    auto HandlerImpl = detail::getSyclObjImpl(Handler);
+    const CGType Type = HandlerImpl->MCGType;
+
+    bool IsKernel = Type == CGType::Kernel;
+    bool KernelUsesAssert = false;
+
+    if (IsKernel)
+      // Kernel only uses assert if it's non interop one
+      KernelUsesAssert = !(Handler.MKernel && Handler.MKernel->isInterop()) &&
+                         ProgramManager::getInstance().kernelUsesAssert(
+                             Handler.MKernelName.c_str());
+
+    auto Event = MIsInorder ? finalizeHandlerInOrder(Handler)
+                            : finalizeHandlerOutOfOrder(Handler);
+
+    auto &PostProcess = *PostProcessorFunc;
 
-      // 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);
+    PostProcess(IsKernel, KernelUsesAssert, Event);
 
-      EventRet = Handler.finalize();
-      EventToBuildDeps = getSyclObjImpl(EventRet);
+    return Event;
+  }
+
+  // template is needed for proper unit testing
+  template <typename HandlerType = handler>
+  event finalizeHandler(HandlerType &Handler,
+                        const optional<SubmitPostProcessF> &PostProcessorFunc) {
+    if (PostProcessorFunc) {
+      return finalizeHandlerPostProcess(Handler, PostProcessorFunc);
     } else {
-      const CGType Type = getSyclObjImpl(Handler)->MCGType;
-      std::lock_guard<std::mutex> Lock{MMutex};
-      // The following code supports barrier synchronization if host task is
-      // involved in the scenario. Native barriers cannot handle host task
-      // dependency so in the case where some commands were not enqueued
-      // (blocked), we track them to prevent barrier from being enqueued
-      // earlier.
-      {
-        std::lock_guard<std::mutex> RequestLock(MMissedCleanupRequestsMtx);
-        for (auto &UpdatedGraph : MMissedCleanupRequests)
-          doUnenqueuedCommandCleanup(UpdatedGraph);
-        MMissedCleanupRequests.clear();
-      }
-      auto &Deps = MGraph.expired() ? MDefaultGraphDeps : MExtGraphDeps;
-      if (Type == CGType::Barrier && !Deps.UnenqueuedCmdEvents.empty()) {
-        Handler.depends_on(Deps.UnenqueuedCmdEvents);
-      }
-      if (Deps.LastBarrier && (Type == CGType::CodeplayHostTask ||
-                               (!Deps.LastBarrier->isEnqueued())))
-        Handler.depends_on(Deps.LastBarrier);
-
-      EventRet = Handler.finalize();
-      EventImplPtr EventRetImpl = getSyclObjImpl(EventRet);
-      if (Type == CGType::CodeplayHostTask)
-        Deps.UnenqueuedCmdEvents.push_back(EventRetImpl);
-      else if (Type == CGType::Barrier || Type == CGType::BarrierWaitlist) {
-        Deps.LastBarrier = EventRetImpl;
-        Deps.UnenqueuedCmdEvents.clear();
-      } else if (!EventRetImpl->isEnqueued()) {
-        Deps.UnenqueuedCmdEvents.push_back(EventRetImpl);
-      }
+      return MIsInorder ? finalizeHandlerInOrder(Handler)
+                        : finalizeHandlerOutOfOrder(Handler);
     }
   }
 

sycl/source/handler.cpp

@@ -532,11 +532,11 @@ event handler::finalize() {
 
       if (DiscardEvent) {
         EnqueueKernel();
-        auto EventImpl = std::make_shared<detail::event_impl>(
-            detail::event_impl::HES_Discarded);
-        MLastEvent = detail::createSyclObjFromImpl<event>(std::move(EventImpl));
+        const auto &EventImpl = detail::getSyclObjImpl(MLastEvent);
+        EventImpl->setStateDiscarded();
       } else {
-        NewEvent = std::make_shared<detail::event_impl>(MQueue);
+        NewEvent = detail::getSyclObjImpl(MLastEvent);
+        NewEvent->setQueue(MQueue);
         NewEvent->setWorkerQueue(MQueue);
         NewEvent->setContextImpl(MQueue->getContextImplPtr());
         NewEvent->setStateIncomplete();
@@ -549,8 +549,6 @@ event handler::finalize() {
           NewEvent->getPreparedDepsEvents() = impl->CGData.MEvents;
           NewEvent->cleanDepEventsThroughOneLevel();
         }
-
-        MLastEvent = detail::createSyclObjFromImpl<event>(std::move(NewEvent));
       }
       return MLastEvent;
     }

sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp

@@ -52,6 +52,8 @@ class LimitedHandler {
   sycl::detail::CGType MCGType;
   std::shared_ptr<MockQueueImpl> MQueue;
   std::shared_ptr<sycl::detail::handler_impl> impl;
+  std::shared_ptr<detail::kernel_impl> MKernel;
+  detail::string MKernelName;
 };
 
 // Needed to use EXPECT_CALL to verify depends_on that originally appends lst
@@ -80,17 +82,16 @@ TEST_F(SchedulerTest, InOrderQueueSyncCheck) {
 
   // Check that tasks submitted to an in-order queue implicitly depend_on the
   // previous task, this is needed to properly sync blocking & blocked tasks.
-  sycl::event Event;
   {
     LimitedHandlerSimulation MockCGH{detail::CGType::CodeplayHostTask, Queue};
     EXPECT_CALL(MockCGH, depends_on(An<const sycl::detail::EventImplPtr &>()))
         .Times(0);
-    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, Event);
+    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, std::nullopt);
   }
   {
     LimitedHandlerSimulation MockCGH{detail::CGType::CodeplayHostTask, Queue};
     EXPECT_CALL(MockCGH, depends_on(An<const sycl::detail::EventImplPtr &>()))
         .Times(1);
-    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, Event);
+    Queue->finalizeHandler<LimitedHandlerSimulation>(MockCGH, std::nullopt);
   }
 }