[SYCLomatic] Add 11 APIs for option --query-api-mapping

clang/examples/DPCT/Driver/cuArray3DGetDescriptor.cu

@@ -0,0 +1,7 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(CUDA_ARRAY3D_DESCRIPTOR *desc, CUarray array) {
+  // Start
+  cuArray3DGetDescriptor(desc /*CUDA_ARRAY3D_DESCRIPTOR **/, array /*CUarray*/);
+  // End
+}

clang/examples/DPCT/Driver/cuArrayGetDescriptor.cu

@@ -0,0 +1,7 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(CUDA_ARRAY_DESCRIPTOR *desc, CUarray array) {
+  // Start
+  cuArrayGetDescriptor(desc /*CUDA_ARRAY_DESCRIPTOR **/, array /*CUarray*/);
+  // End
+}

clang/examples/DPCT/Driver/cuCtxCreate_v3.cu

@@ -0,0 +1,6 @@
+void test(CUcontext *ctx, CUexecAffinityParam *params_array, int num,
+          unsigned int flags, CUdevice device) {
+  // Start
+  cuCtxCreate_v3(ctx /*CUcontext **/, params_array /*CUexecAffinityParam **/, num /*int*/, flags /*unsigned int*/, device /*CUdevice*/);
+  // End
+}

clang/examples/DPCT/Driver/cuCtxCreate_v4.cu

@@ -0,0 +1,6 @@
+void test(CUcontext *ctx, CUctxCreateParams *params_array, unsigned int flags,
+          CUdevice device) {
+  // Start
+  cuCtxCreate_v4(ctx /*CUcontext **/, params_array /*CUctxCreateParams **/, flags /*unsigned int*/, device /*CUdevice*/);
+  // End
+}

clang/examples/DPCT/Driver/cuGetErrorName.cu

@@ -0,0 +1,5 @@
+void test(CUresult r, const char **pstr) {
+  // Start
+  cuGetErrorName(r /*CUresult*/, pstr /*const char ***/);
+  // End
+}

clang/examples/DPCT/Driver/cuMipmappedArrayCreate.cu

@@ -0,0 +1,8 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(CUmipmappedArray *array, CUDA_ARRAY3D_DESCRIPTOR *desc,
+          unsigned int levels) {
+  // Start
+  cuMipmappedArrayCreate(array /*CUmipmappedArray **/, desc /*CUDA_ARRAY3D_DESCRIPTOR **/, levels /*unsigned int*/);
+  // End
+}

clang/examples/DPCT/Guide/surf2DLayeredwrite.cu

@@ -0,0 +1,9 @@
+// Option: --use-experimental-features=bindless_images
+template <typename T>
+__global__ void test(float data, cudaSurfaceObject_t surf, int x, int y,
+                     int layer) {
+  // Start
+  surf2DLayeredwrite<T>(data /*float*/, surf /*cudaSurfaceObject_t*/, x /*int*/,
+                        y /*int*/, layer /*int*/);
+  // End
+}

clang/examples/DPCT/Guide/surf2Dread.cu

@@ -0,0 +1,8 @@
+// Option: --use-experimental-features=bindless_images
+template <typename T>
+__global__ void test(float data, cudaSurfaceObject_t surf, int x, int y) {
+  // Start
+  data /*float*/ =
+      surf2Dread<T>(surf /*cudaSurfaceObject_t*/, x /*int*/, y /*int*/);
+  // End
+}

clang/examples/DPCT/Guide/surf2Dwrite.cu

@@ -0,0 +1,8 @@
+// Option: --use-experimental-features=bindless_images
+template <typename T>
+__global__ void test(float data, cudaSurfaceObject_t surf, int x, int y) {
+  // Start
+  surf2Dwrite<T>(data /*float*/, surf /*cudaSurfaceObject_t*/, x /*int*/,
+                 y /*int*/);
+  // End
+}

clang/examples/DPCT/Guide/surf3Dread.cu

@@ -0,0 +1,9 @@
+// Option: --use-experimental-features=bindless_images
+template <typename T>
+__global__ void test(float data, cudaSurfaceObject_t surf, int x, int y,
+                     int z) {
+  // Start
+  data /*float*/ = surf3Dread<T>(surf /*cudaSurfaceObject_t*/, x /*int*/,
+                                 y /*int*/, z /*int*/);
+  // End
+}

clang/examples/DPCT/Guide/surf3Dwrite.cu

@@ -0,0 +1,9 @@
+// Option: --use-experimental-features=bindless_images
+template <typename T>
+__global__ void test(float data, cudaSurfaceObject_t surf, int x, int y,
+                     int z) {
+  // Start
+  surf3Dwrite<T>(data /*float*/, surf /*cudaSurfaceObject_t*/, x /*int*/,
+                 y /*int*/, z /*int*/);
+  // End
+}

clang/test/dpct/query_api_mapping/Driver/test-after12.4.cu

@@ -0,0 +1,14 @@
+// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2, cuda-10.0, cuda-10.1, cuda-10.2, cuda-11.0, cuda-11.1, cuda-11.2, cuda-11.3, cuda-11.4, cuda-11.5, cuda-11.6, cuda-11.7, cuda-11.8, cuda-12.0, cuda-12.1, cuda-12.2, cuda-12.3, cuda-12.4
+// UNSUPPORTED: v8.0, v9.0, v9.1, v9.2, v10.0, v10.1, v10.2, v11.0, v11.1, v11.2, v11.3, v11.4, v11.5, v11.6, v11.7, v11.8, v12.0, v12.1, v12.2, v12.3, v12.4
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuCtxCreate_v3 | FileCheck %s -check-prefix=CUCTXCREATE_V3
+// CUCTXCREATE_V3:  CUDA API:
+// CUCTXCREATE_V3-NEXT:    cuCtxCreate_v3(ctx /*CUcontext **/, params_array /*CUexecAffinityParam **/, num /*int*/, flags /*unsigned int*/, device /*CUdevice*/);
+// CUCTXCREATE_V3-NEXT:  Is migrated to:
+// CUCTXCREATE_V3-NEXT:    *ctx = dpct::push_device_for_curr_thread(device);
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuCtxCreate_v4 | FileCheck %s -check-prefix=CUCTXCREATE_V4
+// CUCTXCREATE_V4:  CUDA API:
+// CUCTXCREATE_V4-NEXT:    cuCtxCreate_v4(ctx /*CUcontext **/, params_array /*CUctxCreateParams **/, flags /*unsigned int*/, device /*CUdevice*/);
+// CUCTXCREATE_V4-NEXT:  Is migrated to:
+// CUCTXCREATE_V4-NEXT:    *ctx = dpct::push_device_for_curr_thread(device);

clang/test/dpct/query_api_mapping/Driver/test.cu

@@ -9,6 +9,15 @@
 // CUGETERRORSTRING-NEXT:   */
 // CUGETERRORSTRING-NEXT:   *ppc = dpct::get_error_string_dummy(r);
 
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuGetErrorName | FileCheck %s -check-prefix=CUGETERRORNAME
+// CUGETERRORNAME:  CUDA API:
+// CUGETERRORNAME-NEXT:    cuGetErrorName(r /*CUresult*/, pstr /*const char ***/);
+// CUGETERRORNAME-NEXT:  Is migrated to:
+// CUGETERRORNAME-NEXT:    /*
+// CUGETERRORNAME-NEXT:    DPCT1009:0: SYCL reports errors using exceptions and does not use error codes. Please replace the "get_error_string_dummy(...)" with a real error-handling function.
+// CUGETERRORNAME-NEXT:    */
+// CUGETERRORNAME-NEXT:    *pstr = dpct::get_error_string_dummy(r);
+
 /// Initialization
 
 // RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuInit | FileCheck %s -check-prefix=CUINIT
@@ -441,6 +450,24 @@
 // CUTEXREFGETADDRESSMODE-NEXT:   dpct::image_wrapper_base_p t;
 // CUTEXREFGETADDRESSMODE-NEXT:   *pa = t->get_addressing_mode();
 
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuArray3DGetDescriptor | FileCheck %s -check-prefix=CUARRAY3DGETDESCRIPTOR
+// CUARRAY3DGETDESCRIPTOR:  CUDA API:
+// CUARRAY3DGETDESCRIPTOR-NEXT:    cuArray3DGetDescriptor(desc /*CUDA_ARRAY3D_DESCRIPTOR **/, array /*CUarray*/);
+// CUARRAY3DGETDESCRIPTOR-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// CUARRAY3DGETDESCRIPTOR-NEXT:    *desc = array->get_desc();
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuArrayGetDescriptor | FileCheck %s -check-prefix=CUARRAYGETDESCRIPTOR
+// CUARRAYGETDESCRIPTOR:  CUDA API:
+// CUARRAYGETDESCRIPTOR-NEXT:    cuArrayGetDescriptor(desc /*CUDA_ARRAY_DESCRIPTOR **/, array /*CUarray*/);
+// CUARRAYGETDESCRIPTOR-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// CUARRAYGETDESCRIPTOR-NEXT:    *desc = array->get_desc();
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuMipmappedArrayCreate | FileCheck %s -check-prefix=CUMIPMAPPEDARRAYCREATE
+// CUMIPMAPPEDARRAYCREATE:  CUDA API:
+// CUMIPMAPPEDARRAYCREATE-NEXT:    cuMipmappedArrayCreate(array /*CUmipmappedArray **/, desc /*CUDA_ARRAY3D_DESCRIPTOR **/, levels /*unsigned int*/);
+// CUMIPMAPPEDARRAYCREATE-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// CUMIPMAPPEDARRAYCREATE-NEXT:    *array = new dpct::experimental::image_mem_wrapper(*desc, levels);
+
 // RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuTexRefGetFilterMode | FileCheck %s -check-prefix=CUTEXREFGETFILTERMODE
 // CUTEXREFGETFILTERMODE: CUDA API:
 // CUTEXREFGETFILTERMODE-NEXT:   CUtexref t;

clang/test/dpct/query_api_mapping/Guide/test.cu

@@ -80,6 +80,42 @@
 // TEX2DLOD-NEXT: Is migrated to (with the option --use-experimental-features=bindless_images):
 // TEX2DLOD-NEXT:   sycl::ext::oneapi::experimental::sample_mipmap<T>(t, sycl::float2(f1, f2), f3);
 
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=surf2DLayeredwrite | FileCheck %s -check-prefix=SURF2DLAYEREDWRITE
+// SURF2DLAYEREDWRITE:  CUDA API:
+// SURF2DLAYEREDWRITE-NEXT:    surf2DLayeredwrite<T>(data /*float*/, surf /*cudaSurfaceObject_t*/, x /*int*/,
+// SURF2DLAYEREDWRITE-NEXT:                          y /*int*/, layer /*int*/);
+// SURF2DLAYEREDWRITE-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// SURF2DLAYEREDWRITE-NEXT:    sycl::ext::oneapi::experimental::write_image_array(surf, sycl::int2(x / sizeof(data), y), layer, data);
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=surf2Dread | FileCheck %s -check-prefix=SURF2DREAD
+// SURF2DREAD:  CUDA API:
+// SURF2DREAD-NEXT:    data /*float*/ =
+// SURF2DREAD-NEXT:        surf2Dread<T>(surf /*cudaSurfaceObject_t*/, x /*int*/, y /*int*/);
+// SURF2DREAD-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// SURF2DREAD-NEXT:    data /*float*/ =
+// SURF2DREAD-NEXT:        dpct::experimental::fetch_image_by_byte<T>(surf, sycl::int2(x, y));
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=surf2Dwrite | FileCheck %s -check-prefix=SURF2DWRITE
+// SURF2DWRITE:  CUDA API:
+// SURF2DWRITE-NEXT:    surf2Dwrite<T>(data /*float*/, surf /*cudaSurfaceObject_t*/, x /*int*/,
+// SURF2DWRITE-NEXT:                   y /*int*/);
+// SURF2DWRITE-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// SURF2DWRITE-NEXT:    sycl::ext::oneapi::experimental::write_image(surf, sycl::int2(x / sizeof(data), y), data);
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=surf3Dread | FileCheck %s -check-prefix=SURF3DREAD
+// SURF3DREAD:  CUDA API:
+// SURF3DREAD-NEXT:    data /*float*/ = surf3Dread<T>(surf /*cudaSurfaceObject_t*/, x /*int*/,
+// SURF3DREAD-NEXT:                                   y /*int*/, z /*int*/);
+// SURF3DREAD-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// SURF3DREAD-NEXT:    data /*float*/ = dpct::experimental::fetch_image_by_byte<T>(surf, sycl::int3(x, y, z));
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=surf3Dwrite | FileCheck %s -check-prefix=SURF3DWRITE
+// SURF3DWRITE:  CUDA API:
+// SURF3DWRITE-NEXT:    surf3Dwrite<T>(data /*float*/, surf /*cudaSurfaceObject_t*/, x /*int*/,
+// SURF3DWRITE-NEXT:                   y /*int*/, z /*int*/);
+// SURF3DWRITE-NEXT:  Is migrated to (with the option --use-experimental-features=bindless_images):
+// SURF3DWRITE-NEXT:    sycl::ext::oneapi::experimental::write_image(surf, sycl::int3(x / sizeof(data), y, z), data);
+
 // RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=tex3D | FileCheck %s -check-prefix=TEX3D
 // TEX3D: CUDA API:
 // TEX3D-NEXT:   tex3D<T>(t /*cudaTextureObject_t*/, f1 /*float*/, f2 /*float*/, f3 /*float*/);

clang/test/dpct/query_api_mapping/test_all.cu

@@ -679,8 +679,10 @@
 // CHECK-NEXT: cospi
 // CHECK-NEXT: cospif
 // CHECK-NEXT: cuArray3DCreate
+// CHECK-NEXT: cuArray3DGetDescriptor
 // CHECK-NEXT: cuArrayCreate
 // CHECK-NEXT: cuArrayDestroy
+// CHECK-NEXT: cuArrayGetDescriptor
 // CHECK-NEXT: cuCabs
 // CHECK-NEXT: cuCabsf
 // CHECK-NEXT: cuCadd
@@ -702,6 +704,8 @@
 // CHECK-NEXT: cuCsub
 // CHECK-NEXT: cuCsubf
 // CHECK-NEXT: cuCtxCreate
+// CHECK-NEXT: cuCtxCreate_v3
+// CHECK-NEXT: cuCtxCreate_v4
 // CHECK-NEXT: cuCtxDestroy
 // CHECK-NEXT: cuCtxEnablePeerAccess
 // CHECK-NEXT: cuCtxGetApiVersion
@@ -739,6 +743,7 @@
 // CHECK-NEXT: cuFuncGetAttribute
 // CHECK-NEXT: cuFuncSetAttribute
 // CHECK-NEXT: cuFuncSetCacheConfig
+// CHECK-NEXT: cuGetErrorName
 // CHECK-NEXT: cuGetErrorString
 // CHECK-NEXT: cuGraphicsMapResources
 // CHECK-NEXT: cuGraphicsResourceGetMappedPointer
@@ -805,6 +810,7 @@
 // CHECK-NEXT: cuMemsetD32Async
 // CHECK-NEXT: cuMemsetD8
 // CHECK-NEXT: cuMemsetD8Async
+// CHECK-NEXT: cuMipmappedArrayCreate
 // CHECK-NEXT: cuMipmappedArrayDestroy
 // CHECK-NEXT: cuMipmappedArrayGetLevel
 // CHECK-NEXT: cuModuleGetFunction
@@ -2433,6 +2439,11 @@
 // CHECK-NEXT: skipahead_subsequence
 // CHECK-NEXT: sqrt
 // CHECK-NEXT: sqrtf
+// CHECK-NEXT: surf2DLayeredwrite
+// CHECK-NEXT: surf2Dread
+// CHECK-NEXT: surf2Dwrite
+// CHECK-NEXT: surf3Dread
+// CHECK-NEXT: surf3Dwrite
 // CHECK-NEXT: tan
 // CHECK-NEXT: tanf
 // CHECK-NEXT: tanh