[SYCLomatic] Enable migration of cudaStreamCaptureStatus enum and SYCL graphs exp option

clang/include/clang/DPCT/DPCTOptions.inc

@@ -758,6 +758,10 @@ DPCT_ENUM_OPTION(
             "bindless_images", int(ExperimentalFeatures::Exp_BindlessImages),
             "Experimental extension that allows use of bindless images APIs.\n",
             false),
+        DPCT_OPTION_ENUM_VALUE(
+            "graphs", int(ExperimentalFeatures::Exp_Graphs),
+            "Experimental extension that allows use of SYCL Graph APIs.\n",
+            false),
         DPCT_OPTION_ENUM_VALUE(
             "non-uniform-groups",
             int(ExperimentalFeatures::Exp_NonUniformGroups),

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",
@@ -3329,7 +3330,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_.*")))))
@@ -3398,6 +3400,10 @@ void EnumConstantRule::runRule(const MatchFinder::MatchResult &Result) {
       EnumName == "cudaComputeModeExclusiveProcess") {
     handleComputeMode(EnumName, E);
     return;
+  } else if (EnumName == "cudaStreamCaptureStatusInvalidated") {
+    report(E->getBeginLoc(), Diagnostics::API_NOT_MIGRATED, false,
+           "cudaStreamCaptureStatusInvalidated");
+    return;
   } else if (auto ET = dyn_cast<EnumType>(E->getType())) {
     if (auto ETD = ET->getDecl()) {
       auto EnumTypeName = ETD->getName().str();

clang/lib/DPCT/AnalysisInfo.h

@@ -1274,6 +1274,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>();
   }

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",
@@ -1034,6 +1039,23 @@ 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()
+                                          ? "cudaStreamCaptureStatusInvalidated"
+                                          : "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")},

clang/lib/DPCT/ValidateArguments.h

@@ -91,6 +91,7 @@ enum class ExperimentalFeatures : unsigned int {
   Exp_NonUniformGroups,
   Exp_DeviceGlobal,
   Exp_ExperimentalFeaturesEnumSize,
+  Exp_Graphs,
   Exp_All
 };
 enum class HelperFuncPreference : unsigned int { NoQueueDevice = 0 };

clang/test/dpct/cuda-stream-api-cuda10-after.cu

@@ -0,0 +1,44 @@
+// 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 -DBUILD_TEST %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 %}
+
+#ifndef BUILD_TEST
+
+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: /*
+  // CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of cudaStreamCaptureStatusInvalidated is not supported.
+  // CHECK-NEXT: */
+  // CHECK-NEXT: captureStatus = cudaStreamCaptureStatusInvalidated;
+  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));
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of cudaStreamCaptureStatusInvalidated is not supported.
+  // CHECK-NEXT: */
+  // CHECK-NEXT: if (captureStatus == cudaStreamCaptureStatusInvalidated) {
+  if (captureStatus == cudaStreamCaptureStatusInvalidated) {
+    return -1;
+  }
+
+  return 0;
+}
+#endif