[SYCL][Graph] Error on invalid buffer behaviour

sycl/include/sycl/accessor.hpp

@@ -540,6 +540,7 @@ class __SYCL_EXPORT AccessorBaseHost {
   const range<3> &getMemoryRange() const;
   void *getPtr() const noexcept;
   bool isPlaceholder() const;
+  bool isMemoryObjectUsedByGraph() const;
 
   detail::AccHostDataT &getAccData();
 
@@ -1454,6 +1455,18 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
       typename std::iterator_traits<iterator>::difference_type;
   using size_type = std::size_t;
 
+  /// If creating a host_accessor this checks to see if the underlying memory
+  /// object is currently in use by a command_graph, and throws if it is.
+  void throwIfUsedByGraph() const {
+#ifndef __SYCL_DEVICE_ONLY__
+    if (IsHostBuf && AccessorBaseHost::isMemoryObjectUsedByGraph()) {
+      throw sycl::exception(make_error_code(errc::invalid),
+                            "Host accessors cannot be created for buffers "
+                            "which are currently in use by a command graph.");
+    }
+#endif
+  }
+
   // The list of accessor constructors with their arguments
   // -------+---------+-------+----+-----+--------------
   // Dimensions = 0
@@ -1533,6 +1546,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             detail::getSyclObjImpl(BufferRef).get(), AdjustedDim, sizeof(DataT),
             IsPlaceH, BufferRef.OffsetInBytes, BufferRef.IsSubBuffer,
             PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (!AccessorBaseHost::isPlaceholder())
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
@@ -1572,6 +1586,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             detail::getSyclObjImpl(BufferRef).get(), AdjustedDim, sizeof(DataT),
             IsPlaceH, BufferRef.OffsetInBytes, BufferRef.IsSubBuffer,
             PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (!AccessorBaseHost::isPlaceholder())
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
@@ -1607,6 +1622,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     initHostAcc();
@@ -1643,6 +1659,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     initHostAcc();
@@ -1675,6 +1692,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             IsPlaceH, BufferRef.OffsetInBytes, BufferRef.IsSubBuffer,
             PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (!AccessorBaseHost::isPlaceholder())
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
@@ -1710,6 +1728,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             IsPlaceH, BufferRef.OffsetInBytes, BufferRef.IsSubBuffer,
             PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (!AccessorBaseHost::isPlaceholder())
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
@@ -1772,6 +1791,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     initHostAcc();
@@ -1806,6 +1826,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     initHostAcc();
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
@@ -1981,6 +2002,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), IsPlaceH, BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (!AccessorBaseHost::isPlaceholder())
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
@@ -2023,6 +2045,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), IsPlaceH, BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (!AccessorBaseHost::isPlaceholder())
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
@@ -2094,6 +2117,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (BufferRef.isOutOfBounds(AccessOffset, AccessRange,
                                 BufferRef.get_range()))
@@ -2136,6 +2160,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer, PropertyList) {
+    throwIfUsedByGraph();
     preScreenAccessor(PropertyList);
     if (BufferRef.isOutOfBounds(AccessOffset, AccessRange,
                                 BufferRef.get_range()))

sycl/include/sycl/detail/property_helper.hpp

@@ -46,8 +46,10 @@ enum DataLessPropKind {
   GraphNoCycleCheck = 19,
   QueueSubmissionBatched = 20,
   QueueSubmissionImmediate = 21,
+  GraphAssumeDataOutlivesBuffer = 22,
+  GraphAssumeBufferOutlivesGraph = 23,
   // Indicates the last known dataless property.
-  LastKnownDataLessPropKind = 21,
+  LastKnownDataLessPropKind = 23,
   // Exceeding 32 may cause ABI breaking change on some of OSes.
   DataLessPropKindSize = 32
 };

sycl/include/sycl/ext/oneapi/experimental/graph.hpp

@@ -59,13 +59,35 @@ namespace graph {
 
 /// Property passed to command_graph constructor to disable checking for cycles.
 ///
-/// \todo Cycle check not yet implemented.
 class no_cycle_check : public ::sycl::detail::DataLessProperty<
                            ::sycl::detail::GraphNoCycleCheck> {
 public:
   no_cycle_check() = default;
 };
 
+/// Property passed to command_graph constructor to allow buffers to be used
+/// with graphs. Passing this property represents a promise from the user that
+/// the buffer will outlive any graph that it is used in.
+///
+class assume_buffer_outlives_graph
+    : public ::sycl::detail::DataLessProperty<
+          ::sycl::detail::GraphAssumeBufferOutlivesGraph> {
+public:
+  assume_buffer_outlives_graph() = default;
+};
+
+/// Property passed to command_graph constructor to allow buffers created with
+/// host pointers. Passing this property represents a promise from the user that
+/// the host data will outlive the buffer and by extension any graph that it is
+/// used in.
+///
+class assume_data_outlives_buffer
+    : public ::sycl::detail::DataLessProperty<
+          ::sycl::detail::GraphAssumeDataOutlivesBuffer> {
+public:
+  assume_data_outlives_buffer() = default;
+};
+
 } // namespace graph
 
 namespace node {

sycl/source/accessor.cpp

@@ -7,6 +7,7 @@
 //===----------------------------------------------------------------------===//
 
 #include <detail/queue_impl.hpp>
+#include <detail/sycl_mem_obj_t.hpp>
 #include <sycl/accessor.hpp>
 
 namespace sycl {
@@ -68,6 +69,10 @@ void *AccessorBaseHost::getMemoryObject() const { return impl->MSYCLMemObj; }
 
 bool AccessorBaseHost::isPlaceholder() const { return impl->MIsPlaceH; }
 
+bool AccessorBaseHost::isMemoryObjectUsedByGraph() const {
+  return static_cast<detail::SYCLMemObjT *>(impl->MSYCLMemObj)->isUsedInGraph();
+}
+
 LocalAccessorBaseHost::LocalAccessorBaseHost(
     sycl::range<3> Size, int Dims, int ElemSize,
     const property_list &PropertyList) {

sycl/source/detail/graph_impl.cpp

@@ -12,6 +12,7 @@
 #include <detail/program_manager/program_manager.hpp>
 #include <detail/queue_impl.hpp>
 #include <detail/scheduler/commands.hpp>
+#include <detail/sycl_mem_obj_t.hpp>
 #include <sycl/feature_test.hpp>
 #include <sycl/queue.hpp>
 
@@ -117,6 +118,13 @@ void exec_graph_impl::schedule() {
   }
 }
 
+graph_impl::~graph_impl() {
+  clearQueues();
+  for (auto &MemObj : MMemObjs) {
+    MemObj->markNoLongerBeingUsedInGraph();
+  }
+}
+
 std::shared_ptr<node_impl> graph_impl::addSubgraphNodes(
     const std::list<std::shared_ptr<node_impl>> &NodeList) {
   // Find all input and output nodes from the node list
@@ -214,7 +222,27 @@ graph_impl::add(sycl::detail::CG::CGTYPE CGType,
   // A unique set of dependencies obtained by checking requirements and events
   std::set<std::shared_ptr<node_impl>> UniqueDeps;
   const auto &Requirements = CommandGroup->getRequirements();
+  if (!MAllowBuffers && Requirements.size()) {
+    throw sycl::exception(make_error_code(errc::invalid),
+                          "Cannot use buffers in a graph without passing the "
+                          "assume_buffer_outlives_graph property on "
+                          "Graph construction.");
+  }
+
   for (auto &Req : Requirements) {
+    // Track and mark the memory objects being used by the graph.
+    auto MemObj = static_cast<sycl::detail::SYCLMemObjT *>(Req->MSYCLMemObj);
+    if (MemObj->getUserPtr() && !MAllowBuffersHostPointers) {
+      throw sycl::exception(
+          make_error_code(errc::invalid),
+          "Cannot use a buffer which was created with a host pointer in a "
+          "graph without passing the assume_data_outlives_buffer property on "
+          "Graph construction.");
+    }
+    bool WasInserted = MMemObjs.insert(MemObj).second;
+    if (WasInserted) {
+      MemObj->markBeingUsedInGraph();
+    }
     // Look through the graph for nodes which share this requirement
     for (auto NodePtr : MRoots) {
       checkForRequirement(Req, NodePtr, UniqueDeps);
@@ -253,8 +281,10 @@ graph_impl::add(sycl::detail::CG::CGTYPE CGType,
 bool graph_impl::clearQueues() {
   bool AnyQueuesCleared = false;
   for (auto &Queue : MRecordingQueues) {
-    Queue->setCommandGraph(nullptr);
-    AnyQueuesCleared = true;
+    if (auto ValidQueue = Queue.lock(); ValidQueue) {
+      ValidQueue->setCommandGraph(nullptr);
+      AnyQueuesCleared = true;
+    }
   }
   MRecordingQueues.clear();
 

sycl/source/detail/graph_impl.hpp

@@ -26,6 +26,10 @@
 namespace sycl {
 inline namespace _V1 {
 
+namespace detail {
+class SYCLMemObjT;
+}
+
 namespace ext {
 namespace oneapi {
 namespace experimental {
@@ -373,6 +377,14 @@ class graph_impl {
     if (PropList.has_property<property::graph::no_cycle_check>()) {
       MSkipCycleChecks = true;
     }
+    if (PropList.has_property<property::graph::assume_data_outlives_buffer>()) {
+      MAllowBuffersHostPointers = true;
+    }
+    if (PropList
+            .has_property<property::graph::assume_buffer_outlives_graph>()) {
+      MAllowBuffers = true;
+    }
+
     if (SyclDevice.get_info<
             ext::oneapi::experimental::info::device::graph_support>() ==
         info::graph_support_level::unsupported) {
@@ -385,6 +397,8 @@ class graph_impl {
     }
   }
 
+  ~graph_impl();
+
   /// Remove node from list of root nodes.
   /// @param Root Node to remove from list of root nodes.
   void removeRoot(const std::shared_ptr<node_impl> &Root);
@@ -636,13 +650,19 @@ class graph_impl {
   /// @return True if a cycle is detected, false if not.
   bool checkForCycles();
 
+  /// Insert node into list of root nodes.
+  /// @param Root Node to add to list of root nodes.
+  void addRoot(const std::shared_ptr<node_impl> &Root);
+
   /// Context associated with this graph.
   sycl::context MContext;
   /// Device associated with this graph. All graph nodes will execute on this
   /// device.
   sycl::device MDevice;
   /// Unique set of queues which are currently recording to this graph.
-  std::set<std::shared_ptr<sycl::detail::queue_impl>> MRecordingQueues;
+  std::set<std::weak_ptr<sycl::detail::queue_impl>,
+           std::owner_less<std::weak_ptr<sycl::detail::queue_impl>>>
+      MRecordingQueues;
   /// Map of events to their associated recorded nodes.
   std::unordered_map<std::shared_ptr<sycl::detail::event_impl>,
                      std::shared_ptr<node_impl>>
@@ -656,10 +676,17 @@ class graph_impl {
   /// Controls whether we skip the cycle checks in makeEdge, set by the presence
   /// of the no_cycle_check property on construction.
   bool MSkipCycleChecks = false;
+  /// Unique set of SYCL Memory Objects which are currently in use in the graph.
+  std::set<sycl::detail::SYCLMemObjT *> MMemObjs;
 
-  /// Insert node into list of root nodes.
-  /// @param Root Node to add to list of root nodes.
-  void addRoot(const std::shared_ptr<node_impl> &Root);
+  /// Controls whether we allow buffers that are created with host pointers to
+  /// be used in the graph. Set by the presence of the
+  /// assume_data_outlives_buffer property.
+  bool MAllowBuffersHostPointers = false;
+
+  /// Controls whether we allow buffers to be used in the graph. Set by the
+  /// presence of the assume_buffer_outlives_graph property.
+  bool MAllowBuffers = false;
 };
 
 /// Class representing the implementation of command_graph<executable>.

sycl/source/detail/queue_impl.cpp

@@ -147,7 +147,7 @@ event queue_impl::memcpy(const std::shared_ptr<detail::queue_impl> &Self,
 #endif
   // If we have a command graph set we need to capture the copy through normal
   // queue submission rather than execute the copy directly.
-  if (MGraph) {
+  if (MGraph.lock()) {
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvents);
@@ -476,7 +476,7 @@ void queue_impl::wait(const detail::code_location &CodeLoc) {
   TelemetryEvent = instrumentationProlog(CodeLoc, Name, StreamID, IId);
 #endif
 
-  if (MGraph) {
+  if (MGraph.lock()) {
     throw sycl::exception(make_error_code(errc::invalid),
                           "wait cannot be called for a queue which is "
                           "recording to a command graph.");

sycl/source/detail/queue_impl.hpp

@@ -689,7 +689,7 @@ class queue_impl {
 
   std::shared_ptr<ext::oneapi::experimental::detail::graph_impl>
   getCommandGraph() const {
-    return MGraph;
+    return MGraph.lock();
   }
 
 protected:
@@ -868,8 +868,7 @@ class queue_impl {
 
   // Command graph which is associated with this queue for the purposes of
   // recording commands to it.
-  std::shared_ptr<ext::oneapi::experimental::detail::graph_impl> MGraph =
-      nullptr;
+  std::weak_ptr<ext::oneapi::experimental::detail::graph_impl> MGraph{};
 
   friend class sycl::ext::oneapi::experimental::detail::node_impl;
 };