Skip to content

Commit

Permalink
Merge pull request #1809 from emankov/HIPIFY
Browse files Browse the repository at this point in the history 
[HIPIFY][TensorMg][feature] `cuTensorMg` support - Part 1
  • Loading branch information
@emankov
emankov authored Dec 23, 2024
2 parents ca05bb3 + bf96ec6 commit a19ac36
Show file tree
Hide file tree
Showing 4 changed files with 17 additions and 0 deletions.
7 changes: 7 additions & 0 deletions bin/hipify-perl
View file
Original file line number Diff line number Diff line change
Expand Up @@ -6797,6 +6797,7 @@ sub simpleSubstitutions {
subst("curand_poisson.h", "hiprand\/hiprand_kernel.h", "include");
subst("curand_precalc.h", "hiprand\/hiprand_kernel.h", "include");
subst("curand_uniform.h", "hiprand\/hiprand_kernel.h", "include");
subst("cutensorMg.h", "hiptensor.h", "include");
subst("device_functions.h", "hip\/device_functions.h", "include");
subst("device_launch_parameters.h", "", "include");
subst("driver_types.h", "hip\/driver_types.h", "include");
Expand Down Expand Up @@ -10272,8 +10273,10 @@ sub warnHipOnlyUnsupportedFunctions {
"CUTENSOR_STATUS_CUBLAS_ERROR",
"CUTENSOR_R_MIN_TF32",
"CUTENSOR_R_MIN_8U",
"CUTENSOR_R_MIN_8I",
"CUTENSOR_R_MIN_64F",
"CUTENSOR_R_MIN_32U",
"CUTENSOR_R_MIN_32I",
"CUTENSOR_R_MIN_32F",
"CUTENSOR_R_MIN_16F",
"CUTENSOR_R_MIN_16BF",
Expand Down Expand Up @@ -10342,6 +10345,7 @@ sub warnHipOnlyUnsupportedFunctions {
"CUTENSOR_C_16F",
"CUTENSOR_C_16BF",
"CUTENSOR_COMPUTE_TF32",
"CUTENSOR_COMPUTE_3XTF32",
"CUTENSOR_CACHE_MODE_PEDANTIC",
"CUTENSOR_CACHE_MODE_NONE",
"CUTENSOR_AUTOTUNE_MODE_NONE",
Expand Down Expand Up @@ -11724,8 +11728,10 @@ sub warnRocOnlyUnsupportedFunctions {
"CUTENSOR_STATUS_CUBLAS_ERROR",
"CUTENSOR_R_MIN_TF32",
"CUTENSOR_R_MIN_8U",
"CUTENSOR_R_MIN_8I",
"CUTENSOR_R_MIN_64F",
"CUTENSOR_R_MIN_32U",
"CUTENSOR_R_MIN_32I",
"CUTENSOR_R_MIN_32F",
"CUTENSOR_R_MIN_16F",
"CUTENSOR_R_MIN_16BF",
Expand Down Expand Up @@ -11794,6 +11800,7 @@ sub warnRocOnlyUnsupportedFunctions {
"CUTENSOR_C_16F",
"CUTENSOR_C_16BF",
"CUTENSOR_COMPUTE_TF32",
"CUTENSOR_COMPUTE_3XTF32",
"CUTENSOR_CACHE_MODE_PEDANTIC",
"CUTENSOR_CACHE_MODE_NONE",
"CUTENSOR_AUTOTUNE_MODE_NONE",
Expand Down
3 changes: 3 additions & 0 deletions docs/tables/CUTENSOR_API_supported_by_HIP.md
View file
Original file line number Diff line number Diff line change
Expand Up @@ -18,6 +18,7 @@
|`CUTENSOR_COMPUTE_32F`|1.0.1.0| | |2.0.0.0|`HIPTENSOR_COMPUTE_32F`|5.7.0| | | | |
|`CUTENSOR_COMPUTE_32I`|1.0.1.0| | |2.0.0.0|`HIPTENSOR_COMPUTE_32I`|5.7.0| | | | |
|`CUTENSOR_COMPUTE_32U`|1.0.1.0| | |2.0.0.0|`HIPTENSOR_COMPUTE_32U`|5.7.0| | | | |
|`CUTENSOR_COMPUTE_3XTF32`|2.0.0.0| | | | | | | | | |
|`CUTENSOR_COMPUTE_64F`|1.0.1.0| | |2.0.0.0|`HIPTENSOR_COMPUTE_64F`|5.7.0| | | | |
|`CUTENSOR_COMPUTE_8I`|1.0.1.0| | |2.0.0.0|`HIPTENSOR_COMPUTE_8I`|5.7.0| | | | |
|`CUTENSOR_COMPUTE_8U`|1.0.1.0| | |2.0.0.0|`HIPTENSOR_COMPUTE_8U`|5.7.0| | | | |
Expand Down Expand Up @@ -106,8 +107,10 @@
|`CUTENSOR_R_MIN_16BF`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_16F`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_32F`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_32I`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_32U`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_64F`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_8I`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_8U`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_R_MIN_TF32`|1.0.1.0|1.2.0.0| |2.0.0.0| | | | | | |
|`CUTENSOR_STATUS_ALLOC_FAILED`|1.0.1.0| | | |`HIPTENSOR_STATUS_ALLOC_FAILED`|5.7.0| | | | |
Expand Down
1 change: 1 addition & 0 deletions src/CUDA2HIP.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -70,6 +70,7 @@ const std::map <llvm::StringRef, hipCounter> CUDA_INCLUDE_MAP {
{"cudnn.h", {"hipDNN.h", "miopen/miopen.h", CONV_INCLUDE_CUDA_MAIN_H, API_DNN, 0}},
// cuTensor includes
{"cutensor.h", {"hiptensor.h", "", CONV_INCLUDE_CUDA_MAIN_H, API_TENSOR, 0}},
{"cutensorMg.h", {"hiptensor.h", "", CONV_INCLUDE, API_TENSOR, 0}},
// cuFFT includes
{"cufft.h", {"hipfft/hipfft.h", "", CONV_INCLUDE_CUDA_MAIN_H, API_FFT, 0}},
{"cufftXt.h", {"hipfft/hipfftXt.h", "", CONV_INCLUDE, API_FFT, 0}},
Expand Down
6 changes: 6 additions & 0 deletions src/CUDA2HIP_TENSOR_API_types.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -59,6 +59,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TENSOR_TYPE_NAME_MAP {
{"CUTENSOR_COMPUTE_16F", {"HIPTENSOR_COMPUTE_16F", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1}},
{"CUTENSOR_COMPUTE_16BF", {"HIPTENSOR_COMPUTE_16BF", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1}},
{"CUTENSOR_COMPUTE_TF32", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_COMPUTE_3XTF32", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_COMPUTE_32F", {"HIPTENSOR_COMPUTE_32F", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1}},
{"CUTENSOR_COMPUTE_64F", {"HIPTENSOR_COMPUTE_64F", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1}},
{"CUTENSOR_COMPUTE_8U", {"HIPTENSOR_COMPUTE_8U", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1}},
Expand All @@ -73,6 +74,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_TENSOR_TYPE_NAME_MAP {
{"CUTENSOR_C_MIN_64F", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_R_MIN_8U", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_R_MIN_32U", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_R_MIN_8I", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_R_MIN_32I", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_R_MIN_16BF", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_R_MIN_TF32", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
{"CUTENSOR_C_MIN_TF32", {"", "", CONV_NUMERIC_LITERAL, API_TENSOR, 1, UNSUPPORTED}},
Expand Down Expand Up @@ -223,6 +226,7 @@ const std::map<llvm::StringRef, cudaAPIversions> CUDA_TENSOR_TYPE_NAME_VER_MAP {
{"CUTENSOR_COMPUTE_8I", {CUTENSOR_1010, CUDA_0, CUTENSOR_2000 }},
{"CUTENSOR_COMPUTE_32U", {CUTENSOR_1010, CUDA_0, CUTENSOR_2000 }},
{"CUTENSOR_COMPUTE_32I", {CUTENSOR_1010, CUDA_0, CUTENSOR_2000 }},
{"CUTENSOR_COMPUTE_3XTF32", {CUTENSOR_2000, CUDA_0, CUDA_0 }},
{"CUTENSOR_R_MIN_16F", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_C_MIN_16F", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_32F", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
Expand All @@ -231,6 +235,8 @@ const std::map<llvm::StringRef, cudaAPIversions> CUDA_TENSOR_TYPE_NAME_VER_MAP {
{"CUTENSOR_C_MIN_64F", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_8U", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_32U", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_8I", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_32I", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_16BF", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_R_MIN_TF32", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
{"CUTENSOR_C_MIN_TF32", {CUTENSOR_1010, CUTENSOR_1200, CUTENSOR_2000 }},
Expand Down

0 comments on commit a19ac36

Please sign in to comment.