From 788f81568d431ec07f11dc9d37fd156d006cb573 Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Mon, 30 Oct 2023 11:53:50 +0100 Subject: [PATCH] [HIPIFY][6.0.0][hipRAND][doc] Support for ROCm HIP 6.0.0 - Step 32 - hipRAND + Updated the `benchmark_curand_kernel.cpp` test, the regenerated hipify-perl and RAND docs [TODO] + Add synthetic tests on RAND --- bin/hipify-perl | 20 +++++++++---------- docs/tables/CURAND_API_supported_by_HIP.md | 20 +++++++++---------- src/CUDA2HIP_RAND_API_functions.cpp | 12 +++++++---- src/CUDA2HIP_RAND_API_types.cpp | 17 ++++++++++------ .../cuRAND/benchmark_curand_kernel.cpp | 3 +-- 5 files changed, 40 insertions(+), 32 deletions(-) diff --git a/bin/hipify-perl b/bin/hipify-perl index 113318a8..21217aae 100755 --- a/bin/hipify-perl +++ b/bin/hipify-perl @@ -3504,6 +3504,10 @@ sub simpleSubstitutions { subst("curandGenerateSeeds", "hiprandGenerateSeeds", "library"); subst("curandGenerateUniform", "hiprandGenerateUniform", "library"); subst("curandGenerateUniformDouble", "hiprandGenerateUniformDouble", "library"); + subst("curandGetDirectionVectors32", "hiprandGetDirectionVectors32", "library"); + subst("curandGetDirectionVectors64", "hiprandGetDirectionVectors64", "library"); + subst("curandGetScrambleConstants32", "hiprandGetScrambleConstants32", "library"); + subst("curandGetScrambleConstants64", "hiprandGetScrambleConstants64", "library"); subst("curandGetVersion", "hiprandGetVersion", "library"); subst("curandMakeMTGP32Constants", "hiprandMakeMTGP32Constants", "library"); subst("curandMakeMTGP32KernelState", "hiprandMakeMTGP32KernelState", "library"); @@ -4543,6 +4547,8 @@ sub simpleSubstitutions { subst("cufftType_t", "hipfftType_t", "type"); subst("cufftXtCallbackType", "hipfftXtCallbackType", "type"); subst("cufftXtCallbackType_t", "hipfftXtCallbackType_t", "type"); + subst("curandDirectionVectorSet", "hiprandDirectionVectorSet_t", "type"); + subst("curandDirectionVectorSet_t", "hiprandDirectionVectorSet_t", "type"); subst("curandDirectionVectors32_t", "hiprandDirectionVectors32_t", "type"); subst("curandDiscreteDistribution_st", "hiprandDiscreteDistribution_st", "type"); subst("curandDiscreteDistribution_t", "hiprandDiscreteDistribution_t", "type"); @@ -4883,6 +4889,8 @@ sub simpleSubstitutions { subst("CUFFT_UNALIGNED_DATA", "HIPFFT_UNALIGNED_DATA", "numeric_literal"); subst("CUFFT_Z2D", "HIPFFT_Z2D", "numeric_literal"); subst("CUFFT_Z2Z", "HIPFFT_Z2Z", "numeric_literal"); + subst("CURAND_DIRECTION_VECTORS_32_JOEKUO6", "HIPRAND_DIRECTION_VECTORS_32_JOEKUO6", "numeric_literal"); + subst("CURAND_DIRECTION_VECTORS_64_JOEKUO6", "HIPRAND_DIRECTION_VECTORS_64_JOEKUO6", "numeric_literal"); subst("CURAND_RNG_PSEUDO_DEFAULT", "HIPRAND_RNG_PSEUDO_DEFAULT", "numeric_literal"); subst("CURAND_RNG_PSEUDO_MRG32K3A", "HIPRAND_RNG_PSEUDO_MRG32K3A", "numeric_literal"); subst("CURAND_RNG_PSEUDO_MT19937", "HIPRAND_RNG_PSEUDO_MT19937", "numeric_literal"); @@ -4895,6 +4903,8 @@ sub simpleSubstitutions { subst("CURAND_RNG_QUASI_SOBOL32", "HIPRAND_RNG_QUASI_SOBOL32", "numeric_literal"); subst("CURAND_RNG_QUASI_SOBOL64", "HIPRAND_RNG_QUASI_SOBOL64", "numeric_literal"); subst("CURAND_RNG_TEST", "HIPRAND_RNG_TEST", "numeric_literal"); + subst("CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", "HIPRAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", "numeric_literal"); + subst("CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", "HIPRAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", "numeric_literal"); subst("CURAND_STATUS_ALLOCATION_FAILED", "HIPRAND_STATUS_ALLOCATION_FAILED", "numeric_literal"); subst("CURAND_STATUS_ARCH_MISMATCH", "HIPRAND_STATUS_ARCH_MISMATCH", "numeric_literal"); subst("CURAND_STATUS_DOUBLE_PRECISION_REQUIRED", "HIPRAND_STATUS_DOUBLE_PRECISION_REQUIRED", "numeric_literal"); @@ -6970,11 +6980,7 @@ sub warnUnsupportedFunctions { "curandHistogramM2V_st", "curandHistogramM2K_t", "curandHistogramM2K_st", - "curandGetScrambleConstants64", - "curandGetScrambleConstants32", "curandGetProperty", - "curandGetDirectionVectors64", - "curandGetDirectionVectors32", "curandGenerateLongLong", "curandDistribution_t", "curandDistribution_st", @@ -6983,8 +6989,6 @@ sub warnUnsupportedFunctions { "curandDistributionM2Shift_t", "curandDistributionM2Shift_st", "curandDirectionVectors64_t", - "curandDirectionVectorSet_t", - "curandDirectionVectorSet", "cufftXtWorkAreaPolicy_t", "cufftXtWorkAreaPolicy", "cufftXtSubFormat_t", @@ -8563,8 +8567,6 @@ sub warnUnsupportedFunctions { "CUSPARSE_ALG_MERGE_PATH", "CUSPARSE_ALG1", "CUSPARSE_ALG0", - "CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", - "CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", "CURAND_REJECTION", "CURAND_POISSON", "CURAND_ORDERING_QUASI_DEFAULT", @@ -8580,8 +8582,6 @@ sub warnUnsupportedFunctions { "CURAND_HITR", "CURAND_FAST_REJECTION", "CURAND_DISCRETE_GAUSS", - "CURAND_DIRECTION_VECTORS_64_JOEKUO6", - "CURAND_DIRECTION_VECTORS_32_JOEKUO6", "CURAND_DEVICE_API", "CURAND_DEFINITION", "CURAND_CHOOSE_BEST", diff --git a/docs/tables/CURAND_API_supported_by_HIP.md b/docs/tables/CURAND_API_supported_by_HIP.md index 7ac21b2e..3c650311 100644 --- a/docs/tables/CURAND_API_supported_by_HIP.md +++ b/docs/tables/CURAND_API_supported_by_HIP.md @@ -9,8 +9,8 @@ |`CURAND_CHOOSE_BEST`| | | | | | | | | | | |`CURAND_DEFINITION`| | | | | | | | | | | |`CURAND_DEVICE_API`| | | | | | | | | | | -|`CURAND_DIRECTION_VECTORS_32_JOEKUO6`| | | | | | | | | | | -|`CURAND_DIRECTION_VECTORS_64_JOEKUO6`| | | | | | | | | | | +|`CURAND_DIRECTION_VECTORS_32_JOEKUO6`| | | | |`HIPRAND_DIRECTION_VECTORS_32_JOEKUO6`|6.0.0| | | |6.0.0| +|`CURAND_DIRECTION_VECTORS_64_JOEKUO6`| | | | |`HIPRAND_DIRECTION_VECTORS_64_JOEKUO6`|6.0.0| | | |6.0.0| |`CURAND_DISCRETE_GAUSS`| | | | | | | | | | | |`CURAND_FAST_REJECTION`| | | | | | | | | | | |`CURAND_HITR`| | | | | | | | | | | @@ -38,8 +38,8 @@ |`CURAND_RNG_QUASI_SOBOL32`| | | | |`HIPRAND_RNG_QUASI_SOBOL32`|1.5.0| | | | | |`CURAND_RNG_QUASI_SOBOL64`| | | | |`HIPRAND_RNG_QUASI_SOBOL64`|1.5.0| | | | | |`CURAND_RNG_TEST`| | | | |`HIPRAND_RNG_TEST`|1.5.0| | | | | -|`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`| | | | | | | | | | | -|`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`| | | | | | | | | | | +|`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`| | | | |`HIPRAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`|6.0.0| | | |6.0.0| +|`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`| | | | |`HIPRAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`|6.0.0| | | |6.0.0| |`CURAND_STATUS_ALLOCATION_FAILED`| | | | |`HIPRAND_STATUS_ALLOCATION_FAILED`|1.5.0| | | | | |`CURAND_STATUS_ARCH_MISMATCH`| | | | |`HIPRAND_STATUS_ARCH_MISMATCH`|1.5.0| | | | | |`CURAND_STATUS_DOUBLE_PRECISION_REQUIRED`| | | | |`HIPRAND_STATUS_DOUBLE_PRECISION_REQUIRED`|1.5.0| | | | | @@ -53,8 +53,8 @@ |`CURAND_STATUS_SUCCESS`| | | | |`HIPRAND_STATUS_SUCCESS`|1.5.0| | | | | |`CURAND_STATUS_TYPE_ERROR`| | | | |`HIPRAND_STATUS_TYPE_ERROR`|1.5.0| | | | | |`CURAND_STATUS_VERSION_MISMATCH`| | | | |`HIPRAND_STATUS_VERSION_MISMATCH`|1.5.0| | | | | -|`curandDirectionVectorSet`| | | | | | | | | | | -|`curandDirectionVectorSet_t`| | | | | | | | | | | +|`curandDirectionVectorSet`| | | | |`hiprandDirectionVectorSet_t`|6.0.0| | | |6.0.0| +|`curandDirectionVectorSet_t`| | | | |`hiprandDirectionVectorSet_t`|6.0.0| | | |6.0.0| |`curandDirectionVectors32_t`| | | | |`hiprandDirectionVectors32_t`|1.5.0| | | | | |`curandDirectionVectors64_t`| | | | | | | | | | | |`curandDiscreteDistribution_st`| | | | |`hiprandDiscreteDistribution_st`|1.5.0| | | | | @@ -119,11 +119,11 @@ |`curandGenerateSeeds`| | | | |`hiprandGenerateSeeds`|1.5.0| | | | | |`curandGenerateUniform`| | | | |`hiprandGenerateUniform`|1.5.0| | | | | |`curandGenerateUniformDouble`| | | | |`hiprandGenerateUniformDouble`|1.5.0| | | | | -|`curandGetDirectionVectors32`| | | | | | | | | | | -|`curandGetDirectionVectors64`| | | | | | | | | | | +|`curandGetDirectionVectors32`| | | | |`hiprandGetDirectionVectors32`|6.0.0| | | |6.0.0| +|`curandGetDirectionVectors64`| | | | |`hiprandGetDirectionVectors64`|6.0.0| | | |6.0.0| |`curandGetProperty`|8.0| | | | | | | | | | -|`curandGetScrambleConstants32`| | | | | | | | | | | -|`curandGetScrambleConstants64`| | | | | | | | | | | +|`curandGetScrambleConstants32`| | | | |`hiprandGetScrambleConstants32`|6.0.0| | | |6.0.0| +|`curandGetScrambleConstants64`| | | | |`hiprandGetScrambleConstants64`|6.0.0| | | |6.0.0| |`curandGetVersion`| | | | |`hiprandGetVersion`|1.5.0| | | | | |`curandMakeMTGP32Constants`| | | | |`hiprandMakeMTGP32Constants`|1.5.0| | | | | |`curandMakeMTGP32KernelState`| | | | |`hiprandMakeMTGP32KernelState`|1.5.0| | | | | diff --git a/src/CUDA2HIP_RAND_API_functions.cpp b/src/CUDA2HIP_RAND_API_functions.cpp index 19ed9599..6e42ce87 100644 --- a/src/CUDA2HIP_RAND_API_functions.cpp +++ b/src/CUDA2HIP_RAND_API_functions.cpp @@ -40,11 +40,11 @@ const std::map CUDA_RAND_FUNCTION_MAP { {"curandGenerateSeeds", {"hiprandGenerateSeeds", "", CONV_LIB_FUNC, API_RAND, 2}}, {"curandGenerateUniform", {"hiprandGenerateUniform", "", CONV_LIB_FUNC, API_RAND, 2}}, {"curandGenerateUniformDouble", {"hiprandGenerateUniformDouble", "", CONV_LIB_FUNC, API_RAND, 2}}, - {"curandGetDirectionVectors32", {"hiprandGetDirectionVectors32", "", CONV_LIB_FUNC, API_RAND, 2, HIP_UNSUPPORTED}}, - {"curandGetDirectionVectors64", {"hiprandGetDirectionVectors64", "", CONV_LIB_FUNC, API_RAND, 2, HIP_UNSUPPORTED}}, + {"curandGetDirectionVectors32", {"hiprandGetDirectionVectors32", "", CONV_LIB_FUNC, API_RAND, 2}}, + {"curandGetDirectionVectors64", {"hiprandGetDirectionVectors64", "", CONV_LIB_FUNC, API_RAND, 2}}, {"curandGetProperty", {"hiprandGetProperty", "", CONV_LIB_FUNC, API_RAND, 2, HIP_UNSUPPORTED}}, - {"curandGetScrambleConstants32", {"hiprandGetScrambleConstants32", "", CONV_LIB_FUNC, API_RAND, 2, HIP_UNSUPPORTED}}, - {"curandGetScrambleConstants64", {"hiprandGetScrambleConstants64", "", CONV_LIB_FUNC, API_RAND, 2, HIP_UNSUPPORTED}}, + {"curandGetScrambleConstants32", {"hiprandGetScrambleConstants32", "", CONV_LIB_FUNC, API_RAND, 2}}, + {"curandGetScrambleConstants64", {"hiprandGetScrambleConstants64", "", CONV_LIB_FUNC, API_RAND, 2}}, {"curandGetVersion", {"hiprandGetVersion", "", CONV_LIB_FUNC, API_RAND, 2}}, {"curandSetGeneratorOffset", {"hiprandSetGeneratorOffset", "", CONV_LIB_FUNC, API_RAND, 2}}, {"curandSetGeneratorOrdering", {"hiprandSetGeneratorOrdering", "", CONV_LIB_FUNC, API_RAND, 2, HIP_UNSUPPORTED}}, @@ -136,6 +136,10 @@ const std::map HIP_RAND_FUNCTION_VER_MAP { {"hiprand_discrete4", {HIP_1050, HIP_0, HIP_0 }}, {"hiprand_poisson", {HIP_1050, HIP_0, HIP_0 }}, {"hiprand_poisson4", {HIP_1050, HIP_0, HIP_0 }}, + {"hiprandGetDirectionVectors32", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"hiprandGetDirectionVectors64", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"hiprandGetScrambleConstants32", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"hiprandGetScrambleConstants64", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, }; const std::map CUDA_RAND_API_SECTION_MAP { diff --git a/src/CUDA2HIP_RAND_API_types.cpp b/src/CUDA2HIP_RAND_API_types.cpp index 0e484386..8de17aa4 100644 --- a/src/CUDA2HIP_RAND_API_types.cpp +++ b/src/CUDA2HIP_RAND_API_types.cpp @@ -31,8 +31,8 @@ const std::map CUDA_RAND_TYPE_NAME_MAP { {"curandRngType_t", {"hiprandRngType_t", "", CONV_TYPE, API_RAND, 1}}, {"curandGenerator_st", {"hiprandGenerator_st", "", CONV_TYPE, API_RAND, 1}}, {"curandGenerator_t", {"hiprandGenerator_t", "", CONV_TYPE, API_RAND, 1}}, - {"curandDirectionVectorSet", {"hiprandDirectionVectorSet_t", "", CONV_TYPE, API_RAND, 1, HIP_UNSUPPORTED}}, - {"curandDirectionVectorSet_t", {"hiprandDirectionVectorSet_t", "", CONV_TYPE, API_RAND, 1, HIP_UNSUPPORTED}}, + {"curandDirectionVectorSet", {"hiprandDirectionVectorSet_t", "", CONV_TYPE, API_RAND, 1}}, + {"curandDirectionVectorSet_t", {"hiprandDirectionVectorSet_t", "", CONV_TYPE, API_RAND, 1}}, {"curandOrdering", {"hiprandOrdering_t", "", CONV_TYPE, API_RAND, 1, HIP_UNSUPPORTED}}, {"curandOrdering_t", {"hiprandOrdering_t", "", CONV_TYPE, API_RAND, 1, HIP_UNSUPPORTED}}, {"curandDistribution_st", {"hiprandDistribution_st", "", CONV_TYPE, API_RAND, 1, HIP_UNSUPPORTED}}, @@ -112,10 +112,10 @@ const std::map CUDA_RAND_TYPE_NAME_MAP { {"CURAND_ORDERING_QUASI_DEFAULT", {"HIPRAND_ORDERING_QUASI_DEFAULT", "", CONV_NUMERIC_LITERAL, API_RAND, 1, HIP_UNSUPPORTED}}, // RAND choice of direction vector set (enum curandDirectionVectorSet) - {"CURAND_DIRECTION_VECTORS_32_JOEKUO6", {"HIPRAND_DIRECTION_VECTORS_32_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1, HIP_UNSUPPORTED}}, - {"CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1, HIP_UNSUPPORTED}}, - {"CURAND_DIRECTION_VECTORS_64_JOEKUO6", {"HIPRAND_DIRECTION_VECTORS_64_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1, HIP_UNSUPPORTED}}, - {"CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1, HIP_UNSUPPORTED}}, + {"CURAND_DIRECTION_VECTORS_32_JOEKUO6", {"HIPRAND_DIRECTION_VECTORS_32_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1}}, + {"CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1}}, + {"CURAND_DIRECTION_VECTORS_64_JOEKUO6", {"HIPRAND_DIRECTION_VECTORS_64_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1}}, + {"CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", "", CONV_NUMERIC_LITERAL, API_RAND, 1}}, // RAND method (enum curandMethod) {"CURAND_CHOOSE_BEST", {"HIPRAND_CHOOSE_BEST", "", CONV_NUMERIC_LITERAL, API_RAND, 1, HIP_UNSUPPORTED}}, @@ -184,4 +184,9 @@ const std::map HIP_RAND_TYPE_NAME_VER_MAP { {"HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL32", {HIP_1050, HIP_0, HIP_0 }}, {"HIPRAND_RNG_QUASI_SOBOL64", {HIP_1050, HIP_0, HIP_0 }}, {"HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL64", {HIP_1050, HIP_0, HIP_0 }}, + {"hiprandDirectionVectorSet_t", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"HIPRAND_DIRECTION_VECTORS_32_JOEKUO6", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"HIPRAND_DIRECTION_VECTORS_64_JOEKUO6", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, + {"HIPRAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6", {HIP_6000, HIP_0, HIP_0, HIP_LATEST}}, }; diff --git a/tests/unit_tests/libraries/cuRAND/benchmark_curand_kernel.cpp b/tests/unit_tests/libraries/cuRAND/benchmark_curand_kernel.cpp index 47fc9a48..9c5b5517 100644 --- a/tests/unit_tests/libraries/cuRAND/benchmark_curand_kernel.cpp +++ b/tests/unit_tests/libraries/cuRAND/benchmark_curand_kernel.cpp @@ -297,8 +297,7 @@ struct runner CUDA_CALL(cudaMalloc((void **)&directions, size)); // CHECK: hiprandDirectionVectors32_t * h_directions; curandDirectionVectors32_t * h_directions; - // hiprandGetDirectionVectors32 and HIPRAND_DIRECTION_VECTORS_32_JOEKUO6 (of hiprandDirectionVectorSet_t) are yet unsupported by HIP - // CHECK-NOT: CURAND_CALL(hiprandGetDirectionVectors32(&h_directions, HIPRAND_DIRECTION_VECTORS_32_JOEKUO6)); + // CHECK: CURAND_CALL(hiprandGetDirectionVectors32(&h_directions, HIPRAND_DIRECTION_VECTORS_32_JOEKUO6)); CURAND_CALL(curandGetDirectionVectors32(&h_directions, CURAND_DIRECTION_VECTORS_32_JOEKUO6)); // CHECK: CUDA_CALL(hipMemcpy(directions, h_directions, size, hipMemcpyHostToDevice)); CUDA_CALL(cudaMemcpy(directions, h_directions, size, cudaMemcpyHostToDevice));