[SYCL][CUDA] Add the Use Default Stream property

Adding a queue property that uses the default CUDA stream
instead of creating a new CUDA stream.
This is used for interop with CUDA runtime applications that
use the default stream.

Signed-off-by: Ruyman Reyes <[email protected]>

Author: Ruyman Reyes

sycl/include/CL/sycl/detail/properties_traits.def

@@ -9,4 +9,5 @@ __SYCL_PARAM_TRAITS_SPEC(sycl::property::noinit)
 __SYCL_PARAM_TRAITS_SPEC(sycl::property::no_init)
 __SYCL_PARAM_TRAITS_SPEC(sycl::property::context::cuda::use_primary_context)
 __SYCL_PARAM_TRAITS_SPEC(sycl::property::queue::in_order)
+__SYCL_PARAM_TRAITS_SPEC(sycl::property::queue::cuda::use_default_stream)
 __SYCL_PARAM_TRAITS_SPEC(sycl::property::reduction::initialize_to_identity)

sycl/include/CL/sycl/detail/property_helper.hpp

@@ -32,8 +32,9 @@ enum DataLessPropKind {
   BufferUsePinnedHostMemory = 5,
   UsePrimaryContext = 6,
   InitializeToIdentity = 7,
+  UseDefaultStream = 8,
   // Indicates the last known dataless property.
-  LastKnownDataLessPropKind = 7,
+  LastKnownDataLessPropKind = 8,
   // Exceeding 32 may cause ABI breaking change on some of OSes.
   DataLessPropKindSize = 32
 };

sycl/include/CL/sycl/properties/queue_properties.hpp

@@ -17,6 +17,10 @@ namespace queue {
 class in_order : public detail::DataLessProperty<detail::InOrder> {};
 class enable_profiling
     : public detail::DataLessProperty<detail::QueueEnableProfiling> {};
+namespace cuda {
+class use_default_stream
+    : public detail::DataLessProperty<detail::UseDefaultStream> {};
+} // namespace cuda
 } // namespace queue
 } // namespace property
 } // namespace sycl

sycl/source/detail/queue_impl.hpp

@@ -9,6 +9,7 @@
 #pragma once
 
 #include <CL/sycl/context.hpp>
+#include <CL/sycl/detail/cuda_definitions.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/event.hpp>
 #include <CL/sycl/exception.hpp>
@@ -248,6 +249,9 @@ class queue_impl {
     if (MPropList.has_property<property::queue::enable_profiling>()) {
       CreationFlags |= PI_QUEUE_PROFILING_ENABLE;
     }
+    if (MPropList.has_property<property::queue::cuda::use_default_stream>()) {
+      CreationFlags |= __SYCL_PI_CUDA_USE_DEFAULT_STREAM;
+    }
     RT::PiQueue Queue{};
     RT::PiContext Context = MContext->getHandleRef();
     RT::PiDevice Device = MDevice->getHandleRef();

sycl/unittests/pi/cuda/test_queue.cpp

@@ -10,16 +10,18 @@
 
 #include <cuda.h>
 
+#include "TestGetPlatforms.hpp"
 #include "TestGetPlugin.hpp"
 #include <CL/sycl.hpp>
+#include <CL/sycl/backend/cuda.hpp>
 #include <CL/sycl/detail/cuda_definitions.hpp>
 #include <CL/sycl/detail/pi.hpp>
 #include <detail/plugin.hpp>
 #include <pi_cuda.hpp>
 
-using namespace cl::sycl;
+using namespace sycl;
 
-struct CudaTestQueue : public ::testing::Test {
+struct CudaTestQueue : public ::testing::TestWithParam<platform> {
 
 protected:
   detail::plugin plugin = pi::initializeAndGet(backend::cuda);
@@ -149,3 +151,15 @@ TEST_F(CudaTestQueue, PICreateQueueInterop) {
   ASSERT_EQ((plugin.call_nocheck<detail::PiApiKind::piQueueRelease>(queue)),
             PI_SUCCESS);
 }
+
+TEST_P(CudaTestQueue, SYCLQueueDefaultStream) {
+  std::vector<device> CudaDevices = GetParam().get_devices();
+  auto deviceA_ = CudaDevices[0];
+  queue Queue(deviceA_, async_handler{},
+              {property::queue::cuda::use_default_stream{}});
+
+  CUstream CudaStream = get_native<backend::cuda>(Queue);
+  unsigned int flags;
+  cuStreamGetFlags(CudaStream, &flags);
+  ASSERT_EQ(flags, CU_STREAM_DEFAULT);
+}