merge amd-staging into amd-mainline

Change-Id: Ia29cf49cff3b0bb6f5e71024e7ae38e5f235d2ec
ROCm · Aug 9, 2022 · 3b6d104 · 3b6d104
2 parents 2ab347f + 08552e4
commit 3b6d104
Show file tree

Hide file tree

Showing 27 changed files with 1,183 additions and 114 deletions.
diff --git a/CMakeLists.txt b/CMakeLists.txt
@@ -68,7 +68,7 @@ if(LLVM_PACKAGE_VERSION VERSION_GREATER "9.0.1")
 endif()
 
 if(LLVM_PACKAGE_VERSION VERSION_EQUAL "15.0.0" OR LLVM_PACKAGE_VERSION VERSION_GREATER "15.0.0")
-    target_link_libraries(hipify-clang PRIVATE LLVMWindowsDriver)
+    target_link_libraries(hipify-clang PRIVATE LLVMWindowsDriver clangSupport)
 endif()
 
 if(MSVC)
@@ -86,12 +86,6 @@ else()
     set(addr_var )
 endif()
 
-# [ToDo] Remove SWDEV_331863 related guards from CMakeLists.txt and HipifyAction.cpp when the blocker SWDEV_331863 is overcome
-option (SWDEV_331863 "Enables SWDEV-331863 blocker workaround" OFF)
-if(SWDEV_331863)
-    add_definitions(-DSWDEV_331863)
-endif()
-
 set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} ${EXTRA_CFLAGS} ${addr_var}")
 set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} ${EXTRA_CFLAGS} -DHIPIFY_CLANG_RES=\\\"${LLVM_LIBRARY_DIRS}/clang/${LLVM_VERSION_MAJOR}.${LLVM_VERSION_MINOR}.${LLVM_VERSION_PATCH}\\\" ${addr_var}")
 

diff --git a/README.md b/README.md
@@ -45,7 +45,7 @@ After applying all the matchers, the output HIP source is produced.
 
 1. [**LLVM+CLANG**](http://releases.llvm.org) of at least version [4.0.0](http://releases.llvm.org/download.html#4.0.0); the latest stable and recommended release: [**14.0.6**](https://github.com/llvm/llvm-project/releases/tag/llvmorg-14.0.6).
 
-2. [**CUDA**](https://developer.nvidia.com/cuda-downloads) of at least version [8.0](https://developer.nvidia.com/cuda-80-ga2-download-archive), the latest supported version is [**11.7.0**](https://developer.nvidia.com/cuda-downloads).
+2. [**CUDA**](https://developer.nvidia.com/cuda-downloads) of at least version [8.0](https://developer.nvidia.com/cuda-80-ga2-download-archive), the latest supported version is [**11.7.1**](https://developer.nvidia.com/cuda-downloads).
 
 <table align="center">
   <thead>
@@ -160,7 +160,7 @@ After applying all the matchers, the output HIP source is produced.
           <a href="https://github.com/llvm/llvm-project/releases/tag/llvmorg-14.0.2">14.0.2</a>,
           <a href="https://github.com/llvm/llvm-project/releases/tag/llvmorg-14.0.3">14.0.3</a>,
           <a href="https://github.com/llvm/llvm-project/releases/tag/llvmorg-14.0.4">14.0.4</a>
-      <td bgcolor="eefaeb"><a href="https://developer.nvidia.com/cuda-downloads">11.7.0</a></td>
+      <td bgcolor="eefaeb"><a href="https://developer.nvidia.com/cuda-downloads">11.7.1</a></td>
       <td>works only with the patch <br> due to the clang's bug <a href="https://github.com/llvm/llvm-project/issues/54609">54609</a><br>
           <a href="patches/patch_for_clang_14.0.0_bug_54609.zip">patch for 14.0.0</a>**<br>
           <a href="patches/patch_for_clang_14.0.1_bug_54609.zip">patch for 14.0.1</a>**<br>
@@ -171,7 +171,7 @@ After applying all the matchers, the output HIP source is produced.
     <tr align="center">
       <td><a href="https://github.com/llvm/llvm-project/releases/tag/llvmorg-14.0.5">14.0.5</a>,
           <a href="https://github.com/llvm/llvm-project/releases/tag/llvmorg-14.0.6"><b>14.0.6</b></a></td>
-      <td bgcolor="eefaeb"><a href="https://developer.nvidia.com/cuda-downloads"><b>11.7.0</b></a></td>
+      <td bgcolor="eefaeb"><a href="https://developer.nvidia.com/cuda-downloads"><b>11.7.1</b></a></td>
       <td colspan=2 bgcolor="eefaeb"><font color="green"><b>LATEST STABLE CONFIG</b></font></td>
     </tr>
   </tbody>
@@ -385,7 +385,7 @@ Ubuntu 14: LLVM 4.0.0 - 7.1.0, CUDA 7.0 - 9.0, cuDNN 5.0.5 - 7.6.5
 
 Ubuntu 16-18: LLVM 8.0.0 - 14.0.6, CUDA 8.0 - 10.2, cuDNN 5.1.10 - 8.0.5
 
-Ubuntu 20-21: LLVM 9.0.0 - 14.0.6, CUDA 8.0 - 11.7.0, cuDNN 5.1.10 - 8.4.1
+Ubuntu 20-21: LLVM 9.0.0 - 14.0.6, CUDA 8.0 - 11.7.1, cuDNN 5.1.10 - 8.4.1
 
 Minimum build system requirements for the above configurations:
 
@@ -560,8 +560,8 @@ Testing Time: 6.03s
 | 10.0.0 - 11.0.0 | 8.0 - 11.1   | 7.6.5  - 8.0.5 | 2017.15.9.30, 2019.16.8.3                | 3.19.2         | 3.9.1        |
 | 11.0.1 - 11.1.0 | 8.0 - 11.2.2 | 7.6.5  - 8.0.5 | 2017.15.9.31, 2019.16.8.4                | 3.19.3         | 3.9.2        |
 | 12.0.0 - 13.0.1 | 8.0 - 11.5.1 | 7.6.5  - 8.3.2 | 2017.15.9.43, 2019.16.11.9               | 3.22.2         | 3.10.2       |
-| 14.0.0 - 14.0.6 | 8.0 - 11.7.0 | 8.0.5  - 8.4.1 | 2017.15.9.49, 2019.16.11.17, 2022.17.2.6 | 3.23.2         | 3.10.5       |
-| 15.0.0git       | 8.0 - 11.7.0 | 8.0.5  - 8.4.1 | 2017.15.9.49, 2019.16.11.17, 2022.17.2.6 | 3.23.2         | 3.10.5       |
+| 14.0.0 - 14.0.6 | 8.0 - 11.7.1 | 8.0.5  - 8.4.1 | 2017.15.9.49, 2019.16.11.17, 2022.17.2.6 | 3.24.0         | 3.10.6       |
+| 15.0.0git       | 8.0 - 11.7.1 | 8.0.5  - 8.4.1 | 2017.15.9.49, 2019.16.11.17, 2022.17.2.6 | 3.24.0         | 3.10.6       |
 
 *Building with testing support by `Visual Studio 17 2022` on `Windows 10`:*
 

diff --git a/bin/hipify-perl b/bin/hipify-perl
@@ -967,7 +967,7 @@ push(@exclude_filelist, split(',', $exclude_files));
 %exclude_dirhash = map { $_ => 1 } @exclude_dirlist;
 %exclude_filehash = map { $_ => 1 } @exclude_filelist;
 
-@statNames = ("error", "init", "version", "device", "context", "module", "memory", "virtual_memory", "stream_ordered_memory", "addressing", "stream", "event", "external_resource_interop", "stream_memory", "execution", "graph", "occupancy", "texture", "surface", "peer", "graphics", "interactions", "profiler", "openGL", "D3D9", "D3D10", "D3D11", "VDPAU", "EGL", "thread", "complex", "library", "device_library", "device_function", "include", "include_cuda_main_header", "type", "literal", "numeric_literal", "define", "extern_shared", "kernel_launch");
+@statNames = ("error", "init", "version", "device", "context", "module", "memory", "virtual_memory", "stream_ordered_memory", "addressing", "stream", "event", "external_resource_interop", "stream_memory", "execution", "graph", "occupancy", "texture", "surface", "peer", "graphics", "interactions", "profiler", "openGL", "D3D9", "D3D10", "D3D11", "VDPAU", "EGL", "thread", "complex", "library", "device_library", "device_function", "include", "include_cuda_main_header", "include_cuda_main_header_v2", "type", "literal", "numeric_literal", "define", "extern_shared", "kernel_launch");
 
 sub totalStats {
     my %count = %{shift()};
@@ -1599,7 +1599,6 @@ sub rocSubstitutions {
     subst("cublasZtrsv", "rocblas_ztrsv", "library");
     subst("cublasZtrsv_v2", "rocblas_ztrsv", "library");
     subst("cublas.h", "rocblas.h", "include_cuda_main_header");
-    subst("cublas_v2.h", "rocblas.h", "include_cuda_main_header");
     subst("cublasAtomicsMode_t", "rocblas_atomics_mode", "type");
     subst("cublasContext", "_rocblas_handle", "type");
     subst("cublasDataType_t", "rocblas_datatype", "type");
@@ -1935,6 +1934,7 @@ sub simpleSubstitutions {
     subst("cuGraphAddHostNode", "hipGraphAddHostNode", "graph");
     subst("cuGraphAddKernelNode", "hipGraphAddKernelNode", "graph");
     subst("cuGraphChildGraphNodeGetGraph", "hipGraphChildGraphNodeGetGraph", "graph");
+    subst("cuGraphClone", "hipGraphClone", "graph");
     subst("cuGraphCreate", "hipGraphCreate", "graph");
     subst("cuGraphDestroy", "hipGraphDestroy", "graph");
     subst("cuGraphDestroyNode", "hipGraphDestroyNode", "graph");
@@ -1982,6 +1982,7 @@ sub simpleSubstitutions {
     subst("cudaGraphAddMemcpyNodeToSymbol", "hipGraphAddMemcpyNodeToSymbol", "graph");
     subst("cudaGraphAddMemsetNode", "hipGraphAddMemsetNode", "graph");
     subst("cudaGraphChildGraphNodeGetGraph", "hipGraphChildGraphNodeGetGraph", "graph");
+    subst("cudaGraphClone", "hipGraphClone", "graph");
     subst("cudaGraphCreate", "hipGraphCreate", "graph");
     subst("cudaGraphDestroy", "hipGraphDestroy", "graph");
     subst("cudaGraphDestroyNode", "hipGraphDestroyNode", "graph");
@@ -2075,6 +2076,7 @@ sub simpleSubstitutions {
     subst("cudaGetTextureAlignmentOffset", "hipGetTextureAlignmentOffset", "texture");
     subst("cudaGetTextureObjectResourceDesc", "hipGetTextureObjectResourceDesc", "texture");
     subst("cudaGetTextureObjectResourceViewDesc", "hipGetTextureObjectResourceViewDesc", "texture");
+    subst("cudaGetTextureObjectTextureDesc", "hipGetTextureObjectTextureDesc", "texture");
     subst("cudaGetTextureReference", "hipGetTextureReference", "texture");
     subst("cudaUnbindTexture", "hipUnbindTexture", "texture");
     subst("cudaCreateSurfaceObject", "hipCreateSurfaceObject", "surface");
@@ -3238,8 +3240,9 @@ sub simpleSubstitutions {
     subst("caffe2\/operators\/spatial_batch_norm_op.h", "caffe2\/operators\/hip\/spatial_batch_norm_op_miopen.hip", "include");
     subst("channel_descriptor.h", "hip\/channel_descriptor.h", "include");
     subst("cooperative_groups.h", "hip\/hip_cooperative_groups.h", "include");
+    subst("cublas_api.h", "hipblas.h", "include");
     subst("cuda_fp16.h", "hip\/hip_fp16.h", "include");
-    subst("cuda_profiler_api.h", "hip\/hip_profile.h", "include");
+    subst("cuda_profiler_api.h", "hip\/hip_runtime_api.h", "include");
     subst("cuda_runtime_api.h", "hip\/hip_runtime_api.h", "include");
     subst("cuda_texture_types.h", "hip\/hip_texture_types.h", "include");
     subst("cufftXt.h", "hipfftXt.h", "include");
@@ -3266,7 +3269,6 @@ sub simpleSubstitutions {
     subst("cuComplex.h", "hip\/hip_complex.h", "include_cuda_main_header");
     subst("cub\/cub.cuh", "hipcub\/hipcub.hpp", "include_cuda_main_header");
     subst("cublas.h", "hipblas.h", "include_cuda_main_header");
-    subst("cublas_v2.h", "hipblas.h", "include_cuda_main_header");
     subst("cuda.h", "hip\/hip_runtime.h", "include_cuda_main_header");
     subst("cuda_runtime.h", "hip\/hip_runtime.h", "include_cuda_main_header");
     subst("cudnn.h", "hipDNN.h", "include_cuda_main_header");
@@ -4644,6 +4646,7 @@ sub simpleSubstitutions {
     subst("REGISTER_CUDA_OPERATOR", "REGISTER_HIP_OPERATOR", "define");
     subst("REGISTER_CUDA_OPERATOR_CREATOR", "REGISTER_HIP_OPERATOR_CREATOR", "define");
     subst("_CubLog", "_HipcubLog", "define");
+    subst("__CUB_ALIGN_BYTES", "__HIPCUB_ALIGN_BYTES", "define");
     subst("__CUDACC__", "__HIPCC__", "define");
     subst("cudaArrayCubemap", "hipArrayCubemap", "define");
     subst("cudaArrayDefault", "hipArrayDefault", "define");
@@ -6132,7 +6135,6 @@ sub warnUnsupportedFunctions {
         "cudaGraphDebugDotFlagsExtSemasSignalNodeParams",
         "cudaGraphDebugDotFlagsEventNodeParams",
         "cudaGraphDebugDotFlags",
-        "cudaGraphClone",
         "cudaGraphAddMemFreeNode",
         "cudaGraphAddMemAllocNode",
         "cudaGraphAddExternalSemaphoresWaitNode",
@@ -6573,7 +6575,6 @@ sub warnUnsupportedFunctions {
         "cuGraphExecExternalSemaphoresSignalNodeSetParams",
         "cuGraphExecBatchMemOpNodeSetParams",
         "cuGraphDebugDotPrint",
-        "cuGraphClone",
         "cuGraphBatchMemOpNodeSetParams",
         "cuGraphBatchMemOpNodeGetParams",
         "cuGraphAddMemsetNode",
@@ -6678,10 +6679,14 @@ sub warnUnsupportedFunctions {
         "bsrilu02Info",
         "bsric02Info",
         "__curand_umul",
+        "__CUB_LP64__",
+        "_CUB_ASM_PTR_SIZE_",
+        "_CUB_ASM_PTR_",
         "PATCH_LEVEL",
         "MINOR_VERSION",
         "MAX_CUFFT_ERROR",
         "MAJOR_VERSION",
+        "CubVector",
         "CubDebugExit",
         "CUuserObject_st",
         "CUuserObject_flags_enum",
@@ -7688,22 +7693,58 @@ sub warnUnsupportedFunctions {
         "CUDA_ARRAY3D_2DARRAY",
         "CUB_USE_COOPERATIVE_GROUPS",
         "CUB_SUBSCRIPTION_FACTOR",
+        "CUB_STATIC_ASSERT",
         "CUB_SMEM_BANKS",
         "CUB_RUNTIME_ENABLED",
+        "CUB_ROUND_UP_NEAREST",
+        "CUB_ROUND_DOWN_NEAREST",
+        "CUB_QUOTIENT_FLOOR",
+        "CUB_QUOTIENT_CEILING",
         "CUB_PTX_SUBSCRIPTION_FACTOR",
         "CUB_PTX_SMEM_BANKS",
         "CUB_PTX_PREFER_CONFLICT_OVER_PADDING",
         "CUB_PTX_LOG_WARP_THREADS",
         "CUB_PTX_LOG_SMEM_BANKS",
+        "CUB_PREVENT_MACRO_SUBSTITUTION",
         "CUB_PREFER_CONFLICT_OVER_PADDING",
+        "CUB_MSVC_VERSION_FULL",
+        "CUB_MSVC_VERSION",
         "CUB_MAX_DEVICES",
         "CUB_LOG_WARP_THREADS",
         "CUB_LOG_SMEM_BANKS",
         "CUB_IS_HOST_CODE",
         "CUB_IS_DEVICE_CODE",
         "CUB_INCLUDE_HOST_CODE",
         "CUB_INCLUDE_DEVICE_CODE",
-        "CUB_CPP_DIALECT"
+        "CUB_IGNORE_DEPRECATED_DIALECT",
+        "CUB_IGNORE_DEPRECATED_CPP_DIALECT",
+        "CUB_IGNORE_DEPRECATED_CPP_11",
+        "CUB_IGNORE_DEPRECATED_COMPILER",
+        "CUB_IGNORE_DEPRECATED_API",
+        "CUB_HOST_COMPILER_UNKNOWN",
+        "CUB_HOST_COMPILER_MSVC",
+        "CUB_HOST_COMPILER_GCC",
+        "CUB_HOST_COMPILER_CLANG",
+        "CUB_HOST_COMPILER",
+        "CUB_DEVICE_COMPILER_UNKNOWN",
+        "CUB_DEVICE_COMPILER_NVCC",
+        "CUB_DEVICE_COMPILER_MSVC",
+        "CUB_DEVICE_COMPILER_GCC",
+        "CUB_DEVICE_COMPILER_CLANG",
+        "CUB_DEVICE_COMPILER",
+        "CUB_DEPRECATED",
+        "CUB_DEFINE_VECTOR_TYPE",
+        "CUB_DEFINE_DETECT_NESTED_TYPE",
+        "CUB_CPP_DIALECT",
+        "CUB_CPLUSPLUS",
+        "CUB_COMP_DEPR_IMPL1",
+        "CUB_COMP_DEPR_IMPL0",
+        "CUB_COMP_DEPR_IMPL",
+        "CUB_COMPILER_DEPRECATION_SOFT",
+        "CUB_COMPILER_DEPRECATION",
+        "CUB_CAT_",
+        "CUB_CAT",
+        "CUB_ALIGN"
     )
     {
         my $mt = m/($func)/g;
@@ -8153,7 +8194,7 @@ while (@ARGV) {
             transformHostFunctions();
             # TODO: would like to move this code outside loop but it uses $_ which contains the whole file
             unless ($no_output) {
-                my $apiCalls   = $ft{'error'} + $ft{'init'} + $ft{'version'} + $ft{'device'} + $ft{'context'} + $ft{'module'} + $ft{'memory'} + $ft{'virtual_memory'} + $ft{'stream_ordered_memory'} + $ft{'addressing'} + $ft{'stream'} + $ft{'event'} + $ft{'external_resource_interop'} + $ft{'stream_memory'} + $ft{'execution'} + $ft{'graph'} + $ft{'occupancy'} + $ft{'texture'} + $ft{'surface'} + $ft{'peer'} + $ft{'graphics'} + $ft{'interactions'} + $ft{'profiler'} + $ft{'openGL'} + $ft{'D3D9'} + $ft{'D3D10'} + $ft{'D3D11'} + $ft{'VDPAU'} + $ft{'EGL'} + $ft{'thread'} + $ft{'complex'} + $ft{'library'} + $ft{'device_library'} + $ft{'include'} + $ft{'include_cuda_main_header'} + $ft{'type'} + $ft{'literal'} + $ft{'numeric_literal'} + $ft{'define'};
+                my $apiCalls   = $ft{'error'} + $ft{'init'} + $ft{'version'} + $ft{'device'} + $ft{'context'} + $ft{'module'} + $ft{'memory'} + $ft{'virtual_memory'} + $ft{'stream_ordered_memory'} + $ft{'addressing'} + $ft{'stream'} + $ft{'event'} + $ft{'external_resource_interop'} + $ft{'stream_memory'} + $ft{'execution'} + $ft{'graph'} + $ft{'occupancy'} + $ft{'texture'} + $ft{'surface'} + $ft{'peer'} + $ft{'graphics'} + $ft{'interactions'} + $ft{'profiler'} + $ft{'openGL'} + $ft{'D3D9'} + $ft{'D3D10'} + $ft{'D3D11'} + $ft{'VDPAU'} + $ft{'EGL'} + $ft{'thread'} + $ft{'complex'} + $ft{'library'} + $ft{'device_library'} + $ft{'include'} + $ft{'include_cuda_main_header'} + $ft{'include_cuda_main_header_v2'} + $ft{'type'} + $ft{'literal'} + $ft{'numeric_literal'} + $ft{'define'};
                 my $kernStuff  = $hasDeviceCode + $ft{'kernel_launch'} + $ft{'device_function'};
                 my $totalCalls = $apiCalls + $kernStuff;
                 $is_dos = m/\r\n$/;

diff --git a/doc/markdown/CUBLAS_API_supported_by_HIP.md b/doc/markdown/CUBLAS_API_supported_by_HIP.md
@@ -63,7 +63,7 @@
 |`CUBLAS_GEMM_ALGO8_TENSOR_OP`|9.2| | | | | | | |
 |`CUBLAS_GEMM_ALGO9`|9.0| | | | | | | |
 |`CUBLAS_GEMM_ALGO9_TENSOR_OP`|9.2| | | | | | | |
-|`CUBLAS_GEMM_DEFAULT`|8.0| | |`HIPBLAS_GEMM_DEFAULT`|1.8.2| | | |
+|`CUBLAS_GEMM_DEFAULT`|9.0| | |`HIPBLAS_GEMM_DEFAULT`|1.8.2| | | |
 |`CUBLAS_GEMM_DEFAULT_TENSOR_OP`|9.0| | | | | | | |
 |`CUBLAS_GEMM_DFALT`|8.0| | |`HIPBLAS_GEMM_DEFAULT`|1.8.2| | | |
 |`CUBLAS_GEMM_DFALT_TENSOR_OP`|9.0| | | | | | | |

diff --git a/doc/markdown/CUB_API_supported_by_HIP.md b/doc/markdown/CUB_API_supported_by_HIP.md
@@ -4,35 +4,76 @@
 
 |**CUDA**|**A**|**D**|**R**|**HIP**|**A**|**D**|**R**|**E**|
 |:--|:-:|:-:|:-:|:--|:-:|:-:|:-:|:-:|
+|`CUB_ALIGN`| | | | | | | | |
+|`CUB_CAT`| | | | | | | | |
+|`CUB_CAT_`| | | | | | | | |
+|`CUB_COMPILER_DEPRECATION`| | | | | | | | |
+|`CUB_COMPILER_DEPRECATION_SOFT`| | | | | | | | |
+|`CUB_COMP_DEPR_IMPL`| | | | | | | | |
+|`CUB_COMP_DEPR_IMPL0`| | | | | | | | |
+|`CUB_COMP_DEPR_IMPL1`| | | | | | | | |
+|`CUB_CPLUSPLUS`| | | | | | | | |
 |`CUB_CPP_DIALECT`| | | | | | | | |
+|`CUB_DEFINE_DETECT_NESTED_TYPE`| | | | | | | | |
+|`CUB_DEFINE_VECTOR_TYPE`| | | | | | | | |
+|`CUB_DEPRECATED`| | | | | | | | |
+|`CUB_DEVICE_COMPILER`| | | | | | | | |
+|`CUB_DEVICE_COMPILER_CLANG`| | | | | | | | |
+|`CUB_DEVICE_COMPILER_GCC`| | | | | | | | |
+|`CUB_DEVICE_COMPILER_MSVC`| | | | | | | | |
+|`CUB_DEVICE_COMPILER_NVCC`| | | | | | | | |
+|`CUB_DEVICE_COMPILER_UNKNOWN`| | | | | | | | |
+|`CUB_HOST_COMPILER`| | | | | | | | |
+|`CUB_HOST_COMPILER_CLANG`| | | | | | | | |
+|`CUB_HOST_COMPILER_GCC`| | | | | | | | |
+|`CUB_HOST_COMPILER_MSVC`| | | | | | | | |
+|`CUB_HOST_COMPILER_UNKNOWN`| | | | | | | | |
+|`CUB_IGNORE_DEPRECATED_API`| | | | | | | | |
+|`CUB_IGNORE_DEPRECATED_COMPILER`| | | | | | | | |
+|`CUB_IGNORE_DEPRECATED_CPP_11`| | | | | | | | |
+|`CUB_IGNORE_DEPRECATED_CPP_DIALECT`| | | | | | | | |
+|`CUB_IGNORE_DEPRECATED_DIALECT`| | | | | | | | |
 |`CUB_INCLUDE_DEVICE_CODE`| | | | | | | | |
 |`CUB_INCLUDE_HOST_CODE`| | | | | | | | |
 |`CUB_IS_DEVICE_CODE`| | | | | | | | |
 |`CUB_IS_HOST_CODE`| | | | | | | | |
 |`CUB_LOG_SMEM_BANKS`| | | | | | | | |
 |`CUB_LOG_WARP_THREADS`| | | | | | | | |
-|`CUB_MAX`| | | |`CUB_MAX`| | | | |
+|`CUB_MAX`| | | |`CUB_MAX`|4.5.0| | | |
 |`CUB_MAX_DEVICES`| | | | | | | | |
-|`CUB_MIN`| | | |`CUB_MIN`|2.5.0| | | |
+|`CUB_MIN`| | | |`CUB_MIN`|4.5.0| | | |
+|`CUB_MSVC_VERSION`| | | | | | | | |
+|`CUB_MSVC_VERSION_FULL`| | | | | | | | |
 |`CUB_NAMESPACE_BEGIN`| | | |`BEGIN_HIPCUB_NAMESPACE`|2.5.0| | | |
 |`CUB_NAMESPACE_END`| | | |`END_HIPCUB_NAMESPACE`|2.5.0| | | |
 |`CUB_PREFER_CONFLICT_OVER_PADDING`| | | | | | | | |
+|`CUB_PREVENT_MACRO_SUBSTITUTION`| | | | | | | | |
 |`CUB_PTX_ARCH`| | | |`HIPCUB_ARCH`|2.5.0| | | |
 |`CUB_PTX_LOG_SMEM_BANKS`| | | | | | | | |
 |`CUB_PTX_LOG_WARP_THREADS`| | | | | | | | |
 |`CUB_PTX_PREFER_CONFLICT_OVER_PADDING`| | | | | | | | |
 |`CUB_PTX_SMEM_BANKS`| | | | | | | | |
 |`CUB_PTX_SUBSCRIPTION_FACTOR`| | | | | | | | |
 |`CUB_PTX_WARP_THREADS`| | | |`HIPCUB_WARP_THREADS`|2.5.0| | | |
+|`CUB_QUOTIENT_CEILING`| | | | | | | | |
+|`CUB_QUOTIENT_FLOOR`| | | | | | | | |
+|`CUB_ROUND_DOWN_NEAREST`| | | | | | | | |
+|`CUB_ROUND_UP_NEAREST`| | | | | | | | |
 |`CUB_RUNTIME_ENABLED`| | | | | | | | |
 |`CUB_RUNTIME_FUNCTION`| | | |`HIPCUB_RUNTIME_FUNCTION`|2.5.0| | | |
 |`CUB_SMEM_BANKS`| | | | | | | | |
+|`CUB_STATIC_ASSERT`| | | | | | | | |
 |`CUB_STDERR`| | | |`HIPCUB_STDERR`|2.5.0| | | |
 |`CUB_SUBSCRIPTION_FACTOR`| | | | | | | | |
 |`CUB_USE_COOPERATIVE_GROUPS`| | | | | | | | |
 |`CubDebug`| | | |`HipcubDebug`|2.5.0| | | |
 |`CubDebugExit`| | | | | | | | |
+|`CubVector`| | | | | | | | |
+|`_CUB_ASM_PTR_`| | | | | | | | |
+|`_CUB_ASM_PTR_SIZE_`| | | | | | | | |
 |`_CubLog`| | | |`_HipcubLog`|2.5.0| | | |
+|`__CUB_ALIGN_BYTES`| | | |`__HIPCUB_ALIGN_BYTES`|4.5.0| | | |
+|`__CUB_LP64__`| | | | | | | | |
 
 
 \*A - Added; D - Deprecated; R - Removed; E - Experimental
diff --git a/doc/markdown/CUDA_Driver_API_functions_supported_by_HIP.md b/doc/markdown/CUDA_Driver_API_functions_supported_by_HIP.md
@@ -1525,7 +1525,7 @@
 |`cuGraphBatchMemOpNodeGetParams`|11.7| | | | | | | |
 |`cuGraphBatchMemOpNodeSetParams`|11.7| | | | | | | |
 |`cuGraphChildGraphNodeGetGraph`|10.0| | |`hipGraphChildGraphNodeGetGraph`|5.0.0| | | |
-|`cuGraphClone`|10.0| | | | | | | |
+|`cuGraphClone`|10.0| | |`hipGraphClone`|5.0.0| | | |
 |`cuGraphCreate`|10.0| | |`hipGraphCreate`|4.3.0| | | |
 |`cuGraphDebugDotPrint`|11.3| | | | | | | |
 |`cuGraphDestroy`|10.0| | |`hipGraphDestroy`|4.3.0| | | |