[SYCLomatic][Graphics Interop][Driver] Added migration support for CUDA driver Graphics APIs/types used in Blender (#2610)

TejaX-Alaghari · web-flow · commit 3a7fddaa8306 · 2025-01-14T16:57:04.000+08:00
diff --git a/clang/examples/DPCT/Runtime/cuGraphicsMapResources.cu b/clang/examples/DPCT/Runtime/cuGraphicsMapResources.cu
@@ -0,0 +1,9 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(int c, CUgraphicsResource *r, CUstream s) {
+  // Start
+  cuGraphicsMapResources(c /*int*/,
+                         r /*CUgraphicsResource **/,
+                         s /*CUstream*/);
+  // End
+}
diff --git a/clang/examples/DPCT/Runtime/cuGraphicsResourceGetMappedPointer.cu b/clang/examples/DPCT/Runtime/cuGraphicsResourceGetMappedPointer.cu
@@ -0,0 +1,9 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(CUdeviceptr ptr, size_t *s, CUgraphicsResource r) {
+  // Start
+  cuGraphicsResourceGetMappedPointer(&ptr /*CUdeviceptr **/,
+                                      s /*size_t **/,
+                                      r /*CUgraphicsResource*/);
+  // End
+}
diff --git a/clang/examples/DPCT/Runtime/cuGraphicsUnmapResources.cu b/clang/examples/DPCT/Runtime/cuGraphicsUnmapResources.cu
@@ -0,0 +1,9 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(int c, CUgraphicsResource *r, CUstream s) {
+  // Start
+  cuGraphicsUnmapResources(c /*int*/,
+                           r /*CUgraphicsResource **/,
+                           s /*CUstream*/);
+  // End
+}
diff --git a/clang/examples/DPCT/Runtime/cuGraphicsUnregisterResource.cu b/clang/examples/DPCT/Runtime/cuGraphicsUnregisterResource.cu
@@ -0,0 +1,7 @@
+// Option: --use-experimental-features=bindless_images
+
+void test(CUgraphicsResource r) {
+  // Start
+  cuGraphicsUnregisterResource(r /*CUgraphicsResource*/);
+  // End
+}
diff --git a/clang/lib/DPCT/RuleInfra/APINamesTemplateType.inc b/clang/lib/DPCT/RuleInfra/APINamesTemplateType.inc
@@ -474,6 +474,7 @@ TYPE_REWRITE_ENTRY(
         WARNING_FACTORY(Diagnostics::TRY_EXPERIMENTAL_FEATURE, TYPESTR,
                         STR("--use-experimental-features=graph"))))
 
+// Graphics Interop Handle
 TYPE_REWRITE_ENTRY(
     "cudaGraphicsResource",
     TYPE_CONDITIONAL_FACTORY(
@@ -497,6 +498,20 @@ TYPE_REWRITE_ENTRY(
                 Diagnostics::TRY_EXPERIMENTAL_FEATURE, TYPESTR,
                 STR("--use-experimental-features=bindless_images")))))
 
+TYPE_REWRITE_ENTRY(
+    "CUgraphicsResource",
+    TYPE_CONDITIONAL_FACTORY(
+        [](const TypeLoc) { return DpctGlobalInfo::useSYCLCompat(); },
+        WARNING_FACTORY(Diagnostics::UNSUPPORT_SYCLCOMPAT,
+                        STR("CUgraphicsResource")),
+        TYPE_CONDITIONAL_FACTORY(
+            checkEnableBindlessImagesForType(),
+            TYPE_FACTORY(STR(MapNames::getDpctNamespace() +
+                             "experimental::external_mem_wrapper_ptr")),
+            WARNING_FACTORY(
+                Diagnostics::TRY_EXPERIMENTAL_FEATURE, TYPESTR,
+                STR("--use-experimental-features=bindless_images")))))
+
 // External Resource Handle Descs
 TYPE_REWRITE_ENTRY(
     "cudaExternalMemory_t",
diff --git a/clang/lib/DPCT/RuleInfra/TypeLocRewriters.cpp b/clang/lib/DPCT/RuleInfra/TypeLocRewriters.cpp
@@ -343,6 +343,7 @@ void initTypeLocSYCLCompatRewriterMap(
   SYCLCOMPAT_UNSUPPORT("cudaGraphNode_t")
   SYCLCOMPAT_UNSUPPORT("cudaGraphicsResource")
   SYCLCOMPAT_UNSUPPORT("cudaGraphicsResource_t")
+  SYCLCOMPAT_UNSUPPORT("CUgraphicsResource")
   SYCLCOMPAT_UNSUPPORT("cudaExternalMemory_t")
   SYCLCOMPAT_UNSUPPORT("cudaExternalMemoryHandleDesc")
   SYCLCOMPAT_UNSUPPORT("cudaExternalMemoryMipmappedArrayDesc")
diff --git a/clang/lib/DPCT/RulesLang/APINamesGraphicsInterop.inc b/clang/lib/DPCT/RulesLang/APINamesGraphicsInterop.inc
@@ -49,6 +49,17 @@ ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
         ARG("cudaGraphicsMapResources"),
         ARG("--use-experimental-features=bindless_images"))))
 
+ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
+    UseExtBindlessImages,
+    CALL_FACTORY_ENTRY("cuGraphicsMapResources",
+                       CALL(MapNames::getDpctNamespace() +
+                                "experimental::map_resources",
+                            ARG(0), ARG(1), ARG(2))),
+    UNSUPPORT_FACTORY_ENTRY(
+        "cuGraphicsMapResources", Diagnostics::TRY_EXPERIMENTAL_FEATURE,
+        ARG("cuGraphicsMapResources"),
+        ARG("--use-experimental-features=bindless_images"))))
+
 ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
     UseExtBindlessImages,
     MEMBER_CALL_FACTORY_ENTRY(
@@ -59,6 +70,16 @@ ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
         ARG("cudaGraphicsResourceGetMappedPointer"),
         ARG("--use-experimental-features=bindless_images"))))
 
+ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
+    UseExtBindlessImages,
+    MEMBER_CALL_FACTORY_ENTRY(
+        "cuGraphicsResourceGetMappedPointer_v2",
+        ARG(2), true, "get_mapped_pointer", CAST(makeLiteral("void **"), ARG(0)), ARG(1)),
+    UNSUPPORT_FACTORY_ENTRY(
+        "cuGraphicsResourceGetMappedPointer_v2", Diagnostics::TRY_EXPERIMENTAL_FEATURE,
+        ARG("cuGraphicsResourceGetMappedPointer_v2"),
+        ARG("--use-experimental-features=bindless_images"))))
+
 ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
     UseExtBindlessImages,
     ASSIGNABLE_FACTORY(ASSIGN_FACTORY_ENTRY(
@@ -96,6 +117,17 @@ ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
         ARG("cudaGraphicsUnmapResources"),
         ARG("--use-experimental-features=bindless_images"))))
 
+ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
+    UseExtBindlessImages,
+    CALL_FACTORY_ENTRY("cuGraphicsUnmapResources",
+                        CALL(MapNames::getDpctNamespace() +
+                                "experimental::unmap_resources",
+                            ARG(0), ARG(1), ARG(2))),
+    UNSUPPORT_FACTORY_ENTRY(
+        "cuGraphicsUnmapResources", Diagnostics::TRY_EXPERIMENTAL_FEATURE,
+        ARG("cuGraphicsUnmapResources"),
+        ARG("--use-experimental-features=bindless_images"))))
+
 ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
     UseExtBindlessImages,
     DELETER_FACTORY_ENTRY("cudaGraphicsUnregisterResource", ARG(0)),
@@ -104,6 +136,14 @@ ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
         ARG("cudaGraphicsUnregisterResource"),
         ARG("--use-experimental-features=bindless_images"))))
 
+ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
+    UseExtBindlessImages,
+    DELETER_FACTORY_ENTRY("cuGraphicsUnregisterResource", ARG(0)),
+    UNSUPPORT_FACTORY_ENTRY(
+        "cuGraphicsUnregisterResource", Diagnostics::TRY_EXPERIMENTAL_FEATURE,
+        ARG("cuGraphicsUnregisterResource"),
+        ARG("--use-experimental-features=bindless_images"))))
+
 // External Resource APIs
 CONDITIONAL_FACTORY_ENTRY(
     UseExtBindlessImages,
diff --git a/clang/lib/DPCT/RulesLang/RulesLang.cpp b/clang/lib/DPCT/RulesLang/RulesLang.cpp
@@ -355,8 +355,8 @@ void TypeInDeclRule::registerMatcher(MatchFinder &MF) {
                   "cooperative_groups::__v1::grid_group",
                   "cooperative_groups::__v1::thread_block_tile", "cudaGraph_t",
                   "cudaGraphExec_t", "cudaGraphNode_t", "cudaGraphicsResource",
-                  "cudaGraphicsResource_t", "cudaExternalMemory_t",
-                  "cudaExternalMemoryHandleDesc",
+                  "cudaGraphicsResource_t", "CUgraphicsResource",
+                  "cudaExternalMemory_t", "cudaExternalMemoryHandleDesc",
                   "cudaExternalMemoryMipmappedArrayDesc",
                   "cudaExternalMemoryBufferDesc"))))))
           .bind("cudaTypeDefEA"),
diff --git a/clang/lib/DPCT/RulesLang/RulesLangGraphicsInterop.cpp b/clang/lib/DPCT/RulesLang/RulesLangGraphicsInterop.cpp
@@ -50,10 +50,13 @@ void GraphicsInteropRule::registerMatcher(ast_matchers::MatchFinder &MF) {
   auto graphicsInteropAPI = [&]() {
     return hasAnyName(
         "cudaGraphicsD3D11RegisterResource", "cudaGraphicsResourceSetMapFlags",
-        "cudaGraphicsMapResources", "cudaGraphicsResourceGetMappedPointer",
+        "cudaGraphicsMapResources", "cuGraphicsMapResources",
+        "cudaGraphicsResourceGetMappedPointer",
+        "cuGraphicsResourceGetMappedPointer_v2",
         "cudaGraphicsResourceGetMappedMipmappedArray",
         "cudaGraphicsSubResourceGetMappedArray", "cudaGraphicsUnmapResources",
-        "cudaGraphicsUnregisterResource", "cudaImportExternalMemory",
+        "cuGraphicsUnmapResources", "cudaGraphicsUnregisterResource",
+        "cuGraphicsUnregisterResource", "cudaImportExternalMemory",
         "cudaExternalMemoryGetMappedMipmappedArray",
         "cudaExternalMemoryGetMappedBuffer", "cudaDestroyExternalMemory");
   };
diff --git a/clang/lib/DPCT/SrcAPI/APINames.inc b/clang/lib/DPCT/SrcAPI/APINames.inc
@@ -2008,13 +2008,13 @@ ENTRY(cuDeviceCanAccessPeer, cuDeviceCanAccessPeer, true, NO_FLAG, P4, "DPCT1031
 ENTRY(cuDeviceGetP2PAttribute, cuDeviceGetP2PAttribute, false, NO_FLAG, P4, "comment")
 
 // Graphics Interoperability
-ENTRY(cuGraphicsMapResources, cuGraphicsMapResources, false, NO_FLAG, P4, "comment")
+ENTRY(cuGraphicsMapResources, cuGraphicsMapResources, true, NO_FLAG, P4, "successful/DPCT1119")
 ENTRY(cuGraphicsResourceGetMappedMipmappedArray, cuGraphicsResourceGetMappedMipmappedArray, false, NO_FLAG, P4, "comment")
-ENTRY(cuGraphicsResourceGetMappedPointer, cuGraphicsResourceGetMappedPointer_v2, false, NO_FLAG, P4, "comment")
+ENTRY(cuGraphicsResourceGetMappedPointer, cuGraphicsResourceGetMappedPointer_v2, true, NO_FLAG, P4, "successful/DPCT1119")
 ENTRY(cuGraphicsResourceSetMapFlags, cuGraphicsResourceSetMapFlags_v2, false, NO_FLAG, P4, "comment")
 ENTRY(cuGraphicsSubResourceGetMappedArray, cuGraphicsSubResourceGetMappedArray, false, NO_FLAG, P4, "comment")
-ENTRY(cuGraphicsUnmapResources, cuGraphicsUnmapResources, false, NO_FLAG, P4, "comment")
-ENTRY(cuGraphicsUnregisterResource, cuGraphicsUnregisterResource, false, NO_FLAG, P4, "comment")
+ENTRY(cuGraphicsUnmapResources, cuGraphicsUnmapResources, true, NO_FLAG, P4, "successful/DPCT1119")
+ENTRY(cuGraphicsUnregisterResource, cuGraphicsUnregisterResource, true, NO_FLAG, P4, "successful/DPCT1119")
 
 // Driver Entry Point Access
 ENTRY(cuGetProcAddress, cuGetProcAddress, false, NO_FLAG, P7, "comment")
diff --git a/clang/lib/DPCT/SrcAPI/TypeNames.inc b/clang/lib/DPCT/SrcAPI/TypeNames.inc
@@ -38,7 +38,7 @@ ENTRY_TYPE(CUexternalSemaphore, false, NO_FLAG, P4, "comment")
 ENTRY_TYPE(CUgraph, false, NO_FLAG, P4, "comment")
 ENTRY_TYPE(CUgraphExec, false, NO_FLAG, P4, "comment")
 ENTRY_TYPE(CUgraphNode, false, NO_FLAG, P4, "comment")
-ENTRY_TYPE(CUgraphicsResource, false, NO_FLAG, P4, "comment")
+ENTRY_TYPE(CUgraphicsResource, true, NO_FLAG, P4, "successful")
 
 // CUDA Runtime Library
 ENTRY_TYPE(cudaKernelNodeParams, false, NO_FLAG, P4, "comment")
diff --git a/clang/test/dpct/cuGraphics.cu b/clang/test/dpct/cuGraphics.cu
@@ -0,0 +1,58 @@
+// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2
+// UNSUPPORTED: v8.0, v9.0, v9.1, v9.2
+// RUN: dpct --use-experimental-features=bindless_images --format-range=none -out-root %T/cuGraphics %s --cuda-include-path="%cuda-path/include" -- -x cuda --cuda-host-only --std=c++14
+// RUN: FileCheck --input-file %T/cuGraphics/cuGraphics.dp.cpp --match-full-lines %s
+// RUN: %if build_lit %{icpx -c -DNO_BUILD_TEST -fsycl %T/cuGraphics/cuGraphics.dp.cpp -o %T/cuGraphics/cuGraphics.dp.o %}
+
+#include <cuda.h>
+
+int main() {
+  // CHECK: dpct::experimental::external_mem_wrapper_ptr resource;
+  // CHECK-NEXT: dpct::experimental::external_mem_wrapper_ptr *resources;
+  // CHECK-NEXT: dpct::experimental::external_mem_wrapper_ptr **resources_ptr;
+  CUgraphicsResource resource;
+  CUgraphicsResource *resources;
+  CUgraphicsResource **resources_ptr;
+
+  // CHECK: dpct::experimental::external_mem_wrapper_ptr resources_arr[10];
+  CUgraphicsResource resources_arr[10];
+
+  // CHECK: dpct::experimental::external_mem_wrapper_ptr resource1, *resources1, **resources_ptr1;
+  CUgraphicsResource resource1, *resources1, **resources_ptr1;
+
+  resources_arr[0] = resource;
+  resources_arr[1] = resource1;
+
+  CUdeviceptr pDevPtr;
+  size_t pSize;
+
+  CUstream stream;
+  cuStreamCreate(&stream, 0);
+
+#ifdef _WIN32
+  // CHECK-WINDOWS: dpct::experimental::map_resources(2, resources_arr, stream);
+  cuGraphicsMapResources(2, resources_arr, stream);
+
+  // CHECK-WINDOWS: dpct::experimental::map_resources(1, &resource, stream);
+  cuGraphicsMapResources(1, &resource, stream);
+#endif // _WIN32
+
+  // CHECK: resource->get_mapped_pointer((void **)&pDevPtr, &pSize);
+  cuGraphicsResourceGetMappedPointer(&pDevPtr, &pSize, resource);
+
+#ifdef _WIN32
+  // CHECK-WINDOWS: dpct::experimental::unmap_resources(2, resources_arr, stream);
+  cuGraphicsUnmapResources(2, resources_arr, stream);
+
+  // CHECK-WINDOWS: dpct::experimental::unmap_resources(1, &resource, stream);
+  cuGraphicsUnmapResources(1, &resource, stream);
+#endif // _WIN32
+
+  // CHECK: delete resource;
+  cuGraphicsUnregisterResource(resource);
+
+  // CHECK: delete resource1;
+  cuGraphicsUnregisterResource(resource1);
+
+  return 0;
+}
diff --git a/clang/test/dpct/cuGraphics_default_option.cu b/clang/test/dpct/cuGraphics_default_option.cu
@@ -0,0 +1,54 @@
+// 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 -out-root %T/cuGraphics_default_option %s --cuda-include-path="%cuda-path/include" -- -x cuda --cuda-host-only --std=c++14
+// RUN: FileCheck --input-file %T/cuGraphics_default_option/cuGraphics_default_option.dp.cpp --match-full-lines %s
+// RUN: %if build_lit %{icpx -c -DNO_BUILD_TEST -fsycl %T/cuGraphics_default_option/cuGraphics_default_option.dp.cpp -o %T/cuGraphics_default_option/cudaGraphicsResource_test.dp.o %}
+
+#ifndef NO_BUILD_TEST
+#include <cuda.h>
+
+int main() {
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of CUgraphicsResource is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  CUgraphicsResource resource, *resources;
+
+  CUdeviceptr devPtr;
+  size_t size;
+
+  CUstream stream;
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cuGraphicsMapResources is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  cuGraphicsMapResources(2, resources, stream);
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cuGraphicsMapResources is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  cuGraphicsMapResources(1, &resource, stream);
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cuGraphicsResourceGetMappedPointer_v2 is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  cuGraphicsResourceGetMappedPointer(&devPtr, &size, resource);
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cuGraphicsUnmapResources is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  cuGraphicsUnmapResources(2, resources, stream);
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cuGraphicsUnmapResources is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  cuGraphicsUnmapResources(1, &resource, stream);
+
+  // CHECK: /*
+  // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cuGraphicsUnregisterResource is not supported, please try to remigrate with option: --use-experimental-features=bindless_images.
+  // CHECK-NEXT: */
+  cuGraphicsUnregisterResource(resource);
+
+  return 0;
+}
+
+#endif
diff --git a/clang/test/dpct/cudaGraphics.cu b/clang/test/dpct/cudaGraphics.cu
@@ -50,6 +50,9 @@ int main() {
   resources_arr[0] = resource;
   resources_arr[1] = resource1;
 
+  void *devPtr;
+  size_t size;
+
   // CHECK: int mapFlags = 0;
   // CHECK-NEXT: int mapFlags1 = 0;
   // CHECK-NEXT: int mapFlags2 = 0;
@@ -94,6 +97,9 @@ int main() {
   // CHECK: *array_ptr = resource->get_sub_resource_mapped_array(arrayIndex, mipLevel);
   cudaGraphicsSubResourceGetMappedArray(array_ptr, resource, arrayIndex, mipLevel);
 
+  // CHECK: resource->get_mapped_pointer(&devPtr, &size);
+  cudaGraphicsResourceGetMappedPointer(&devPtr, &size, resource);
+
 #ifdef _WIN32
   // CHECK-WINDOWS: dpct::experimental::unmap_resources(1, &resource);
   cudaGraphicsUnmapResources(1, &resource);
diff --git a/clang/test/dpct/know_unsupported_type.cu b/clang/test/dpct/know_unsupported_type.cu
@@ -46,11 +46,6 @@ int main(int argc, char **argv) {
     // CHECK-NEXT: CUgraphNode cugn;
     CUgraphNode cugn;
     // CHECK: /*
-    // CHECK-NEXT: DPCT1082:{{[0-9]+}}: Migration of CUgraphicsResource type is not supported.
-    // CHECK-NEXT: */
-    // CHECK-NEXT: CUgraphicsResource cugr;
-    CUgraphicsResource cugr;
-    // CHECK: /*
     // CHECK-NEXT: DPCT1082:{{[0-9]+}}: Migration of nvmlDevice_t type is not supported.
     // CHECK-NEXT: */
     // CHECK-NEXT: nvmlDevice_t nvmld;
diff --git a/clang/test/dpct/query_api_mapping/Driver/test_graphics_interop.cu b/clang/test/dpct/query_api_mapping/Driver/test_graphics_interop.cu
@@ -0,0 +1,32 @@
+// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2
+// UNSUPPORTED: v8.0, v9.0, v9.1, v9.2
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuGraphicsMapResources | FileCheck %s -check-prefix=CU_GRAPHICS_MAP_RESOURCES
+// CU_GRAPHICS_MAP_RESOURCES: CUDA API:
+// CU_GRAPHICS_MAP_RESOURCES-NEXT:    cuGraphicsMapResources(c /*int*/,
+// CU_GRAPHICS_MAP_RESOURCES-NEXT:                           r /*CUgraphicsResource **/,
+// CU_GRAPHICS_MAP_RESOURCES-NEXT:                           s /*CUstream*/);
+// CU_GRAPHICS_MAP_RESOURCES-NEXT: Is migrated to (with the option --use-experimental-features=bindless_images):
+// CU_GRAPHICS_MAP_RESOURCES-NEXT:    dpct::experimental::map_resources(c, r, s);
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuGraphicsResourceGetMappedPointer | FileCheck %s -check-prefix=CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER
+// CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER: CUDA API:
+// CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER-NEXT:    cuGraphicsResourceGetMappedPointer(&ptr /*CUdeviceptr **/,
+// CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER-NEXT:                                       s /*size_t **/,
+// CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER-NEXT:                                       r /*CUgraphicsResource*/);
+// CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER-NEXT: Is migrated to (with the option --use-experimental-features=bindless_images):
+// CU_GRAPHICS_RESOURCE_GET_MAPPED_POINTER-NEXT:    r->get_mapped_pointer((void **)&ptr, s);
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuGraphicsUnmapResources | FileCheck %s -check-prefix=CU_GRAPHICS_UNMAP_RESOURCES
+// CU_GRAPHICS_UNMAP_RESOURCES: CUDA API:
+// CU_GRAPHICS_UNMAP_RESOURCES-NEXT:    cuGraphicsUnmapResources(c /*int*/,
+// CU_GRAPHICS_UNMAP_RESOURCES-NEXT:                             r /*CUgraphicsResource **/,
+// CU_GRAPHICS_UNMAP_RESOURCES-NEXT:                             s /*CUstream*/);
+// CU_GRAPHICS_UNMAP_RESOURCES-NEXT: Is migrated to (with the option --use-experimental-features=bindless_images):
+// CU_GRAPHICS_UNMAP_RESOURCES-NEXT:    dpct::experimental::unmap_resources(c, r, s);
+
+// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cuGraphicsUnregisterResource | FileCheck %s -check-prefix=CU_GRAPHICS_UNREGISTER_RESOURCE
+// CU_GRAPHICS_UNREGISTER_RESOURCE: CUDA API:
+// CU_GRAPHICS_UNREGISTER_RESOURCE-NEXT:    cuGraphicsUnregisterResource(r /*CUgraphicsResource*/);
+// CU_GRAPHICS_UNREGISTER_RESOURCE-NEXT: Is migrated to (with the option --use-experimental-features=bindless_images):
+// CU_GRAPHICS_UNREGISTER_RESOURCE-NEXT:    delete r;
diff --git a/clang/test/dpct/query_api_mapping/test_all.cu b/clang/test/dpct/query_api_mapping/test_all.cu
@@ -643,6 +643,10 @@
 // CHECK-NEXT: cuFuncSetAttribute
 // CHECK-NEXT: cuFuncSetCacheConfig
 // CHECK-NEXT: cuGetErrorString
+// CHECK-NEXT: cuGraphicsMapResources
+// CHECK-NEXT: cuGraphicsResourceGetMappedPointer
+// CHECK-NEXT: cuGraphicsUnmapResources
+// CHECK-NEXT: cuGraphicsUnregisterResource
 // CHECK-NEXT: cuInit
 // CHECK-NEXT: cuLaunchKernel
 // CHECK-NEXT: cuMemAddressFree
