diff --git a/projects/hiprand/CHANGELOG.md b/projects/hiprand/CHANGELOG.md index f6b64753afa..a37ef929066 100644 --- a/projects/hiprand/CHANGELOG.md +++ b/projects/hiprand/CHANGELOG.md @@ -3,6 +3,13 @@ Documentation for hipRAND is available at [https://rocm.docs.amd.com/projects/hipRAND/en/latest/](https://rocm.docs.amd.com/projects/hipRAND/en/latest/). +## hipRAND 3.1.0 for ROCm 7.1 + +### Resolved issues + +* Updated error handling for several hipRAND unit tests to accomodate the new hipGetLastError behaviour that was introduced in ROCm 7.0. +As of ROCm 7.0, the internal error state is cleared on each call to `hipGetLastError` rather than on every HIP API call. + ## hipRAND 3.0.0 for ROCm 7.0 ### Added diff --git a/projects/hiprand/CMakeLists.txt b/projects/hiprand/CMakeLists.txt index 8230fe2daf7..989b443df43 100644 --- a/projects/hiprand/CMakeLists.txt +++ b/projects/hiprand/CMakeLists.txt @@ -41,7 +41,7 @@ endif() # hipRAND project # project(hipRAND CXX) -set(hipRAND_VERSION "3.0.0") +set(hipRAND_VERSION "3.1.0") # Build options option(BUILD_ADDRESS_SANITIZER "Build with address sanitizer enabled" OFF) diff --git a/projects/hiprand/test/test_common.hpp b/projects/hiprand/test/test_common.hpp index 8a40590ad31..f5d7f8ff9f3 100644 --- a/projects/hiprand/test/test_common.hpp +++ b/projects/hiprand/test/test_common.hpp @@ -25,7 +25,21 @@ #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) +// GoogleTest-compatible HIP_CHECK macro. FAIL is called to log the Google Test trace. +// The lambda is invoked immediately as assertions that generate a fatal failure can +// only be used in void-returning functions. +#define HIP_CHECK(condition) \ + { \ + hipError_t error = condition; \ + if(error != hipSuccess) \ + { \ + [error]() \ + { FAIL() << "HIP error " << error << ": " << hipGetErrorString(error); }(); \ + exit(error); \ + } \ + } + +#define HIPRAND_CHECK(state) ASSERT_EQ(state, HIPRAND_STATUS_SUCCESS) #define HIP_CHECK_NON_VOID(condition) \ { \ diff --git a/projects/hiprand/test/test_hiprand_api.cpp b/projects/hiprand/test/test_hiprand_api.cpp index 44afb687793..2280c10550a 100644 --- a/projects/hiprand/test/test_hiprand_api.cpp +++ b/projects/hiprand/test/test_hiprand_api.cpp @@ -26,8 +26,6 @@ #include -#define HIPRAND_CHECK(state) ASSERT_EQ(state, HIPRAND_STATUS_SUCCESS) - constexpr hiprandRngType_t hiprand_rng_types[] = {HIPRAND_RNG_PSEUDO_XORWOW, HIPRAND_RNG_PSEUDO_MRG32K3A, HIPRAND_RNG_PSEUDO_MTGP32, @@ -570,7 +568,7 @@ INSTANTIATE_TEST_SUITE_P(hiprand, hiprand_ordering, ::testing::ValuesIn(hiprand_ void hiprand_generate_uniform_host_test_func(hiprandRngType_t rng_type) { hiprandGenerator_t generator = 0; - HIP_CHECK(hiprandCreateGeneratorHost(&generator, rng_type)); + HIPRAND_CHECK(hiprandCreateGeneratorHost(&generator, rng_type)); constexpr size_t output_size = 8192; std::vector output_host(output_size); diff --git a/projects/hiprand/test/test_hiprand_kernel.cpp b/projects/hiprand/test/test_hiprand_kernel.cpp index 51bf3750e62..f8eb32180f0 100644 --- a/projects/hiprand/test/test_hiprand_kernel.cpp +++ b/projects/hiprand/test/test_hiprand_kernel.cpp @@ -95,7 +95,7 @@ void load_sobol_vectors_to_gpu(const unsigned int dimensions, constexpr unsigned int vec_size = sizeof(T) * 8; - HIP_CHECK(get_sobol_directions(&h_directions)); + HIPRAND_CHECK(get_sobol_directions(&h_directions)); HIP_CHECK(hipMallocHelper(reinterpret_cast(direction_vectors), sizeof(T) * dimensions * vec_size)); @@ -114,7 +114,7 @@ void load_scrambled_sobol_constants_and_vectors_to_gpu(const unsigned int d const T* h_constants; - HIP_CHECK(get_sobol_constants(&h_constants)); + HIPRAND_CHECK(get_sobol_constants(&h_constants)); HIP_CHECK( hipMallocHelper(reinterpret_cast(scramble_constants), sizeof(T) * dimensions)); @@ -765,10 +765,10 @@ void hiprand_kernel_h_hiprand_mtgp_init_test() mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); HIP_CHECK(hipFree(states)); @@ -997,10 +997,10 @@ void hiprand_kernel_h_hiprand_mtgp_test() mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); const size_t output_size = 8192; @@ -1155,10 +1155,10 @@ void hiprand_kernel_h_hiprand_uniform_mtgp_test() mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); const size_t output_size = 8192; @@ -1328,10 +1328,10 @@ void hiprand_kernel_h_hiprand_normal_mtgp_test() mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); const size_t output_size = 8192; @@ -1522,10 +1522,10 @@ void hiprand_kernel_h_hiprand_log_normal_mtgp_test() mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); const size_t output_size = 8192; @@ -1829,10 +1829,10 @@ void hiprand_kernel_h_hiprand_poisson_mtgp_test(double lambda) mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); const size_t output_size = 8192; @@ -1990,10 +1990,10 @@ void hiprand_kernel_h_hiprand_discrete_mtgp_test(double lambda) mtgp32_kernel_params_t* k; HIP_CHECK(hipMallocHelper((void**)&k, states_size * sizeof(mtgp32_kernel_params_t))); - HIP_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); + HIPRAND_CHECK(hiprandMakeMTGP32Constants(mtgp32dc_params_fast_11213, k)); unsigned long long seed = 0; - HIP_CHECK( + HIPRAND_CHECK( hiprandMakeMTGP32KernelState(states, mtgp32dc_params_fast_11213, k, states_size, seed)); const size_t output_size = 8192; diff --git a/projects/rocrand/CHANGELOG.md b/projects/rocrand/CHANGELOG.md index d6d4d9204e1..be562e63fb9 100644 --- a/projects/rocrand/CHANGELOG.md +++ b/projects/rocrand/CHANGELOG.md @@ -3,6 +3,13 @@ Documentation for rocRAND is available at [https://rocm.docs.amd.com/projects/rocRAND/en/latest/](https://rocm.docs.amd.com/projects/rocRAND/en/latest/) +## rocRAND 4.1.0 for ROCm 7.1 + +### Resolved issues + +* Updated error handling for several rocRAND unit tests to accomodate the new hipGetLastError behaviour that was introduced in ROCm 7.0. +As of ROCm 7.0, the internal error state is cleared on each call to `hipGetLastError` rather than on every HIP API call. + ## rocRAND 4.0.0 for ROCm 7.0 ### Added diff --git a/projects/rocrand/CMakeLists.txt b/projects/rocrand/CMakeLists.txt index 57309c33e1d..9e6319a9494 100644 --- a/projects/rocrand/CMakeLists.txt +++ b/projects/rocrand/CMakeLists.txt @@ -178,7 +178,7 @@ if(BUILD_ADDRESS_SANITIZER AND BUILD_SHARED_LIBS) endif() # Set version variables -rocm_setup_version( VERSION "4.0.0" ) +rocm_setup_version( VERSION "4.1.0" ) set ( rocrand_VERSION ${rocRAND_VERSION} ) # Old-style version number used within the library's API. rocrand_get_version should be modified. math(EXPR rocrand_VERSION_NUMBER "${rocRAND_VERSION_MAJOR} * 100000 + ${rocRAND_VERSION_MINOR} * 100 + ${rocRAND_VERSION_PATCH}") diff --git a/projects/rocrand/test/internal/test_rocrand_config_dispatch.cpp b/projects/rocrand/test/internal/test_rocrand_config_dispatch.cpp index 1d1b45cdada..97bedd2d35c 100644 --- a/projects/rocrand/test/internal/test_rocrand_config_dispatch.cpp +++ b/projects/rocrand/test/internal/test_rocrand_config_dispatch.cpp @@ -121,11 +121,11 @@ TEST(rocrand_config_dispatch_tests, get_config_on_host_and_device) HIP_CHECK(hipMallocHelper(&d_grid_size, sizeof(*d_grid_size))); rocrand_impl::host::generator_config config{}; - const hipError_t error = rocrand_impl::host::get_generator_config( + const hipError_t err = rocrand_impl::host::get_generator_config( stream, ROCRAND_ORDERING_PSEUDO_DEFAULT, config); - HIP_CHECK(error); + HIP_CHECK(err); hipLaunchKernelGGL(write_config, dim3(config.blocks), diff --git a/projects/rocrand/test/internal/test_rocrand_generator_type.cpp b/projects/rocrand/test/internal/test_rocrand_generator_type.cpp index 64b50750d97..ebf968ce182 100644 --- a/projects/rocrand/test/internal/test_rocrand_generator_type.cpp +++ b/projects/rocrand/test/internal/test_rocrand_generator_type.cpp @@ -29,7 +29,19 @@ #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) +// GoogleTest-compatible HIP_CHECK macro. FAIL is called to log the Google Test trace. +// The lambda is invoked immediately as assertions that generate a fatal failure can +// only be used in void-returning functions. +#define HIP_CHECK(condition) \ + { \ + hipError_t error = condition; \ + if(error != hipSuccess) \ + { \ + [error]() \ + { FAIL() << "HIP error " << error << ": " << hipGetErrorString(error); }(); \ + exit(error); \ + } \ + } using namespace rocrand_impl::host; diff --git a/projects/rocrand/test/test_common.hpp b/projects/rocrand/test/test_common.hpp index 427d8141a59..1698213d674 100644 --- a/projects/rocrand/test/test_common.hpp +++ b/projects/rocrand/test/test_common.hpp @@ -30,7 +30,19 @@ #include #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) +// GoogleTest-compatible HIP_CHECK macro. FAIL is called to log the Google Test trace. +// The lambda is invoked immediately as assertions that generate a fatal failure can +// only be used in void-returning functions. +#define HIP_CHECK(condition) \ + { \ + hipError_t error = condition; \ + if(error != hipSuccess) \ + { \ + [error]() \ + { FAIL() << "HIP error " << error << ": " << hipGetErrorString(error); }(); \ + exit(error); \ + } \ + } #define HIP_CHECK_NON_VOID(condition) \ { \ diff --git a/projects/rocrand/test/test_rocrand_kernel_sobol64.cpp b/projects/rocrand/test/test_rocrand_kernel_sobol64.cpp index 5779ce2f50b..2b05da628ce 100644 --- a/projects/rocrand/test/test_rocrand_kernel_sobol64.cpp +++ b/projects/rocrand/test/test_rocrand_kernel_sobol64.cpp @@ -29,7 +29,20 @@ #include -#define HIP_CHECK(state) ASSERT_EQ(state, hipSuccess) +// GoogleTest-compatible HIP_CHECK macro. FAIL is called to log the Google Test trace. +// The lambda is invoked immediately as assertions that generate a fatal failure can +// only be used in void-returning functions. +#define HIP_CHECK(condition) \ + { \ + hipError_t error = condition; \ + if(error != hipSuccess) \ + { \ + [error]() \ + { FAIL() << "HIP error " << error << ": " << hipGetErrorString(error); }(); \ + exit(error); \ + } \ + } + #define ROCRAND_CHECK(state) ASSERT_EQ(state, ROCRAND_STATUS_SUCCESS) template