[SYCL] Fix depends_on handling with pi commands (#5901)

Fixes two related issues:
1) if pi task is blocked by host task or host accessor it can not be enqueued and piEvent is not present in its event_impl. When we schedule new pi task with explicit (depends_on) dependency on the first one - its is absent in MDeps since we have no usual memory dependencies and present in MPreparedDepsEvents. MPreparedDepsEvents is used in enqueueImp for obtaining piEvents. Any events from MPreparedDepsEvents w/o pi events will be just skipped.
AddDep always call processDepEvent which distributes events to MPreparedDepsEvents (pi event expected) and MPreparedHostDepsEvents (no pi event) so replacement of MDeps in enqueueCommand should be valid.
2) if we have kernel w/o usual memory dependencies (MDeps & MUsers are empty) blocked kernel will be just "cleanup" and its execution will be skipped.

Signed-off-by: Tikhomirova, Kseniya <[email protected]>

Author: KseniyaTikhomirova

sycl/source/detail/event_impl.cpp

@@ -342,10 +342,16 @@ event_impl::get_info<info::event::command_execution_status>() {
   if (MState == HES_Discarded)
     return info::event_command_status::ext_oneapi_unknown;
 
-  if (!MHostEvent && MEvent) {
-    return get_event_info<info::event::command_execution_status>(
-        this->getHandleRef(), this->getPlugin());
+  if (!MHostEvent) {
+    // Command is enqueued and PiEvent is ready
+    if (MEvent)
+      return get_event_info<info::event::command_execution_status>(
+          this->getHandleRef(), this->getPlugin());
+    // Command is blocked and not enqueued, PiEvent is not assigned yet
+    else if (MCommand)
+      return sycl::info::event_command_status::submitted;
   }
+
   return MHostEvent && MState.load() != HES_Complete
              ? sycl::info::event_command_status::submitted
              : info::event_command_status::complete;

sycl/source/detail/event_impl.hpp

@@ -212,6 +212,11 @@ class event_impl {
   }
   bool needsCleanupAfterWait() { return MNeedsCleanupAfterWait; }
 
+  /// Returns worker queue for command.
+  ///
+  /// @return a reference to MWorkerQueue.
+  QueueImplPtr &getWorkerQueue() { return MWorkerQueue; };
+
   /// Checks if an event is in a fully intialized state. Default-constructed
   /// events will return true only after having initialized its native event,
   /// while other events will assume that they are fully initialized at
@@ -243,6 +248,8 @@ class event_impl {
   std::weak_ptr<queue_impl> MQueue;
   const bool MIsProfilingEnabled = false;
 
+  QueueImplPtr MWorkerQueue;
+
   /// Dependency events prepared for waiting by backend.
   std::vector<EventImplPtr> MPreparedDepsEvents;
   std::vector<EventImplPtr> MPreparedHostDepsEvents;

sycl/source/detail/queue_impl.hpp

@@ -120,9 +120,7 @@ class queue_impl {
     }
     if (!MHostQueue) {
       const QueueOrder QOrder =
-          MPropList.has_property<property::queue::in_order>()
-              ? QueueOrder::Ordered
-              : QueueOrder::OOO;
+          MIsInorder ? QueueOrder::Ordered : QueueOrder::OOO;
       MQueues.push_back(createQueue(QOrder));
     }
   }
@@ -202,6 +200,8 @@ class queue_impl {
   /// \return true if this queue has discard_events support.
   bool has_discard_events_support() const { return MHasDiscardEventsSupport; }
 
+  bool isInOrder() const { return MIsInorder; }
+
   /// Queries SYCL queue for information.
   ///
   /// The return type depends on information being queried.

sycl/source/detail/scheduler/commands.cpp

@@ -203,17 +203,34 @@ static std::string commandToName(Command::CommandType Type) {
 }
 #endif
 
-static std::vector<RT::PiEvent>
-getPiEvents(const std::vector<EventImplPtr> &EventImpls) {
+std::vector<RT::PiEvent>
+Command::getPiEvents(const std::vector<EventImplPtr> &EventImpls) const {
   std::vector<RT::PiEvent> RetPiEvents;
   for (auto &EventImpl : EventImpls) {
-    if (EventImpl->getHandleRef() != nullptr)
-      RetPiEvents.push_back(EventImpl->getHandleRef());
+    if (EventImpl->getHandleRef() == nullptr)
+      continue;
+
+    // Do not add redundant event dependencies for in-order queues.
+    // At this stage dependency is definitely pi task and need to check if
+    // current one is a host task. In this case we should not skip pi event due
+    // to different sync mechanisms for different task types on in-order queue.
+    const QueueImplPtr &WorkerQueue = getWorkerQueue();
+    if (EventImpl->getWorkerQueue() == WorkerQueue &&
+        WorkerQueue->isInOrder() && !isHostTask())
+      continue;
+
+    RetPiEvents.push_back(EventImpl->getHandleRef());
   }
 
   return RetPiEvents;
 }
 
+bool Command::isHostTask() const {
+  return (MType == CommandType::RUN_CG) /* host task has this type also */ &&
+         ((static_cast<const ExecCGCommand *>(this))->getCG().getType() ==
+          CG::CGTYPE::CodeplayHostTask);
+}
+
 static void flushCrossQueueDeps(const std::vector<EventImplPtr> &EventImpls,
                                 const QueueImplPtr &Queue) {
   for (auto &EventImpl : EventImpls) {
@@ -240,7 +257,8 @@ class DispatchHostTask {
     // sophisticated waiting mechanism to allow to utilize this thread for any
     // other available job and resume once all required events are ready.
     for (auto &PluginWithEvents : RequiredEventsPerPlugin) {
-      std::vector<RT::PiEvent> RawEvents = getPiEvents(PluginWithEvents.second);
+      std::vector<RT::PiEvent> RawEvents =
+          MThisCmd->getPiEvents(PluginWithEvents.second);
       try {
         PluginWithEvents.first->call<PiApiKind::piEventsWait>(RawEvents.size(),
                                                               RawEvents.data());
@@ -393,10 +411,12 @@ void Command::waitForEvents(QueueImplPtr Queue,
 Command::Command(CommandType Type, QueueImplPtr Queue)
     : MQueue(std::move(Queue)),
       MEvent(std::make_shared<detail::event_impl>(MQueue)),
+      MWorkerQueue(MEvent->getWorkerQueue()),
       MPreparedDepsEvents(MEvent->getPreparedDepsEvents()),
       MPreparedHostDepsEvents(MEvent->getPreparedHostDepsEvents()),
       MType(Type) {
   MSubmittedQueue = MQueue;
+  MWorkerQueue = MQueue;
   MEvent->setCommand(this);
   MEvent->setContextImpl(MQueue->getContextImplPtr());
   MEvent->setStateIncomplete();
@@ -600,12 +620,6 @@ Command *Command::processDepEvent(EventImplPtr DepEvent, const DepDesc &Dep,
 
   Command *ConnectionCmd = nullptr;
 
-  // Do not add redundant event dependencies for in-order queues.
-  if (Dep.MDepCommand && Dep.MDepCommand->getWorkerQueue() == WorkerQueue &&
-      WorkerQueue->has_property<property::queue::in_order>() &&
-      getType() != CommandType::HOST_TASK)
-    return nullptr;
-
   ContextImplPtr DepEventContext = DepEvent->getContextImpl();
   // If contexts don't match we'll connect them using host task
   if (DepEventContext != WorkerContext && !WorkerContext->is_host()) {
@@ -621,14 +635,14 @@ const ContextImplPtr &Command::getWorkerContext() const {
   return MQueue->getContextImplPtr();
 }
 
-const QueueImplPtr &Command::getWorkerQueue() const { return MQueue; }
+const QueueImplPtr &Command::getWorkerQueue() const {
+  assert(MWorkerQueue && "MWorkerQueue must not be nullptr");
+  return MWorkerQueue;
+}
 
 bool Command::producesPiEvent() const { return true; }
 
-bool Command::supportsPostEnqueueCleanup() const {
-  // Isolated commands are cleaned up separately
-  return !MUsers.empty() || !MDeps.empty();
-}
+bool Command::supportsPostEnqueueCleanup() const { return true; }
 
 Command *Command::addDep(DepDesc NewDep, std::vector<Command *> &ToCleanUp) {
   Command *ConnectionCmd = nullptr;
@@ -1298,6 +1312,9 @@ MemCpyCommand::MemCpyCommand(Requirement SrcReq,
   if (!MSrcQueue->is_host()) {
     MEvent->setContextImpl(MSrcQueue->getContextImplPtr());
   }
+
+  MWorkerQueue = MQueue->is_host() ? MSrcQueue : MQueue;
+
   emitInstrumentationDataProxy();
 }
 
@@ -1335,10 +1352,6 @@ const ContextImplPtr &MemCpyCommand::getWorkerContext() const {
   return getWorkerQueue()->getContextImplPtr();
 }
 
-const QueueImplPtr &MemCpyCommand::getWorkerQueue() const {
-  return MQueue->is_host() ? MSrcQueue : MQueue;
-}
-
 bool MemCpyCommand::producesPiEvent() const {
   // TODO remove this workaround once the batching issue is addressed in Level
   // Zero plugin.
@@ -1481,6 +1494,8 @@ MemCpyCommandHost::MemCpyCommandHost(Requirement SrcReq,
     MEvent->setContextImpl(MSrcQueue->getContextImplPtr());
   }
 
+  MWorkerQueue = MQueue->is_host() ? MSrcQueue : MQueue;
+
   emitInstrumentationDataProxy();
 }
 
@@ -1518,10 +1533,6 @@ const ContextImplPtr &MemCpyCommandHost::getWorkerContext() const {
   return getWorkerQueue()->getContextImplPtr();
 }
 
-const QueueImplPtr &MemCpyCommandHost::getWorkerQueue() const {
-  return MQueue->is_host() ? MSrcQueue : MQueue;
-}
-
 pi_int32 MemCpyCommandHost::enqueueImp() {
   const QueueImplPtr &Queue = getWorkerQueue();
   waitForPreparedHostEvents();

sycl/source/detail/scheduler/commands.hpp

@@ -199,19 +199,28 @@ class Command {
 
   /// Get the queue this command will be submitted to. Could differ from MQueue
   /// for memory copy commands.
-  virtual const QueueImplPtr &getWorkerQueue() const;
+  const QueueImplPtr &getWorkerQueue() const;
 
   /// Returns true iff the command produces a PI event on non-host devices.
   virtual bool producesPiEvent() const;
 
   /// Returns true iff this command can be freed by post enqueue cleanup.
   virtual bool supportsPostEnqueueCleanup() const;
 
+  /// Collect PI events from EventImpls and filter out some of them in case of
+  /// in order queue
+  std::vector<RT::PiEvent>
+  getPiEvents(const std::vector<EventImplPtr> &EventImpls) const;
+
+  bool isHostTask() const;
+
 protected:
   QueueImplPtr MQueue;
   QueueImplPtr MSubmittedQueue;
   EventImplPtr MEvent;
 
+  QueueImplPtr &MWorkerQueue;
+
   /// Dependency events prepared for waiting by backend.
   /// See processDepEvent for details.
   std::vector<EventImplPtr> &MPreparedDepsEvents;
@@ -252,6 +261,10 @@ class Command {
     return MPreparedHostDepsEvents;
   }
 
+  const std::vector<EventImplPtr> &getPreparedDepsEvents() const {
+    return MPreparedDepsEvents;
+  }
+
   /// Contains list of dependencies(edges)
   std::vector<DepDesc> MDeps;
   /// Contains list of commands that depend on the command.
@@ -492,7 +505,6 @@ class MemCpyCommand : public Command {
   const Requirement *getRequirement() const final { return &MDstReq; }
   void emitInstrumentationData() final;
   const ContextImplPtr &getWorkerContext() const final;
-  const QueueImplPtr &getWorkerQueue() const final;
   bool producesPiEvent() const final;
 
 private:
@@ -517,7 +529,6 @@ class MemCpyCommandHost : public Command {
   const Requirement *getRequirement() const final { return &MDstReq; }
   void emitInstrumentationData() final;
   const ContextImplPtr &getWorkerContext() const final;
-  const QueueImplPtr &getWorkerQueue() const final;
 
 private:
   pi_int32 enqueueImp() final;

sycl/source/detail/scheduler/graph_processor.cpp

@@ -58,23 +58,20 @@ bool Scheduler::GraphProcessor::enqueueCommand(
     return false;
   }
 
-  // Recursively enqueue all the dependencies first and
-  // exit immediately if any of the commands cannot be enqueued.
-  for (DepDesc &Dep : Cmd->MDeps) {
-    if (!enqueueCommand(Dep.MDepCommand, EnqueueResult, ToCleanUp, Blocking))
-      return false;
+  // Recursively enqueue all the implicit + explicit backend level dependencies
+  // first and exit immediately if any of the commands cannot be enqueued.
+  for (const EventImplPtr &Event : Cmd->getPreparedDepsEvents()) {
+    if (Command *DepCmd = static_cast<Command *>(Event->getCommand()))
+      if (!enqueueCommand(DepCmd, EnqueueResult, ToCleanUp, Blocking))
+        return false;
   }
 
-  // Asynchronous host operations (amongst dependencies of an arbitrary command)
-  // are not supported (see Command::processDepEvent method). This impacts
-  // operation of host-task feature a lot with hangs and long-runs. Hence we
-  // have this workaround here.
-  // This workaround is safe as long as the only asynchronous host operation we
-  // have is a host task.
-  // This may iterate over some of dependencies in Cmd->MDeps. Though, the
-  // enqueue operation is idempotent and the second call will result in no-op.
-  // TODO remove the workaround when proper fix for host-task dispatching is
-  // implemented.
+  // Recursively enqueue all the implicit + explicit host dependencies and
+  // exit immediately if any of the commands cannot be enqueued.
+  // Host task execution is asynchronous. In current implementation enqueue for
+  // this command will wait till host task completion by waitInternal call on
+  // MHostDepsEvents. TO FIX: implement enqueue of blocked commands on host task
+  // completion stage and eliminate this event waiting in enqueue.
   for (const EventImplPtr &Event : Cmd->getPreparedHostDepsEvents()) {
     if (Command *DepCmd = static_cast<Command *>(Event->getCommand()))
       if (!enqueueCommand(DepCmd, EnqueueResult, ToCleanUp, Blocking))

sycl/source/detail/scheduler/scheduler.cpp

@@ -157,12 +157,6 @@ EventImplPtr Scheduler::addCG(std::unique_ptr<detail::CG> CommandGroup,
         CleanUp();
         std::rethrow_exception(std::current_exception());
       }
-
-      // If there are no memory dependencies decouple and free the command.
-      // Though, dismiss ownership of native kernel command group as it's
-      // resources may be in use by backend and synchronization point here is
-      // at native kernel execution finish.
-      CleanUp();
     }
   }
   cleanupCommands(ToCleanUp);

sycl/unittests/scheduler/CMakeLists.txt

@@ -22,4 +22,5 @@ add_sycl_unittest(SchedulerTests OBJECT
     LeafLimitDiffContexts.cpp
     InOrderQueueSyncCheck.cpp
     RunOnHostIntelCG.cpp
+    EnqueueWithDependsOnDeps.cpp
 )