Rewrite stream init dependence LIT test as unit test

Signed-off-by: Mikhail Lychkov <[email protected]>

Author: mlychkov

sycl/include/CL/sycl/handler.hpp

@@ -65,6 +65,9 @@ template <typename T_Src, int Dims_Src, cl::sycl::access::mode AccessMode_Src,
           cl::sycl::access::placeholder IsPlaceholder_Dst>
 class __copyAcc2Acc;
 
+// For unit testing purposes
+class MockHandler;
+
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 
@@ -1951,6 +1954,8 @@ class __SYCL_EXPORT handler {
   friend void detail::associateWithHandler(handler &,
                                            detail::AccessorBaseHost *,
                                            access::target);
+
+  friend class ::MockHandler;
 };
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)

sycl/source/CMakeLists.txt

@@ -134,6 +134,7 @@ set(SYCL_SOURCES
     "detail/scheduler/commands.cpp"
     "detail/scheduler/leaves_collection.cpp"
     "detail/scheduler/scheduler.cpp"
+    "detail/scheduler/scheduler_helpers.cpp"
     "detail/scheduler/graph_processor.cpp"
     "detail/scheduler/graph_builder.cpp"
     "detail/spec_constant_impl.cpp"

sycl/source/detail/scheduler/scheduler_helpers.cpp

@@ -0,0 +1,55 @@
+//==-- scheduler_helpers.cpp - SYCL Scheduler helper functions --*- C++ -*-===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <CL/sycl/queue.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
+#include <detail/scheduler/scheduler_helpers.hpp>
+#include <detail/stream_impl.hpp>
+
+__SYCL_INLINE_NAMESPACE(cl) {
+namespace sycl {
+namespace detail {
+
+void initStream(StreamImplPtr Stream, QueueImplPtr Queue) {
+  Scheduler::StreamBuffers *StrBufs{};
+
+  {
+    std::lock_guard<std::recursive_mutex> lock(
+        Scheduler::getInstance().StreamBuffersPoolMutex);
+
+    auto StreamBuf =
+        Scheduler::getInstance().StreamBuffersPool.find(Stream.get());
+    assert((StreamBuf != Scheduler::getInstance().StreamBuffersPool.end()) &&
+           "Stream is unexpectedly not found in pool.");
+
+    StrBufs = StreamBuf->second;
+  }
+
+  assert(StrBufs && "No buffers for a stream.");
+
+  // Real size of full flush buffer is saved only in buffer_impl field of
+  // FlushBuf object.
+  size_t FlushBufSize = getSyclObjImpl(StrBufs->FlushBuf)->get_count();
+
+  auto Q = createSyclObjFromImpl<queue>(Queue);
+  Q.submit([&](handler &cgh) {
+    auto FlushBufAcc =
+        StrBufs->FlushBuf.get_access<access::mode::discard_write,
+                                     access::target::host_buffer>(
+            cgh, range<1>(FlushBufSize), id<1>(0));
+    cgh.codeplay_host_task([=] {
+      char *FlushBufPtr = FlushBufAcc.get_pointer();
+      std::memset(FlushBufPtr, 0, FlushBufAcc.get_size());
+    });
+  });
+}
+
+} // namespace detail
+} // namespace sycl
+} // __SYCL_INLINE_NAMESPACE(cl)

sycl/source/detail/scheduler/scheduler_helpers.hpp

@@ -8,46 +8,21 @@
 
 #pragma once
 
-#include <CL/sycl/queue.hpp>
-#include <detail/scheduler/scheduler.hpp>
+#include <CL/sycl/detail/defines_elementary.hpp>
+
+#include <memory>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 namespace detail {
 
-void initStream(StreamImplPtr Stream, QueueImplPtr Queue) {
-  Scheduler::StreamBuffers *StrBufs{};
-
-  {
-    std::lock_guard<std::mutex> lock(
-        Scheduler::getInstance().StreamBuffersPoolMutex);
-
-    auto StreamBuf =
-        Scheduler::getInstance().StreamBuffersPool.find(Stream.get());
-    assert((StreamBuf != Scheduler::getInstance().StreamBuffersPool.end()) &&
-           "Stream is unexpectedly not found in pool.");
-
-    StrBufs = StreamBuf->second;
-  }
-
-  assert(StrBufs && "No buffers for a stream.");
-
-  // Real size of full flush buffer is saved only in buffer_impl field of
-  // FlushBuf object.
-  size_t FlushBufSize = getSyclObjImpl(StrBufs->FlushBuf)->get_count();
-
-  auto Q = createSyclObjFromImpl<queue>(Queue);
-  Q.submit([&](handler &cgh) {
-    auto FlushBufAcc =
-        StrBufs->FlushBuf.get_access<access::mode::discard_write,
-                                     access::target::host_buffer>(
-            cgh, range<1>(FlushBufSize), id<1>(0));
-    cgh.codeplay_host_task([=] {
-      char *FlushBufPtr = FlushBufAcc.get_pointer();
-      std::memset(FlushBufPtr, 0, FlushBufAcc.get_size());
-    });
-  });
-}
+class stream_impl;
+class queue_impl;
+
+using StreamImplPtr = std::shared_ptr<detail::stream_impl>;
+using QueueImplPtr = std::shared_ptr<detail::queue_impl>;
+
+void initStream(StreamImplPtr Stream, QueueImplPtr Queue);
 
 } // namespace detail
 } // namespace sycl

sycl/test/on-device/basic_tests/stream/stream_flush_buf_init_dep.cpp

@@ -9,5 +9,6 @@ add_sycl_unittest(SchedulerTests OBJECT
     LinkedAllocaDependencies.cpp
     LeavesCollection.cpp
     NoUnifiedHostMemory.cpp
+    StreamInitDependencyOnHost.cpp
     utils.cpp
 )

sycl/unittests/scheduler/CMakeLists.txt

@@ -130,6 +130,12 @@ class MockScheduler : public cl::sycl::detail::Scheduler {
                    const cl::sycl::detail::QueueImplPtr &Queue) {
     return MGraphBuilder.insertMemoryMove(Record, Req, Queue);
   }
+
+  cl::sycl::detail::Command *
+  addCG(std::unique_ptr<cl::sycl::detail::CG> CommandGroup,
+        cl::sycl::detail::QueueImplPtr Queue) {
+    return MGraphBuilder.addCG(std::move(CommandGroup), Queue);
+  }
 };
 
 void addEdge(cl::sycl::detail::Command *User, cl::sycl::detail::Command *Dep,

sycl/unittests/scheduler/SchedulerTestUtils.hpp

@@ -0,0 +1,122 @@
+//==------ LinkedAllocaDependencies.cpp --- Scheduler unit tests -----------==//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include "SchedulerTest.hpp"
+#include "SchedulerTestUtils.hpp"
+
+#include <detail/scheduler/scheduler_helpers.hpp>
+
+using namespace cl::sycl;
+
+class MockHandler : public sycl::handler {
+public:
+  MockHandler(shared_ptr_class<detail::queue_impl> Queue, bool IsHost)
+      : sycl::handler(Queue, IsHost) {}
+
+  void setType(detail::CG::CGTYPE Type) {
+    static_cast<sycl::handler *>(this)->MCGType = Type;
+  }
+
+  void addStream(const detail::StreamImplPtr &Stream) {
+    sycl::handler::addStream(Stream);
+  }
+
+  unique_ptr_class<detail::CG> finalize() {
+    auto CGH = static_cast<sycl::handler *>(this);
+    unique_ptr_class<detail::CG> CommandGroup;
+    switch (CGH->MCGType) {
+    case detail::CG::KERNEL:
+    case detail::CG::RUN_ON_HOST_INTEL: {
+      CommandGroup.reset(new detail::CGExecKernel(
+          std::move(CGH->MNDRDesc), std::move(CGH->MHostKernel),
+          std::move(CGH->MKernel), std::move(CGH->MArgsStorage),
+          std::move(CGH->MAccStorage), std::move(CGH->MSharedPtrStorage),
+          std::move(CGH->MRequirements), std::move(CGH->MEvents),
+          std::move(CGH->MArgs), std::move(CGH->MKernelName),
+          std::move(CGH->MOSModuleHandle), std::move(CGH->MStreamStorage),
+          CGH->MCGType, CGH->MCodeLoc));
+      break;
+    }
+    default:
+      throw runtime_error("Unhandled type of command group",
+                          PI_INVALID_OPERATION);
+    }
+
+    return CommandGroup;
+  }
+};
+
+using CmdTypeTy = cl::sycl::detail::Command::CommandType;
+
+static bool ValidateDepCommandsTree(const detail::Command *Cmd,
+                                    std::queue<CmdTypeTy> DepCmdsTypes,
+                                    const detail::SYCLMemObjI *MemObj) {
+  if (DepCmdsTypes.empty())
+    return true;
+  else if (!Cmd)
+    return false;
+
+  CmdTypeTy DepCmdType = DepCmdsTypes.front();
+  DepCmdsTypes.pop();
+
+  for (const detail::DepDesc &Dep : Cmd->MDeps) {
+    if (Dep.MDepCommand && (Dep.MDepCommand->getType() == DepCmdType) &&
+        Dep.MDepRequirement && (Dep.MDepRequirement->MSYCLMemObj == MemObj) &&
+        ValidateDepCommandsTree(Dep.MDepCommand, DepCmdsTypes, MemObj))
+      return true;
+  }
+
+  return false;
+}
+
+TEST_F(SchedulerTest, StreamInitDependencyOnHost) {
+  cl::sycl::queue HQueue(host_selector{});
+  detail::QueueImplPtr HQueueImpl = detail::getSyclObjImpl(HQueue);
+
+  // Emulating processing of command group function
+  MockHandler MockCGH(HQueueImpl, true);
+  MockCGH.setType(detail::CG::KERNEL);
+
+  // Emulating construction of stream object inside command group
+  detail::StreamImplPtr StreamImpl =
+      std::make_shared<detail::stream_impl>(1024, 200, MockCGH);
+  detail::GlobalBufAccessorT FlushBufAcc =
+      StreamImpl->accessGlobalFlushBuf(MockCGH);
+  MockCGH.addStream(StreamImpl);
+
+  detail::SYCLMemObjI *FlushBufMemObjPtr =
+      detail::getSyclObjImpl(FlushBufAcc)->MSYCLMemObj;
+  ASSERT_TRUE(!!FlushBufMemObjPtr)
+      << "Memory object for stream flush buffer not initialized";
+
+  unique_ptr_class<detail::CG> MainCG = MockCGH.finalize();
+
+  // Emulate call of Scheduler::addCG
+  vector_class<detail::StreamImplPtr> Streams =
+      static_cast<detail::CGExecKernel *>(MainCG.get())->getStreams();
+  ASSERT_EQ(Streams.size(), 1u) << "Invalid number of stream objects";
+
+  initStream(Streams[0], HQueueImpl);
+
+  MockScheduler MS;
+  detail::Command *NewCmd = MS.addCG(std::move(MainCG), HQueueImpl);
+  ASSERT_TRUE(!!NewCmd) << "Failed to add command group into scheduler";
+  ASSERT_GT(NewCmd->MDeps.size(), 0u)
+      << "No deps appeared in the new exec kernel command";
+
+  // Searching in dependencies for CG execution command that initializes flush
+  // buffer of a stream that is supposed to be used inside NewCmd's CG.
+  // Tree of dependencies should look like:
+  // [MAIN_CG] -> [EMPTY_NODE {FlushBufMemObj}] -> [FILL_CG {FlushBufMemObj}] ->
+  //     [[ALLOC_TASK {FlushBufMemObj}]
+  std::queue<CmdTypeTy> DepCmdsTypes({CmdTypeTy::EMPTY_TASK,
+                                      CmdTypeTy::RUN_CG, // FILL_CG
+                                      CmdTypeTy::ALLOCA});
+  ASSERT_TRUE(ValidateDepCommandsTree(NewCmd, DepCmdsTypes, FlushBufMemObjPtr))
+      << "Dependency on stream flush buffer initialization not found";
+}