[DO NOT MERGE] Graph cleanup experiments

A very rough implementation of cleaning up command nodes after they're
enqueued and stop being leaves, alloca commands excluded. Handles only a
subset of cases.

Author: sergey-semenov

sycl/source/detail/event_impl.cpp

@@ -235,8 +235,6 @@ void event_impl::wait_and_throw(
 
 void event_impl::cleanupCommand(
     std::shared_ptr<cl::sycl::detail::event_impl> Self) const {
-  if (MCommand && !SYCLConfig<SYCL_DISABLE_EXECUTION_GRAPH_CLEANUP>::get())
-    detail::Scheduler::getInstance().cleanupFinishedCommands(std::move(Self));
 }
 
 template <>

sycl/source/detail/scheduler/commands.cpp

@@ -260,8 +260,9 @@ class DispatchHostTask {
     // of empty command.
     // Also, it's possible to have record deallocated prior to enqueue process.
     // Thus we employ read-lock of graph.
+    std::vector<Command *> EnqueuedCmds;
+    Scheduler &Sched = Scheduler::getInstance();
     {
-      Scheduler &Sched = Scheduler::getInstance();
       Scheduler::ReadLockT Lock(Sched.MGraphLock);
 
       std::vector<DepDesc> Deps = MThisCmd->MDeps;
@@ -272,8 +273,10 @@ class DispatchHostTask {
       EmptyCmd->MEnqueueStatus = EnqueueResultT::SyclEnqueueReady;
 
       for (const DepDesc &Dep : Deps)
-        Scheduler::enqueueLeavesOfReqUnlocked(Dep.MDepRequirement);
+        Scheduler::enqueueLeavesOfReqUnlocked(Dep.MDepRequirement,
+                                              EnqueuedCmds);
     }
+    Sched.cleanupCommands(EnqueuedCmds);
   }
 };
 
@@ -614,7 +617,7 @@ void Command::emitInstrumentation(uint16_t Type, const char *Txt) {
 #endif
 }
 
-bool Command::enqueue(EnqueueResultT &EnqueueResult, BlockingT Blocking) {
+bool Command::enqueue(EnqueueResultT &EnqueueResult, BlockingT Blocking, std::vector<Command *> &EnqueuedCommands) {
   // Exit if already enqueued
   if (MEnqueueStatus == EnqueueResultT::SyclEnqueueSuccess)
     return true;
@@ -683,6 +686,8 @@ bool Command::enqueue(EnqueueResultT &EnqueueResult, BlockingT Blocking) {
     // Consider the command is successfully enqueued if return code is
     // CL_SUCCESS
     MEnqueueStatus = EnqueueResultT::SyclEnqueueSuccess;
+    if (MLeafCounter == 0 && (!MDeps.empty() || !MUsers.empty()))
+      EnqueuedCommands.push_back(this);
   }
 
   // Emit this correlation signal before the task end

sycl/source/detail/scheduler/commands.hpp

@@ -124,7 +124,7 @@ class Command {
   /// \param Blocking if this argument is true, function will wait for the
   ///        command to be unblocked before calling enqueueImp.
   /// \return true if the command is enqueued.
-  virtual bool enqueue(EnqueueResultT &EnqueueResult, BlockingT Blocking);
+  virtual bool enqueue(EnqueueResultT &EnqueueResult, BlockingT Blocking, std::vector<Command *> &EnqueuedCommands);
 
   bool isFinished();
 

sycl/source/detail/scheduler/graph_builder.cpp

@@ -194,6 +194,8 @@ MemObjRecord *Scheduler::GraphBuilder::getOrInsertMemObjRecord(
           ToEnqueue.push_back(ConnectionCmd);
         Dependency->addUser(Dependant);
         --(Dependency->MLeafCounter);
+        if (Dependency->MLeafCounter == 0 && Dependency->isSuccessfullyEnqueued())
+          cleanupCommand(Dependency);
       };
 
   const ContextImplPtr &InteropCtxPtr = Req->MSYCLMemObj->getInteropContext();
@@ -225,17 +227,25 @@ MemObjRecord *Scheduler::GraphBuilder::getOrInsertMemObjRecord(
   return MemObject->MRecord.get();
 }
 
-void Scheduler::GraphBuilder::updateLeaves(const std::set<Command *> &Cmds,
-                                           MemObjRecord *Record,
-                                           access::mode AccessMode) {
+void Scheduler::GraphBuilder::updateLeaves(
+    const std::set<Command *> &Cmds, MemObjRecord *Record,
+    access::mode AccessMode, std::vector<Command *> *CommandsToCleanUp) {
 
   const bool ReadOnlyReq = AccessMode == access::mode::read;
   if (ReadOnlyReq)
     return;
 
   for (Command *Cmd : Cmds) {
+    bool WasLeaf = Cmd->MLeafCounter > 0;
     Cmd->MLeafCounter -= Record->MReadLeaves.remove(Cmd);
     Cmd->MLeafCounter -= Record->MWriteLeaves.remove(Cmd);
+    if (Cmd->MLeafCounter == 0 && Cmd->isSuccessfullyEnqueued()) {
+      if (CommandsToCleanUp) {
+        if (WasLeaf)
+          CommandsToCleanUp->push_back(Cmd);
+      } else
+        cleanupCommand(Cmd);
+    }
   }
 }
 
@@ -963,14 +973,23 @@ Scheduler::GraphBuilder::addCG(std::unique_ptr<detail::CG> CommandGroup,
   // Node dependencies can be modified further when adding the node to leaves,
   // iterate over their copy.
   // FIXME employ a reference here to eliminate copying of a vector
+  // Updating leaves might also clean up some of the dep commands, so update
+  // their users first.
+  // FIXME there's probably a better way of handling cleanup & leaf/dep update
+  // here considering that some of the updated might be destroyed by cleanup
+  // immediately after.
   std::vector<DepDesc> Deps = NewCmd->MDeps;
+  std::vector<Command *> CommandsToCleanUp;
   for (DepDesc &Dep : Deps) {
     Dep.MDepCommand->addUser(NewCmd.get());
     const Requirement *Req = Dep.MDepRequirement;
     MemObjRecord *Record = getMemObjRecord(Req->MSYCLMemObj);
-    updateLeaves({Dep.MDepCommand}, Record, Req->MAccessMode);
+    updateLeaves({Dep.MDepCommand}, Record, Req->MAccessMode,
+                 &CommandsToCleanUp);
     addNodeToLeaves(Record, NewCmd.get(), Req->MAccessMode, ToEnqueue);
   }
+  for (Command *Cmd : CommandsToCleanUp)
+    cleanupCommand(Cmd);
 
   // Register all the events as dependencies
   for (detail::EventImplPtr e : Events) {
@@ -993,9 +1012,13 @@ void Scheduler::GraphBuilder::decrementLeafCountersForRecord(
     MemObjRecord *Record) {
   for (Command *Cmd : Record->MReadLeaves) {
     --(Cmd->MLeafCounter);
+    if (Cmd->MLeafCounter == 0 && Cmd->isSuccessfullyEnqueued())
+      cleanupCommand(Cmd);
   }
   for (Command *Cmd : Record->MWriteLeaves) {
     --(Cmd->MLeafCounter);
+    if (Cmd->MLeafCounter == 0 && Cmd->isSuccessfullyEnqueued())
+      cleanupCommand(Cmd);
   }
 }
 
@@ -1096,6 +1119,52 @@ void Scheduler::GraphBuilder::cleanupCommandsForRecord(
   handleVisitedNodes(MVisitedCmds);
 }
 
+
+void Scheduler::GraphBuilder::cleanupCommand(Command *Cmd) {
+  if (SYCLConfig<SYCL_DISABLE_EXECUTION_GRAPH_CLEANUP>::get())
+    return;
+  assert(Cmd->MLeafCounter == 0 && Cmd->isSuccessfullyEnqueued());
+  // Isolated command nodes are cleaned up by scheduler instead.
+  assert(Cmd->MDeps.size() != 0 || Cmd->MUsers.size() != 0);
+  Command::CommandType CmdT = Cmd->getType();
+  // Allocas have to be kept alive until memory objects are released.
+  if (CmdT == Command::ALLOCA || CmdT == Command::ALLOCA_SUB_BUF)
+    return;
+
+  // FIXME handle host tasks
+  if (CmdT == Command::RUN_CG) {
+    auto *ExecCGCmd = static_cast<ExecCGCommand *>(Cmd);
+    if (ExecCGCmd->getCG().getType() == CG::CGTYPE::CodeplayHostTask) {
+      return;
+    }
+  }
+  assert(CmdT != Command::ALLOCA && CmdT != Command::ALLOCA_SUB_BUF);
+
+  for (Command *UserCmd : Cmd->MUsers) {
+    for (DepDesc &Dep : UserCmd->MDeps) {
+      // Link the users of the command to the alloca command(s) instead
+      if (Dep.MDepCommand == Cmd) {
+        // ... unless the user is the alloca itself.
+        if (Dep.MAllocaCmd == UserCmd) {
+          Dep.MDepCommand = nullptr;
+        }
+        else {
+          Dep.MDepCommand = Dep.MAllocaCmd;
+          Dep.MDepCommand->MUsers.insert(UserCmd);
+        }
+      }
+    }
+  }
+  // Update dependency users
+  for (DepDesc &Dep : Cmd->MDeps) {
+    Command *DepCmd = Dep.MDepCommand;
+    DepCmd->MUsers.erase(Cmd);
+  }
+  Cmd->getEvent()->setCommand(nullptr);
+  Cmd->getEvent()->cleanupDependencyEvents();
+  delete Cmd;
+}
+
 void Scheduler::GraphBuilder::cleanupFinishedCommands(
     Command *FinishedCmd,
     std::vector<std::shared_ptr<stream_impl>> &StreamsToDeallocate) {

sycl/source/detail/scheduler/graph_processor.cpp

@@ -23,15 +23,15 @@ static Command *getCommand(const EventImplPtr &Event) {
 
 void Scheduler::GraphProcessor::waitForEvent(EventImplPtr Event,
                                              ReadLockT &GraphReadLock,
-                                             bool LockTheLock) {
+                                             std::vector<Command *> &EnqueuedCmds, bool LockTheLock) {
   Command *Cmd = getCommand(Event);
   // Command can be nullptr if user creates cl::sycl::event explicitly or the
   // event has been waited on by another thread
   if (!Cmd)
     return;
 
   EnqueueResultT Res;
-  bool Enqueued = enqueueCommand(Cmd, Res, BLOCKING);
+  bool Enqueued = enqueueCommand(Cmd, Res, EnqueuedCmds, BLOCKING);
   if (!Enqueued && EnqueueResultT::SyclEnqueueFailed == Res.MResult)
     // TODO: Reschedule commands.
     throw runtime_error("Enqueue process failed.", PI_INVALID_OPERATION);
@@ -47,7 +47,7 @@ void Scheduler::GraphProcessor::waitForEvent(EventImplPtr Event,
 
 bool Scheduler::GraphProcessor::enqueueCommand(Command *Cmd,
                                                EnqueueResultT &EnqueueResult,
-                                               BlockingT Blocking) {
+                                               std::vector<Command *> &EnqueuedCommands, BlockingT Blocking) {
   if (!Cmd || Cmd->isSuccessfullyEnqueued())
     return true;
 
@@ -60,7 +60,7 @@ bool Scheduler::GraphProcessor::enqueueCommand(Command *Cmd,
   // 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, Blocking))
+    if (!enqueueCommand(Dep.MDepCommand, EnqueueResult, EnqueuedCommands, Blocking))
       return false;
   }
 
@@ -76,7 +76,7 @@ bool Scheduler::GraphProcessor::enqueueCommand(Command *Cmd,
   // implemented.
   for (const EventImplPtr &Event : Cmd->getPreparedHostDepsEvents()) {
     if (Command *DepCmd = static_cast<Command *>(Event->getCommand()))
-      if (!enqueueCommand(DepCmd, EnqueueResult, Blocking))
+      if (!enqueueCommand(DepCmd, EnqueueResult, EnqueuedCommands, Blocking))
         return false;
   }
 
@@ -93,7 +93,7 @@ bool Scheduler::GraphProcessor::enqueueCommand(Command *Cmd,
   // on completion of C and starts cleanup process. This thread is still in the
   // middle of enqueue of B. The other thread modifies dependency list of A by
   // removing C out of it. Iterators become invalid.
-  return Cmd->enqueue(EnqueueResult, Blocking);
+  return Cmd->enqueue(EnqueueResult, Blocking, EnqueuedCommands);
 }
 
 } // namespace detail