From 0894aa1c4aa0c769c34e1e4142247c2c831bfab7 Mon Sep 17 00:00:00 2001 From: Stanley Tsang Date: Fri, 18 Dec 2020 15:27:53 -0500 Subject: [PATCH 1/5] Adding HMM unit test option. --- CMakeLists.txt | 2 +- test/test_common.hpp | 43 ++++++++++++++++++++++ test/test_hiprand_api.cpp | 20 +++++----- test/test_hiprand_cpp_wrapper.cpp | 12 +++--- test/test_hiprand_kernel.cpp | 22 +++++------ test/test_rocrand_cpp_wrapper.cpp | 12 +++--- test/test_rocrand_generate.cpp | 6 +-- test/test_rocrand_generate_log_normal.cpp | 6 +-- test/test_rocrand_generate_normal.cpp | 6 +-- test/test_rocrand_generate_poisson.cpp | 4 +- test/test_rocrand_generate_uniform.cpp | 6 +-- test/test_rocrand_kernel_mrg32k3a.cpp | 15 ++++---- test/test_rocrand_kernel_mtgp32.cpp | 23 ++++++------ test/test_rocrand_kernel_philox4x32_10.cpp | 17 ++++----- test/test_rocrand_kernel_sobol32.cpp | 23 ++++++------ test/test_rocrand_kernel_xorwow.cpp | 17 ++++----- test/test_rocrand_mrg32k3a_prng.cpp | 25 ++++++------- test/test_rocrand_mtgp32_prng.cpp | 17 ++++----- test/test_rocrand_philox_prng.cpp | 13 +++---- test/test_rocrand_sobol32_qrng.cpp | 15 ++++---- test/test_rocrand_xorwow_prng.cpp | 13 +++---- 21 files changed, 176 insertions(+), 141 deletions(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 7cdda2188..d9b1a439a 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -64,7 +64,7 @@ endif() # Use target ID syntax if supported for AMDGPU_TARGETS if(TARGET_ID_SUPPORT) - set(AMDGPU_TARGETS gfx900:xnack-;gfx906:xnack-;gfx908:xnack- CACHE STRING "List of specific machine types for library to target") + set(AMDGPU_TARGETS gfx900:xnack-;gfx906:xnack-;gfx908:xnack-;gfx908:xnack+ CACHE STRING "List of specific machine types for library to target") else() set(AMDGPU_TARGETS gfx900;gfx906;gfx908 CACHE STRING "List of specific machine types for library to target") endif() diff --git a/test/test_common.hpp b/test/test_common.hpp index f3d00862f..9327c0ee2 100644 --- a/test/test_common.hpp +++ b/test/test_common.hpp @@ -21,9 +21,19 @@ #ifndef TEST_COMMON_HPP_ #define TEST_COMMON_HPP_ +#include + #define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) #define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#define HIP_CHECK_NON_VOID(condition) \ +{ \ + hipError_t error = condition; \ + if(error != hipSuccess){ \ + std::cout << "HIP error: " << error << " line: " << __LINE__ << std::endl; \ + exit(error); \ + } \ +} const rocrand_rng_type rng_types[] = { ROCRAND_RNG_PSEUDO_PHILOX4_32_10, ROCRAND_RNG_PSEUDO_MRG32K3A, @@ -32,4 +42,37 @@ const rocrand_rng_type rng_types[] = { ROCRAND_RNG_QUASI_SOBOL32 }; +bool supports_hmm() +{ + hipDeviceProp_t device_prop; + int device_id; + HIP_CHECK_NON_VOID(hipGetDevice(&device_id)); + HIP_CHECK_NON_VOID(hipGetDeviceProperties(&device_prop, device_id)); + if (device_prop.managedMemory == 1) return true; + + return false; +} + +bool use_hmm() +{ + return std::getenv("ROCRAND_USE_HMM"); +} + +// Helper for HMM allocations: if device supports managedMemory, and HMM is requested through +// ROCRAND_MALLOC_MANAGED environment variable +template +hipError_t hipMallocHelper(T** devPtr, size_t size) +{ + if (supports_hmm() && use_hmm()) + { + return hipMallocManaged((void**)devPtr, size); + } + else + { + return hipMalloc((void**)devPtr, size); + } + return hipSuccess; +} + + #endif // TEST_COMMON_HPP_ diff --git a/test/test_hiprand_api.cpp b/test/test_hiprand_api.cpp index 3ea2fdfe1..c5fd70f76 100644 --- a/test/test_hiprand_api.cpp +++ b/test/test_hiprand_api.cpp @@ -24,9 +24,11 @@ #include #include -#define HIP_CHECK(x) ASSERT_EQ(x, hipSuccess) +#include "test_common.hpp" #define HIPRAND_CHECK(state) ASSERT_EQ(state, HIPRAND_STATUS_SUCCESS) + + template void hiprand_generate_test_func() { @@ -36,7 +38,7 @@ void hiprand_generate_test_func() const size_t output_size = 8192; unsigned int * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(unsigned int)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -101,7 +103,7 @@ void hiprand_generate_uniform_test_func() const size_t output_size = 8192; float * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(float)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -166,7 +168,7 @@ void hiprand_generate_uniform_double_test_func() const size_t output_size = 8192; double * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(double)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -231,7 +233,7 @@ void hiprand_generate_normal_test_func() const size_t output_size = 8192; float * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(float)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -304,7 +306,7 @@ void hiprand_generate_normal_double_test_func() const size_t output_size = 8192; double * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(double)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -377,7 +379,7 @@ void hiprand_generate_lognormal_test_func() const size_t output_size = 8192; float * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(float)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -454,7 +456,7 @@ void hiprand_generate_lognormal_double_test_func() const size_t output_size = 8192; double * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(double)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -532,7 +534,7 @@ void hiprand_generate_poisson_test_func() const size_t output_size = 8192; unsigned int * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(unsigned int)) ); HIP_CHECK(hipDeviceSynchronize()); diff --git a/test/test_hiprand_cpp_wrapper.cpp b/test/test_hiprand_cpp_wrapper.cpp index c5ebf80fb..c9042eca3 100644 --- a/test/test_hiprand_cpp_wrapper.cpp +++ b/test/test_hiprand_cpp_wrapper.cpp @@ -24,7 +24,7 @@ #include #include -#define HIP_CHECK(x) ASSERT_EQ(x, hipSuccess) +#include "test_common.hpp" TEST(hiprand_cpp_wrapper, hiprand_error) { @@ -230,7 +230,7 @@ void hiprand_uniform_int_dist_template() const size_t output_size = 8192; IntType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(IntType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -287,7 +287,7 @@ void hiprand_uniform_real_dist_template() const size_t output_size = 8192; RealType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(RealType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -363,7 +363,7 @@ void hiprand_normal_dist_template() const size_t output_size = 8192; RealType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(RealType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -461,7 +461,7 @@ void hiprand_lognormal_dist_template() const size_t output_size = 8192; RealType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(RealType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -568,7 +568,7 @@ void hiprand_poisson_dist_template(const double lambda) const size_t output_size = 8192; IntType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(IntType)) ); HIP_CHECK(hipDeviceSynchronize()); diff --git a/test/test_hiprand_kernel.cpp b/test/test_hiprand_kernel.cpp index 22912338f..dead95324 100644 --- a/test/test_hiprand_kernel.cpp +++ b/test/test_hiprand_kernel.cpp @@ -26,12 +26,12 @@ #include + #define QUALIFIERS __forceinline__ __host__ __device__ #include #include -#define HIP_CHECK(x) ASSERT_EQ(x, hipSuccess) -#define HIPRAND_CHECK(state) ASSERT_EQ(state, HIPRAND_STATUS_SUCCESS) +#include "test_common.hpp" template __global__ @@ -235,7 +235,7 @@ void hiprand_kernel_h_hiprand_init_test() const size_t states_size = 256; state_type * states; - HIP_CHECK(hipMalloc((void **)&states, states_size * sizeof(state_type))); + HIP_CHECK(hipMallocHelper((void **)&states, states_size * sizeof(state_type))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -283,7 +283,7 @@ TEST(hiprand_kernel_h_philox4x32_10, hiprand_init_nvcc) const size_t states_size = 256; state_type * states; - HIP_CHECK(hipMalloc((void **)&states, states_size * sizeof(state_type))); + HIP_CHECK(hipMallocHelper((void **)&states, states_size * sizeof(state_type))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -338,7 +338,7 @@ TEST(hiprand_kernel_h_philox4x32_10, hiprand_skip_nvcc) const size_t states_size = 256; state_type * states; - HIP_CHECK(hipMalloc((void **)&states, states_size * sizeof(state_type))); + HIP_CHECK(hipMallocHelper((void **)&states, states_size * sizeof(state_type))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -380,7 +380,7 @@ void hiprand_kernel_h_hiprand_test() const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -441,7 +441,7 @@ void hiprand_kernel_h_hiprand_uniform_test() const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -502,7 +502,7 @@ void hiprand_kernel_h_hiprand_normal_test() const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -571,7 +571,7 @@ void hiprand_kernel_h_hiprand_log_normal_test() const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -644,7 +644,7 @@ void hiprand_kernel_h_hiprand_poisson_test(double lambda) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -690,7 +690,7 @@ void hiprand_kernel_h_hiprand_discrete_test(double lambda) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hiprandDiscreteDistribution_t discrete_distribution; diff --git a/test/test_rocrand_cpp_wrapper.cpp b/test/test_rocrand_cpp_wrapper.cpp index 87ebb6a08..366190403 100644 --- a/test/test_rocrand_cpp_wrapper.cpp +++ b/test/test_rocrand_cpp_wrapper.cpp @@ -24,7 +24,7 @@ #include #include -#define HIP_CHECK(x) ASSERT_EQ(x, hipSuccess) +#include "test_common.hpp" TEST(rocrand_cpp_wrapper, rocrand_error) { @@ -230,7 +230,7 @@ void rocrand_uniform_int_dist_template() const size_t output_size = 8192; IntType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(IntType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -287,7 +287,7 @@ void rocrand_uniform_real_dist_template() const size_t output_size = 8192; RealType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(RealType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -363,7 +363,7 @@ void rocrand_normal_dist_template() const size_t output_size = 8192; RealType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(RealType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -461,7 +461,7 @@ void rocrand_lognormal_dist_template() const size_t output_size = 8192; RealType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(RealType)) ); HIP_CHECK(hipDeviceSynchronize()); @@ -568,7 +568,7 @@ void rocrand_poisson_dist_template(const double lambda) const size_t output_size = 8192; IntType * output; HIP_CHECK( - hipMalloc((void **)&output, + hipMallocHelper((void **)&output, output_size * sizeof(IntType)) ); HIP_CHECK(hipDeviceSynchronize()); diff --git a/test/test_rocrand_generate.cpp b/test/test_rocrand_generate.cpp index 7a7763703..1246e79c7 100644 --- a/test/test_rocrand_generate.cpp +++ b/test/test_rocrand_generate.cpp @@ -42,7 +42,7 @@ TEST_P(rocrand_generate_tests, int_test) const size_t size = 12563; unsigned int * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -80,7 +80,7 @@ TEST_P(rocrand_generate_tests, char_test) const size_t size = 12563; unsigned char * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(unsigned char))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(unsigned char))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -118,7 +118,7 @@ TEST_P(rocrand_generate_tests, short_test) const size_t size = 12563; unsigned short * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(unsigned short))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(unsigned short))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes diff --git a/test/test_rocrand_generate_log_normal.cpp b/test/test_rocrand_generate_log_normal.cpp index 72399ab3c..dc57c261e 100644 --- a/test/test_rocrand_generate_log_normal.cpp +++ b/test/test_rocrand_generate_log_normal.cpp @@ -44,7 +44,7 @@ TEST_P(rocrand_generate_log_normal_tests, float_test) float mean = 5.0f; float stddev = 2.0f; float * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -84,7 +84,7 @@ TEST_P(rocrand_generate_log_normal_tests, double_test) double mean = 5.0; double stddev = 2.0; double * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(double))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(double))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -124,7 +124,7 @@ TEST_P(rocrand_generate_log_normal_tests, half_test) half mean = 5.0f; half stddev = 2.0f; half * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(half))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(half))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes diff --git a/test/test_rocrand_generate_normal.cpp b/test/test_rocrand_generate_normal.cpp index 65717de4c..a41b569fe 100644 --- a/test/test_rocrand_generate_normal.cpp +++ b/test/test_rocrand_generate_normal.cpp @@ -44,7 +44,7 @@ TEST_P(rocrand_generate_normal_tests, float_test) float mean = 5.0f; float stddev = 2.0f; float * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -84,7 +84,7 @@ TEST_P(rocrand_generate_normal_tests, double_test) double mean = 5.0; double stddev = 2.0; double * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(double))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(double))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -124,7 +124,7 @@ TEST_P(rocrand_generate_normal_tests, half_test) half mean = 5.0f; half stddev = 2.0f; half * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(half))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(half))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes diff --git a/test/test_rocrand_generate_poisson.cpp b/test/test_rocrand_generate_poisson.cpp index 8f22f7220..55ba56516 100644 --- a/test/test_rocrand_generate_poisson.cpp +++ b/test/test_rocrand_generate_poisson.cpp @@ -43,7 +43,7 @@ TEST_P(rocrand_generate_poisson_tests, uint_test) const size_t size = 12563; double lambda = 100.0; unsigned int * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); ROCRAND_CHECK( @@ -82,7 +82,7 @@ TEST_P(rocrand_generate_poisson_tests, out_of_range_test) const size_t size = 256; double lambda = 0.0; unsigned int * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&data, size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); EXPECT_EQ( diff --git a/test/test_rocrand_generate_uniform.cpp b/test/test_rocrand_generate_uniform.cpp index 820365124..75096318e 100644 --- a/test/test_rocrand_generate_uniform.cpp +++ b/test/test_rocrand_generate_uniform.cpp @@ -42,7 +42,7 @@ TEST_P(rocrand_generate_uniform_tests, float_test) const size_t size = 12563; float * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(float))); + HIP_CHECK(hipMallocHelper(&data, size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -80,7 +80,7 @@ TEST_P(rocrand_generate_uniform_tests, double_test) const size_t size = 12563; double * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(double))); + HIP_CHECK(hipMallocHelper(&data, size * sizeof(double))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes @@ -118,7 +118,7 @@ TEST_P(rocrand_generate_uniform_tests, half_test) const size_t size = 12563; half * data; - HIP_CHECK(hipMalloc((void **)&data, size * sizeof(half))); + HIP_CHECK(hipMallocHelper(&data, size * sizeof(half))); HIP_CHECK(hipDeviceSynchronize()); // Any sizes diff --git a/test/test_rocrand_kernel_mrg32k3a.cpp b/test/test_rocrand_kernel_mrg32k3a.cpp index 3c895fa3a..9e707fdb4 100644 --- a/test/test_rocrand_kernel_mrg32k3a.cpp +++ b/test/test_rocrand_kernel_mrg32k3a.cpp @@ -30,8 +30,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" template __global__ @@ -189,7 +188,7 @@ TEST(rocrand_kernel_mrg32k3a, rocrand) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -225,7 +224,7 @@ TEST(rocrand_kernel_mrg32k3a, rocrand_uniform) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -261,7 +260,7 @@ TEST(rocrand_kernel_mrg32k3a, rocrand_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -305,7 +304,7 @@ TEST(rocrand_kernel_mrg32k3a, rocrand_log_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -357,7 +356,7 @@ TEST_P(rocrand_kernel_mrg32k3a_poisson, rocrand_poisson) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -404,7 +403,7 @@ TEST_P(rocrand_kernel_mrg32k3a_poisson, rocrand_discrete) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); rocrand_discrete_distribution discrete_distribution; diff --git a/test/test_rocrand_kernel_mtgp32.cpp b/test/test_rocrand_kernel_mtgp32.cpp index 248fd496e..7572dffc1 100644 --- a/test/test_rocrand_kernel_mtgp32.cpp +++ b/test/test_rocrand_kernel_mtgp32.cpp @@ -30,8 +30,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" template __global__ @@ -195,13 +194,13 @@ TEST(rocrand_kernel_mtgp32, rocrand) typedef rocrand_state_mtgp32 state_type; state_type * states; - hipMalloc(&states, sizeof(state_type) * 8); + hipMallocHelper(&states, sizeof(state_type) * 8); ROCRAND_CHECK(rocrand_make_state_mtgp32(states, mtgp32dc_params_fast_11213, 8, 0)); const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -237,13 +236,13 @@ TEST(rocrand_kernel_mtgp32, rocrand_uniform) typedef rocrand_state_mtgp32 state_type; state_type * states; - hipMalloc(&states, sizeof(state_type) * 8); + hipMallocHelper(&states, sizeof(state_type) * 8); ROCRAND_CHECK(rocrand_make_state_mtgp32(states, mtgp32dc_params_fast_11213, 8, 0)); const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -279,13 +278,13 @@ TEST(rocrand_kernel_mtgp32, rocrand_normal) typedef rocrand_state_mtgp32 state_type; state_type * states; - hipMalloc(&states, sizeof(state_type) * 8); + hipMallocHelper(&states, sizeof(state_type) * 8); ROCRAND_CHECK(rocrand_make_state_mtgp32(states, mtgp32dc_params_fast_11213, 8, 0)); const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -329,13 +328,13 @@ TEST(rocrand_kernel_mtgp32, rocrand_log_normal) typedef rocrand_state_mtgp32 state_type; state_type * states; - hipMalloc(&states, sizeof(state_type) * 8); + hipMallocHelper(&states, sizeof(state_type) * 8); ROCRAND_CHECK(rocrand_make_state_mtgp32(states, mtgp32dc_params_fast_11213, 8, 0)); const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -387,13 +386,13 @@ TEST_P(rocrand_kernel_mtgp32_poisson, rocrand_poisson) const double lambda = GetParam(); state_type * states; - hipMalloc(&states, sizeof(state_type) * 8); + hipMallocHelper(&states, sizeof(state_type) * 8); ROCRAND_CHECK(rocrand_make_state_mtgp32(states, mtgp32dc_params_fast_11213, 8, 0)); const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( diff --git a/test/test_rocrand_kernel_philox4x32_10.cpp b/test/test_rocrand_kernel_philox4x32_10.cpp index 4c6eac4e3..dc1e430b3 100644 --- a/test/test_rocrand_kernel_philox4x32_10.cpp +++ b/test/test_rocrand_kernel_philox4x32_10.cpp @@ -30,8 +30,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" template __global__ @@ -217,7 +216,7 @@ TEST(rocrand_kernel_philox4x32_10, rocrand_init) const size_t states_size = 256; state_type * states; - HIP_CHECK(hipMalloc((void **)&states, states_size * sizeof(state_type))); + HIP_CHECK(hipMallocHelper((void **)&states, states_size * sizeof(state_type))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -270,7 +269,7 @@ TEST(rocrand_kernel_philox4x32_10, rocrand) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -306,7 +305,7 @@ TEST(rocrand_kernel_philox4x32_10, rocrand_uniform) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -342,7 +341,7 @@ TEST(rocrand_kernel_philox4x32_10, rocrand_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -386,7 +385,7 @@ TEST(rocrand_kernel_philox4x32_10, rocrand_log_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -438,7 +437,7 @@ TEST_P(rocrand_kernel_philox4x32_10_poisson, rocrand_poisson) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -485,7 +484,7 @@ TEST_P(rocrand_kernel_philox4x32_10_poisson, rocrand_discrete) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); rocrand_discrete_distribution discrete_distribution; diff --git a/test/test_rocrand_kernel_sobol32.cpp b/test/test_rocrand_kernel_sobol32.cpp index ab270a557..0e8422d2b 100644 --- a/test/test_rocrand_kernel_sobol32.cpp +++ b/test/test_rocrand_kernel_sobol32.cpp @@ -30,8 +30,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" template __global__ @@ -152,11 +151,11 @@ TEST(rocrand_kernel_sobol32, rocrand) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); unsigned int * m_vector; - HIP_CHECK(hipMalloc(&m_vector, sizeof(unsigned int) * 8 * 32)); + HIP_CHECK(hipMallocHelper(&m_vector, sizeof(unsigned int) * 8 * 32)); HIP_CHECK(hipMemcpy(m_vector, h_sobol32_direction_vectors, sizeof(unsigned int) * 8 * 32, hipMemcpyHostToDevice)); HIP_CHECK(hipDeviceSynchronize()); @@ -194,11 +193,11 @@ TEST(rocrand_kernel_sobol32, rocrand_uniform) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); unsigned int * m_vector; - HIP_CHECK(hipMalloc(&m_vector, sizeof(unsigned int) * 8 * 32)); + HIP_CHECK(hipMallocHelper(&m_vector, sizeof(unsigned int) * 8 * 32)); HIP_CHECK(hipMemcpy(m_vector, h_sobol32_direction_vectors, sizeof(unsigned int) * 8 * 32, hipMemcpyHostToDevice)); HIP_CHECK(hipDeviceSynchronize()); @@ -236,11 +235,11 @@ TEST(rocrand_kernel_sobol32, rocrand_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); unsigned int * m_vector; - HIP_CHECK(hipMalloc(&m_vector, sizeof(unsigned int) * 8 * 32)); + HIP_CHECK(hipMallocHelper(&m_vector, sizeof(unsigned int) * 8 * 32)); HIP_CHECK(hipMemcpy(m_vector, h_sobol32_direction_vectors, sizeof(unsigned int) * 8 * 32, hipMemcpyHostToDevice)); HIP_CHECK(hipDeviceSynchronize()); @@ -286,11 +285,11 @@ TEST(rocrand_kernel_sobol32, rocrand_log_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); unsigned int * m_vector; - HIP_CHECK(hipMalloc(&m_vector, sizeof(unsigned int) * 8 * 32)); + HIP_CHECK(hipMallocHelper(&m_vector, sizeof(unsigned int) * 8 * 32)); HIP_CHECK(hipMemcpy(m_vector, h_sobol32_direction_vectors, sizeof(unsigned int) * 8 * 32, hipMemcpyHostToDevice)); HIP_CHECK(hipDeviceSynchronize()); @@ -343,13 +342,13 @@ TEST_P(rocrand_kernel_sobol32_poisson, rocrand_poisson) const double lambda = GetParam(); unsigned int * m_vector; - HIP_CHECK(hipMalloc(&m_vector, sizeof(unsigned int) * 8 * 32)); + HIP_CHECK(hipMallocHelper(&m_vector, sizeof(unsigned int) * 8 * 32)); HIP_CHECK(hipMemcpy(m_vector, h_sobol32_direction_vectors, sizeof(unsigned int) * 8 * 32, hipMemcpyHostToDevice)); HIP_CHECK(hipDeviceSynchronize()); const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( diff --git a/test/test_rocrand_kernel_xorwow.cpp b/test/test_rocrand_kernel_xorwow.cpp index c6d868c78..8371d45b3 100644 --- a/test/test_rocrand_kernel_xorwow.cpp +++ b/test/test_rocrand_kernel_xorwow.cpp @@ -30,8 +30,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" template __global__ @@ -212,7 +211,7 @@ TEST(rocrand_kernel_xorwow, rocrand_init) const size_t states_size = 256; state_type * states; - HIP_CHECK(hipMalloc((void **)&states, states_size * sizeof(state_type))); + HIP_CHECK(hipMallocHelper((void **)&states, states_size * sizeof(state_type))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -255,7 +254,7 @@ TEST(rocrand_kernel_xorwow, rocrand) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -291,7 +290,7 @@ TEST(rocrand_kernel_xorwow, rocrand_uniform) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -327,7 +326,7 @@ TEST(rocrand_kernel_xorwow, rocrand_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -371,7 +370,7 @@ TEST(rocrand_kernel_xorwow, rocrand_log_normal) const size_t output_size = 8192; float * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(float))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(float))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -423,7 +422,7 @@ TEST_P(rocrand_kernel_xorwow_poisson, rocrand_poisson) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); hipLaunchKernelGGL( @@ -470,7 +469,7 @@ TEST_P(rocrand_kernel_xorwow_poisson, rocrand_discrete) const size_t output_size = 8192; unsigned int * output; - HIP_CHECK(hipMalloc((void **)&output, output_size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&output, output_size * sizeof(unsigned int))); HIP_CHECK(hipDeviceSynchronize()); rocrand_discrete_distribution discrete_distribution; diff --git a/test/test_rocrand_mrg32k3a_prng.cpp b/test/test_rocrand_mrg32k3a_prng.cpp index 6311d3e91..e70a71ba8 100644 --- a/test/test_rocrand_mrg32k3a_prng.cpp +++ b/test/test_rocrand_mrg32k3a_prng.cpp @@ -27,8 +27,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" using rocrand_device::detail::mad_u64_u32; @@ -57,10 +56,10 @@ TEST(rocrand_mrg32k3a_prng_tests, mad_u64_u32_test) unsigned int * y; unsigned long long * z; unsigned long long * r; - HIP_CHECK(hipMalloc((void **)&x, size * sizeof(unsigned int))); - HIP_CHECK(hipMalloc((void **)&y, size * sizeof(unsigned int))); - HIP_CHECK(hipMalloc((void **)&z, size * sizeof(unsigned long long))); - HIP_CHECK(hipMalloc((void **)&r, size * sizeof(unsigned long long))); + HIP_CHECK(hipMallocHelper((void **)&x, size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&y, size * sizeof(unsigned int))); + HIP_CHECK(hipMallocHelper((void **)&z, size * sizeof(unsigned long long))); + HIP_CHECK(hipMallocHelper((void **)&r, size * sizeof(unsigned long long))); unsigned int h_x[size]; unsigned int h_y[size]; @@ -106,7 +105,7 @@ TEST(rocrand_mrg32k3a_prng_tests, uniform_uint_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_mrg32k3a g; ROCRAND_CHECK(g.generate(data, size)); @@ -131,7 +130,7 @@ TEST(rocrand_mrg32k3a_prng_tests, uniform_float_test) { const size_t size = 1313; float * data; - hipMalloc(&data, sizeof(float) * size); + hipMallocHelper(&data, sizeof(float) * size); rocrand_mrg32k3a g; ROCRAND_CHECK(g.generate(data, size)); @@ -158,7 +157,7 @@ TEST(rocrand_mrg32k3a_prng_tests, normal_float_test) { const size_t size = 1314; float * data; - hipMalloc(&data, sizeof(float) * size); + hipMallocHelper(&data, sizeof(float) * size); rocrand_mrg32k3a g; ROCRAND_CHECK(g.generate_normal(data, size, 2.0f, 5.0f)); @@ -192,7 +191,7 @@ TEST(rocrand_mrg32k3a_prng_tests, poisson_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_mrg32k3a g; ROCRAND_CHECK(g.generate_poisson(data, size, 5.5)); @@ -230,7 +229,7 @@ TEST(rocrand_mrg32k3a_prng_tests, state_progress_test) // Device data const size_t size = 1025; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generator rocrand_mrg32k3a g0; @@ -272,7 +271,7 @@ TEST(rocrand_mrg32k3a_prng_tests, same_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_mrg32k3a g0, g1; @@ -316,7 +315,7 @@ TEST(rocrand_mrg32k3a_prng_tests, different_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper((void**)&data, sizeof(unsigned int) * size)); // Generators rocrand_mrg32k3a g0, g1; diff --git a/test/test_rocrand_mtgp32_prng.cpp b/test/test_rocrand_mtgp32_prng.cpp index 3c5ccae33..a0eebadcb 100644 --- a/test/test_rocrand_mtgp32_prng.cpp +++ b/test/test_rocrand_mtgp32_prng.cpp @@ -28,14 +28,13 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" TEST(rocrand_mtgp32_prng_tests, uniform_uint_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_mtgp32 g; ROCRAND_CHECK(g.generate(data, size)); @@ -60,7 +59,7 @@ TEST(rocrand_mtgp32_prng_tests, uniform_float_test) { const size_t size = 1313; float * data; - hipMalloc(&data, sizeof(float) * size); + hipMallocHelper(&data, sizeof(float) * size); rocrand_mtgp32 g; ROCRAND_CHECK(g.generate(data, size)); @@ -87,7 +86,7 @@ TEST(rocrand_mtgp32_prng_tests, normal_float_test) { const size_t size = 1313; float * data; - hipMalloc(&data, sizeof(float) * size); + hipMallocHelper(&data, sizeof(float) * size); rocrand_mtgp32 g; ROCRAND_CHECK(g.generate_normal(data, size, 2.0f, 5.0f)); @@ -121,7 +120,7 @@ TEST(rocrand_mtgp32_prng_tests, poisson_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_mtgp32 g; ROCRAND_CHECK(g.generate_poisson(data, size, 5.5)); @@ -159,7 +158,7 @@ TEST(rocrand_mtgp32_prng_tests, state_progress_test) // Device data const size_t size = 1025; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generator rocrand_mtgp32 g0; @@ -201,7 +200,7 @@ TEST(rocrand_mtgp32_prng_tests, same_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_mtgp32 g0, g1; @@ -245,7 +244,7 @@ TEST(rocrand_mtgp32_prng_tests, different_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_mtgp32 g0, g1; diff --git a/test/test_rocrand_philox_prng.cpp b/test/test_rocrand_philox_prng.cpp index b6ba4fa84..ccd3893dc 100644 --- a/test/test_rocrand_philox_prng.cpp +++ b/test/test_rocrand_philox_prng.cpp @@ -27,14 +27,13 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" TEST(rocrand_philox_prng_tests, uniform_uint_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * (size + 1))); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * (size + 1))); rocrand_philox4x32_10 g; ROCRAND_CHECK(g.generate(data+1, size)); @@ -59,7 +58,7 @@ TEST(rocrand_philox_prng_tests, uniform_float_test) { const size_t size = 1313; float * data; - HIP_CHECK(hipMalloc(&data, sizeof(float) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(float) * size)); rocrand_philox4x32_10 g; ROCRAND_CHECK(g.generate(data, size)); @@ -89,7 +88,7 @@ TEST(rocrand_philox_prng_tests, state_progress_test) // Device data const size_t size = 1025; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generator rocrand_philox4x32_10 g0; @@ -130,7 +129,7 @@ TEST(rocrand_philox_prng_tests, same_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_philox4x32_10 g0, g1; @@ -173,7 +172,7 @@ TEST(rocrand_philox_prng_tests, different_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_philox4x32_10 g0, g1; diff --git a/test/test_rocrand_sobol32_qrng.cpp b/test/test_rocrand_sobol32_qrng.cpp index f5f6d56d1..84ebf87c2 100644 --- a/test/test_rocrand_sobol32_qrng.cpp +++ b/test/test_rocrand_sobol32_qrng.cpp @@ -28,14 +28,13 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" TEST(rocrand_sobol32_qrng_tests, uniform_uint_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_sobol32 g; ROCRAND_CHECK(g.generate(data, size)); @@ -60,7 +59,7 @@ TEST(rocrand_sobol32_qrng_tests, uniform_float_test) { const size_t size = 1313; float * data; - HIP_CHECK(hipMalloc(&data, sizeof(float) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(float) * size)); rocrand_sobol32 g; ROCRAND_CHECK(g.generate(data, size)); @@ -87,7 +86,7 @@ TEST(rocrand_sobol32_qrng_tests, normal_float_test) { const size_t size = 1313; float * data; - HIP_CHECK(hipMalloc(&data, sizeof(float) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(float) * size)); rocrand_sobol32 g; ROCRAND_CHECK(g.generate_normal(data, size, 2.0f, 5.0f)); @@ -121,7 +120,7 @@ TEST(rocrand_sobol32_qrng_tests, poisson_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_sobol32 g; ROCRAND_CHECK(g.generate_poisson(data, size, 5.5)); @@ -156,7 +155,7 @@ TEST(rocrand_sobol32_qrng_tests, dimesions_test) { const size_t size = 12345; float * data; - HIP_CHECK(hipMalloc(&data, sizeof(float) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(float) * size)); rocrand_sobol32 g; @@ -179,7 +178,7 @@ TEST(rocrand_sobol32_qrng_tests, state_progress_test) // Device data const size_t size = 1025; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generator rocrand_sobol32 g0; diff --git a/test/test_rocrand_xorwow_prng.cpp b/test/test_rocrand_xorwow_prng.cpp index 06ec72ec0..132bb6335 100644 --- a/test/test_rocrand_xorwow_prng.cpp +++ b/test/test_rocrand_xorwow_prng.cpp @@ -27,8 +27,7 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) -#define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) +#include "test_common.hpp" TEST(rocrand_xorwow_prng_tests, init_test) { @@ -65,7 +64,7 @@ TEST(rocrand_xorwow_prng_tests, uniform_uint_test) { const size_t size = 1313; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); rocrand_xorwow g; ROCRAND_CHECK(g.generate(data, size)); @@ -90,7 +89,7 @@ TEST(rocrand_xorwow_prng_tests, uniform_float_test) { const size_t size = 1313; float * data; - hipMalloc(&data, sizeof(float) * size); + hipMallocHelper(&data, sizeof(float) * size); rocrand_xorwow g; ROCRAND_CHECK(g.generate(data, size)); @@ -120,7 +119,7 @@ TEST(rocrand_xorwow_prng_tests, state_progress_test) // Device data const size_t size = 1025; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generator rocrand_xorwow g0; @@ -162,7 +161,7 @@ TEST(rocrand_xorwow_prng_tests, same_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_xorwow g0, g1; @@ -206,7 +205,7 @@ TEST(rocrand_xorwow_prng_tests, different_seed_test) // Device side data const size_t size = 1024; unsigned int * data; - HIP_CHECK(hipMalloc(&data, sizeof(unsigned int) * size)); + HIP_CHECK(hipMallocHelper(&data, sizeof(unsigned int) * size)); // Generators rocrand_xorwow g0, g1; From 05b58a61b1563e42035a643a7d355d9594954377 Mon Sep 17 00:00:00 2001 From: Stanley Tsang Date: Thu, 28 Jan 2021 00:54:14 +0000 Subject: [PATCH 2/5] Temporarily disabling managed memory check --- test/test_common.hpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/test/test_common.hpp b/test/test_common.hpp index 9327c0ee2..4d366d873 100644 --- a/test/test_common.hpp +++ b/test/test_common.hpp @@ -63,7 +63,7 @@ bool use_hmm() template hipError_t hipMallocHelper(T** devPtr, size_t size) { - if (supports_hmm() && use_hmm()) + if (use_hmm()) { return hipMallocManaged((void**)devPtr, size); } From 508123466cecffb09dbb13ca514875240c04645b Mon Sep 17 00:00:00 2001 From: Stanley Tsang Date: Wed, 2 Jun 2021 18:02:33 -0500 Subject: [PATCH 3/5] Removing gfx908 xnack+ target --- CMakeLists.txt | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 43ce08bd1..80eae1421 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -64,7 +64,7 @@ endif() # Use target ID syntax if supported for AMDGPU_TARGETS if(TARGET_ID_SUPPORT) - set(AMDGPU_TARGETS gfx900:xnack-;gfx906:xnack-;gfx908:xnack-;gfx908:xnack+ CACHE STRING "List of specific machine types for library to target") + set(AMDGPU_TARGETS gfx900:xnack-;gfx906:xnack-;gfx908:xnack- CACHE STRING "List of specific machine types for library to target") else() set(AMDGPU_TARGETS gfx900;gfx906;gfx908 CACHE STRING "List of specific machine types for library to target") endif() From 3ba7dedee809bdbd5cce634ec94f928a434ac8f4 Mon Sep 17 00:00:00 2001 From: Stanley Tsang Date: Wed, 2 Jun 2021 19:27:28 -0500 Subject: [PATCH 4/5] Updating HMM env check --- test/test_common.hpp | 11 ++++++++++- 1 file changed, 10 insertions(+), 1 deletion(-) diff --git a/test/test_common.hpp b/test/test_common.hpp index 4d366d873..da588eeb0 100644 --- a/test/test_common.hpp +++ b/test/test_common.hpp @@ -55,7 +55,16 @@ bool supports_hmm() bool use_hmm() { - return std::getenv("ROCRAND_USE_HMM"); + if (getenv("ROCRAND_USE_HMM") == nullptr) + { + return false; + } + + if (strcmp(getenv("ROCRAND_USE_HMM"), "1") == 0) + { + return true; + } + return false; } // Helper for HMM allocations: if device supports managedMemory, and HMM is requested through From 0e2fd40fde088474ffd82c5b76bcaf780bab3a7d Mon Sep 17 00:00:00 2001 From: Stanley Tsang Date: Wed, 2 Jun 2021 20:29:53 -0500 Subject: [PATCH 5/5] Updating comment --- test/test_common.hpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/test/test_common.hpp b/test/test_common.hpp index da588eeb0..68c30e71a 100644 --- a/test/test_common.hpp +++ b/test/test_common.hpp @@ -67,8 +67,8 @@ bool use_hmm() return false; } -// Helper for HMM allocations: if device supports managedMemory, and HMM is requested through -// ROCRAND_MALLOC_MANAGED environment variable +// Helper for HMM allocations: if HMM is requested through +// setting environment variable ROCRAND_USE_HMM=1 template hipError_t hipMallocHelper(T** devPtr, size_t size) {