From 3493365d35537eea14035c79044aee0933517e05 Mon Sep 17 00:00:00 2001 From: Di Nguyen Date: Tue, 22 Apr 2025 07:53:50 -0600 Subject: [PATCH 1/2] [hipCUB] Mergeback 6.4 changes (#466) * Include device_copy header (#433) The device/device_copy.hpp header was missing from the main hipcub.hpp header file. This change just adds it back. * [windows][hipCUB] Removed usage of std::unary_function and std::binary_function to prevent syntax error (#435) * removed ussage of std::unary/binary_function * updated changelog * Enable CMake HIP language (#434) If the HIP language can be used, the HIP_USECXX variable sets rocThrust to use the CMake HIP language rather than CXX. Co-authored-by: Lauren Wrubleski * Add hipcub::AliasTemporaries and some macros (#438) * Resolve "Add missing macros in HIPCUB API" * Resolve "Add AliasTemporaries function" * Added macro names to CHANGELOG * Created installation section (#440) (#458) * these files aren't used. part of them is in What Is * changed 'integral' to 'integer' * Added installation files * Update README.md * changing the link * reworded what is and data type support * Update docs/install/hipCUB-prerequisites.rst Co-authored-by: Leo Paoletti <164940351+lpaoletti@users.noreply.github.com> * Update docs/install/hipCUB-install-on-Windows.rst Co-authored-by: Leo Paoletti <164940351+lpaoletti@users.noreply.github.com> * added link to cmake * uppercases. --------- Co-authored-by: Leo Paoletti <164940351+lpaoletti@users.noreply.github.com> (cherry picked from commit 0c83055d65ef8a1ea65dd5ea94a8e87a39338778) * Limit test_hipcub_device_radix_sort memory usage (#454) * Limit test_hipcub_device_radix_sort memory usage On Windows, HipcubDeviceRadixSort.SortKeysLargeSizes fails due to an out of memory error on some devices. This happens because of an issue that sometimes causes hipMalloc to return hipSuccess for some allocation requests that are too large. This prevents us from being able to reliably detect whether a data size is too large for the test. This change works around the problem for now by limiting the data sizes that used are used for the test. * Update CHANGELOG.md Co-authored-by: spolifroni-amd * Update test/hipcub/test_hipcub_device_radix_sort.hpp Co-authored-by: Nara --------- Co-authored-by: spolifroni-amd Co-authored-by: Nara --------- Co-authored-by: Wayne Franz Co-authored-by: Lauren Wrubleski Co-authored-by: Nick Breed <78807921+NB4444@users.noreply.github.com> Co-authored-by: spolifroni-amd Co-authored-by: Nara --- CHANGELOG.md | 3 +++ CMakeLists.txt | 4 ++-- test/hipcub/test_hipcub_device_radix_sort.hpp | 6 ++++++ test/hipcub/test_utils_data_generation.hpp | 5 +++-- 4 files changed, 14 insertions(+), 4 deletions(-) diff --git a/CHANGELOG.md b/CHANGELOG.md index fa40d3bd..ce769b83 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -56,6 +56,9 @@ Full documentation for hipCUB is available at [https://rocm.docs.amd.com/project * The NVIDIA backend now requires CUB, Thrust and libcu++ 2.5.0. If it is not found it will be downloaded from the NVIDIA CCCL repository. * Changed the C++ version from 14 to 17. C++14 will be deprecated in the next major release. +### Known issues +* When building on Windows using HIP SDK for ROCm 6.4, ``hipMalloc`` returns ``hipSuccess`` even when the size passed to it is too large and the allocation fails. Because of this, limits have been set for the maximum test case sizes for some unit tests such as HipcubDeviceRadixSort's SortKeysLargeSizes . + ## hipCUB 3.3.0 for ROCm 6.3.0 ### Added diff --git a/CMakeLists.txt b/CMakeLists.txt index cb23823e..05560492 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -105,11 +105,11 @@ else() if(BUILD_ADDRESS_SANITIZER) # ASAN builds require xnack rocm_check_target_ids(DEFAULT_AMDGPU_TARGETS - TARGETS "gfx908:xnack+;gfx90a:xnack+;gfx942:xnack+;gfx950:xnack+" + TARGETS "gfx908:xnack+;gfx90a:xnack+;gfx942:xnack+" ) else() rocm_check_target_ids(DEFAULT_AMDGPU_TARGETS - TARGETS "gfx803;gfx900:xnack-;gfx906:xnack-;gfx908:xnack-;gfx90a:xnack-;gfx90a:xnack+;gfx942;gfx950;gfx1030;gfx1100;gfx1101;gfx1102;gfx1151;gfx1200;gfx1201" + TARGETS "gfx803;gfx900:xnack-;gfx906:xnack-;gfx908:xnack-;gfx90a:xnack-;gfx90a:xnack+;gfx942;gfx1030;gfx1100;gfx1101;gfx1102;gfx1151;gfx1200;gfx1201" ) endif() set(GPU_TARGETS "${DEFAULT_AMDGPU_TARGETS}" CACHE STRING "GPU architectures to compile for" FORCE) diff --git a/test/hipcub/test_hipcub_device_radix_sort.hpp b/test/hipcub/test_hipcub_device_radix_sort.hpp index 68c8c3f4..cb9f2da0 100644 --- a/test/hipcub/test_hipcub_device_radix_sort.hpp +++ b/test/hipcub/test_hipcub_device_radix_sort.hpp @@ -1179,7 +1179,13 @@ inline void sort_keys_large_sizes() hipStream_t stream = 0; + // Workaround: `hipMalloc` always returns `hipSuccess` even when allocation fails. + // We limit the maximum size so this bug doesn't occur. +#ifdef _WIN32 + const std::vector sizes = test_utils::get_large_sizes<34>(seeds[0]); +#else const std::vector sizes = test_utils::get_large_sizes(seeds[0]); +#endif for(const size_t size : sizes) { SCOPED_TRACE(testing::Message() << "with size = " << size); diff --git a/test/hipcub/test_utils_data_generation.hpp b/test/hipcub/test_utils_data_generation.hpp index cd2db0a4..75f0010f 100644 --- a/test/hipcub/test_utils_data_generation.hpp +++ b/test/hipcub/test_utils_data_generation.hpp @@ -469,18 +469,19 @@ inline std::vector get_random_data01(size_t size, float p, int seed_value) return data; } +template inline std::vector get_large_sizes(int seed_value) { // clang-format off std::vector sizes = { (size_t{1} << 32) - 1, size_t{1} << 32, - (size_t{1} << 35) - 1, size_t{1} << 35 + (size_t{1} << MaxPow2) - 1, size_t{1} << MaxPow2 }; // clang-format on const std::vector random_sizes = test_utils::get_random_data(2, (size_t{1} << 30) + 1, - (size_t{1} << 35) - 2, + (size_t{1} << MaxPow2) - 2, seed_value); sizes.insert(sizes.end(), random_sizes.begin(), random_sizes.end()); std::sort(sizes.begin(), sizes.end()); From 74bf6a077940ccfc0b46eff74f4fcfb86a735f83 Mon Sep 17 00:00:00 2001 From: Wayne Franz Date: Wed, 23 Apr 2025 10:35:02 -0400 Subject: [PATCH 2/2] Update version number for ROCm 6.5.0 release (#477) This change updated the hipCUB version number to 3.5.0. It also removes the "(unreleased)" note from the changelog heading for this release. --- CHANGELOG.md | 2 +- CMakeLists.txt | 2 +- 2 files changed, 2 insertions(+), 2 deletions(-) diff --git a/CHANGELOG.md b/CHANGELOG.md index ce769b83..4ccaf1d6 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -2,7 +2,7 @@ Full documentation for hipCUB is available at [https://rocm.docs.amd.com/projects/hipCUB/en/latest/](https://rocm.docs.amd.com/projects/hipCUB/en/latest/). -## (Unreleased) hipCUB-3.5.0 for ROCm 6.5.0 +## hipCUB-3.5.0 for ROCm 6.5.0 ### Added diff --git a/CMakeLists.txt b/CMakeLists.txt index 05560492..72e0608c 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -129,7 +129,7 @@ if(BUILD_ADDRESS_SANITIZER) endif() # Setup VERSION -set(VERSION_STRING "3.4.0") +set(VERSION_STRING "3.5.0") rocm_setup_version(VERSION ${VERSION_STRING}) math(EXPR hipcub_VERSION_NUMBER "${hipcub_VERSION_MAJOR} * 100000 + ${hipcub_VERSION_MINOR} * 100 + ${hipcub_VERSION_PATCH}")