Merge pull request #1623 from emankov/HIPIFY

[HIPIFY][perl] Fix cuda2roc headers substitution under the `--roc` option
ROCm · Aug 19, 2024 · 548a2df · 548a2df
2 parents 54ba47b + e273b27
commit 548a2df
Show file tree

Hide file tree

Showing 2 changed files with 33 additions and 5 deletions.
diff --git a/bin/hipify-perl b/bin/hipify-perl
@@ -3100,8 +3100,35 @@ sub rocSubstitutions {
     subst("cusparseZsctr", "rocsparse_zsctr", "library");
     subst("__half", "rocblas_half", "device_type");
     subst("__nv_bfloat16", "rocblas_bfloat16", "device_type");
+    subst("cublas_api.h", "rocblas.h", "include");
+    subst("curand_discrete.h", "rocrand\/rocrand_discrete.h", "include");
+    subst("curand_discrete2.h", "rocrand\/rocrand_discrete.h", "include");
+    subst("curand_globals.h", "rocrand\/rocrand_common.h", "include");
+    subst("curand_kernel.h", "rocrand\/rocrand_kernel.h", "include");
+    subst("curand_lognormal.h", "rocrand\/rocrand_log_normal.h", "include");
+    subst("curand_mrg32k3a.h", "rocrand\/rocrand_mrg32k3a.h", "include");
+    subst("curand_mtgp32.h", "rocrand\/rocrand_mtgp32.h", "include");
+    subst("curand_mtgp32_host.h", "rocrand\/rocrand_mtgp32.h", "include");
+    subst("curand_mtgp32_kernel.h", "rocrand\/rocrand_mtgp32.h", "include");
+    subst("curand_mtgp32dc_p_11213.h", "rocrand\/rocrand_mtgp32_11213.h", "include");
+    subst("curand_normal.h", "rocrand\/rocrand_normal.h", "include");
+    subst("curand_normal_static.h", "rocrand\/rocrand_normal.h", "include");
+    subst("curand_philox4x32_x.h", "rocrand\/rocrand_philox4x32_10.h", "include");
+    subst("curand_poisson.h", "rocrand\/rocrand_poisson.h", "include");
+    subst("curand_precalc.h", "rocrand\/rocrand_xorwow_precomputed.h", "include");
+    subst("curand_uniform.h", "rocrand\/rocrand_uniform.h", "include");
     subst("cublas.h", "rocblas.h", "include_cuda_main_header");
+    subst("cudnn.h", "miopen\/miopen.h", "include_cuda_main_header");
+    subst("curand.h", "rocrand\/rocrand.h", "include_cuda_main_header");
+    subst("cusolverDn.h", "rocsolver\/rocsolver.h", "include_cuda_main_header");
+    subst("cusolverMg.h", "rocsolver\/rocsolver.h", "include_cuda_main_header");
+    subst("cusolverRf.h", "rocsolver\/rocsolver.h", "include_cuda_main_header");
+    subst("cusolverSp.h", "rocsolver\/rocsolver.h", "include_cuda_main_header");
+    subst("cusolverSp_LOWLEVEL_PREVIEW.h", "rocsolver\/rocsolver.h", "include_cuda_main_header");
+    subst("cusolver_common.h", "rocsolver\/rocsolver.h", "include_cuda_main_header");
+    subst("cusparse.h", "rocsparse.h", "include_cuda_main_header");
     subst("cublas_v2.h", "rocblas.h", "include_cuda_main_header_v2");
+    subst("cusparse_v2.h", "rocsparse.h", "include_cuda_main_header_v2");
     subst("bsric02Info", "_rocsparse_mat_info", "type");
     subst("bsric02Info_t", "rocsparse_mat_info", "type");
     subst("bsrilu02Info", "_rocsparse_mat_info", "type");

diff --git a/src/CUDA2HIP_Perl.cpp b/src/CUDA2HIP_Perl.cpp
@@ -373,15 +373,16 @@ namespace perl {
     bool bTranslateToRoc = TranslateToRoc;
     TranslateToRoc = true;
     for (int i = 0; i < NUM_CONV_TYPES; ++i) {
-      if (i == CONV_INCLUDE_CUDA_MAIN_H || i == CONV_INCLUDE_CUDA_MAIN_V2_H) {
+      if (i == CONV_INCLUDE_CUDA_MAIN_H || i == CONV_INCLUDE_CUDA_MAIN_V2_H || i == CONV_INCLUDE) {
         for (auto &ma : CUDA_INCLUDE_MAP) {
           if (i == ma.second.type) {
+            if (!Statistics::isToRoc(ma.second)) continue;
+            string sROC = ma.second.rocName.str();
+            if (sROC.empty()) continue;
             string sCUDA = ma.first.str();
-            if (sCUDA != "cublas.h" && sCUDA != "cublas_v2.h") continue;
-            string sHIP = ma.second.rocName.str();
             sCUDA = regex_replace(sCUDA, regex("/"), "\\/");
-            sHIP = regex_replace(sHIP, regex("/"), "\\/");
-            *streamPtr.get() << tab << "subst(\"" << sCUDA << "\", \"" << sHIP << "\", \"" << counterNames[ma.second.type] << "\");" << endl;
+            sROC = regex_replace(sROC, regex("/"), "\\/");
+            *streamPtr.get() << tab << "subst(\"" << sCUDA << "\", \"" << sROC << "\", \"" << counterNames[ma.second.type] << "\");" << endl;
           }
         }
       } else {