[SYCLomatic] Enable migration of cudaStreamCaptureStatus enum and SYC…

…L graphs exp option Signed-off-by: Ahmed, Daiyaan <[email protected]>
oneapi-src · Apr 23, 2024 · 4b6d51d · 4b6d51d
1 parent 7a56e37
commit 4b6d51d
Show file tree

Hide file tree

Showing 6 changed files with 65 additions and 3 deletions.
diff --git a/clang/include/clang/DPCT/DPCTOptions.inc b/clang/include/clang/DPCT/DPCTOptions.inc
@@ -427,6 +427,9 @@ DPCT_ENUM_OPTION(DPCT_OPT_TYPE(static llvm::cl::bits<ExperimentalFeatures>), Exp
         false),
              DPCT_OPT_ENUM("bindless_images", int(ExperimentalFeatures::Exp_BindlessImages),
         "Experimental extension that allows use of bindless images APIs.\n",
+        false),
+             DPCT_OPT_ENUM("graphs", int(ExperimentalFeatures::Exp_Graphs),
+        "Experimental extension that allows use of SYCL Graph APIs.\n",
         false),
              DPCT_OPT_ENUM("non-uniform-groups", int(ExperimentalFeatures::Exp_NonUniformGroups),
         "Experimental extension that allows use of non-uniform groups.\n",

diff --git a/clang/lib/DPCT/ASTTraversal.cpp b/clang/lib/DPCT/ASTTraversal.cpp
@@ -1707,8 +1707,9 @@ void TypeInDeclRule::registerMatcher(MatchFinder &MF) {
               "CUstream_st", "thrust::complex", "thrust::device_vector",
               "thrust::device_ptr", "thrust::device_reference",
               "thrust::host_vector", "cublasHandle_t", "CUevent_st", "__half",
-              "half", "__half2", "half2", "cudaMemoryAdvise", "cudaError_enum",
-              "cudaDeviceProp", "cudaPitchedPtr", "thrust::counting_iterator",
+              "half", "__half2", "half2", "cudaMemoryAdvise",
+              "cudaStreamCaptureStatus", "cudaError_enum", "cudaDeviceProp",
+              "cudaPitchedPtr", "thrust::counting_iterator",
               "thrust::transform_iterator", "thrust::permutation_iterator",
               "thrust::iterator_difference", "cusolverDnHandle_t",
               "cusolverDnParams_t", "gesvdjInfo_t", "syevjInfo_t",
@@ -3318,7 +3319,8 @@ void EnumConstantRule::registerMatcher(MatchFinder &MF) {
           to(enumConstantDecl(anyOf(
               hasType(enumDecl(hasAnyName(
                   "cudaComputeMode", "cudaMemcpyKind", "cudaMemoryAdvise",
-                  "cudaDeviceAttr", "libraryPropertyType_t", "cudaDataType_t",
+                  "cudaStreamCaptureStatus", "cudaDeviceAttr",
+                  "libraryPropertyType_t", "cudaDataType_t",
                   "cublasComputeType_t", "CUmem_advise_enum", "cufftType_t",
                   "cufftType", "cudaMemoryType", "CUctx_flags_enum"))),
               matchesName("CUDNN_.*"), matchesName("CUSOLVER_.*")))))

diff --git a/clang/lib/DPCT/AnalysisInfo.h b/clang/lib/DPCT/AnalysisInfo.h
@@ -1247,6 +1247,9 @@ class DpctGlobalInfo {
   static bool useExtBindlessImages() {
     return getUsingExperimental<ExperimentalFeatures::Exp_BindlessImages>();
   }
+  static bool useExtGraphs() {
+    return getUsingExperimental<ExperimentalFeatures::Exp_Graphs>();
+  }
   static bool useExpNonUniformGroups() {
     return getUsingExperimental<ExperimentalFeatures::Exp_NonUniformGroups>();
   }

diff --git a/clang/lib/DPCT/MapNames.cpp b/clang/lib/DPCT/MapNames.cpp
@@ -364,6 +364,11 @@ void MapNames::setExplicitNamespaceMap() {
       {"cusparseHandle_t",
        std::make_shared<TypeNameRule>(getDpctNamespace() + "queue_ptr")},
       {"cudaMemoryAdvise", std::make_shared<TypeNameRule>("int")},
+      {"cudaStreamCaptureStatus",
+       std::make_shared<TypeNameRule>(
+           DpctGlobalInfo::useExtGraphs()
+               ? getClNamespace() + "ext::oneapi::experimental::queue_state"
+               : "int")},
       {"CUmem_advise", std::make_shared<TypeNameRule>("int")},
       {"cudaPos", std::make_shared<TypeNameRule>(getClNamespace() + "id<3>")},
       {"cudaExtent",
@@ -1028,6 +1033,25 @@ void MapNames::setExplicitNamespaceMap() {
        std::make_shared<EnumNameRule>("0")},
       {"cudaMemAdviseSetAccessedBy", std::make_shared<EnumNameRule>("0")},
       {"cudaMemAdviseUnsetAccessedBy", std::make_shared<EnumNameRule>("0")},
+      // enum cudaStreamCaptureStatus
+      {"cudaStreamCaptureStatusNone",
+       std::make_shared<EnumNameRule>(
+           DpctGlobalInfo::useExtGraphs()
+               ? getClNamespace() +
+                     "ext::oneapi::experimental::queue_state::executing"
+               : "0")},
+      {"cudaStreamCaptureStatusActive",
+       std::make_shared<EnumNameRule>(
+           DpctGlobalInfo::useExtGraphs()
+               ? getClNamespace() +
+                     "ext::oneapi::experimental::queue_state::recording"
+               : "0")},
+      {"cudaStreamCaptureStatusInvalidated",
+       std::make_shared<EnumNameRule>(
+           DpctGlobalInfo::useExtGraphs()
+               ? getClNamespace() +
+                     "ext::oneapi::experimental::queue_state::executing"
+               : "0")},
       // enum CUmem_advise_enum
       {"CU_MEM_ADVISE_SET_READ_MOSTLY", std::make_shared<EnumNameRule>("0")},
       {"CU_MEM_ADVISE_UNSET_READ_MOSTLY", std::make_shared<EnumNameRule>("0")},

diff --git a/clang/lib/DPCT/ValidateArguments.h b/clang/lib/DPCT/ValidateArguments.h
@@ -90,6 +90,7 @@ enum class ExperimentalFeatures : unsigned int {
   Exp_BindlessImages,
   Exp_NonUniformGroups,
   Exp_ExperimentalFeaturesEnumSize,
+  Exp_Graphs,
   Exp_All
 };
 enum class HelperFuncPreference : unsigned int { NoQueueDevice = 0 };

diff --git a/clang/test/dpct/cuda-stream-api-cuda10-after.cu b/clang/test/dpct/cuda-stream-api-cuda10-after.cu
@@ -0,0 +1,29 @@
+// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2
+// UNSUPPORTED: v8.0, v9.0, v9.1, v9.2
+// RUN: dpct --format-range=none --use-experimental-features=graphs -out-root %T/cuda-stream-api-cuda10-after %s --cuda-include-path="%cuda-path/include" -- -x cuda --cuda-host-only --std=c++14
+// RUN: FileCheck --input-file %T/cuda-stream-api-cuda10-after/cuda-stream-api-cuda10-after.dp.cpp --match-full-lines %s
+// RUN: %if build_lit %{icpx -c -fsycl %T/cuda-stream-api-cuda10-after/cuda-stream-api-cuda10-after.dp.cpp -o %T/cuda-stream-api-cuda10-after/cuda-stream-api-cuda10-after.dp.o %}
+
+template <typename T>
+// CHECK: void my_error_checker(T ReturnValue, char const *const FuncName) {
+void my_error_checker(T ReturnValue, char const *const FuncName) {
+}
+
+#define MY_ERROR_CHECKER(CALL) my_error_checker((CALL), #CALL)
+
+int main() {
+  cudaStream_t s0;
+  // CHECK: sycl::ext::oneapi::experimental::queue_state captureStatus = sycl::ext::oneapi::experimental::queue_state::executing;
+  // CHECK-NEXT: captureStatus = sycl::ext::oneapi::experimental::queue_state::recording;
+  // CHECK-NEXT: captureStatus = sycl::ext::oneapi::experimental::queue_state::executing;
+  cudaStreamCaptureStatus captureStatus = cudaStreamCaptureStatusNone;
+  captureStatus = cudaStreamCaptureStatusActive;
+  captureStatus = cudaStreamCaptureStatusInvalidated;
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1027:{{[0-9]+}}: The call to cudaStreamIsCapturing was replaced with 0 because SYCL currently does not support capture operations on queues.
+  // CHECK-NEXT: */
+  // CHECK: MY_ERROR_CHECKER(0);
+  MY_ERROR_CHECKER(cudaStreamIsCapturing(s0, &captureStatus));
+
+  return 0;
+}