diff --git a/projects/hipcub/test/hipcub/common_test_header.hpp b/projects/hipcub/test/hipcub/common_test_header.hpp index 093bdafea540..737c29142cda 100755 --- a/projects/hipcub/test/hipcub/common_test_header.hpp +++ b/projects/hipcub/test/hipcub/common_test_header.hpp @@ -160,6 +160,8 @@ hipError_t hipMallocHelper(T** devPtr, size_t size) { if (use_hmm()) { + if (size == 0) + std::cerr << "Warning: attempting to allocate a buffer of size 0 with hipMallocManaged." << std::endl; return hipMallocManaged((void**)devPtr, size); } else diff --git a/projects/hipcub/test/hipcub/test_hipcub_device_segmented_radix_sort.hpp b/projects/hipcub/test/hipcub/test_hipcub_device_segmented_radix_sort.hpp index d32259dc4215..470410994aed 100644 --- a/projects/hipcub/test/hipcub/test_hipcub_device_segmented_radix_sort.hpp +++ b/projects/hipcub/test/hipcub/test_hipcub_device_segmented_radix_sort.hpp @@ -266,8 +266,12 @@ inline void sort_keys_empty_data() offsets[0] = 0; offsets[1] = 0; - key_type* d_keys; - HIP_CHECK(test_common_utils::hipMallocHelper(&d_keys, size * sizeof(key_type))); + key_type* d_keys = nullptr; + // hipMallocManaged will not allocate buffers of size 0. + // If HMM is enabled and size is 0, leave d_keys set to nullptr. + if (!(size == 0 && test_common_utils::use_hmm())) + HIP_CHECK(test_common_utils::hipMallocHelper(&d_keys, size * sizeof(key_type))); + HIP_CHECK(hipMemcpy(d_keys, keys_input.data(), size * sizeof(key_type), @@ -337,8 +341,9 @@ inline void sort_keys_empty_data() hipMemcpyDeviceToHost)); HIP_CHECK(hipFree(d_temporary_storage)); - HIP_CHECK(hipFree(d_keys)); HIP_CHECK(hipFree(d_offsets)); + if (d_keys) + HIP_CHECK(hipFree(d_keys)); // Output should not have changed ASSERT_NO_FATAL_FAILURE(test_utils::assert_eq(keys_output, keys_input));