Merge pull request #1242 from emankov/HIPIFY

[HIPIFY] Sync with CUDA 12.3.0 - Step 6 - Runtime API functions
ROCm · Dec 29, 2023 · e4ace40 · e4ace40
2 parents 9679b39 + 6c7e320
commit e4ace40
Show file tree

Hide file tree

Showing 6 changed files with 81 additions and 11 deletions.
diff --git a/bin/hipify-perl b/bin/hipify-perl
@@ -3268,6 +3268,7 @@ sub simpleSubstitutions {
     subst("cudaStreamIsCapturing", "hipStreamIsCapturing", "stream");
     subst("cudaStreamQuery", "hipStreamQuery", "stream");
     subst("cudaStreamSynchronize", "hipStreamSynchronize", "stream");
+    subst("cudaStreamUpdateCaptureDependencies", "hipStreamUpdateCaptureDependencies", "stream");
     subst("cudaStreamWaitEvent", "hipStreamWaitEvent", "stream");
     subst("cudaThreadExchangeStreamCaptureMode", "hipThreadExchangeStreamCaptureMode", "stream");
     subst("cuEventCreate", "hipEventCreateWithFlags", "event");
@@ -8352,11 +8353,14 @@ sub warnUnsupportedFunctions {
         "cudaSyncPolicyBlockingSync",
         "cudaSyncPolicyAuto",
         "cudaSurfaceFormatMode",
+        "cudaStreamUpdateCaptureDependencies_v2",
         "cudaStreamSetAttribute",
         "cudaStreamLegacy",
         "cudaStreamGetId",
+        "cudaStreamGetCaptureInfo_v3",
         "cudaStreamGetAttribute",
         "cudaStreamCopyAttributes",
+        "cudaStreamBeginCaptureToGraph",
         "cudaStreamAttributeSynchronizationPolicy",
         "cudaStreamAttributePriority",
         "cudaStreamAttributeMemSyncDomainMap",
@@ -8463,8 +8467,11 @@ sub warnUnsupportedFunctions {
         "cudaGraphicsCubeFaceNegativeY",
         "cudaGraphicsCubeFaceNegativeX",
         "cudaGraphicsCubeFace",
+        "cudaGraphRemoveDependencies_v2",
         "cudaGraphNodeSetParams",
         "cudaGraphNodeParams",
+        "cudaGraphNodeGetDependentNodes_v2",
+        "cudaGraphNodeGetDependencies_v2",
         "cudaGraphKernelNodePortProgrammatic",
         "cudaGraphKernelNodePortLaunchCompletion",
         "cudaGraphKernelNodePortDefault",
@@ -8477,6 +8484,7 @@ sub warnUnsupportedFunctions {
         "cudaGraphInstantiateMultipleDevicesNotSupported",
         "cudaGraphInstantiateInvalidStructure",
         "cudaGraphInstantiateError",
+        "cudaGraphGetEdges_v2",
         "cudaGraphExecUpdateResultInfo_st",
         "cudaGraphExecUpdateResultInfo",
         "cudaGraphExecUpdateErrorAttributesChanged",
@@ -8491,11 +8499,14 @@ sub warnUnsupportedFunctions {
         "cudaGraphDebugDotFlagsConditionalNodeParams",
         "cudaGraphConditionalNodeType",
         "cudaGraphConditionalHandleFlags",
+        "cudaGraphConditionalHandleCreate",
         "cudaGraphConditionalHandle",
         "cudaGraphCondTypeWhile",
         "cudaGraphCondTypeIf",
         "cudaGraphCondAssignDefault",
+        "cudaGraphAddNode_v2",
         "cudaGraphAddNode",
+        "cudaGraphAddDependencies_v2",
         "cudaGetTextureObjectTextureDesc_v2",
         "cudaGetSurfaceReference",
         "cudaGetSurfaceObjectResourceDesc",
@@ -8521,6 +8532,7 @@ sub warnUnsupportedFunctions {
         "cudaGLMapFlags",
         "cudaGLMapBufferObjectAsync",
         "cudaGLMapBufferObject",
+        "cudaFuncGetName",
         "cudaFuncAttributeRequiredClusterWidth",
         "cudaFuncAttributeRequiredClusterHeight",
         "cudaFuncAttributeRequiredClusterDepth",
@@ -8967,6 +8979,7 @@ sub warnUnsupportedFunctions {
         "cuGraphicsD3D11RegisterResource",
         "cuGraphicsD3D10RegisterResource",
         "cuGraphRemoveDependencies_v2",
+        "cuGraphNodeGetDependentNodes_v2",
         "cuGraphNodeGetDependencies_v2",
         "cuGraphGetEdges_v2",
         "cuGraphExecMemsetNodeSetParams",

diff --git a/docs/tables/CUDA_Driver_API_functions_supported_by_HIP.md b/docs/tables/CUDA_Driver_API_functions_supported_by_HIP.md
@@ -1872,6 +1872,7 @@
 |`cuGraphNodeGetDependencies`|10.0| | | |`hipGraphNodeGetDependencies`|5.0.0| | | | |
 |`cuGraphNodeGetDependencies_v2`|12.3| | | | | | | | | |
 |`cuGraphNodeGetDependentNodes`|10.0| | | |`hipGraphNodeGetDependentNodes`|5.0.0| | | | |
+|`cuGraphNodeGetDependentNodes_v2`|12.3| | | | | | | | | |
 |`cuGraphNodeGetEnabled`|11.6| | | |`hipGraphNodeGetEnabled`|5.5.0| | | | |
 |`cuGraphNodeGetType`|10.0| | | |`hipGraphNodeGetType`|5.0.0| | | | |
 |`cuGraphNodeSetEnabled`|11.6| | | |`hipGraphNodeSetEnabled`|5.5.0| | | | |

diff --git a/docs/tables/CUDA_Runtime_API_functions_supported_by_HIP.md b/docs/tables/CUDA_Runtime_API_functions_supported_by_HIP.md
@@ -66,6 +66,7 @@
 |`cudaStreamAddCallback`| | | | |`hipStreamAddCallback`|1.6.0| | | | |
 |`cudaStreamAttachMemAsync`| | | | |`hipStreamAttachMemAsync`|3.7.0| | | | |
 |`cudaStreamBeginCapture`|10.0| | | |`hipStreamBeginCapture`|4.3.0| | | | |
+|`cudaStreamBeginCaptureToGraph`|12.3| | | | | | | | | |
 |`cudaStreamCopyAttributes`|11.0| | | | | | | | | |
 |`cudaStreamCreate`| | | | |`hipStreamCreate`|1.6.0| | | | |
 |`cudaStreamCreateWithFlags`| | | | |`hipStreamCreateWithFlags`|1.6.0| | | | |
@@ -74,13 +75,16 @@
 |`cudaStreamEndCapture`|10.0| | | |`hipStreamEndCapture`|4.3.0| | | | |
 |`cudaStreamGetAttribute`|11.0| | | | | | | | | |
 |`cudaStreamGetCaptureInfo`|10.1| | | |`hipStreamGetCaptureInfo`|5.0.0| | | | |
+|`cudaStreamGetCaptureInfo_v3`|12.3| | | | | | | | | |
 |`cudaStreamGetFlags`| | | | |`hipStreamGetFlags`|1.6.0| | | | |
 |`cudaStreamGetId`|12.0| | | | | | | | | |
 |`cudaStreamGetPriority`| | | | |`hipStreamGetPriority`|2.0.0| | | | |
 |`cudaStreamIsCapturing`|10.0| | | |`hipStreamIsCapturing`|5.0.0| | | | |
 |`cudaStreamQuery`| | | | |`hipStreamQuery`|1.6.0| | | | |
 |`cudaStreamSetAttribute`|11.0| | | | | | | | | |
 |`cudaStreamSynchronize`| | | | |`hipStreamSynchronize`|1.6.0| | | | |
+|`cudaStreamUpdateCaptureDependencies`|11.3| | | |`hipStreamUpdateCaptureDependencies`|5.0.0| | | | |
+|`cudaStreamUpdateCaptureDependencies_v2`|12.3| | | | | | | | | |
 |`cudaStreamWaitEvent`| | | | |`hipStreamWaitEvent`|1.6.0| | | | |
 |`cudaThreadExchangeStreamCaptureMode`|10.1| | | |`hipThreadExchangeStreamCaptureMode`|5.2.0| | | | |
 
@@ -115,6 +119,7 @@
 |**CUDA**|**A**|**D**|**C**|**R**|**HIP**|**A**|**D**|**C**|**R**|**E**|
 |:--|:-:|:-:|:-:|:-:|:--|:-:|:-:|:-:|:-:|:-:|
 |`cudaFuncGetAttributes`| | | | |`hipFuncGetAttributes`|1.9.0| | | | |
+|`cudaFuncGetName`|12.3| | | | | | | | | |
 |`cudaFuncSetAttribute`|9.0| | | |`hipFuncSetAttribute`|3.9.0| | | | |
 |`cudaFuncSetCacheConfig`| | | | |`hipFuncSetCacheConfig`|1.6.0| | | | |
 |`cudaFuncSetSharedMemConfig`| | | | |`hipFuncSetSharedMemConfig`|3.9.0| | | | |
@@ -412,6 +417,7 @@
 |`cudaDeviceSetGraphMemAttribute`|11.4| | | |`hipDeviceSetGraphMemAttribute`|5.3.0| | | | |
 |`cudaGraphAddChildGraphNode`|10.0| | | |`hipGraphAddChildGraphNode`|5.0.0| | | | |
 |`cudaGraphAddDependencies`|10.0| | | |`hipGraphAddDependencies`|4.5.0| | | | |
+|`cudaGraphAddDependencies_v2`|12.3| | | | | | | | | |
 |`cudaGraphAddEmptyNode`|10.0| | | |`hipGraphAddEmptyNode`|4.5.0| | | | |
 |`cudaGraphAddEventRecordNode`|11.1| | | |`hipGraphAddEventRecordNode`|5.0.0| | | | |
 |`cudaGraphAddEventWaitNode`|11.1| | | |`hipGraphAddEventWaitNode`|5.0.0| | | | |
@@ -427,8 +433,10 @@
 |`cudaGraphAddMemcpyNodeToSymbol`|11.1| | | |`hipGraphAddMemcpyNodeToSymbol`|5.0.0| | | | |
 |`cudaGraphAddMemsetNode`|10.0| | | |`hipGraphAddMemsetNode`|4.3.0| | | | |
 |`cudaGraphAddNode`|12.2| | | | | | | | | |
+|`cudaGraphAddNode_v2`|12.3| | | | | | | | | |
 |`cudaGraphChildGraphNodeGetGraph`|10.0| | | |`hipGraphChildGraphNodeGetGraph`|5.0.0| | | | |
 |`cudaGraphClone`|10.0| | | |`hipGraphClone`|5.0.0| | | | |
+|`cudaGraphConditionalHandleCreate`|12.3| | | | | | | | | |
 |`cudaGraphCreate`|10.0| | | |`hipGraphCreate`|4.3.0| | | | |
 |`cudaGraphDebugDotPrint`|11.3| | | |`hipGraphDebugDotPrint`|5.5.0| | | | |
 |`cudaGraphDestroy`|10.0| | | |`hipGraphDestroy`|4.3.0| | | | |
@@ -458,6 +466,7 @@
 |`cudaGraphExternalSemaphoresWaitNodeGetParams`|11.2| | | |`hipGraphExternalSemaphoresWaitNodeGetParams`|6.0.0| | | | |
 |`cudaGraphExternalSemaphoresWaitNodeSetParams`|11.2| | | |`hipGraphExternalSemaphoresWaitNodeSetParams`|6.0.0| | | | |
 |`cudaGraphGetEdges`|10.0| | | |`hipGraphGetEdges`|5.0.0| | | | |
+|`cudaGraphGetEdges_v2`|12.3| | | | | | | | | |
 |`cudaGraphGetNodes`|10.0| | | |`hipGraphGetNodes`|4.5.0| | | | |
 |`cudaGraphGetRootNodes`|10.0| | | |`hipGraphGetRootNodes`|4.5.0| | | | |
 |`cudaGraphHostNodeGetParams`|10.0| | | |`hipGraphHostNodeGetParams`|5.0.0| | | | |
@@ -482,13 +491,16 @@
 |`cudaGraphMemsetNodeSetParams`|11.0| | | |`hipGraphMemsetNodeSetParams`|4.5.0| | | | |
 |`cudaGraphNodeFindInClone`|11.0| | | |`hipGraphNodeFindInClone`|5.0.0| | | | |
 |`cudaGraphNodeGetDependencies`|11.0| | | |`hipGraphNodeGetDependencies`|5.0.0| | | | |
+|`cudaGraphNodeGetDependencies_v2`|12.3| | | | | | | | | |
 |`cudaGraphNodeGetDependentNodes`|11.0| | | |`hipGraphNodeGetDependentNodes`|5.0.0| | | | |
+|`cudaGraphNodeGetDependentNodes_v2`|12.3| | | | | | | | | |
 |`cudaGraphNodeGetEnabled`|11.6| | | |`hipGraphNodeGetEnabled`|5.5.0| | | | |
 |`cudaGraphNodeGetType`|11.0| | | |`hipGraphNodeGetType`|5.0.0| | | | |
 |`cudaGraphNodeSetEnabled`|11.6| | | |`hipGraphNodeSetEnabled`|5.5.0| | | | |
 |`cudaGraphNodeSetParams`|12.2| | | | | | | | | |
 |`cudaGraphReleaseUserObject`|11.3| | | |`hipGraphReleaseUserObject`|5.3.0| | | | |
 |`cudaGraphRemoveDependencies`|11.0| | | |`hipGraphRemoveDependencies`|5.0.0| | | | |
+|`cudaGraphRemoveDependencies_v2`|12.3| | | | | | | | | |
 |`cudaGraphRetainUserObject`|11.3| | | |`hipGraphRetainUserObject`|5.3.0| | | | |
 |`cudaGraphUpload`|11.1| | | |`hipGraphUpload`|5.3.0| | | | |
 |`cudaUserObjectCreate`|11.3| | | |`hipUserObjectCreate`|5.3.0| | | | |

diff --git a/src/CUDA2HIP_Driver_API_functions.cpp b/src/CUDA2HIP_Driver_API_functions.cpp
@@ -459,7 +459,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   {"cuStreamBeginCapture",                                 {"hipStreamBeginCapture",                                   "", CONV_STREAM, API_DRIVER, SEC::STREAM}},
   {"cuStreamBeginCapture_v2",                              {"hipStreamBeginCapture",                                   "", CONV_STREAM, API_DRIVER, SEC::STREAM}},
   {"cuStreamBeginCapture_ptsz",                            {"hipStreamBeginCapture_ptsz",                              "", CONV_STREAM, API_DRIVER, SEC::STREAM, HIP_UNSUPPORTED}},
-  //
+  // cudaStreamBeginCaptureToGraph
   {"cuStreamBeginCaptureToGraph",                          {"hipStreamBeginCaptureToGraph",                            "", CONV_STREAM, API_DRIVER, SEC::STREAM, HIP_UNSUPPORTED}},
   // cudaStreamCopyAttributes
   {"cuStreamCopyAttributes",                               {"hipStreamCopyAttributes",                                 "", CONV_STREAM, API_DRIVER, SEC::STREAM, HIP_UNSUPPORTED}},
@@ -477,11 +477,11 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   // cudaStreamGetCaptureInfo
   {"cuStreamGetCaptureInfo",                               {"hipStreamGetCaptureInfo",                                 "", CONV_STREAM, API_DRIVER, SEC::STREAM}},
   {"cuStreamGetCaptureInfo_v2",                            {"hipStreamGetCaptureInfo_v2",                              "", CONV_STREAM, API_DRIVER, SEC::STREAM}},
-  //
+  // cudaStreamGetCaptureInfo_v3
   {"cuStreamGetCaptureInfo_v3",                            {"hipStreamGetCaptureInfo_v3",                              "", CONV_STREAM, API_DRIVER, SEC::STREAM, HIP_UNSUPPORTED}},
-  //
+  // cudaStreamUpdateCaptureDependencies
   {"cuStreamUpdateCaptureDependencies",                    {"hipStreamUpdateCaptureDependencies",                      "", CONV_STREAM, API_DRIVER, SEC::STREAM}},
-  //
+  // cudaStreamUpdateCaptureDependencies_v2
   {"cuStreamUpdateCaptureDependencies_v2",                 {"hipStreamUpdateCaptureDependencies_v2",                   "", CONV_STREAM, API_DRIVER, SEC::STREAM, HIP_UNSUPPORTED}},
   // no analogue
   {"cuStreamGetCtx",                                       {"hipStreamGetContext",                                     "", CONV_STREAM, API_DRIVER, SEC::STREAM, HIP_UNSUPPORTED}},
@@ -596,7 +596,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   // no analogue
   // NOTE: Not equal to cudaLaunchKernelExC due to different signatures
   {"cuLaunchKernelEx",                                     {"hipLaunchKernelEx",                                       "", CONV_EXECUTION, API_DRIVER, SEC::EXECUTION, HIP_UNSUPPORTED}},
-  //
+  // cudaFuncGetName
   {"cuFuncGetName",                                        {"hipFuncGetName",                                          "", CONV_EXECUTION, API_DRIVER, SEC::EXECUTION, HIP_UNSUPPORTED}},
 
   // 23. Execution Control [DEPRECATED]
@@ -627,7 +627,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   {"cuGraphAddChildGraphNode",                             {"hipGraphAddChildGraphNode",                               "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphAddDependencies
   {"cuGraphAddDependencies",                               {"hipGraphAddDependencies",                                 "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
-  //
+  // cudaGraphAddDependencies_v2
   {"cuGraphAddDependencies_v2",                            {"hipGraphAddDependencies_v2",                              "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
   // cudaGraphAddEmptyNode
   {"cuGraphAddEmptyNode",                                  {"hipGraphAddEmptyNode",                                    "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
@@ -661,7 +661,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   {"cuGraphExecDestroy",                                   {"hipGraphExecDestroy",                                     "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphGetEdges
   {"cuGraphGetEdges",                                      {"hipGraphGetEdges",                                        "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
-  //
+  // cudaGraphGetEdges_v2
   {"cuGraphGetEdges_v2",                                   {"hipGraphGetEdges_v2",                                     "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
   // cudaGraphGetNodes
   {"cuGraphGetNodes",                                      {"hipGraphGetNodes",                                        "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
@@ -700,10 +700,12 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   {"cuGraphNodeFindInClone",                               {"hipGraphNodeFindInClone",                                 "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphNodeGetDependencies
   {"cuGraphNodeGetDependencies",                           {"hipGraphNodeGetDependencies",                             "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
-  //
+  // cudaGraphNodeGetDependencies_v2
   {"cuGraphNodeGetDependencies_v2",                        {"hipGraphNodeGetDependencies_v2",                          "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
   // cudaGraphNodeGetDependentNodes
   {"cuGraphNodeGetDependentNodes",                         {"hipGraphNodeGetDependentNodes",                           "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
+  // cudaGraphNodeGetDependentNodes_v2
+  {"cuGraphNodeGetDependentNodes_v2",                      {"hipGraphNodeGetDependentNodes_v2",                        "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
   // cudaGraphNodeGetEnabled
   {"cuGraphNodeGetEnabled",                                {"hipGraphNodeGetEnabled",                                  "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphNodeGetType
@@ -712,7 +714,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   {"cuGraphNodeSetEnabled",                                {"hipGraphNodeSetEnabled",                                  "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphRemoveDependencies
   {"cuGraphRemoveDependencies",                            {"hipGraphRemoveDependencies",                              "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
-  //
+  // cudaGraphRemoveDependencies_v2
   {"cuGraphRemoveDependencies_v2",                         {"hipGraphRemoveDependencies_v2",                           "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
   // no analogue
   // NOTE: Not equal to cudaGraphExecMemcpyNodeSetParams due to different signatures:
@@ -805,13 +807,13 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DRIVER_FUNCTION_MAP {
   {"cuGraphExecGetFlags",                                  {"hipGraphExecGetFlags",                                    "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphAddNode
   {"cuGraphAddNode",                                       {"hipGraphAddNode",                                         "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
-  //
+  // cudaGraphAddNode_v2
   {"cuGraphAddNode_v2",                                    {"hipGraphAddNode_v2",                                      "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
   // cudaGraphNodeSetParams
   {"cuGraphNodeSetParams",                                 {"hipGraphNodeSetParams",                                   "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
   // cudaGraphExecNodeSetParams
   {"cuGraphExecNodeSetParams",                             {"hipGraphExecNodeSetParams",                               "", CONV_GRAPH, API_DRIVER, SEC::GRAPH}},
-  //
+  // cudaGraphConditionalHandleCreate
   {"cuGraphConditionalHandleCreate",                       {"hipGraphConditionalHandleCreate",                         "", CONV_GRAPH, API_DRIVER, SEC::GRAPH, HIP_UNSUPPORTED}},
 
   // 25. Occupancy
@@ -1431,6 +1433,7 @@ const std::map<llvm::StringRef, cudaAPIversions> CUDA_DRIVER_FUNCTION_VER_MAP {
   {"cuGraphRemoveDependencies_v2",                         {CUDA_123, CUDA_0,   CUDA_0  }},
   {"cuGraphAddNode_v2",                                    {CUDA_123, CUDA_0,   CUDA_0  }},
   {"cuGraphConditionalHandleCreate",                       {CUDA_123, CUDA_0,   CUDA_0  }},
+  {"cuGraphNodeGetDependentNodes_v2",                      {CUDA_123, CUDA_0,   CUDA_0  }},
 };
 
 const std::map<llvm::StringRef, hipAPIversions> HIP_DRIVER_FUNCTION_VER_MAP {