[SYCL][Graph] Implement graph creation and finalization for CUDA backend (#279)

* [SYCL][Graph] Implement graph creation and finalization for CUDA backend

Implements CommandBuffer creation, retain, release and finalize functions
for the cuda backend.
These functions rely on the cuda graph feature which is part of cuda runtime.
Consequently, this PR adds the required linking dependencies.
Fixes a commandbuffer bug in the cuda pluging initialization.
Reports cuda backend as supported for the graph extension.
Adds a tests that creates and finalizes an empty graph.

Author: mfrancepillois

sycl/plugins/cuda/pi_cuda.cpp

@@ -1204,8 +1204,7 @@ pi_result piPluginInit(pi_plugin *PluginInit) {
 #define _PI_API(api)                                                           \
   (PluginInit->PiFunctionTable).api = (decltype(&::api))(&api);
 #include <sycl/detail/pi.def>
-#undef _PI_API
-
+#undef _PI_API  
   return PI_SUCCESS;
 }
 

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

@@ -106,6 +106,30 @@ static void setCopyParams(const void *SrcPtr, const CUmemorytype_enum SrcType,
   Params.Depth = 1;
 }
 
+ur_exp_command_buffer_handle_t_::ur_exp_command_buffer_handle_t_(
+    ur_context_handle_t hContext, ur_device_handle_t hDevice)
+    : Context(hContext),
+      Device(hDevice), cudaGraph{nullptr}, cudaGraphExec{nullptr}, RefCount{1} {
+  urContextRetain(hContext);
+  urDeviceRetain(hDevice);
+}
+
+// The ur_exp_command_buffer_handle_t_ destructor release all the memory objects
+// allocated for command_buffer managment
+ur_exp_command_buffer_handle_t_::~ur_exp_command_buffer_handle_t_() {
+  // Release the memory allocated to the Context stored in the command_buffer
+  urContextRelease(Context);
+
+  // Release the device
+  urDeviceRelease(Device);
+
+  // Release the memory allocated to the CudaGraph
+  cuGraphDestroy(cudaGraph);
+
+  // Release the memory allocated to the CudaGraphExec
+  cuGraphExecDestroy(cudaGraphExec);
+}
+
 UR_APIEXPORT ur_result_t UR_APICALL urCommandBufferCreateExp(
     ur_context_handle_t hContext, ur_device_handle_t hDevice,
     const ur_exp_command_buffer_desc_t *hCommandBufferDesc,

sycl/plugins/unified_runtime/ur/adapters/cuda/command_buffer.hpp

@@ -7,6 +7,7 @@
 //===----------------------------------------------------------------------===//
 
 #include <ur/ur.hpp>
+#include <ur_api.h>
 
 #include "context.hpp"
 #include <cuda.h>

sycl/plugins/unified_runtime/ur/adapters/cuda/device.cpp

@@ -610,6 +610,8 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo(ur_device_handle_t hDevice,
 
     std::string SupportedExtensions = "cl_khr_fp64 cl_khr_subgroups ";
     SupportedExtensions += "pi_ext_intel_devicelib_assert ";
+    // Return supported for the UR command-buffer experimental feature
+    SupportedExtensions += ("ur_exp_command_buffer ");
     SupportedExtensions += " ";
 
     int Major = 0;

sycl/test-e2e/Graph/cuda_backend_create_and_finalize_empty_graph.cpp

@@ -0,0 +1,35 @@
+// REQUIRES: cuda, gpu
+// RUN: %{build} -o %t.out
+// RUN: %{run} %t.out
+
+// Tests the ability to finalize a empty command graph
+// without submitting the graph.
+
+#include "graph_common.hpp"
+
+int GetCudaBackend(const sycl::device &Dev) {
+  // Return 1 if the device backend is "cuda" or 0 else.
+  // 0 does not prevent another device to be picked as a second choice
+  return Dev.get_backend() == backend::ext_oneapi_cuda;
+}
+
+int main() {
+  sycl::device CudaDev{GetCudaBackend};
+  queue Queue{CudaDev};
+
+  // Skip the test if no cuda backend found
+  if (CudaDev.get_backend() != backend::ext_oneapi_cuda)
+    return 0;
+
+  std::error_code ErrorCode = make_error_code(sycl::errc::success);
+  // This should not throw an exception
+  try {
+    exp_ext::command_graph Graph{Queue.get_context(), CudaDev};
+    auto GraphExec = Graph.finalize();
+  } catch (const sycl::exception &e) {
+    ErrorCode = e.code();
+  }
+  assert(ErrorCode == sycl::errc::success);
+
+  return 0;
+}

sycl/test-e2e/Graph/device_query.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: level_zero, gpu
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out