Merge branch 'hardened-tests' into 'develop_stream'

Add hardened assertions in CI and fix undefined behaviour in tests Closes #778 See merge request amd/libraries/rocPRIM!700
ROCm · Jul 23, 2024 · 46faf89 · 46faf89
2 parents 39c9d9b + 4120f91
commit 46faf89
Show file tree

Hide file tree

Showing 4 changed files with 27 additions and 13 deletions.
diff --git a/.gitlab-ci.yml b/.gitlab-ci.yml
@@ -150,13 +150,20 @@ build:cmake-minimum-apt:
   extends:
     - .gpus:rocm-gpus
     - .rules:build
+  variables:
+    EXTRA_CMAKE_CXX_FLAGS: ""
   script:
     - mkdir -p $BUILD_DIR
     - cd $BUILD_DIR
+    - | # Add hardened libc++ assertions for tests only
+        if [[ $BUILD_TARGET == "TEST" ]]; then
+            echo "Configuring with hardened libc++!"
+            EXTRA_CMAKE_CXX_FLAGS+=" -D_GLIBCXX_ASSERTIONS=ON"
+        fi
     - cmake
       -G Ninja
       -D CMAKE_CXX_COMPILER="$AMDCLANG"
-      -D CMAKE_CXX_FLAGS="-Wall -Wextra -Werror"
+      -D CMAKE_CXX_FLAGS="-Wall -Wextra -Werror $EXTRA_CMAKE_CXX_FLAGS"
       -D CMAKE_BUILD_TYPE="$BUILD_TYPE"
       -D BUILD_$BUILD_TARGET=ON
       -D BUILD_EXAMPLE=ON

diff --git a/CHANGELOG.md b/CHANGELOG.md
@@ -33,6 +33,7 @@ Documentation for rocPRIM is available at
 * Fixed a bug in the generation of input data for benchmarks, which caused incorrect performance to be reported in specific cases. It may affect the reported performance for one-byte types (`uint8_t` and `int8_t`) and instantiations of `custom_type`. Specifically, device binary search, device histogram, device merge and warp sort are affected.
 * Fixed a bug for `rocprim::merge_path_search` where using `unsigned` offsets would output wrong results.
 * Fixed a bug for `rocprim::thread_load` and `rocprim::thread_store` where `float` and `double` were not casted to the correct type resulting in wrong results.
+* Fix tests failing when compiling with `-D_GLIBCXX_ASSERTIONS=ON`.
 
 ### Deprecations
 

diff --git a/test/rocprim/test_device_batch_memcpy.cpp b/test/rocprim/test_device_batch_memcpy.cpp
@@ -60,7 +60,7 @@ struct DeviceBatchMemcpyParams
 };
 
 template<class Params>
-struct DeviceBatchMemcpyTests : public ::testing::Test
+struct RocprimDeviceBatchMemcpyTests : public ::testing::Test
 {
     using value_type                      = typename Params::value_type;
     using size_type                       = typename Params::size_type;
@@ -108,9 +108,9 @@ typedef ::testing::Types<
     // size_type: int64_t
     DeviceBatchMemcpyParams<uint8_t, int64_t, true, true, 1024, 64 * 1024>,
     DeviceBatchMemcpyParams<uint8_t, int64_t, true, true, 1024, 128 * 1024>>
-    DeviceBatchMemcpyTestsParams;
+    RocprimDeviceBatchMemcpyTestsParams;
 
-TYPED_TEST_SUITE(DeviceBatchMemcpyTests, DeviceBatchMemcpyTestsParams);
+TYPED_TEST_SUITE(RocprimDeviceBatchMemcpyTests, RocprimDeviceBatchMemcpyTestsParams);
 
 // Used for generating offsets. We generate a permutation map and then derive
 // offsets via a sum scan over the sizes in the order of the permutation. This
@@ -310,7 +310,7 @@ void check_result(ContainerMemCpy& /*h_input_for_memcpy*/,
     }
 }
 
-TYPED_TEST(DeviceBatchMemcpyTests, SizeAndTypeVariation)
+TYPED_TEST(RocprimDeviceBatchMemcpyTests, SizeAndTypeVariation)
 {
     using value_type         = typename TestFixture::value_type;
     using buffer_size_type   = typename TestFixture::size_type;
@@ -349,13 +349,16 @@ TYPED_TEST(DeviceBatchMemcpyTests, SizeAndTypeVariation)
 
     auto iter = h_buffer_num_elements.begin();
 
-    iter = test_utils::generate_random_data_n(iter, num_tlev, 1, wlev_min_elems - 1, rng);
-    iter = test_utils::generate_random_data_n(iter,
-                                              num_wlev,
-                                              wlev_min_elems,
-                                              blev_min_elems - 1,
-                                              rng);
-    iter = test_utils::generate_random_data_n(iter, num_blev, blev_min_elems, max_elems, rng);
+    if(num_tlev > 0)
+        iter = test_utils::generate_random_data_n(iter, num_tlev, 1, wlev_min_elems - 1, rng);
+    if(num_wlev > 0)
+        iter = test_utils::generate_random_data_n(iter,
+                                                  num_wlev,
+                                                  wlev_min_elems,
+                                                  blev_min_elems - 1,
+                                                  rng);
+    if(num_blev > 0)
+        iter = test_utils::generate_random_data_n(iter, num_blev, blev_min_elems, max_elems, rng);
 
     const byte_offset_type total_num_elements = std::accumulate(h_buffer_num_elements.begin(),
                                                                 h_buffer_num_elements.end(),

diff --git a/test/rocprim/test_intrinsics.cpp b/test/rocprim/test_intrinsics.cpp
@@ -109,7 +109,10 @@ struct test_type_helper
     /// Initialize some random data for this type, of \p n elements and with random seed <tt>seed</tt>.
     static std::vector<T> get_random_data(size_t n, seed_type seed)
     {
-        return test_utils::get_random_data<T>(n, static_cast<T>(-100), static_cast<T>(100), seed);
+        return test_utils::get_random_data<T>(n,
+                                              test_utils::saturate_cast<T>(-100),
+                                              test_utils::saturate_cast<T>(100),
+                                              seed);
     }
 };