[SYCL][CUDA] Fix get_native interop for device (#6649)

This patch fixes: #6635

In #6483, the implementation of `get_native` for device for the CUDA plugin was mistakenly moved to the experimental interface header, and so it was no longer available for the regular interface, causing build issues.

For the CUDA plugin there is currently two interfaces for the CUDA interop, the "legacy" one which is used by projects such as oneMKL and oneDNN, and the "experimental" one, defined in the `sycl/ext/oneapi/experimental/backend/cuda.hpp` header  which implements the interop as described in the CUDA backend specification proposed here: KhronosGroup/SYCL-Docs#197

Author: npmiller

sycl/include/sycl/backend.hpp

@@ -101,6 +101,7 @@ struct BufferInterop<backend::opencl, DataT, Dimensions, AllocatorT> {
   }
 };
 
+#if SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO
 template <backend BackendName, typename DataT, int Dimensions,
           typename AllocatorT>
 auto get_native_buffer(const buffer<DataT, Dimensions, AllocatorT, void> &Obj)
@@ -115,6 +116,7 @@ auto get_native_buffer(const buffer<DataT, Dimensions, AllocatorT, void> &Obj)
         PI_ERROR_INVALID_OPERATION);
   return Obj.template getNative<BackendName>();
 }
+#endif
 } // namespace detail
 
 template <backend BackendName, class SyclObjectT>
@@ -147,6 +149,7 @@ auto get_native(const buffer<DataT, Dimensions, AllocatorT> &Obj)
   return detail::get_native_buffer<BackendName>(Obj);
 }
 
+#if SYCL_BACKEND_OPENCL
 template <>
 inline backend_return_t<backend::opencl, event>
 get_native<backend::opencl, event>(const event &Obj) {
@@ -164,6 +167,23 @@ get_native<backend::opencl, event>(const event &Obj) {
   }
   return ReturnValue;
 }
+#endif
+
+#if SYCL_EXT_ONEAPI_BACKEND_CUDA
+template <>
+inline backend_return_t<backend::ext_oneapi_cuda, device>
+get_native<backend::ext_oneapi_cuda, device>(const device &Obj) {
+  // TODO use SYCL 2020 exception when implemented
+  if (Obj.get_backend() != backend::ext_oneapi_cuda) {
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_ERROR_INVALID_OPERATION);
+  }
+  // CUDA uses a 32-bit int instead of an opaque pointer like other backends,
+  // so we need a specialization with static_cast instead of reinterpret_cast.
+  return static_cast<backend_return_t<backend::ext_oneapi_cuda, device>>(
+      Obj.getNative());
+}
+#endif
 
 // Native handle of an accessor should be accessed through interop_handler
 template <backend BackendName, typename DataT, int Dimensions,

sycl/include/sycl/ext/oneapi/experimental/backend/cuda.hpp

@@ -73,20 +73,6 @@ inline device make_device<backend::ext_oneapi_cuda>(
   return ext::oneapi::cuda::make_device(NativeHandle);
 }
 
-template <>
-backend_return_t<backend::ext_oneapi_cuda, device>
-get_native<backend::ext_oneapi_cuda, device>(const device &Obj) {
-  // TODO use SYCL 2020 exception when implemented
-  if (Obj.get_backend() != backend::ext_oneapi_cuda) {
-    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
-                              PI_ERROR_INVALID_OPERATION);
-  }
-  // CUDA uses a 32-bit int instead of an opaque pointer like other backends,
-  // so we need a specialization with static_cast instead of reinterpret_cast.
-  return static_cast<backend_return_t<backend::ext_oneapi_cuda, device>>(
-      Obj.getNative());
-}
-
 // CUDA event specialization
 template <>
 inline event make_event<backend::ext_oneapi_cuda>(

sycl/test/basic_tests/interop-cuda-experimental.cpp renamed to sycl/test/basic_tests/interop-cuda.cpp

@@ -1,12 +1,18 @@
 // REQUIRES: cuda
 // RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out
 // RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note -D__SYCL_INTERNAL_API %s -o %t.out
+//
+/// Also test the experimental CUDA interop interface
+// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note -DSYCL_EXT_ONEAPI_BACKEND_CUDA_EXPERIMENTAL %s -o %t.out
+// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note -D__SYCL_INTERNAL_API -DSYCL_EXT_ONEAPI_BACKEND_CUDA_EXPERIMENTAL %s -o %t.out
 // expected-no-diagnostics
 
-// Test for experimental CUDA interop API
+// Test for legacy and experimental CUDA interop API
 
-#define SYCL_EXT_ONEAPI_BACKEND_CUDA_EXPERIMENTAL 1
+#ifdef SYCL_EXT_ONEAPI_BACKEND_CUDA_EXPERIMENTAL
 #include <sycl/ext/oneapi/experimental/backend/cuda.hpp>
+#endif
+
 #include <sycl/sycl.hpp>
 
 using namespace sycl;
@@ -73,6 +79,7 @@ int main() {
   // behavior of these template functions is defined by the SYCL backend
   // specification document.
 
+#ifdef SYCL_EXT_ONEAPI_BACKEND_CUDA_EXPERIMENTAL
   backend_input_t<backend::ext_oneapi_cuda, device> InteropDeviceInput{
       cu_device};
   device InteropDevice =
@@ -85,6 +92,7 @@ int main() {
   event InteropEvent = make_event<backend::ext_oneapi_cuda>(cu_event, Context);
 
   queue InteropQueue = make_queue<backend::ext_oneapi_cuda>(cu_queue, Context);
+#endif
 
   return 0;
 }