[SYCL][Graph] Backend integration and feature additions for SYCL Graphs (3/4) (#10033)

# Backend integration and feature additions for SYCL Graphs
This is the third patch of a series that adds support for an
[experimental command graph
extension](#5626)

A snapshot of the complete work can be seen in draft PR #9375 which has
support all the specification defined ways of
adding nodes and edges to the graph, including both Explicit and Record
& Replay graph construction. The two types of nodes currently
implemented are kernel execution and memcpy commands.

See https://github.com/reble/llvm#implementation-status for the status
of our total work.

## Scope
This third patch focuses on integrating the graphs runtime with the
backend support added in #9992 as well as any remaining runtime features
and bug fixes, and should include no ABI-breaking changes:
* Graphs runtime changes to use PI/UR command-buffers.
* Various improvements to the Graphs runtime classes.
* New memory manager methods for appending copies to a command-buffer.
* Changes to the Scheduler and related CG classes to enable Graphs.
* Device info query for command-graph support.
* Minor changes to some runtime classes to enable Graphs.

## Following Split PRs
Future follow-up PRs with the remainder of our work on the extension
will include:
* Add end-to-end tests for SYCL Graph extension. (4/4)
* NFC changes - Design doc and codeowner update.

## Authors
Co-authored-by: Pablo Reble <[email protected]>
Co-authored-by: Julian Miller <[email protected]>
Co-authored-by: Ben Tracy <[email protected]>
Co-authored-by: Ewan Crawford <[email protected]>
Co-authored-by: Maxime France-Pillois
<[email protected]>

Author: 4 people

sycl/include/sycl/detail/cg.hpp

@@ -497,6 +497,17 @@ class CGCopyFromDeviceGlobal : public CG {
   size_t getOffset() { return MOffset; }
 };
 
+/// "Execute command-buffer" command group class.
+class CGExecCommandBuffer : public CG {
+public:
+  sycl::detail::pi::PiExtCommandBuffer MCommandBuffer;
+
+  CGExecCommandBuffer(sycl::detail::pi::PiExtCommandBuffer CommandBuffer,
+                      CG::StorageInitHelper CGData)
+      : CG(CGTYPE::ExecCommandBuffer, std::move(CGData)),
+        MCommandBuffer(CommandBuffer) {}
+};
+
 } // namespace detail
 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
 } // namespace sycl

sycl/include/sycl/detail/pi.h

@@ -2284,7 +2284,7 @@ __SYCL_EXPORT pi_result piextCommandBufferMemBufferCopyRect(
 /// \param sync_point The sync_point associated with this memory operation.
 __SYCL_EXPORT pi_result piextCommandBufferMemBufferRead(
     pi_ext_command_buffer command_buffer, pi_mem buffer, size_t offset,
-    size_t size, void *dst, pi_uint32 num_events_in_wait_list,
+    size_t size, void *dst, pi_uint32 num_sync_points_in_wait_list,
     const pi_ext_sync_point *sync_point_wait_list,
     pi_ext_sync_point *sync_point);
 

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

@@ -221,8 +221,6 @@ class __SYCL_EXPORT executable_command_graph {
 
   int MTag;
   std::shared_ptr<detail::exec_graph_impl> impl;
-
-  friend class modifiable_command_graph;
 };
 } // namespace detail
 
@@ -248,7 +246,10 @@ class command_graph : public detail::modifiable_command_graph {
 template <>
 class command_graph<graph_state::executable>
     : public detail::executable_command_graph {
-private:
+
+protected:
+  friend command_graph<graph_state::executable>
+  detail::modifiable_command_graph::finalize(const sycl::property_list &) const;
   using detail::executable_command_graph::executable_command_graph;
 };
 

sycl/include/sycl/handler.hpp

@@ -1542,6 +1542,12 @@ class __SYCL_EXPORT handler {
     setType(detail::CG::CodeplayHostTask);
   }
 
+  /// @brief Get the command graph if any associated with this handler. It can
+  /// come from either the associated queue or from being set explicitly through
+  /// the appropriate constructor.
+  std::shared_ptr<ext::oneapi::experimental::detail::graph_impl>
+  getCommandGraph() const;
+
 public:
   handler(const handler &) = delete;
   handler(handler &&) = delete;

sycl/include/sycl/info/ext_oneapi_device_traits.def

@@ -9,6 +9,10 @@ __SYCL_PARAM_TRAITS_TEMPLATE_SPEC(ext::oneapi::experimental,device, max_work_gro
 __SYCL_PARAM_TRAITS_SPEC(ext::oneapi::experimental, device, architecture,
                          ext::oneapi::experimental::architecture,
                          PI_EXT_ONEAPI_DEVICE_INFO_IP_VERSION)
+__SYCL_PARAM_TRAITS_SPEC(
+    ext::oneapi::experimental, device, graph_support,
+    ext::oneapi::experimental::info::graph_support_level,
+    0 /* No PI device code needed */)
 #ifdef __SYCL_PARAM_TRAITS_TEMPLATE_SPEC_NEEDS_UNDEF
 #undef __SYCL_PARAM_TRAITS_TEMPLATE_SPEC
 #undef __SYCL_PARAM_TRAITS_TEMPLATE_SPEC_NEEDS_UNDEF

sycl/include/sycl/info/info_desc.hpp

@@ -185,9 +185,14 @@ template <typename T, T param> struct compatibility_param_traits {};
   } /*namespace info */                                                        \
   } /*namespace Namespace */
 
-namespace ext::oneapi::experimental::info::device {
+namespace ext::oneapi::experimental::info {
+
+enum class graph_support_level { unsupported = 0, native, emulated };
+
+namespace device {
 template <int Dimensions> struct max_work_groups;
-} // namespace ext::oneapi::experimental::info::device
+} // namespace device
+} // namespace ext::oneapi::experimental::info
 #include <sycl/info/ext_codeplay_device_traits.def>
 #include <sycl/info/ext_intel_device_traits.def>
 #include <sycl/info/ext_oneapi_device_traits.def>

sycl/include/sycl/queue.hpp

@@ -76,6 +76,16 @@ static event submitAssertCapture(queue &, event &, queue *,
 #endif
 } // namespace detail
 
+namespace ext {
+namespace oneapi {
+namespace experimental {
+// State of a queue with regards to graph recording,
+// returned by info::queue::state
+enum class queue_state { executing, recording };
+} // namespace experimental
+} // namespace oneapi
+} // namespace ext
+
 /// Encapsulates a single SYCL queue which schedules kernels on a SYCL device.
 ///
 /// A SYCL queue can be used to submit command groups to be executed by the SYCL
@@ -283,6 +293,9 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   /// \return SYCL device this queue was constructed with.
   device get_device() const;
 
+  /// \return State the queue is currently in.
+  ext::oneapi::experimental::queue_state ext_oneapi_get_state() const;
+
   /// \return true if this queue is a SYCL host queue.
   __SYCL2020_DEPRECATED(
       "is_host() is deprecated as the host device is no longer supported.")

sycl/plugins/level_zero/pi_level_zero.cpp

@@ -1118,46 +1118,46 @@ pi_result piextCommandBufferMemBufferCopyRect(
 
 pi_result piextCommandBufferMemBufferRead(
     pi_ext_command_buffer CommandBuffer, pi_mem Buffer, size_t Offset,
-    size_t Size, void *Dst, pi_uint32 NumEventsInWaitList,
+    size_t Size, void *Dst, pi_uint32 NumSyncPointsInWaitList,
     const pi_ext_sync_point *SyncPointWaitList, pi_ext_sync_point *SyncPoint) {
-  return pi2ur::piextCommandBufferMemBufferRead(CommandBuffer, Buffer, Offset,
-                                                Size, Dst, NumEventsInWaitList,
-                                                SyncPointWaitList, SyncPoint);
+  return pi2ur::piextCommandBufferMemBufferRead(
+      CommandBuffer, Buffer, Offset, Size, Dst, NumSyncPointsInWaitList,
+      SyncPointWaitList, SyncPoint);
 }
 
 pi_result piextCommandBufferMemBufferReadRect(
     pi_ext_command_buffer CommandBuffer, pi_mem Buffer,
     pi_buff_rect_offset BufferOffset, pi_buff_rect_offset HostOffset,
     pi_buff_rect_region Region, size_t BufferRowPitch, size_t BufferSlicePitch,
     size_t HostRowPitch, size_t HostSlicePitch, void *Ptr,
-    pi_uint32 NumEventsInWaitList, const pi_ext_sync_point *SyncPointWaitList,
-    pi_ext_sync_point *SyncPoint) {
+    pi_uint32 NumSyncPointsInWaitList,
+    const pi_ext_sync_point *SyncPointWaitList, pi_ext_sync_point *SyncPoint) {
   return pi2ur::piextCommandBufferMemBufferReadRect(
       CommandBuffer, Buffer, BufferOffset, HostOffset, Region, BufferRowPitch,
-      BufferSlicePitch, HostRowPitch, HostSlicePitch, Ptr, NumEventsInWaitList,
-      SyncPointWaitList, SyncPoint);
+      BufferSlicePitch, HostRowPitch, HostSlicePitch, Ptr,
+      NumSyncPointsInWaitList, SyncPointWaitList, SyncPoint);
 }
 
 pi_result piextCommandBufferMemBufferWrite(
     pi_ext_command_buffer CommandBuffer, pi_mem Buffer, size_t Offset,
-    size_t Size, const void *Ptr, pi_uint32 NumEventsInWaitList,
+    size_t Size, const void *Ptr, pi_uint32 NumSyncPointsInWaitList,
     const pi_ext_sync_point *SyncPointWaitList, pi_ext_sync_point *SyncPoint) {
-  return pi2ur::piextCommandBufferMemBufferWrite(CommandBuffer, Buffer, Offset,
-                                                 Size, Ptr, NumEventsInWaitList,
-                                                 SyncPointWaitList, SyncPoint);
+  return pi2ur::piextCommandBufferMemBufferWrite(
+      CommandBuffer, Buffer, Offset, Size, Ptr, NumSyncPointsInWaitList,
+      SyncPointWaitList, SyncPoint);
 }
 
 pi_result piextCommandBufferMemBufferWriteRect(
     pi_ext_command_buffer CommandBuffer, pi_mem Buffer,
     pi_buff_rect_offset BufferOffset, pi_buff_rect_offset HostOffset,
     pi_buff_rect_region Region, size_t BufferRowPitch, size_t BufferSlicePitch,
     size_t HostRowPitch, size_t HostSlicePitch, const void *Ptr,
-    pi_uint32 NumEventsInWaitList, const pi_ext_sync_point *SyncPointWaitList,
-    pi_ext_sync_point *SyncPoint) {
+    pi_uint32 NumSyncPointsInWaitList,
+    const pi_ext_sync_point *SyncPointWaitList, pi_ext_sync_point *SyncPoint) {
   return pi2ur::piextCommandBufferMemBufferWriteRect(
       CommandBuffer, Buffer, BufferOffset, HostOffset, Region, BufferRowPitch,
-      BufferSlicePitch, HostRowPitch, HostSlicePitch, Ptr, NumEventsInWaitList,
-      SyncPointWaitList, SyncPoint);
+      BufferSlicePitch, HostRowPitch, HostSlicePitch, Ptr,
+      NumSyncPointsInWaitList, SyncPointWaitList, SyncPoint);
 }
 
 pi_result piextEnqueueCommandBuffer(pi_ext_command_buffer CommandBuffer,

sycl/plugins/opencl/pi_opencl.cpp

@@ -2726,6 +2726,12 @@ pi_result piPluginInit(pi_plugin *PluginInit) {
   _PI_CL(piextCommandBufferMemBufferCopy, piextCommandBufferMemBufferCopy)
   _PI_CL(piextCommandBufferMemBufferCopyRect,
          piextCommandBufferMemBufferCopyRect)
+  _PI_CL(piextCommandBufferMemBufferRead, piextCommandBufferMemBufferRead)
+  _PI_CL(piextCommandBufferMemBufferReadRect,
+         piextCommandBufferMemBufferReadRect)
+  _PI_CL(piextCommandBufferMemBufferWrite, piextCommandBufferMemBufferWrite)
+  _PI_CL(piextCommandBufferMemBufferWriteRect,
+         piextCommandBufferMemBufferWriteRect)
   _PI_CL(piextEnqueueCommandBuffer, piextEnqueueCommandBuffer)
 
   _PI_CL(piextKernelSetArgMemObj, piextKernelSetArgMemObj)

sycl/plugins/unified_runtime/ur/adapters/level_zero/command_buffer.cpp

@@ -535,9 +535,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urCommandBufferAppendMembufferCopyExp(
     uint32_t NumSyncPointsInWaitList,
     const ur_exp_command_buffer_sync_point_t *SyncPointWaitList,
     ur_exp_command_buffer_sync_point_t *SyncPoint) {
-  (void)SrcOffset;
-  (void)DstOffset;
-
   auto SrcBuffer = ur_cast<ur_mem_handle_t>(SrcMem);
   auto DstBuffer = ur_cast<ur_mem_handle_t>(DstMem);
 
@@ -553,8 +550,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urCommandBufferAppendMembufferCopyExp(
                                  CommandBuffer->Device));
 
   return enqueueCommandBufferMemCopyHelper(
-      UR_COMMAND_MEM_BUFFER_COPY, CommandBuffer, ZeHandleDst, ZeHandleSrc, Size,
-      NumSyncPointsInWaitList, SyncPointWaitList, SyncPoint);
+      UR_COMMAND_MEM_BUFFER_COPY, CommandBuffer, ZeHandleDst + DstOffset,
+      ZeHandleSrc + SrcOffset, Size, NumSyncPointsInWaitList, SyncPointWaitList,
+      SyncPoint);
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL urCommandBufferAppendMembufferCopyRectExp(