diff --git a/projects/rocthrust/CHANGELOG.md b/projects/rocthrust/CHANGELOG.md index ebe8d247a7d..8e269efe406 100644 --- a/projects/rocthrust/CHANGELOG.md +++ b/projects/rocthrust/CHANGELOG.md @@ -10,6 +10,25 @@ Documentation for rocThrust available at * Updated the required version of Google Benchmark from 1.8.0 to 1.9.0. * Drop `c++14` support for rocthrust. * Renamed `cpp14_required.h` to `cpp_version_check.h` +* Refactored `test_header.hpp` into separte modules `test_param_fixtures.hpp`, `test_real_assertions.hpp`, `test_imag_assertions.hpp`, and `test_utils.hpp`. + * This is done to prevent unit tests from having access to modules that they're not testing. This will improve the accuracy of code coverage reports. + +### Added +* Additional unit tests for: + * binary_search + * complex + * c99math + * catrig + * ccosh + * cexp + * clog + * csin + * csqrt + * ctan +* Added `test_param_fixtures.hpp` to store all the parameters for typed test suites. +* Added `test_real_assertions.hpp` to handle unit test assertions for real numbers. +* Added `test_imag_assertions.hpp` to handle unit test assertions for imaginary numbers. +* `clang++` is now used to compile google benchmarks on Windows. * Added gfx950 support. * Merged changes from upstream CCCL/thrust 2.6.0 @@ -17,6 +36,8 @@ Documentation for rocThrust available at * `device_malloc_allocator.h` has been removed. This header file was unused and should not impact users. * Removed C++14 support, only C++17 is supported. +* `test_header.hpp` has been removed. The `HIP_CHECK` function, as well as the `test` and `inter_run_bwr` namespaces, have been moved to `test_utils.hpp`. +* `test_assertions.hpp` has been split into `test_real_assertions.hpp` and `test_imag_assertions.hpp`. ### Upcoming changes diff --git a/projects/rocthrust/test/CMakeLists.txt b/projects/rocthrust/test/CMakeLists.txt index 62ee5974fb5..faf30056c2a 100644 --- a/projects/rocthrust/test/CMakeLists.txt +++ b/projects/rocthrust/test/CMakeLists.txt @@ -135,6 +135,7 @@ if(BUILD_TEST) add_rocthrust_test("allocator") add_rocthrust_test("allocator_aware_policies") add_rocthrust_test("async_copy") + add_rocthrust_test("async_for_each") add_rocthrust_test("async_reduce") add_rocthrust_test("async_scan") add_rocthrust_test("async_sort") @@ -145,6 +146,7 @@ if(BUILD_TEST) add_rocthrust_test("binary_search_vector_descending") add_rocthrust_test("complex") add_rocthrust_test("complex_transform") + add_rocthrust_test("complex_various") add_rocthrust_test("constant_iterator") add_rocthrust_test("copy") add_rocthrust_test("copy_n") diff --git a/projects/rocthrust/test/test_adjacent_difference.cpp b/projects/rocthrust/test/test_adjacent_difference.cpp index 2d481b9e277..f56e4c0d5bf 100644 --- a/projects/rocthrust/test/test_adjacent_difference.cpp +++ b/projects/rocthrust/test/test_adjacent_difference.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" #if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP diff --git a/projects/rocthrust/test/test_advance.cpp b/projects/rocthrust/test/test_advance.cpp index 821434b334b..538c9bec498 100644 --- a/projects/rocthrust/test/test_advance.cpp +++ b/projects/rocthrust/test/test_advance.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(AdvanceVectorTests, VectorSignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_allocator.cpp b/projects/rocthrust/test/test_allocator.cpp index 34836a2225d..0cbd1cebd9a 100644 --- a/projects/rocthrust/test/test_allocator.cpp +++ b/projects/rocthrust/test/test_allocator.cpp @@ -27,7 +27,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" // The `thrust::device_malloc_allocator` is deprecated, considered removed and is only // here for compatibility, once it is removed, the `THRUST_SUPPRESS_DEPRECATED_PUSH` diff --git a/projects/rocthrust/test/test_allocator_aware_policies.cpp b/projects/rocthrust/test/test_allocator_aware_policies.cpp index 2be3de233b2..d599df8ff89 100644 --- a/projects/rocthrust/test/test_allocator_aware_policies.cpp +++ b/projects/rocthrust/test/test_allocator_aware_policies.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template struct test_allocator_t diff --git a/projects/rocthrust/test/test_async_copy.cpp b/projects/rocthrust/test/test_async_copy.cpp index 90060882c45..8ce8fa7db38 100644 --- a/projects/rocthrust/test/test_async_copy.cpp +++ b/projects/rocthrust/test/test_async_copy.cpp @@ -22,7 +22,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(AsyncCopyTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_async_for_each.cpp b/projects/rocthrust/test/test_async_for_each.cpp index 385457bb7ee..544a691f537 100644 --- a/projects/rocthrust/test/test_async_for_each.cpp +++ b/projects/rocthrust/test/test_async_for_each.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" #define DEFINE_ASYNC_FOR_EACH_CALLABLE(name, ...) \ struct THRUST_PP_CAT2(name, _fn) \ @@ -58,10 +59,9 @@ void test_async_for_each() for (auto size : get_sizes()) { SCOPED_TRACE(testing::Message() << "with size = " << size); - for (size_t seed_index = 0; seed_index < random_seeds_count + seed_size; seed_index++) + for (auto seed : get_seeds()) { - unsigned int seed_value = seed_index < random_seeds_count ? rand() : seeds[seed_index - random_seeds_count]; - thrust::host_vector h0_data = get_random_data(size, T(-1000), T(1000), seed_value); + thrust::host_vector h0_data = get_random_data(size, T(-1000), T(1000), seed); thrust::device_vector d0_data(h0_data); thrust::for_each(h0_data.begin(), h0_data.end(), UnaryFunction{}); diff --git a/projects/rocthrust/test/test_async_reduce.cpp b/projects/rocthrust/test/test_async_reduce.cpp index 5308fef48fd..0a435e327a9 100644 --- a/projects/rocthrust/test/test_async_reduce.cpp +++ b/projects/rocthrust/test/test_async_reduce.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(AsyncReduceTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_async_scan.cpp b/projects/rocthrust/test/test_async_scan.cpp index 20018c562c6..e01dee7a879 100644 --- a/projects/rocthrust/test/test_async_scan.cpp +++ b/projects/rocthrust/test/test_async_scan.cpp @@ -20,7 +20,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(AsyncScanTests, NumericalTestsParams) diff --git a/projects/rocthrust/test/test_async_sort.cpp b/projects/rocthrust/test/test_async_sort.cpp index 021b297d387..25410a266bc 100644 --- a/projects/rocthrust/test/test_async_sort.cpp +++ b/projects/rocthrust/test/test_async_sort.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(AsyncSortTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_async_transform.cpp b/projects/rocthrust/test/test_async_transform.cpp index 07696b2b6d9..00b58acb177 100644 --- a/projects/rocthrust/test/test_async_transform.cpp +++ b/projects/rocthrust/test/test_async_transform.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template struct divide_by_2 diff --git a/projects/rocthrust/test/test_binary_search.cpp b/projects/rocthrust/test/test_binary_search.cpp index 67cccadea7c..350a4ebf0d4 100644 --- a/projects/rocthrust/test/test_binary_search.cpp +++ b/projects/rocthrust/test/test_binary_search.cpp @@ -22,9 +22,11 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" -TESTS_DEFINE(BinarySearchTestsInKernel, NumericalTestsParams); +TESTS_DEFINE(SingleValueTests, NumericalTestsParams); template struct init_scalar @@ -44,170 +46,462 @@ struct init_tuple } }; -struct custom_less +template +bool host_compare(const T& lhs, const T& rhs) +{ + return lhs < rhs; +} + +template +__device__ bool device_compare(const T& lhs, const T& rhs) +{ + return lhs < rhs; +} + +template +__global__ THRUST_HIP_LAUNCH_BOUNDS_DEFAULT void +single_value_kernel(T* device_search, T* device_input, size_t* device_output, const size_t N, DeviceFunc f) { - template - __device__ inline bool operator()(T a, T b) + constexpr size_t items_per_block = items_per_thread * block_size; + const size_t offset = (blockIdx.x * items_per_block) + (threadIdx.x * items_per_thread); + + for (size_t i = 0; i < items_per_thread; i++) { - return a < b; + device_output[offset + i] = f(device_search, device_search + N, device_input[offset + i]); } -}; +} -template -__global__ THRUST_HIP_LAUNCH_BOUNDS_DEFAULT void lower_bound_kernel(size_t n, T* input, ptrdiff_t* output) +template +void RunSingleValueTest(const ExpectedFunction& ef, const ThrustDeviceFunction& df, const ThrustHostFunction& hf) { - output[0] = thrust::lower_bound(thrust::device, input, input + n, T(0), custom_less()) - input; - output[1] = thrust::lower_bound(thrust::device, input, input + n, T(1)) - input; - output[2] = thrust::lower_bound(thrust::device, input, input + n, T(2)) - input; - output[3] = thrust::lower_bound(thrust::device, input, input + n, T(3)) - input; - output[4] = thrust::lower_bound(thrust::device, input, input + n, T(4), custom_less()) - input; - output[5] = thrust::lower_bound(thrust::device, input, input + n, T(5)) - input; - output[6] = thrust::lower_bound(thrust::device, input, input + n, T(6)) - input; - output[7] = thrust::lower_bound(thrust::device, input, input + n, T(7)) - input; - output[8] = thrust::lower_bound(thrust::device, input, input + n, T(8)) - input; - output[9] = thrust::lower_bound(thrust::device, input, input + n, T(9), custom_less()) - input; + constexpr size_t grid_size = 1024; + constexpr size_t items_per_thread = 12; + constexpr size_t block_size = 32; + constexpr size_t items_per_block = items_per_thread * block_size; + constexpr size_t size = items_per_block * grid_size; + + double maxi = static_cast(std::numeric_limits::max()); + double mini = static_cast(std::numeric_limits::min()); + + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution dis(mini, maxi); + + T* host_search = new T[size]; + T* host_input = new T[size]; + for (size_t i = 0; i < size; i++) + { + host_search[i] = static_cast(dis(gen)); + host_input[i] = static_cast(dis(gen)); + } + + std::sort(host_search, host_search + size); + + size_t* host_expected = new size_t[size]; + for (size_t i = 0; i < size; i++) + { + host_expected[i] = ef(host_search, host_search + size, host_input[i]); + } + + size_t* host_thrust_output = new size_t[size]; + for (size_t i = 0; i < size; i++) + { + host_thrust_output[i] = hf(host_search, host_search + size, host_input[i]); + } + + T* device_search; + T* device_input; + HIP_CHECK(hipMalloc(&device_search, sizeof(T) * size)); + HIP_CHECK(hipMalloc(&device_input, sizeof(size_t) * size)); + HIP_CHECK(hipMemcpy(device_search, host_search, sizeof(T) * size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(device_input, host_input, sizeof(T) * size, hipMemcpyHostToDevice)); + + size_t* device_output; + HIP_CHECK(hipMalloc(&device_output, sizeof(size_t) * size)); + + hipLaunchKernelGGL( + HIP_KERNEL_NAME(single_value_kernel), + dim3(grid_size), + dim3(block_size), + 0, + 0, + device_search, + device_input, + device_output, + size, + df); + + size_t* device_thrust_output = new size_t[size]; + HIP_CHECK(hipMemcpy(device_thrust_output, device_output, sizeof(size_t) * size, hipMemcpyDeviceToHost)); + + for (size_t i = 0; i < size; i++) + { + ASSERT_EQ(host_expected[i], device_thrust_output[i]); + ASSERT_EQ(host_expected[i], host_thrust_output[i]); + } + + delete[] host_search; + delete[] host_expected; + delete[] host_thrust_output; + delete[] device_thrust_output; + + HIP_CHECK(hipFree(device_search)); + HIP_CHECK(hipFree(device_input)); + HIP_CHECK(hipFree(device_output)); } -TYPED_TEST(BinarySearchTestsInKernel, TestLowerBound) +TYPED_TEST(SingleValueTests, LowerBound) { using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + return std::lower_bound(begin, end, value) - begin; + }, + [=] __device__(T * begin, T * end, const T& value) { + return thrust::lower_bound(thrust::device, begin, end, value) - begin; + }, + [=](T* begin, T* end, const T& value) { + return thrust::lower_bound(begin, end, value) - begin; + }); +} + +TYPED_TEST(SingleValueTests, LowerBoundWithCustomComp) +{ + using T = typename TestFixture::input_type; SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); - thrust::device_vector d_input(5); - d_input[0] = 0; - d_input[1] = 2; - d_input[2] = 5; - d_input[3] = 7; - d_input[4] = 8; + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + return std::lower_bound(begin, end, value, host_compare) - begin; + }, + [=] __device__(T * begin, T * end, const T& value) { + return thrust::lower_bound(thrust::device, begin, end, value, device_compare) - begin; + }, + [=](T* begin, T* end, const T& value) { + return thrust::lower_bound(begin, end, value, host_compare) - begin; + }); +} - thrust::device_vector d_output(10); +TYPED_TEST(SingleValueTests, UpperBound) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); - hipLaunchKernelGGL( - HIP_KERNEL_NAME(lower_bound_kernel), - dim3(1), - dim3(1), - 0, - 0, - size_t(d_input.size()), - thrust::raw_pointer_cast(d_input.data()), - thrust::raw_pointer_cast(d_output.data())); - - thrust::host_vector output = d_output; - ASSERT_EQ(output[0], 0); - ASSERT_EQ(output[1], 1); - ASSERT_EQ(output[2], 1); - ASSERT_EQ(output[3], 2); - ASSERT_EQ(output[4], 2); - ASSERT_EQ(output[5], 2); - ASSERT_EQ(output[6], 3); - ASSERT_EQ(output[7], 3); - ASSERT_EQ(output[8], 4); - ASSERT_EQ(output[9], 5); -} - -template -__global__ THRUST_HIP_LAUNCH_BOUNDS_DEFAULT void upper_bound_kernel(size_t n, T* input, ptrdiff_t* output) -{ - output[0] = thrust::upper_bound(thrust::device, input, input + n, T(0)) - input; - output[1] = thrust::upper_bound(thrust::device, input, input + n, T(1)) - input; - output[2] = thrust::upper_bound(thrust::device, input, input + n, T(2)) - input; - output[3] = thrust::upper_bound(thrust::device, input, input + n, T(3)) - input; - output[4] = thrust::upper_bound(thrust::device, input, input + n, T(4)) - input; - output[5] = thrust::upper_bound(thrust::device, input, input + n, T(5)) - input; - output[6] = thrust::upper_bound(thrust::device, input, input + n, T(6)) - input; - output[7] = thrust::upper_bound(thrust::device, input, input + n, T(7)) - input; - output[8] = thrust::upper_bound(thrust::device, input, input + n, T(8)) - input; - output[9] = thrust::upper_bound(thrust::device, input, input + n, T(9)) - input; -} - -TYPED_TEST(BinarySearchTestsInKernel, TestUpperBound) + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + return std::upper_bound(begin, end, value) - begin; + }, + [=] __device__(T * begin, T * end, const T& value) { + return thrust::upper_bound(thrust::device, begin, end, value) - begin; + }, + [=](T* begin, T* end, const T& value) { + return thrust::upper_bound(begin, end, value) - begin; + }); +} + +TYPED_TEST(SingleValueTests, UpperBoundWithCustomComp) { using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + return std::upper_bound(begin, end, value, host_compare) - begin; + }, + [=] __device__(T * begin, T * end, const T& value) { + return thrust::upper_bound(thrust::device, begin, end, value, device_compare) - begin; + }, + [=](T* begin, T* end, const T& value) { + return thrust::upper_bound(begin, end, value, host_compare) - begin; + }); +} - thrust::device_vector d_input(5); - d_input[0] = 0; - d_input[1] = 2; - d_input[2] = 5; - d_input[3] = 7; - d_input[4] = 8; +TYPED_TEST(SingleValueTests, TestSingleValueBinarySearch) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); - thrust::device_vector d_output(10); + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + return std::binary_search(begin, end, value); + }, + [=] __device__(T * begin, T * end, const T& value) { + return thrust::binary_search(thrust::device, begin, end, value); + }, + [=](T* begin, T* end, const T& value) { + return thrust::binary_search(begin, end, value); + }); +} - hipLaunchKernelGGL( - HIP_KERNEL_NAME(upper_bound_kernel), - dim3(1), - dim3(1), - 0, - 0, - size_t(d_input.size()), - thrust::raw_pointer_cast(d_input.data()), - thrust::raw_pointer_cast(d_output.data())); - - thrust::host_vector output = d_output; - ASSERT_EQ(output[0], 1); - ASSERT_EQ(output[1], 1); - ASSERT_EQ(output[2], 2); - ASSERT_EQ(output[3], 2); - ASSERT_EQ(output[4], 2); - ASSERT_EQ(output[5], 3); - ASSERT_EQ(output[6], 3); - ASSERT_EQ(output[7], 4); - ASSERT_EQ(output[8], 5); - ASSERT_EQ(output[9], 5); -} - -template -__global__ THRUST_HIP_LAUNCH_BOUNDS_DEFAULT void binary_search_kernel(size_t n, T* input, bool* output) -{ - output[0] = thrust::binary_search(thrust::device, input, input + n, T(0)); - output[1] = thrust::binary_search(thrust::device, input, input + n, T(1)); - output[2] = thrust::binary_search(thrust::device, input, input + n, T(2)); - output[3] = thrust::binary_search(thrust::device, input, input + n, T(3)); - output[4] = thrust::binary_search(thrust::device, input, input + n, T(4)); - output[5] = thrust::binary_search(thrust::device, input, input + n, T(5)); - output[6] = thrust::binary_search(thrust::device, input, input + n, T(6)); - output[7] = thrust::binary_search(thrust::device, input, input + n, T(7)); - output[8] = thrust::binary_search(thrust::device, input, input + n, T(8)); - output[9] = thrust::binary_search(thrust::device, input, input + n, T(9)); -} - -TYPED_TEST(BinarySearchTestsInKernel, TestBinarySearch) +TYPED_TEST(SingleValueTests, TestSingleValueBinarySearchWithCustomComp) { using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + return std::binary_search(begin, end, value, host_compare); + }, + [=] __device__(T * begin, T * end, const T& value) { + return thrust::binary_search(thrust::device, begin, end, value, device_compare); + }, + [=](T* begin, T* end, const T& value) { + return thrust::binary_search(begin, end, value, host_compare); + }); +} +TYPED_TEST(SingleValueTests, EqualRange) +{ + using T = typename TestFixture::input_type; SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); - thrust::device_vector d_input(5); - d_input[0] = 0; - d_input[1] = 2; - d_input[2] = 5; - d_input[3] = 7; - d_input[4] = 8; + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + auto out = std::equal_range(begin, end, value); + return out.second - out.first; + }, + [=] __device__(T * begin, T * end, const T& value) { + auto out = thrust::equal_range(thrust::device, begin, end, value); + return out.second - out.first; + }, + [=](T* begin, T* end, const T& value) { + auto out = thrust::equal_range(begin, end, value); + return out.second - out.first; + }); +} + +TYPED_TEST(SingleValueTests, EqualRangeWithCustomComp) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); - thrust::device_vector d_output(10); + RunSingleValueTest( + [=](T* begin, T* end, const T& value) { + auto out = std::equal_range(begin, end, value, host_compare); + return out.second - out.first; + }, + [=] __device__(T * begin, T * end, const T& value) { + auto out = thrust::equal_range(thrust::device, begin, end, value, device_compare); + return out.second - out.first; + }, + [=](T* begin, T* end, const T& value) { + auto out = thrust::equal_range(begin, end, value, host_compare); + return out.second - out.first; + }); +} + +TESTS_DEFINE(VectorTests, NumericalTestsParams); + +template +__global__ THRUST_HIP_LAUNCH_BOUNDS_DEFAULT void +multiple_value_kernel(T* device_search_arr, T* device_input, size_t* device_output, const DeviceFunc& f) +{ + constexpr size_t items_per_block = items_per_thread * block_size; + const size_t offset = (items_per_block * blockIdx.x) + (items_per_thread * threadIdx.x); + + f(device_search_arr, + device_search_arr + items_per_block, + device_input + offset, + device_input + offset + items_per_thread, + device_output + offset); +} + +template +void RunVectorTest(const ExpectedFunction& ef, const ThrustDeviceFunction& df, const ThrustHostFunction& hf) +{ + constexpr size_t grid_size = 1024; + constexpr size_t items_per_thread = 12; + constexpr size_t block_size = 32; + constexpr size_t items_per_block = items_per_thread * block_size; + constexpr size_t size = items_per_block * grid_size; + + double maxi = static_cast(std::numeric_limits::max()); + double mini = static_cast(std::numeric_limits::min()); + + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution dis(mini, maxi); + + T* host_search_arr = new T[items_per_block]; + for (size_t i = 0; i < items_per_block; i++) + { + host_search_arr[i] = static_cast(dis(gen)); + } + + std::sort(host_search_arr, host_search_arr + items_per_block); + + T* host_input = new T[size]; + for (size_t grid_idx = 0; grid_idx < grid_size; grid_idx++) + { + size_t offset = grid_idx * items_per_block; + for (size_t i = 0; i < items_per_block; i++) + { + host_input[offset + i] = static_cast(dis(gen)); + } + } + + size_t* thrust_host_output = new size_t[size]; + size_t* thrust_device_output = new size_t[size]; + size_t* expected_output = new size_t[size]; + + for (size_t bIdx = 0; bIdx < grid_size; bIdx++) + { + size_t offset = bIdx * items_per_block; + + // getting thrust host output + hf(host_search_arr, + host_search_arr + items_per_block, + host_input + offset, + host_input + offset + items_per_block, + thrust_host_output + offset); + // getting expected output + for (size_t i = 0; i < items_per_block; i++) + { + expected_output[offset + i] = ef(host_search_arr, host_search_arr + items_per_block, host_input[offset + i]); + } + } + + // getting thrust device output + T *device_search_arr, *device_input; + size_t* device_output; + HIP_CHECK(hipMalloc(&device_search_arr, sizeof(T) * items_per_block)); + HIP_CHECK(hipMalloc(&device_input, sizeof(T) * size)); + HIP_CHECK(hipMalloc(&device_output, sizeof(size_t) * size)); + + HIP_CHECK(hipMemcpy(device_search_arr, host_search_arr, sizeof(T) * items_per_block, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(device_input, host_input, sizeof(T) * size, hipMemcpyHostToDevice)); hipLaunchKernelGGL( - HIP_KERNEL_NAME(binary_search_kernel), - dim3(1), - dim3(1), + HIP_KERNEL_NAME(multiple_value_kernel), + dim3(grid_size), + dim3(block_size), 0, 0, - size_t(d_input.size()), - thrust::raw_pointer_cast(d_input.data()), - thrust::raw_pointer_cast(d_output.data())); - - thrust::host_vector output = d_output; - ASSERT_EQ(output[0], true); - ASSERT_EQ(output[1], false); - ASSERT_EQ(output[2], true); - ASSERT_EQ(output[3], false); - ASSERT_EQ(output[4], false); - ASSERT_EQ(output[5], true); - ASSERT_EQ(output[6], false); - ASSERT_EQ(output[7], true); - ASSERT_EQ(output[8], true); - ASSERT_EQ(output[9], false); + device_search_arr, + device_input, + device_output, + df); + HIP_CHECK(hipMemcpy(thrust_device_output, device_output, sizeof(size_t) * size, hipMemcpyDeviceToHost)); + + for (size_t i = 0; i < size; i++) + { + ASSERT_EQ(expected_output[i], thrust_host_output[i]); + ASSERT_EQ(expected_output[i], thrust_device_output[i]); + } + + delete[] host_search_arr; + delete[] host_input; + delete[] thrust_host_output; + delete[] thrust_device_output; + delete[] expected_output; + + HIP_CHECK(hipFree(device_search_arr)); + HIP_CHECK(hipFree(device_input)); + HIP_CHECK(hipFree(device_output)); +} + +TYPED_TEST(VectorTests, LowerBound) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunVectorTest( + [=](T* begin, T* end, const T& value) { + return std::lower_bound(begin, end, value) - begin; + }, + [=] __device__(T * s_begin, T * s_end, T * i_begin, T * i_end, size_t * out) { + thrust::lower_bound(thrust::device, s_begin, s_end, i_begin, i_end, out); + }, + [=](T* s_begin, T* s_end, T* i_begin, T* i_end, size_t* out) { + thrust::lower_bound(s_begin, s_end, i_begin, i_end, out); + }); +} + +TYPED_TEST(VectorTests, LowerBoundWithCustomComp) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunVectorTest( + [=](T* begin, T* end, const T& value) { + return std::lower_bound(begin, end, value, host_compare) - begin; + }, + [=] __device__(T * s_begin, T * s_end, T * i_begin, T * i_end, size_t * out) { + thrust::lower_bound(thrust::device, s_begin, s_end, i_begin, i_end, out, device_compare); + }, + [=](T* s_begin, T* s_end, T* i_begin, T* i_end, size_t* out) { + thrust::lower_bound(s_begin, s_end, i_begin, i_end, out, host_compare); + }); +} + +TYPED_TEST(VectorTests, UpperBound) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunVectorTest( + [=](T* begin, T* end, const T& value) { + return std::upper_bound(begin, end, value) - begin; + }, + [=] __device__(T * s_begin, T * s_end, T * i_begin, T * i_end, size_t * out) { + thrust::upper_bound(thrust::device, s_begin, s_end, i_begin, i_end, out); + }, + [=](T* s_begin, T* s_end, T* i_begin, T* i_end, size_t* out) { + thrust::upper_bound(s_begin, s_end, i_begin, i_end, out); + }); +} + +TYPED_TEST(VectorTests, UpperBoundWithCustomComp) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunVectorTest( + [=](T* begin, T* end, const T& value) { + return std::upper_bound(begin, end, value, host_compare) - begin; + }, + [=] __device__(T * s_begin, T * s_end, T * i_begin, T * i_end, size_t * out) { + thrust::upper_bound(thrust::device, s_begin, s_end, i_begin, i_end, out, device_compare); + }, + [=](T* s_begin, T* s_end, T* i_begin, T* i_end, size_t* out) { + thrust::upper_bound(s_begin, s_end, i_begin, i_end, out, host_compare); + }); +} + +TYPED_TEST(VectorTests, BinarySearch) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunVectorTest( + [=](T* begin, T* end, const T& value) { + return std::binary_search(begin, end, value); + }, + [=] __device__(T * s_begin, T * s_end, T * i_begin, T * i_end, size_t * out) { + thrust::binary_search(thrust::device, s_begin, s_end, i_begin, i_end, out); + }, + [=](T* s_begin, T* s_end, T* i_begin, T* i_end, size_t* out) { + thrust::binary_search(s_begin, s_end, i_begin, i_end, out); + }); +} + +TYPED_TEST(VectorTests, BinarySearchWithCustomComp) +{ + using T = typename TestFixture::input_type; + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + RunVectorTest( + [=](T* begin, T* end, const T& value) { + return std::binary_search(begin, end, value, host_compare); + }, + [=] __device__(T * s_begin, T * s_end, T * i_begin, T * i_end, size_t * out) { + thrust::binary_search(thrust::device, s_begin, s_end, i_begin, i_end, out, device_compare); + }, + [=](T* s_begin, T* s_end, T* i_begin, T* i_end, size_t* out) { + thrust::binary_search(s_begin, s_end, i_begin, i_end, out, host_compare); + }); } TESTS_DEFINE(BinarySearchTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_binary_search_descending.cpp b/projects/rocthrust/test/test_binary_search_descending.cpp index 77e2a666a50..ee7f817e43b 100644 --- a/projects/rocthrust/test/test_binary_search_descending.cpp +++ b/projects/rocthrust/test/test_binary_search_descending.cpp @@ -20,7 +20,10 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" + TESTS_DEFINE(BinarySearchDescendingTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_binary_search_vector.cpp b/projects/rocthrust/test/test_binary_search_vector.cpp index c634fc10257..ab64123f2d8 100644 --- a/projects/rocthrust/test/test_binary_search_vector.cpp +++ b/projects/rocthrust/test/test_binary_search_vector.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(BinarySearchVectorTestsInKernel, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_binary_search_vector_descending.cpp b/projects/rocthrust/test/test_binary_search_vector_descending.cpp index 8f5bd5a47cd..7a8a0361c06 100644 --- a/projects/rocthrust/test/test_binary_search_vector_descending.cpp +++ b/projects/rocthrust/test/test_binary_search_vector_descending.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(BinarySearchVectorDescendingTests, FullTestsParams); TESTS_DEFINE(BinarySearchVectorDescendingIntegerTests, SignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_complex.cpp b/projects/rocthrust/test/test_complex.cpp index 36d1e060f9a..3639d6f7340 100644 --- a/projects/rocthrust/test/test_complex.cpp +++ b/projects/rocthrust/test/test_complex.cpp @@ -17,7 +17,44 @@ #include -#include "test_header.hpp" +#include "test_imag_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" + +#define CHECK_CORRECT(stComplex, ttComplex, lreal, limag, rreal, rimag) \ + do \ + { \ + SCOPED_TRACE(testing::Message() \ + << "\nLHS --- Real: " << lreal << " Imag: " << limag << std::endl \ + << "RHS --- Real: " << rreal << " Imag: " << rimag << std::endl \ + << "AC: " << rreal * lreal << " BD: " << limag * rimag << std::endl); \ + if (std::isinf(stComplex.real())) \ + { \ + ASSERT_TRUE(std::isinf(ttComplex.real())); \ + } \ + else if (std::isnan(stComplex.real())) \ + { \ + ASSERT_TRUE(std::isnan(ttComplex.real())); \ + } \ + else \ + { \ + auto eps = stComplex.real() < 0.01 ? 1e-2 : abs(stComplex.real() * 0.05); \ + ASSERT_NEAR(stComplex.real(), ttComplex.real(), eps); \ + } \ + if (std::isinf(stComplex.imag())) \ + { \ + ASSERT_TRUE(std::isinf(ttComplex.imag())); \ + } \ + else if (std::isnan(stComplex.imag())) \ + { \ + ASSERT_TRUE(std::isnan(ttComplex.imag())); \ + } \ + else \ + { \ + auto eps = stComplex.imag() < 0.01 ? 1e-2 : abs(stComplex.imag() * 0.05); \ + ASSERT_NEAR(stComplex.imag(), ttComplex.imag(), eps); \ + } \ + } while (0) TESTS_DEFINE(ComplexTests, FloatTestsParams); @@ -322,3 +359,473 @@ TYPED_TEST(ComplexTests, TestComplexStreamOperators) ASSERT_NEAR_COMPLEX(a, b); } } + +TESTS_PAIRS_DEFINE(ComplexPairsTests, PairsTestsParams) + +TYPED_TEST(ComplexPairsTests, TestAssignOperator) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + constexpr size_t test_it = 123456; + + const double tmini = static_cast(std::numeric_limits::min()); + const double tmaxi = static_cast(std::numeric_limits::max()); + + const double umini = static_cast(std::numeric_limits::min()); + const double umaxi = static_cast(std::numeric_limits::max()); + + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution tdis(tmini, tmaxi); + std::uniform_real_distribution udis(umini, umaxi); + + for (size_t i = 0; i < test_it; i++) + { + thrust::complex thrustNum; + std::complex expectedNum; + T treal = tdis(gen); + + thrustNum = treal; + expectedNum = treal; + + ASSERT_EQ(thrustNum.imag(), expectedNum.imag()); + ASSERT_EQ(thrustNum.real(), expectedNum.real()); + + U ureal = static_cast(udis(gen)); + U uimag = static_cast(udis(gen)); + thrust::complex thrustOrigin(static_cast(ureal), static_cast(uimag)); + + thrustNum = thrustOrigin; + ASSERT_EQ(thrustNum.imag(), static_cast(thrustOrigin.imag())); + ASSERT_EQ(thrustNum.real(), static_cast(thrustOrigin.real())); + + ureal = static_cast(udis(gen)); + uimag = static_cast(udis(gen)); + std::complex stdOriginU(static_cast(ureal), static_cast(uimag)); + + thrustNum = stdOriginU; + expectedNum = stdOriginU; + + ASSERT_EQ(thrustNum.imag(), expectedNum.imag()); + ASSERT_EQ(thrustNum.real(), expectedNum.real()); + + treal = tdis(gen); + T timag = tdis(gen); + std::complex stdOriginT(static_cast(treal), static_cast(timag)); + + thrustNum = stdOriginT; + expectedNum = stdOriginT; + + ASSERT_EQ(thrustNum.imag(), expectedNum.imag()); + ASSERT_EQ(thrustNum.real(), expectedNum.real()); + } +} + +template +void run_compound_tests( + const StdComplexOp& standard_complex_operator, + const StdScalarOp& standard_scalar_operator, + const ThrustComplexOp& thrust_complex_operator, + const ThrustScalarOp& thrust_scalar_operator, + T mini = -500000, + T maxi = 500000) +{ + constexpr T test_it = 15; + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution dis(mini, maxi); + + T inc = (maxi - mini) / test_it; + + for (T treal = mini; treal <= maxi; treal += inc) + { + for (T timag = mini; timag <= maxi; timag += inc) + { + for (U ureal = mini; ureal <= maxi; ureal += inc) + { + for (U uimag = mini; uimag <= maxi; uimag += inc) + { + std::complex stComplex(treal, timag); + std::complex suComplex(ureal, uimag); + + thrust::complex ttComplex(treal, timag); + thrust::complex tuComplex(ureal, uimag); + + standard_complex_operator(stComplex, suComplex); + thrust_complex_operator(ttComplex, tuComplex); + + CHECK_CORRECT(stComplex, ttComplex, treal, timag, ureal, uimag); + + stComplex = std::complex(treal, timag); + ttComplex = thrust::complex(treal, timag); + + standard_scalar_operator(stComplex, ureal); + thrust_scalar_operator(ttComplex, ureal); + + CHECK_CORRECT(stComplex, ttComplex, treal, timag, ureal, uimag); + + stComplex = std::complex(treal, timag); + ttComplex = thrust::complex(treal, timag); + + standard_scalar_operator(stComplex, uimag); + thrust_scalar_operator(ttComplex, uimag); + + CHECK_CORRECT(stComplex, ttComplex, treal, timag, ureal, uimag); + } + } + } + } +} + +/** + * Due to hipcc compiler issue, std complex functions for + * complex number is not working properly. To get + * around this we will have to calculate the real and + * imaginary parts separately manually. + */ + +TYPED_TEST(ComplexPairsTests, TestCompoundPlusOperator) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_compound_tests( + [=](std::complex& lhs, const std::complex& rhs) { + using type = typename thrust::detail::promoted_numerical_type::type; + lhs = std::complex(lhs.real() + rhs.real(), lhs.imag() + rhs.imag()); + }, + [=](std::complex& lhs, const U& rhs) { + using type = typename thrust::detail::promoted_numerical_type::type; + lhs = std::complex(lhs.real() + rhs, lhs.imag()); + }, + [=](thrust::complex& lhs, const thrust::complex& rhs) { + lhs += rhs; + }, + [=](thrust::complex& lhs, const U& rhs) { + lhs += rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestCompoundMinusOperator) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_compound_tests( + [=](std::complex& lhs, const std::complex& rhs) { + using type = typename thrust::detail::promoted_numerical_type::type; + lhs = std::complex(lhs.real() - rhs.real(), lhs.imag() - rhs.imag()); + }, + [=](std::complex& lhs, const U& rhs) { + using type = typename thrust::detail::promoted_numerical_type::type; + lhs = std::complex(lhs.real() - rhs, lhs.imag()); + }, + [=](thrust::complex& lhs, const thrust::complex& rhs) { + lhs -= rhs; + }, + [=](thrust::complex& lhs, const U& rhs) { + lhs -= rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestCompoundMultiplyOperator) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_compound_tests( + [=](std::complex& lhs, const std::complex& rhs) { + // (a + bi)(c + di) = ac + i(ad + bc) - bd + using type = typename thrust::detail::promoted_numerical_type::type; + + type a = static_cast(lhs.real()); + type b = static_cast(lhs.imag()); + type c = static_cast(rhs.real()); + type d = static_cast(rhs.imag()); + + type ac = a * c; + type bd = b * d; + type real = ac - bd; + type imag = (a * d) + (b * c); + + lhs = std::complex(real, imag); + }, + [=](std::complex& lhs, const U& rhs) { + lhs = std::complex(lhs.real() * rhs, lhs.imag() * rhs); + }, + [=](thrust::complex& lhs, const thrust::complex& rhs) { + lhs *= rhs; + }, + [=](thrust::complex& lhs, const U& rhs) { + lhs *= rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestCompoundDivisionOperator) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_compound_tests( + [=](std::complex& lhs, const std::complex& rhs) { + using type = typename thrust::detail::promoted_numerical_type::type; + + // (a + bi) / (c + di) = ((ac + bd) + (bc -ad)i) / (c^2 + d^2) + type a = static_cast(lhs.real()); + type b = static_cast(lhs.imag()); + type c = static_cast(rhs.real()); + type d = static_cast(rhs.imag()); + + type ac = a * c; + type bd = b * d; + type bc = b * c; + type ad = a * d; + type denom = c * c + d * d; + type real = (ac + bd) / denom; + type imag = (bc - ad) / denom; + + lhs = std::complex(real, imag); + }, + [=](std::complex& lhs, const U& rhs) { + lhs /= rhs; + }, + [=](thrust::complex& lhs, const thrust::complex& rhs) { + lhs /= rhs; + }, + [=](thrust::complex& lhs, const U& rhs) { + lhs /= rhs; + }); +} + +template +void run_equality_operator_tests(const EqualOp& f, const bool scalar = false) +{ + constexpr T test_size = 123456; + + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution dis(std::numeric_limits::min(), std::numeric_limits::max()); + + for (size_t i = 0; i < test_size; i++) + { + T treal = dis(gen); + T timag = dis(gen); + + U ureal; + U uimag; + + if (i % 2) + { + ureal = dis(gen); + uimag = dis(gen); + } + else + { + ureal = treal; + uimag = timag; + } + + thrust::complex tcomplex; + thrust::complex ucomplex; + if (scalar) + { + tcomplex = thrust::complex(treal); + ucomplex = thrust::complex(ureal); + } + else + { + tcomplex = thrust::complex(treal, timag); + ucomplex = thrust::complex(ureal, uimag); + } + + ASSERT_EQ((treal == ureal && (scalar ? true : timag == uimag)), f(tcomplex, ucomplex)); + } +} + +TYPED_TEST(ComplexPairsTests, TestEqualOperator_ThrustComplex_ThrustComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_equality_operator_tests([=](const thrust::complex& lhs, const thrust::complex& rhs) { + return lhs == rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestEqualOperator_ThrustComplex_StdComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_equality_operator_tests([=](const thrust::complex& lhs, const thrust::complex& rhs) { + std::complex rhs_(rhs.real(), rhs.imag()); + return lhs == rhs_; + }); +} + +TYPED_TEST(ComplexPairsTests, TestEqualOperator_StdComplex_ThrustComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_equality_operator_tests([=](const thrust::complex& lhs, const thrust::complex& rhs) { + std::complex lhs_(lhs.real(), lhs.imag()); + return lhs_ == rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestEqualOperator_Scalar_ThrustComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_equality_operator_tests( + [=](const thrust::complex& lhs, const thrust::complex& rhs) { + return lhs.real() == rhs; + }, + true); +} + +TYPED_TEST(ComplexPairsTests, TestEqualOperator_ThrustComplex_Scalar) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_equality_operator_tests( + [=](const thrust::complex& lhs, const thrust::complex& rhs) { + return lhs == rhs.real(); + }, + true); +} + +template +void run_inequality_operator_tests(const InequalOp& f, const bool scalar = false) +{ + constexpr T test_size = 123456; + + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution dis(std::numeric_limits::min(), std::numeric_limits::max()); + + for (size_t i = 0; i < test_size; i++) + { + T treal = dis(gen); + T timag = dis(gen); + + U ureal; + U uimag; + + if (i % 2) + { + ureal = dis(gen); + uimag = dis(gen); + } + else + { + ureal = treal; + uimag = timag; + } + + thrust::complex tcomplex; + thrust::complex ucomplex; + if (scalar) + { + tcomplex = thrust::complex(treal); + ucomplex = thrust::complex(ureal); + } + else + { + tcomplex = thrust::complex(treal, timag); + ucomplex = thrust::complex(ureal, uimag); + } + + ASSERT_EQ(!(treal == ureal && (scalar ? true : timag == uimag)), f(tcomplex, ucomplex)); + } +} + +TYPED_TEST(ComplexPairsTests, TestInequalOperator_ThrustComplex_ThrustComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_inequality_operator_tests([=](const thrust::complex& lhs, const thrust::complex& rhs) { + return lhs != rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestInequalOperator_ThrustComplex_StdComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_inequality_operator_tests([=](const thrust::complex& lhs, const thrust::complex& rhs) { + std::complex rhs_(rhs.real(), rhs.imag()); + return lhs != rhs_; + }); +} + +TYPED_TEST(ComplexPairsTests, TestInequalOperator_StdComplex_ThrustComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_inequality_operator_tests([=](const thrust::complex& lhs, const thrust::complex& rhs) { + std::complex lhs_(lhs.real(), lhs.imag()); + return lhs_ != rhs; + }); +} + +TYPED_TEST(ComplexPairsTests, TestInequalOperator_Scalar_ThrustComplex) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_inequality_operator_tests( + [=](const thrust::complex& lhs, const thrust::complex& rhs) { + return lhs.real() != rhs; + }, + true); +} + +TYPED_TEST(ComplexPairsTests, TestInequalOperator_ThrustComplex_Scalar) +{ + using T = typename TestFixture::first_type; + using U = typename TestFixture::second_type; + + SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); + + run_inequality_operator_tests( + [=](const thrust::complex& lhs, const thrust::complex& rhs) { + return lhs != rhs.real(); + }, + true); +} diff --git a/projects/rocthrust/test/test_complex_transform.cpp b/projects/rocthrust/test/test_complex_transform.cpp index b639b7c72c8..bdc92114af6 100644 --- a/projects/rocthrust/test/test_complex_transform.cpp +++ b/projects/rocthrust/test/test_complex_transform.cpp @@ -21,7 +21,9 @@ #include -#include "test_header.hpp" +#include "test_imag_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ComplexTransformTests, FloatTestsParams); diff --git a/projects/rocthrust/test/test_complex_various.cpp b/projects/rocthrust/test/test_complex_various.cpp new file mode 100644 index 00000000000..437f88832f9 --- /dev/null +++ b/projects/rocthrust/test/test_complex_various.cpp @@ -0,0 +1,1137 @@ +/* + * Copyright 2008-2013 NVIDIA Corporation + * Modifications Copyright© 2019-2025 Advanced Micro Devices, Inc. All rights reserved. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include + +#include +#include +#include + +#include + +#define M_PI 3.1415926535897932 + +#define CHECK_CORRECT(std_complex, thrust_complex, real_eps, imag_eps, real_val, imag_val) \ + do \ + { \ + SCOPED_TRACE( \ + testing::Message() \ + << std::endl \ + << "Input Real Value: " << real_val << " Input Imaginary Value: " << imag_val << std::endl \ + << "Std Output Real Value: " << std_complex.real() << " Std Output Imag Value: " << std_complex.imag() \ + << std::endl \ + << "Thrust Output Real Value: " << thrust_complex.real() \ + << " Thrust Output Imag Value: " << thrust_complex.imag() << std::endl); \ + if (std::isinf(std_complex.real())) \ + { \ + ASSERT_TRUE(std::isinf(thrust_complex.real())); \ + } \ + else if (std::isnan(std_complex.real())) \ + { \ + ASSERT_TRUE(std::isnan(thrust_complex.real())); \ + } \ + else \ + { \ + ASSERT_NEAR(std_complex.real(), thrust_complex.real(), real_eps); \ + } \ + if (std::isinf(std_complex.imag())) \ + { \ + ASSERT_TRUE(std::isinf(thrust_complex.imag())); \ + } \ + else if (std::isnan(std_complex.imag())) \ + { \ + ASSERT_TRUE(std::isnan(thrust_complex.imag())); \ + } \ + else \ + { \ + ASSERT_NEAR(std_complex.imag(), thrust_complex.imag(), imag_eps); \ + } \ + } while (0) + +#define GET_EPS(x) ((std::abs(x) < 0.05) ? 1e-1 : std::abs(x) * 0.05) + +// This test suite aims to test vairous different implementations +// in the thrust/detail/complex directory + +using FloatDouble = ::testing::Types; + +template +class VariousComplexTest : public ::testing::Test +{ +public: + using T = Type; +}; + +template +void run_rng_test(const StdFunc& sf, const ThrustFunc& tf) +{ + constexpr size_t test_size = 10000; + + std::random_device rd; + std::mt19937 gen(rd()); + std::uniform_real_distribution dis(std::numeric_limits::min(), std::numeric_limits::max()); + + for (size_t i = 0; i < test_size; i++) + { + T r_num = dis(gen); + T expected, actual; + if constexpr (SingleParam) + { + expected = sf(r_num); + actual = tf(r_num); + } + else + { + T r_num_ = dis(gen); + expected = sf(r_num, r_num_); + actual = tf(r_num, r_num_); + } + + if (std::isinf(expected)) + { + ASSERT_TRUE(std::isinf(actual)); + } + else if (std::isnan(expected)) + { + ASSERT_TRUE(std::isnan(actual)); + } + else + { + ASSERT_NEAR(expected, actual, expected * 0.01); + } + } +} + +template +void run_trig_tests( + const StdFunc& std_func, const ThrustFunc& thrust_func, double rmini, double rmaxi, double imini, double imaxi) +{ + // To run N tests but with N^2 algorithim, we must sqrt(N) + // so that total tests is still N + const double test_size = std::sqrt(100000); + + const double real_inc = (rmaxi - rmini) / test_size; + const double imag_inc = (imaxi - imini) / test_size; + + for (double real = rmini; real <= rmaxi; real += real_inc) + { + for (double imag = imini; imag <= imaxi; imag += imag_inc) + { + thrust::complex thrust_complex(real, imag); + std::complex std_complex(real, imag); + + thrust::complex thrust_out = thrust_func(thrust_complex); + std::complex std_out = std_func(std_complex); + + T real_eps = GET_EPS(std_out.real()); + T imag_eps = GET_EPS(std_out.imag()); + + CHECK_CORRECT(std_out, thrust_out, real_eps, imag_eps, real, imag); + } + } +} + +// =================================================================== +// +// C99 MATH +// +// =================================================================== + +TYPED_TEST_SUITE(VariousComplexTest, FloatDouble); +TYPED_TEST(VariousComplexTest, getInf) +{ + using T = typename TestFixture::T; + + T inf = thrust::detail::complex::infinity(); + ASSERT_TRUE(std::isinf(inf)); +} + +#if defined _MSC_VER + +TYPED_TEST(VariousComplexTest, isinf) +{ + using T = typename TestFixture::T; + T inf = thrust::detail::complex::infinity(); + ASSERT_EQ(std::isinf(inf), thrust::detail::complex::isinf(inf)); + + run_rng_test( + [](const T& x) { + return std::isinf(x); + }, + [](const T& x) { + return thrust::detail::complex::isinf(x); + }); +} + +TYPED_TEST(VariousComplexTest, isnan) +{ + using T = typename TestFixture::T; + T nan = std::numeric_limits::quiet_NaN(); + ASSERT_EQ(std::isnan(nan), thrust::detail::complex::isnan(nan)); + + run_rng_test( + [](const T& x) { + return std::isnan(x); + }, + [](const T& x) { + return thrust::detail::complex::isnan(x); + }); +} + +TYPED_TEST(VariousComplexTest, signbit) +{ + using T = typename TestFixture::T; + run_rng_test( + [](const T& x) { + return std::signbit(x); + }, + [](const T& x) { + return thrust::detail::complex::signbit(x); + }); +} + +TYPED_TEST(VariousComplexTest, isfinite) +{ + using T = typename TestFixture::T; + run_rng_test( + [](const T& x) { + return std::isfinite(x); + }, + [](const T& x) { + return thrust::detail::complex::isfinite(x); + }); +} + +TEST(VariousComplexTest, copysign) +{ + run_rng_test( + [](const double& x, const double& y) { + return std::copysign(x, y); + }, + [](const double& x, const double& y) { + return thrust::detail::complex::copysign(x, y); + }); +} + +TEST(VariousComplexTest, copysignf) +{ + run_rng_test( + [](const float x, const float y) { + return std::copysignf(x, y); + }, + [](const float x, const float y) { + return thrust::detail::complex::copysignf(x, y); + }); +} + +# if !defined(__CUDACC__) && !defined(_NVHPC_CUDA) + +TYPED_TEST(VariousComplexTest, log1p) +{ + using T = typename TestFixture::T; + run_rng_test( + [](const T& x) { + return std::log1p(x); + }, + [](const T& x) { + return thrust::detail::complex::log1p(x); + }); +} + +TYPED_TEST(VariousComplexTest, log1pf) +{ + using T = typename TestFixture::T; + run_rng_test( + [](const T& x) { + return std::log1pf(x); + }, + [](const T& x) { + return thrust::detail::complex::log1pf(x); + }); +} +# endif + +# if _MSC_VER <= 1500 && !defined(__clang__) + +TEST(VariousComplexTest, hypot) +{ + run_rng_test( + [](const double& x, const double& y) { + return std::hypot(x, y); + }, + [](const double& x, const double& y) { + return thrust::detail::complex::hypot(x, y); + }); +} + +TEST(VariousComplexTest, hypotf) +{ + run_rng_test( + [](const float& x, const float& y) { + return std::hypotf(x, y); + }, + [](const float& x, const float& y) { + return thrust::detail::complex::hypotf(x, y); + }); +} + +# endif // _MSC_VER <= 1500 + +#endif + +/** + * Due to hipcc compiler issue, std trig functions for + * complex number is not working properly. To get + * around this we will have to calculate the real and + * imaginary parts separately manually. + */ + +// =================================================================== +// +// CATRIG & CATRIGF +// +// =================================================================== + +#ifndef _WIN32 + +TYPED_TEST(VariousComplexTest, asinh) +{ + /** + * Due to hipcc compiler issues, negative range for the real is not working + * for std::asinh. + */ + + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::asinh(x); + }, + [](thrust::complex& x) { + return thrust::asinh(x); + }, + 0, + 10000, + -1, + 1); +} + +TYPED_TEST(VariousComplexTest, asinh_special) +{ + // test special cases for asinh + using T = typename TestFixture::T; + + T qNan = std::numeric_limits::quiet_NaN(); + T posInf = std::numeric_limits::infinity(); + + thrust::complex zero_zero(0, 0); + thrust::complex out = thrust::asinh(zero_zero); + ASSERT_EQ(0, out.real()); + ASSERT_EQ(0, out.imag()); + + thrust::complex x_inf(1000, posInf); + out = thrust::asinh(x_inf); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_NEAR(M_PI / 2, out.imag(), M_PI / 2 * 0.001); + + thrust::complex x_nan(1000, qNan); + out = thrust::asinh(x_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex inf_x(posInf, 10000); + out = thrust::asinh(inf_x); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_EQ(0, out.imag()); + + thrust::complex inf_inf(posInf, posInf); + out = thrust::asinh(inf_inf); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_NEAR(M_PI / 4, out.imag(), M_PI / 4 * 0.001); + + thrust::complex inf_nan(posInf, qNan); + out = thrust::asinh(inf_nan); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_zero(qNan, 0); + out = thrust::asinh(nan_zero); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_EQ(0, out.imag()); + + thrust::complex nan_x(qNan, 1000); + out = thrust::asinh(nan_x); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_inf(qNan, posInf); + out = thrust::asinh(nan_inf); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_nan(qNan, qNan); + out = thrust::asinh(nan_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); +} + +TYPED_TEST(VariousComplexTest, asin) +{ + /** + * Due to hipcc compiler issues, negative range for the real and imaginary is not working + * for std::asin. Also a poistive range greater than 1000 for the imaginary cause issues too + */ + + using T = typename TestFixture::T; + + run_trig_tests( + [](std::complex& x) { + return std::asin(x); + }, + [](thrust::complex& x) { + return thrust::asin(x); + }, + -1, + 1, + 0, + 1000); +} + +TYPED_TEST(VariousComplexTest, acosh) +{ + using T = typename TestFixture::T; + T max_range = std::sqrt(std::numeric_limits::max()); + + run_trig_tests( + [](std::complex& x) { + return std::acosh(x); + }, + [](thrust::complex& x) { + return thrust::acosh(x); + }, + 1, + max_range, + -max_range, + max_range); +} + +TYPED_TEST(VariousComplexTest, acosh_special) +{ + // test special cases for acosh + using T = typename TestFixture::T; + + T qNan = std::numeric_limits::quiet_NaN(); + T posInf = std::numeric_limits::infinity(); + T negInf = -std::numeric_limits::infinity(); + + thrust::complex zero_zero(0, 0); + thrust::complex out = thrust::acosh(zero_zero); + ASSERT_EQ(0, out.real()); + ASSERT_NEAR(M_PI * 0.5, out.imag(), M_PI * 0.005); + + thrust::complex x_inf(1000, posInf); + out = thrust::acosh(x_inf); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_NEAR(M_PI / 2, out.imag(), M_PI / 2 * 0.001); + + thrust::complex x_nan(1000, qNan); + out = thrust::acosh(x_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex neg_inf_x(negInf, 10000); + out = thrust::acosh(neg_inf_x); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_NEAR(M_PI, out.imag(), M_PI * 0.001); + + thrust::complex inf_x(posInf, 10000); + out = thrust::acosh(inf_x); + ASSERT_TRUE(std::isinf(out.real())); + ASSERT_EQ(0, out.imag()); + + thrust::complex neg_inf_pos_inf(negInf, posInf); + out = thrust::acosh(neg_inf_pos_inf); + ASSERT_EQ(posInf, out.real()); + ASSERT_NEAR(M_PI * 0.75, out.imag(), M_PI * 0.0075); + + thrust::complex pos_inf_nan(posInf, qNan); + out = thrust::acosh(pos_inf_nan); + ASSERT_EQ(posInf, out.real()); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex neg_inf_nan(negInf, qNan); + out = thrust::acosh(neg_inf_nan); + ASSERT_EQ(posInf, out.real()); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_x(qNan, 1000); + out = thrust::acosh(nan_x); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_inf(qNan, posInf); + out = thrust::acosh(nan_inf); + ASSERT_EQ(posInf, out.real()); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_nan(qNan, qNan); + out = thrust::acosh(nan_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); +} + +TYPED_TEST(VariousComplexTest, acos) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::acos(x); + }, + [](thrust::complex& x) { + return thrust::acos(x); + }, + -1, + 1, + -1000, + 1000); +} + +TYPED_TEST(VariousComplexTest, acos_special) +{ + // test special cases for acos + using T = typename TestFixture::T; + + T qNan = std::numeric_limits::quiet_NaN(); + T posInf = std::numeric_limits::infinity(); + T negInf = -std::numeric_limits::infinity(); + + thrust::complex zero_zero(0, 0); + thrust::complex out = thrust::acos(zero_zero); + ASSERT_NEAR(M_PI * 0.5, out.real(), M_PI * 0.005); + ASSERT_EQ(0, out.imag()); + + thrust::complex zero_nan(0, qNan); + out = thrust::acos(zero_nan); + ASSERT_NEAR(M_PI * 0.5, out.real(), M_PI * 0.005); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex x_inf(1000, posInf); + out = thrust::acos(x_inf); + ASSERT_NEAR(M_PI * 0.5, out.real(), M_PI * 0.005); + ASSERT_EQ(negInf, out.imag()); + + thrust::complex x_nan(1000, qNan); + out = thrust::acos(x_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex neg_inf_x(negInf, 10000); + out = thrust::acos(neg_inf_x); + ASSERT_NEAR(M_PI, out.real(), M_PI * 0.001); + ASSERT_EQ(negInf, out.imag()); + + thrust::complex pos_inf_x(posInf, 10000); + out = thrust::acos(pos_inf_x); + ASSERT_EQ(0, out.real()); + ASSERT_EQ(negInf, out.imag()); + + thrust::complex neg_inf_pos_inf(negInf, posInf); + out = thrust::acos(neg_inf_pos_inf); + ASSERT_NEAR(M_PI * 0.75, out.real(), M_PI * 0.0075); + ASSERT_EQ(negInf, out.imag()); + + thrust::complex pos_inf_pos_inf(posInf, posInf); + out = thrust::acos(pos_inf_pos_inf); + ASSERT_NEAR(M_PI * 0.25, out.real(), M_PI * 0.0025); + ASSERT_EQ(negInf, out.imag()); + + thrust::complex pos_inf_nan(posInf, qNan); + out = thrust::acos(pos_inf_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isinf(out.imag())); + + thrust::complex nan_x(qNan, 1000); + out = thrust::acos(nan_x); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_inf(qNan, posInf); + out = thrust::acos(nan_inf); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_EQ(negInf, out.imag()); + + thrust::complex nan_nan(qNan, qNan); + out = thrust::acos(nan_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); +} + +TYPED_TEST(VariousComplexTest, atanh) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::atanh(x); + }, + [](thrust::complex& x) { + return thrust::atanh(x); + }, + -1, + 1, + -1000, + 1000); +} + +TYPED_TEST(VariousComplexTest, catanh) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::atanh(x); + }, + [](thrust::complex& x) { + return thrust::detail::complex::catanh(x); + }, + -1, + 1, + -1000, + 1000); +} + +TYPED_TEST(VariousComplexTest, atanh_special) +{ + // test special cases for atanh + using T = typename TestFixture::T; + + T qNan = std::numeric_limits::quiet_NaN(); + T posInf = std::numeric_limits::infinity(); + + thrust::complex zero_zero(0, 0); + thrust::complex out = thrust::atanh(zero_zero); + ASSERT_EQ(0, out.real()); + ASSERT_EQ(0, out.imag()); + + thrust::complex zero_nan(0, qNan); + out = thrust::atanh(zero_nan); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex one_zero(1, 0); + out = thrust::atanh(one_zero); + ASSERT_EQ(posInf, out.real()); + ASSERT_EQ(0, out.imag()); + + thrust::complex x_inf(1000, posInf); + out = thrust::atanh(x_inf); + ASSERT_EQ(0, out.real()); + ASSERT_NEAR(M_PI * 0.5, out.imag(), M_PI * 0.005); + + thrust::complex x_nan(1000, qNan); + out = thrust::atanh(x_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex pos_inf_x(posInf, 10000); + out = thrust::atanh(pos_inf_x); + ASSERT_EQ(0, out.real()); + ASSERT_NEAR(M_PI * 0.5, out.imag(), M_PI * 0.005); + + thrust::complex pos_inf_nan(posInf, qNan); + out = thrust::atanh(pos_inf_nan); + ASSERT_EQ(0, out.real()); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_x(qNan, 1000); + out = thrust::atanh(nan_x); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); + + thrust::complex nan_inf(qNan, posInf); + out = thrust::atanh(nan_inf); + ASSERT_EQ(0, out.real()); + ASSERT_NEAR(M_PI * 0.5, out.imag(), M_PI * 0.005); + + thrust::complex nan_nan(qNan, qNan); + out = thrust::atanh(nan_nan); + ASSERT_TRUE(std::isnan(out.real())); + ASSERT_TRUE(std::isnan(out.imag())); +} + +TYPED_TEST(VariousComplexTest, atan) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::atan(x); + }, + [](thrust::complex& x) { + return thrust::atan(x); + }, + -1, + 1, + -1000, + 1000); +} + +TYPED_TEST(VariousComplexTest, catan) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::atan(x); + }, + [](thrust::complex& x) { + return thrust::detail::complex::catan(x); + }, + -1, + 1, + -1000, + 1000); +} + +TEST(VariousComplexTest, acos_long_double) +{ + run_trig_tests( + [](std::complex& x) { + return std::acos(x); + }, + [](thrust::complex& x) { + return thrust::acos(x); + }, + -1, + 1, + 0, + 1); +} + +TEST(VariousComplexTest, asin_long_double) +{ + run_trig_tests( + [](std::complex& x) { + return std::asin(x); + }, + [](thrust::complex& x) { + return thrust::asin(x); + }, + -1, + 1, + 0, + 1); +} + +TEST(VariousComplexTest, atan_long_double) +{ + run_trig_tests( + [](std::complex& x) { + return std::atan(x); + }, + [](thrust::complex& x) { + return thrust::atan(x); + }, + -1, + 1, + 0, + 1); +} + +TEST(VariousComplexTest, acosh_long_double) +{ + run_trig_tests( + [](std::complex& x) { + return std::acosh(x); + }, + [](thrust::complex& x) { + return thrust::acosh(x); + }, + 1, + 100000, + 0, + 10000); +} + +TEST(VariousComplexTest, asinh_long_double) +{ + run_trig_tests( + [](std::complex& x) { + return std::asinh(x); + }, + [](thrust::complex& x) { + return thrust::asinh(x); + }, + -1, + 1, + 0, + 1); +} + +TEST(VariousComplexTest, atanh_long_double) +{ + run_trig_tests( + [](std::complex& x) { + return std::atanh(x); + }, + [](thrust::complex& x) { + return thrust::atanh(x); + }, + -1, + 1, + 0, + 1); +} + +// =================================================================== +// +// CCOSH & CCOSHF +// +// =================================================================== + +TYPED_TEST(VariousComplexTest, cosh) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + // cos(a + bi) = cosh(a) * cos(b) + isinh(a) * sin(b) + double real = std::cosh((double) x.real()) * std::cos((double) x.imag()); + double imag = std::sinh((double) x.real()) * std::sin((double) x.imag()); + + return std::complex((T) real, (T) imag); + }, + [](thrust::complex& x) { + return thrust::cosh(x); + }, + -710, + 710, + std::numeric_limits::min(), + std::numeric_limits::max()); + //-710 and 710 since cosh and sinh is quadratic and anything above 710 will result + // in out of bounds even for double! +} + +TYPED_TEST(VariousComplexTest, cos) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + // cos(a + bi) = cos(a) * cosh(b) − isin(a) * sinh(b) + double real = std::cos((double) x.real()) * std::cosh((double) x.imag()); + double imag = std::sin((double) x.real()) * std::sinh((double) x.imag()); + + SCOPED_TRACE(testing::Message() << real << " " << imag << std::endl); + + return std::complex((T) real, (T) -imag); + }, + [](thrust::complex& x) { + return thrust::cos(x); + }, + std::numeric_limits::min(), + std::numeric_limits::max(), + -710, + 710); + //-710 and 710 since cosh and sinh is quadratic and anything above 710 will result + // in out of bounds even for double! +} + +// =================================================================== +// +// CEXP & CEXPF +// +// =================================================================== + +TYPED_TEST(VariousComplexTest, exp) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + return std::exp(x); + }, + [](thrust::complex& x) { + return thrust::exp(x); + }, + std::numeric_limits::min(), + std::numeric_limits::max(), + std::numeric_limits::min(), + std::numeric_limits::max()); +} + +// // =================================================================== +// // +// // CLOG & CLOGF +// // +// // =================================================================== + +template +std::complex stdLog(std::complex& x) +{ + double t_real = x.real(); + double t_imag = x.imag(); + + double r = std::sqrt(std::pow(t_real, 2) + std::pow(t_imag, 2)); + double theta = std::atan2(t_imag, t_real); + + double real = std::log(r); + double imag = theta; + + return std::complex((DataType) real, (DataType) imag); +} + +TYPED_TEST(VariousComplexTest, log) +{ + using T = typename TestFixture::T; + T max_range = std::sqrt(std::numeric_limits::max() / 5); + + run_trig_tests( + stdLog, + [](thrust::complex& x) { + return thrust::log(x); + }, + -max_range, + max_range, + -max_range, + max_range); +} + +TEST(VariousComplexTest, logf_special_cases) +{ + using T = float; + + auto run_test = [=](T real, T imag) { + std::complex std_in(real, imag); + thrust::complex thrust_in(real, imag); + + std::complex std_out = stdLog(std_in); + thrust::complex thrust_out = thrust::log(thrust_in); + + T real_eps = GET_EPS(std_out.real()); + T imag_eps = GET_EPS(std_out.imag()); + + CHECK_CORRECT(std_out, thrust_out, real_eps, imag_eps, real, imag); + }; + + union converted + { + float out; + uint32_t in; + } c; + + // Testing clogf for case of ay > 1e34 + run_test(2e34f, 3e34f); + // checking cases where real is 1, and imaginary is less than 1 + run_test(1.0f, 5e-20f); + run_test(1.0f, 0.5); + + // checking cases where real or imag is less than 1e-6 + run_test(5e-7f, 0.5); + run_test(0.5f, 5e-7f); + run_test(0.0, 0.5); + run_test(0.5, 0.0); + + // checking cases where real or imag is greater than 1e6 + run_test(5e7f, 0.5); + run_test(0.5f, 5e7f); + + c.in = 0x7f800000; + run_test(c.out, 0.5); // inf + run_test(0.5, c.out); // inf + + // checking case where r <= 0.7 + run_test(0.55, 0.55); + + // checking NAN cases + c.in = 0xffc00001; + run_test(c.out, c.out); +} + +TYPED_TEST(VariousComplexTest, log10) +{ + using T = typename TestFixture::T; + T max_range = std::sqrt(std::numeric_limits::max() / 5); + + auto std_log10 = [](std::complex& x) { + double t_real = x.real(); + double t_imag = x.imag(); + + double r = std::sqrt(std::pow(t_real, 2) + std::pow(t_imag, 2)); + double theta = std::atan2(t_imag, t_real); + + double real = std::log10(r); + double imag = std::log10(std::exp(1.0)) * theta; + + return std::complex((T) real, (T) imag); + }; + + run_trig_tests( + std_log10, + [](thrust::complex& x) { + return thrust::log10(x); + }, + -max_range, + max_range, + -max_range, + max_range); + + run_trig_tests( + std_log10, + [](thrust::complex& x) { + return thrust::log10(x); + }, + -5, + 5, + -5, + 5); +} + +// // =================================================================== +// // +// // CSIN & CSINF +// // +// // =================================================================== + +TYPED_TEST(VariousComplexTest, sinh) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + /** + * Due to hipcc compiler issue, std::sinh(complex) + * is not working properly. Instead we + * will have to calculate this manually + */ + + // sinh(a + bi) = sinh(a) * cos(b) + icosh(a) * sin(b) + + double real = std::sinh((double) x.real()) * std::cos((double) x.imag()); + double imag = std::cosh((double) x.real()) * std::sin((double) x.imag()); + + return std::complex((T) real, (T) imag); + }, + [](thrust::complex& x) { + return thrust::sinh(x); + }, + std::numeric_limits::min(), + std::numeric_limits::max(), + -710, + 710); + //-710 and 710 since cosh and sinh is quadratic and anything above 710 will result + // in out of bounds even for double! +} + +TYPED_TEST(VariousComplexTest, sin) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + // sin(a + bi) = sin(a) * cosh(b) + icos(a) * sinh(b) + + double real = std::sin((double) x.real()) * std::cosh((double) x.imag()); + double imag = std::cos((double) x.real()) * std::sinh((double) x.imag()); + + return std::complex((T) real, (T) imag); + }, + [](thrust::complex& x) { + return thrust::sin(x); + }, + std::numeric_limits::min(), + std::numeric_limits::max(), + -710, + 710); + //-710 and 710 since cosh and sinh is quadratic and anything above 710 will result + // in out of bounds even for double! +} + +// // =================================================================== +// // +// // CSQRT & CSQRTF +// // +// // =================================================================== + +TYPED_TEST(VariousComplexTest, sqrt) +{ + using T = typename TestFixture::T; + + T max_range = std::sqrt(std::numeric_limits::max() / 5); + + run_trig_tests( + [](std::complex& x) { + return std::sqrt(x); + }, + [](thrust::complex& x) { + return thrust::sqrt(x); + }, + -max_range, + max_range, + -max_range, + max_range); +} + +// // =================================================================== +// // +// // CTAN & CTANF +// // +// // =================================================================== + +TYPED_TEST(VariousComplexTest, tanh) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + // tanh(a + bi) = sinh(2a) / (cosh(b) + cos(2a) + isin(2b) / (cosh(b) + cos(2a) + + double t_real = (double) x.real(); + double t_imag = (double) x.imag(); + + double denom = std::cosh(2.0 * t_real) + std::cos(2 * t_imag); + + T real = std::sinh(2.0 * t_real) / denom; + T imag = std::sin(2.0 * t_imag) / denom; + + return std::complex((T) real, (T) imag); + }, + [](thrust::complex& x) { + return thrust::tanh(x); + }, + -710 / 2, + 710 / 2, + std::numeric_limits::min() / 2, + std::numeric_limits::max() / 2); +} + +TYPED_TEST(VariousComplexTest, tan) +{ + using T = typename TestFixture::T; + run_trig_tests( + [](std::complex& x) { + // tan(a + bi) = sin(2a) / (cos(b) + cosh(2a)) + isinh(2b) / (cos(b) + cosh(2a)) + + double t_real = (double) x.real(); + double t_imag = (double) x.imag(); + + double denom = std::cos(2.0 * t_real) + std::cosh(2 * t_imag); + + T real = std::sin(2.0 * t_real) / denom; + T imag = std::sinh(2.0 * t_imag) / denom; + + return std::complex((T) real, (T) imag); + }, + [](thrust::complex& x) { + return thrust::tan(x); + }, + std::numeric_limits::min() / 2, + std::numeric_limits::max() / 2, + -710 / 2, + 710 / 2); + //-710 / 2 and 710 / 2 since cosh and sinh is quadratic and anything above 710 will result + // in out of bounds even for double! We also have to acount the 2 multiplier); +} +#endif // _WIN32 diff --git a/projects/rocthrust/test/test_constant_iterator.cpp b/projects/rocthrust/test/test_constant_iterator.cpp index 09186ccfdff..75974ee311a 100644 --- a/projects/rocthrust/test/test_constant_iterator.cpp +++ b/projects/rocthrust/test/test_constant_iterator.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" using namespace thrust; diff --git a/projects/rocthrust/test/test_copy.cpp b/projects/rocthrust/test/test_copy.cpp index 9cc00493017..2758db8f03d 100644 --- a/projects/rocthrust/test/test_copy.cpp +++ b/projects/rocthrust/test/test_copy.cpp @@ -26,7 +26,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(CopyTests, FullWithLargeTypesTestsParams) TESTS_DEFINE(CopyIntegerTests, IntegerTestsParams) @@ -430,7 +432,7 @@ TYPED_TEST(CopyIntegerTests, TestCopyIf) TEST(CopyLargeTypesTests, TestCopyIfStencilLargeType) { - using T = test::large_data; + using T = large_data; SCOPED_TRACE(testing::Message() << "with device_id= " << test::set_device_from_ctest()); diff --git a/projects/rocthrust/test/test_copy_n.cpp b/projects/rocthrust/test/test_copy_n.cpp index a5217e3b28b..01bbfd3058c 100644 --- a/projects/rocthrust/test/test_copy_n.cpp +++ b/projects/rocthrust/test/test_copy_n.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(CopyNTests, FullTestsParams); TESTS_DEFINE(CopyNPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_count.cpp b/projects/rocthrust/test/test_count.cpp index 0672ca40d19..521ef95903b 100644 --- a/projects/rocthrust/test/test_count.cpp +++ b/projects/rocthrust/test/test_count.cpp @@ -22,7 +22,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(CountTests, FullTestsParams); TESTS_DEFINE(CountPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_counting_iterator.cpp b/projects/rocthrust/test/test_counting_iterator.cpp index 5c915614d83..8fc5830d82f 100644 --- a/projects/rocthrust/test/test_counting_iterator.cpp +++ b/projects/rocthrust/test/test_counting_iterator.cpp @@ -22,7 +22,8 @@ #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(CountingIteratorTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_dereference.cpp b/projects/rocthrust/test/test_dereference.cpp index 0fd5b10c5f8..6813b46f7d5 100644 --- a/projects/rocthrust/test/test_dereference.cpp +++ b/projects/rocthrust/test/test_dereference.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template __global__ THRUST_HIP_LAUNCH_BOUNDS_DEFAULT void diff --git a/projects/rocthrust/test/test_device_delete.cpp b/projects/rocthrust/test/test_device_delete.cpp index 1433697478a..c0c3e1c529a 100644 --- a/projects/rocthrust/test/test_device_delete.cpp +++ b/projects/rocthrust/test/test_device_delete.cpp @@ -20,7 +20,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" struct Foo { diff --git a/projects/rocthrust/test/test_device_ptr.cpp b/projects/rocthrust/test/test_device_ptr.cpp index 2ad7bd090bb..cf4cbfa549c 100644 --- a/projects/rocthrust/test/test_device_ptr.cpp +++ b/projects/rocthrust/test/test_device_ptr.cpp @@ -20,7 +20,8 @@ #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" #define HIP_CHECK_HMM(condition) \ { \ diff --git a/projects/rocthrust/test/test_device_reference.cpp b/projects/rocthrust/test/test_device_reference.cpp index c95815c018f..a67b0217acf 100644 --- a/projects/rocthrust/test/test_device_reference.cpp +++ b/projects/rocthrust/test/test_device_reference.cpp @@ -18,7 +18,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(DeviceReferenceTests, NumericalTestsParams); TESTS_DEFINE(DeviceReferenceIntegerTests, IntegerTestsParams); diff --git a/projects/rocthrust/test/test_discard_iterator.cpp b/projects/rocthrust/test/test_discard_iterator.cpp index 30a5d81a96f..b8f5da7427d 100644 --- a/projects/rocthrust/test/test_discard_iterator.cpp +++ b/projects/rocthrust/test/test_discard_iterator.cpp @@ -17,7 +17,8 @@ #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" using namespace thrust; diff --git a/projects/rocthrust/test/test_distance.cpp b/projects/rocthrust/test/test_distance.cpp index 76b573500ae..4079a983ac9 100644 --- a/projects/rocthrust/test/test_distance.cpp +++ b/projects/rocthrust/test/test_distance.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(DistanceTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_equal.cpp b/projects/rocthrust/test/test_equal.cpp index 4cf1b0544f7..1d71d5e2215 100644 --- a/projects/rocthrust/test/test_equal.cpp +++ b/projects/rocthrust/test/test_equal.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(EqualTests, FullTestsParams); TESTS_DEFINE(EqualsPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_fill.cpp b/projects/rocthrust/test/test_fill.cpp index b35ff7c713f..b4dc4e461c6 100644 --- a/projects/rocthrust/test/test_fill.cpp +++ b/projects/rocthrust/test/test_fill.cpp @@ -23,7 +23,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(FillTests, FullTestsParams); TESTS_DEFINE(FillPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_find.cpp b/projects/rocthrust/test/test_find.cpp index 088efa08ffd..e62af20e1d9 100644 --- a/projects/rocthrust/test/test_find.cpp +++ b/projects/rocthrust/test/test_find.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(FindTestsVector, FullTestsParams); TESTS_DEFINE(FindTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_for_each.cpp b/projects/rocthrust/test/test_for_each.cpp index f5a23246b97..900fe7832fc 100644 --- a/projects/rocthrust/test/test_for_each.cpp +++ b/projects/rocthrust/test/test_for_each.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ForEachTests, SignedIntegerTestsParams) TESTS_DEFINE(ForEachVectorTests, FullTestsParams) diff --git a/projects/rocthrust/test/test_gather.cpp b/projects/rocthrust/test/test_gather.cpp index 1448e890fd8..2b86e01196b 100644 --- a/projects/rocthrust/test/test_gather.cpp +++ b/projects/rocthrust/test/test_gather.cpp @@ -22,7 +22,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(GatherTests, FullTestsParams); TESTS_DEFINE(PrimitiveGatherTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_generate.cpp b/projects/rocthrust/test/test_generate.cpp index 83ebc8d42c3..bbdda493889 100644 --- a/projects/rocthrust/test/test_generate.cpp +++ b/projects/rocthrust/test/test_generate.cpp @@ -19,8 +19,9 @@ #include #include -#include "test_header.hpp" - +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" THRUST_DISABLE_MSVC_POSSIBLE_LOSS_OF_DATA_WARNING_BEGIN using VectorParams = ::testing::Types>, Params>>; diff --git a/projects/rocthrust/test/test_assertions.hpp b/projects/rocthrust/test/test_imag_assertions.hpp similarity index 55% rename from projects/rocthrust/test/test_assertions.hpp rename to projects/rocthrust/test/test_imag_assertions.hpp index 068c15d0920..195b657be6c 100644 --- a/projects/rocthrust/test/test_assertions.hpp +++ b/projects/rocthrust/test/test_imag_assertions.hpp @@ -1,46 +1,12 @@ -/* - * Copyright 2008-2013 NVIDIA Corporation - * Modifications Copyright© 2019-2025 Advanced Micro Devices, Inc. All rights reserved. - * - * Licensed under the Apache License, Version 2.0 (the "License"); - * you may not use this file except in compliance with the License. - * You may obtain a copy of the License at - * - * http://www.apache.org/licenses/LICENSE-2.0 - * - * Unless required by applicable law or agreed to in writing, software - * distributed under the License is distributed on an "AS IS" BASIS, - * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. - * See the License for the specific language governing permissions and - * limitations under the License. - */ - -#include -#include +#pragma once #include #include #include -template -testing::AssertionResult -CmpHelperEQQuite(const char* lhs_expression, const char* rhs_expression, const T1& lhs, const T2& rhs) -{ - if (lhs == rhs) - { - return testing::AssertionSuccess(); - } - - testing::Message msg; - msg << "Expressions during equality check:"; - msg << "\n " << lhs_expression; - msg << "\n " << rhs_expression; - - return testing::AssertionFailure() << msg; -} - -#define ASSERT_EQ_QUIET(val1, val2) ASSERT_PRED_FORMAT2(CmpHelperEQQuite, val1, val2) +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template testing::AssertionResult ComplexCompare( @@ -158,51 +124,3 @@ testing::AssertionResult ComplexVectorNearPredFormat( #define ASSERT_NEAR_COMPLEX_VECTOR(val1, val2) \ ASSERT_NEAR_COMPLEX_VECTOR_ERROR( \ val1, val2, thrust::complex(std::numeric_limits::epsilon(), std::numeric_limits::epsilon())) - -template -testing::AssertionResult bitwise_equal(const char* a_expr, const char* b_expr, const T& a, const T& b) -{ - if (std::memcmp(&a, &b, sizeof(T)) == 0) - { - return testing::AssertionSuccess(); - } - - // googletest's operator<< doesn't see the above overload for int128_t - std::stringstream a_str; - std::stringstream b_str; - a_str << std::hexfloat << a; - b_str << std::hexfloat << b; - - return testing::AssertionFailure() - << "Expected strict/bitwise equality of these values: " << std::endl - << " " << a_expr << ": " << std::hexfloat << a_str.str() << std::endl - << " " << b_expr << ": " << std::hexfloat << b_str.str() << std::endl; -} - -#define ASSERT_BITWISE_EQ(a, b) ASSERT_PRED_FORMAT2(bitwise_equal, a, b) - -template -void assert_bit_eq(IterA result_begin, IterA result_end, IterB expected_begin, IterB expected_end) -{ - using value_a_t = typename std::iterator_traits::value_type; - using value_b_t = typename std::iterator_traits::value_type; - - ASSERT_EQ(std::distance(result_begin, result_end), std::distance(expected_begin, expected_end)); - auto result_it = result_begin; - auto expected_it = expected_begin; - for (size_t index = 0; result_it != result_end; ++result_it, ++expected_it, ++index) - { - // The cast is needed, because the argument can be an std::vector iterator, which's operator* - // returns a proxy object that must be converted to bool - const auto result = static_cast(*result_it); - const auto expected = static_cast(*expected_it); - - ASSERT_BITWISE_EQ(result, expected) << "where index = " << index; - } -} - -template -void assert_bit_eq(const thrust::host_vector& result, const thrust::host_vector& expected) -{ - assert_bit_eq(result.begin(), result.end(), expected.begin(), expected.end()); -} diff --git a/projects/rocthrust/test/test_inner_product.cpp b/projects/rocthrust/test/test_inner_product.cpp index 4799a57a85c..3151691f35f 100644 --- a/projects/rocthrust/test/test_inner_product.cpp +++ b/projects/rocthrust/test/test_inner_product.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(InnerProductTests, FullTestsParams); TESTS_DEFINE(PrimitiveInnerProductTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_is_partitioned.cpp b/projects/rocthrust/test/test_is_partitioned.cpp index 9bb3b73b8e3..2516573a5ec 100644 --- a/projects/rocthrust/test/test_is_partitioned.cpp +++ b/projects/rocthrust/test/test_is_partitioned.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(IsPartitionedTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_is_sorted.cpp b/projects/rocthrust/test/test_is_sorted.cpp index 0f3a6006a45..ea5938489fa 100644 --- a/projects/rocthrust/test/test_is_sorted.cpp +++ b/projects/rocthrust/test/test_is_sorted.cpp @@ -18,7 +18,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(IsSortedTests, FullTestsParams); TESTS_DEFINE(IsSortedVectorTests, VectorSignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_is_sorted_until.cpp b/projects/rocthrust/test/test_is_sorted_until.cpp index 6dfec9acd5f..6476a68e2b7 100644 --- a/projects/rocthrust/test/test_is_sorted_until.cpp +++ b/projects/rocthrust/test/test_is_sorted_until.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(IsSortedUntilTests, FullTestsParams); TESTS_DEFINE(IsSortedUntilVectorTests, VectorSignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_max_element.cpp b/projects/rocthrust/test/test_max_element.cpp index 05c4cd44446..2ed4c465d28 100644 --- a/projects/rocthrust/test/test_max_element.cpp +++ b/projects/rocthrust/test/test_max_element.cpp @@ -18,7 +18,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(MaxElementTests, FullTestsParams); TESTS_DEFINE(MaxElementPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_memory.cpp b/projects/rocthrust/test/test_memory.cpp index bfe03df4c16..705fbfec004 100644 --- a/projects/rocthrust/test/test_memory.cpp +++ b/projects/rocthrust/test/test_memory.cpp @@ -26,7 +26,8 @@ #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TEST(HipThrustMemory, VoidMalloc) { diff --git a/projects/rocthrust/test/test_merge.cpp b/projects/rocthrust/test/test_merge.cpp index a5be3c799fd..46e7099ade7 100644 --- a/projects/rocthrust/test/test_merge.cpp +++ b/projects/rocthrust/test/test_merge.cpp @@ -25,7 +25,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(MergeTests, FullTestsParams); TESTS_DEFINE(PrimitiveMergeTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_merge_by_key.cpp b/projects/rocthrust/test/test_merge_by_key.cpp index 6f54804b969..ac5a78285b8 100644 --- a/projects/rocthrust/test/test_merge_by_key.cpp +++ b/projects/rocthrust/test/test_merge_by_key.cpp @@ -24,7 +24,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template > struct ParamsMerge diff --git a/projects/rocthrust/test/test_min_element.cpp b/projects/rocthrust/test/test_min_element.cpp index cb522936902..fbe2b418417 100644 --- a/projects/rocthrust/test/test_min_element.cpp +++ b/projects/rocthrust/test/test_min_element.cpp @@ -18,7 +18,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(MinElementTests, FullTestsParams); TESTS_DEFINE(MinElementPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_minmax_element.cpp b/projects/rocthrust/test/test_minmax_element.cpp index b22e7243d25..650526f0c22 100644 --- a/projects/rocthrust/test/test_minmax_element.cpp +++ b/projects/rocthrust/test/test_minmax_element.cpp @@ -18,7 +18,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(MinmaxElementTests, FullTestsParams); TESTS_DEFINE(MinMaxElementPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_mismatch.cpp b/projects/rocthrust/test/test_mismatch.cpp index 001e6a4c39f..59aa12122fc 100644 --- a/projects/rocthrust/test/test_mismatch.cpp +++ b/projects/rocthrust/test/test_mismatch.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(MismatchTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_mr_disjoint_pool.cpp b/projects/rocthrust/test/test_mr_disjoint_pool.cpp index ee55c386a0e..5b6143d37af 100644 --- a/projects/rocthrust/test/test_mr_disjoint_pool.cpp +++ b/projects/rocthrust/test/test_mr_disjoint_pool.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" struct alloc_id { diff --git a/projects/rocthrust/test/test_mr_new.cpp b/projects/rocthrust/test/test_mr_new.cpp index 7f146cc70a9..4d7deda0ad7 100644 --- a/projects/rocthrust/test/test_mr_new.cpp +++ b/projects/rocthrust/test/test_mr_new.cpp @@ -18,7 +18,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template void TestAlignment(MemoryResource memres, std::size_t size, std::size_t alignment) diff --git a/projects/rocthrust/test/test_mr_pool.cpp b/projects/rocthrust/test/test_mr_pool.cpp index 5c020bef60e..56fe94bc7ee 100644 --- a/projects/rocthrust/test/test_mr_pool.cpp +++ b/projects/rocthrust/test/test_mr_pool.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template struct reference diff --git a/projects/rocthrust/test/test_mr_pool_options.cpp b/projects/rocthrust/test/test_mr_pool_options.cpp index de6e4761565..3d96c4a9e00 100644 --- a/projects/rocthrust/test/test_mr_pool_options.cpp +++ b/projects/rocthrust/test/test_mr_pool_options.cpp @@ -17,7 +17,8 @@ #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TEST(MrPoolOptionsTests, TestPoolOptionsBasicValidity) { diff --git a/projects/rocthrust/test/test_optional.cpp b/projects/rocthrust/test/test_optional.cpp index 16c4fa9e39a..cd00b3a9d30 100644 --- a/projects/rocthrust/test/test_optional.cpp +++ b/projects/rocthrust/test/test_optional.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TEST(OptionalTests, IsTriviallyCopyable) { diff --git a/projects/rocthrust/test/test_pair.cpp b/projects/rocthrust/test/test_pair.cpp index 3b54eaf8c8c..ec11f551e55 100644 --- a/projects/rocthrust/test/test_pair.cpp +++ b/projects/rocthrust/test/test_pair.cpp @@ -22,7 +22,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PairTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_pair_reduce.cpp b/projects/rocthrust/test/test_pair_reduce.cpp index 7c121f8af60..6e6f1e2e852 100644 --- a/projects/rocthrust/test/test_pair_reduce.cpp +++ b/projects/rocthrust/test/test_pair_reduce.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PairReduceTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_pair_scan.cpp b/projects/rocthrust/test/test_pair_scan.cpp index 483dd91045a..8c187a834cd 100644 --- a/projects/rocthrust/test/test_pair_scan.cpp +++ b/projects/rocthrust/test/test_pair_scan.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PairScanVariablesTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_pair_sort.cpp b/projects/rocthrust/test/test_pair_sort.cpp index 511cd4013fd..b9937383729 100644 --- a/projects/rocthrust/test/test_pair_sort.cpp +++ b/projects/rocthrust/test/test_pair_sort.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PairSortTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_pair_transform.cpp b/projects/rocthrust/test/test_pair_transform.cpp index af4b89b682f..f274ac38f34 100644 --- a/projects/rocthrust/test/test_pair_transform.cpp +++ b/projects/rocthrust/test/test_pair_transform.cpp @@ -20,7 +20,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PairTransformTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_parallel_for.cpp b/projects/rocthrust/test/test_parallel_for.cpp index ea2a493ba72..08e56a5803d 100644 --- a/projects/rocthrust/test/test_parallel_for.cpp +++ b/projects/rocthrust/test/test_parallel_for.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" #if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP # define PARALLEL_FOR thrust::hip_rocprim::parallel_for diff --git a/projects/rocthrust/test/test_header.hpp b/projects/rocthrust/test/test_param_fixtures.hpp similarity index 66% rename from projects/rocthrust/test/test_header.hpp rename to projects/rocthrust/test/test_param_fixtures.hpp index d2f0d4717fa..ce3d16a4088 100644 --- a/projects/rocthrust/test/test_header.hpp +++ b/projects/rocthrust/test/test_param_fixtures.hpp @@ -1,19 +1,26 @@ -/* - * Copyright 2008-2013 NVIDIA Corporation - * Modifications Copyright© 2019-2025 Advanced Micro Devices, Inc. All rights reserved. - * - * Licensed under the Apache License, Version 2.0 (the "License"); - * you may not use this file except in compliance with the License. - * You may obtain a copy of the License at - * - * http://www.apache.org/licenses/LICENSE-2.0 - * - * Unless required by applicable law or agreed to in writing, software - * distributed under the License is distributed on an "AS IS" BASIS, - * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. - * See the License for the specific language governing permissions and - * limitations under the License. - */ +// Copyright (C) 2024-2025 Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in +// all copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +// THE SOFTWARE. + +#pragma once + +#include #include #include @@ -29,145 +36,9 @@ #include "bitwise_repro/bwr_db.hpp" #include -// HIP API -#if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP -# include -# include - -// 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. -# ifndef HIP_CHECK -# define HIP_CHECK(condition) \ - do \ - { \ - hipError_t error = condition; \ - if (error != hipSuccess) \ - { \ - [error]() { \ - FAIL() << "HIP error " << error << ": " << hipGetErrorString(error); \ - }(); \ - exit(error); \ - } \ - } while (0) -# endif - -#endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP - -#include -#include -#include - -#include "test_assertions.hpp" -#include "test_utils.hpp" - -namespace test -{ - -inline char* get_env(const char* name) -{ - char* env; -#ifdef _MSC_VER - size_t len; - errno_t err = _dupenv_s(&env, &len, name); - if (err) - { - return nullptr; - } -#else - env = std::getenv(name); -#endif - return env; -} - -inline void clean_env(char* name) -{ -#ifdef _MSC_VER - if (name != nullptr) - { - free(name); - } -#endif - (void) name; -} - -inline int set_device_from_ctest() -{ - static const std::string rg0 = "CTEST_RESOURCE_GROUP_0"; - char* env = get_env(rg0.c_str()); - int device = 0; - if (env != nullptr) - { - std::string amdgpu_target(env); - std::transform( - amdgpu_target.cbegin(), - amdgpu_target.cend(), - amdgpu_target.begin(), - // Feeding std::toupper plainly results in implicitly truncating conversions between int and char triggering - // warnings. - [](unsigned char c) { - return static_cast(std::toupper(c)); - }); - char* env_reqs = get_env((rg0 + "_" + amdgpu_target).c_str()); - std::string reqs(env_reqs); - device = std::atoi(reqs.substr(reqs.find(':') + 1, reqs.find(',') - (reqs.find(':') + 1)).c_str()); - clean_env(env_reqs); - HIP_CHECK(hipSetDevice(device)); - } - clean_env(env); - return device; -} -} // namespace test - -// If enabled, set up the database for inter-run bitwise reproducibility testing. -// Inter-run testing is enabled through the following environment variables: -// ROCTHRUST_BWR_PATH - path to the database (or where it should be created) -// ROCTHRUST_BWR_GENERATE - if set to 1, info about any function calls not -// found in the database will be inserted. No errors will be reported in this mode. -namespace inter_run_bwr -{ -// Disable this testing by default. -bool enabled = false; - -// This code doesn't need to be visible outside this file. -namespace -{ -const static std::string path_env = "ROCTHRUST_BWR_PATH"; -const static std::string generate_env = "ROCTHRUST_BWR_GENERATE"; - -// Check the environment variables to see if the database should be -// instantiated, and if so, what mode it should be in. -std::unique_ptr create_db() -{ - // Get the path to the database from an environment variable. - const char* db_path = std::getenv(path_env.c_str()); - const char* db_mode = std::getenv(generate_env.c_str()); - if (db_path) - { - // Check if we are allowed to insert rows into the database if - // we encounter calls that aren't already recorded. - BitwiseReproDB::Mode mode = BitwiseReproDB::Mode::test_mode; - if (db_mode && std::stoi(db_mode) > 0) - { - mode = BitwiseReproDB::Mode::generate_mode; - } - - enabled = true; - return std::make_unique(db_path, mode); - } - else if (db_mode) - { - throw std::runtime_error("ROCTHRUST_BWR_GENERATE is defined, but no database path was given.\n" - "Please set ROCTHRUST_BWR_PATH to the database path."); - } - - return nullptr; -} -} // namespace - -// Create/open the run-to-run bitwise reproducibility database. -std::unique_ptr db = create_db(); -} // namespace inter_run_bwr +/** + * test_param_fixtures contains all parameters for typed_test suites + */ // Input type parameter template > @@ -197,8 +68,7 @@ struct Params, ExecutionPolicy> TYPED_TEST_SUITE(x, y); // Set of test parameter types -namespace test -{ + class large_data { public: @@ -241,7 +111,6 @@ bool __host__ __device__ operator==(T const& lhs, large_data const& rhs) { return static_cast(lhs).data[0] == rhs.data[0]; } -} // namespace test // Host and device vectors of all type as a test parameter using FullTestsParams = ::testing::Types< @@ -285,7 +154,7 @@ using FullWithLargeTypesTestsParams = ::testing::Types< Params, std::decay_t>, Params, std::decay_t>, Params>, - Params>>; + Params>>; // Host and device vectors of signed type using VectorSignedTestsParams = @@ -436,3 +305,25 @@ using AllInOutTestsParams = ::testing::Types< ParamsInOut, ParamsInOut, ParamsInOut>; + +// --------------------Pairs test parameters-------- +template + +struct ParamsPairs +{ + using first_type = T; + using second_type = U; +}; + +#define TESTS_PAIRS_DEFINE(x, y) \ + template \ + class x : public ::testing::Test \ + { \ + public: \ + using first_type = typename ParamsPairs::first_type; \ + using second_type = typename ParamsPairs::second_type; \ + }; \ + TYPED_TEST_SUITE(x, y); + +using PairsTestsParams = ::testing:: + Types, ParamsPairs, ParamsPairs, ParamsPairs>; diff --git a/projects/rocthrust/test/test_partition.cpp b/projects/rocthrust/test/test_partition.cpp index e3627fc4f3e..2d072611ed7 100644 --- a/projects/rocthrust/test/test_partition.cpp +++ b/projects/rocthrust/test/test_partition.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PartitionTests, FullTestsParams); TESTS_DEFINE(PartitionVectorTests, VectorSignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_partition_point.cpp b/projects/rocthrust/test/test_partition_point.cpp index a985552122f..9e13848d7de 100644 --- a/projects/rocthrust/test/test_partition_point.cpp +++ b/projects/rocthrust/test/test_partition_point.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PartitionPointVectorTests, VectorSignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_permutation_iterator.cpp b/projects/rocthrust/test/test_permutation_iterator.cpp index f6c7b869b83..26454cf001b 100644 --- a/projects/rocthrust/test/test_permutation_iterator.cpp +++ b/projects/rocthrust/test/test_permutation_iterator.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(PermutationIteratorTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_random.cpp b/projects/rocthrust/test/test_random.cpp index 754657233ab..3e98b68930b 100644 --- a/projects/rocthrust/test/test_random.cpp +++ b/projects/rocthrust/test/test_random.cpp @@ -24,7 +24,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" #if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP diff --git a/projects/rocthrust/test/test_real_assertions.hpp b/projects/rocthrust/test/test_real_assertions.hpp new file mode 100644 index 00000000000..2d722b02021 --- /dev/null +++ b/projects/rocthrust/test/test_real_assertions.hpp @@ -0,0 +1,104 @@ +// Copyright (C) 2024-2025 Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in +// all copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +// THE SOFTWARE. + +/** + * test_real_assertions will contain assertions for normal numbers only, + * complex assertions will be in a seperate file + */ + +#pragma once + +#include + +#include +#include + +#include + +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" + +template +testing::AssertionResult +CmpHelperEQQuite(const char* lhs_expression, const char* rhs_expression, const T1& lhs, const T2& rhs) +{ + if (lhs == rhs) + { + return testing::AssertionSuccess(); + } + + testing::Message msg; + msg << "Expressions during equality check:"; + msg << "\n " << lhs_expression; + msg << "\n " << rhs_expression; + + return testing::AssertionFailure() << msg; +} + +#define ASSERT_EQ_QUIET(val1, val2) ASSERT_PRED_FORMAT2(CmpHelperEQQuite, val1, val2) + +template +testing::AssertionResult bitwise_equal(const char* a_expr, const char* b_expr, const T& a, const T& b) +{ + if (std::memcmp(&a, &b, sizeof(T)) == 0) + { + return testing::AssertionSuccess(); + } + + // googletest's operator<< doesn't see the above overload for int128_t + std::stringstream a_str; + std::stringstream b_str; + a_str << std::hexfloat << a; + b_str << std::hexfloat << b; + + return testing::AssertionFailure() + << "Expected strict/bitwise equality of these values: " << std::endl + << " " << a_expr << ": " << std::hexfloat << a_str.str() << std::endl + << " " << b_expr << ": " << std::hexfloat << b_str.str() << std::endl; +} + +#define ASSERT_BITWISE_EQ(a, b) ASSERT_PRED_FORMAT2(bitwise_equal, a, b) + +template +void assert_bit_eq(IterA result_begin, IterA result_end, IterB expected_begin, IterB expected_end) +{ + using value_a_t = typename std::iterator_traits::value_type; + using value_b_t = typename std::iterator_traits::value_type; + + ASSERT_EQ(std::distance(result_begin, result_end), std::distance(expected_begin, expected_end)); + auto result_it = result_begin; + auto expected_it = expected_begin; + for (size_t index = 0; result_it != result_end; ++result_it, ++expected_it, ++index) + { + // The cast is needed, because the argument can be an std::vector iterator, which's operator* + // returns a proxy object that must be converted to bool + const auto result = static_cast(*result_it); + const auto expected = static_cast(*expected_it); + + ASSERT_BITWISE_EQ(result, expected) << "where index = " << index; + } +} + +template +void assert_bit_eq(const thrust::host_vector& result, const thrust::host_vector& expected) +{ + assert_bit_eq(result.begin(), result.end(), expected.begin(), expected.end()); +} + diff --git a/projects/rocthrust/test/test_reduce.cpp b/projects/rocthrust/test/test_reduce.cpp index e71c6b86c20..e7da07d922d 100644 --- a/projects/rocthrust/test/test_reduce.cpp +++ b/projects/rocthrust/test/test_reduce.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ReduceTests, FullTestsParams); TESTS_DEFINE(ReduceIntegerTests, UnsignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_reduce_by_key.cpp b/projects/rocthrust/test/test_reduce_by_key.cpp index 4a53bcd60f5..8b04282d1a1 100644 --- a/projects/rocthrust/test/test_reduce_by_key.cpp +++ b/projects/rocthrust/test/test_reduce_by_key.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ReduceByKeysIntegralTests, IntegerTestsParams); TESTS_DEFINE(ReduceByKeysTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_remove.cpp b/projects/rocthrust/test/test_remove.cpp index 0a7a97f65e2..490e4bb111e 100644 --- a/projects/rocthrust/test/test_remove.cpp +++ b/projects/rocthrust/test/test_remove.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(RemoveTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_replace.cpp b/projects/rocthrust/test/test_replace.cpp index cbb2cbbf886..5b7ee495bb0 100644 --- a/projects/rocthrust/test/test_replace.cpp +++ b/projects/rocthrust/test/test_replace.cpp @@ -20,7 +20,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ReplaceTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_reproducibility.cpp b/projects/rocthrust/test/test_reproducibility.cpp index c1252a0fea3..e270a50f304 100644 --- a/projects/rocthrust/test/test_reproducibility.cpp +++ b/projects/rocthrust/test/test_reproducibility.cpp @@ -22,7 +22,10 @@ #include #include "bitwise_repro/bwr_utils.hpp" -#include "test_header.hpp" + +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" using ReproducibilityTestParams = ::testing::Types, std::decay_t>, diff --git a/projects/rocthrust/test/test_reverse_iterator.cpp b/projects/rocthrust/test/test_reverse_iterator.cpp index bd59b4cc26b..07c9fb986fb 100644 --- a/projects/rocthrust/test/test_reverse_iterator.cpp +++ b/projects/rocthrust/test/test_reverse_iterator.cpp @@ -19,7 +19,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ReverseIteratorTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_scan.cpp b/projects/rocthrust/test/test_scan.cpp index 3f2bf804686..a70d2c6ba07 100644 --- a/projects/rocthrust/test/test_scan.cpp +++ b/projects/rocthrust/test/test_scan.cpp @@ -22,7 +22,9 @@ #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ScanTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_scan_by_key.cpp b/projects/rocthrust/test/test_scan_by_key.cpp index 67dfa270ff5..a00d63d8aa8 100644 --- a/projects/rocthrust/test/test_scan_by_key.cpp +++ b/projects/rocthrust/test/test_scan_by_key.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ScanByKeyTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_scatter.cpp b/projects/rocthrust/test/test_scatter.cpp index 324b95c7ade..9f716b21cc0 100644 --- a/projects/rocthrust/test/test_scatter.cpp +++ b/projects/rocthrust/test/test_scatter.cpp @@ -23,7 +23,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ScatterTests, FullTestsParams); TESTS_DEFINE(ScatterPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_sequence.cpp b/projects/rocthrust/test/test_sequence.cpp index 995f4fc3811..8993e7825c6 100644 --- a/projects/rocthrust/test/test_sequence.cpp +++ b/projects/rocthrust/test/test_sequence.cpp @@ -21,7 +21,8 @@ #include #include -#include "test_header.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SequenceTests, FullTestsParams); TESTS_DEFINE(PrimitiveSequenceTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_difference.cpp b/projects/rocthrust/test/test_set_difference.cpp index 5bfd264e789..03bdc221c25 100644 --- a/projects/rocthrust/test/test_set_difference.cpp +++ b/projects/rocthrust/test/test_set_difference.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetDifferenceTests, FullTestsParams); TESTS_DEFINE(SetDifferencePrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_difference_by_key.cpp b/projects/rocthrust/test/test_set_difference_by_key.cpp index cf30ebf066c..f9be7aba6a6 100644 --- a/projects/rocthrust/test/test_set_difference_by_key.cpp +++ b/projects/rocthrust/test/test_set_difference_by_key.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(setDifferenceByKeyTests, FullTestsParams); TESTS_DEFINE(SetDifferenceByKeyPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_difference_by_key_descending.cpp b/projects/rocthrust/test/test_set_difference_by_key_descending.cpp index 9e65fa9908a..25517e2f7e9 100644 --- a/projects/rocthrust/test/test_set_difference_by_key_descending.cpp +++ b/projects/rocthrust/test/test_set_difference_by_key_descending.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetDifferenceByKeyDescendingTests, FullTestsParams); TESTS_DEFINE(SetDifferenceByKeyDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_difference_descending.cpp b/projects/rocthrust/test/test_set_difference_descending.cpp index d6d15564e05..d63f7c7a7d9 100644 --- a/projects/rocthrust/test/test_set_difference_descending.cpp +++ b/projects/rocthrust/test/test_set_difference_descending.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetDifferenceDescendingTests, FullTestsParams); TESTS_DEFINE(SetDifferenceDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_intersection.cpp b/projects/rocthrust/test/test_set_intersection.cpp index ad8b855bd37..cf992de7a17 100644 --- a/projects/rocthrust/test/test_set_intersection.cpp +++ b/projects/rocthrust/test/test_set_intersection.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetIntersectionTests, FullTestsParams); TESTS_DEFINE(SetIntersectionPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_intersection_by_key.cpp b/projects/rocthrust/test/test_set_intersection_by_key.cpp index 4c6567afee6..0c086cac461 100644 --- a/projects/rocthrust/test/test_set_intersection_by_key.cpp +++ b/projects/rocthrust/test/test_set_intersection_by_key.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetIntersectionByKeyTests, FullTestsParams); TESTS_DEFINE(SetIntersectionByKeyPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_intersection_by_key_descending.cpp b/projects/rocthrust/test/test_set_intersection_by_key_descending.cpp index 6f310e7b36c..aaa4312c3e4 100644 --- a/projects/rocthrust/test/test_set_intersection_by_key_descending.cpp +++ b/projects/rocthrust/test/test_set_intersection_by_key_descending.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetIntersectionByKeyDescendingTests, FullTestsParams); TESTS_DEFINE(SetIntersectionByKeyDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_intersection_descending.cpp b/projects/rocthrust/test/test_set_intersection_descending.cpp index 169fe37eff1..16a958a840b 100644 --- a/projects/rocthrust/test/test_set_intersection_descending.cpp +++ b/projects/rocthrust/test/test_set_intersection_descending.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetIntersectionDescendingTests, FullTestsParams); TESTS_DEFINE(SetIntersectionDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_intersection_key_value.cpp b/projects/rocthrust/test/test_set_intersection_key_value.cpp index deba5bc015d..ff15adc3545 100644 --- a/projects/rocthrust/test/test_set_intersection_key_value.cpp +++ b/projects/rocthrust/test/test_set_intersection_key_value.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetIntersectionKeyValueTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_symmetric_difference.cpp b/projects/rocthrust/test/test_set_symmetric_difference.cpp index 87a5b3fd885..1a421fdfac2 100644 --- a/projects/rocthrust/test/test_set_symmetric_difference.cpp +++ b/projects/rocthrust/test/test_set_symmetric_difference.cpp @@ -23,7 +23,9 @@ #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetSymmetricDifferenceTests, FullTestsParams); TESTS_DEFINE(SetSymmetricDifferencePrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_symmetric_difference_by_key.cpp b/projects/rocthrust/test/test_set_symmetric_difference_by_key.cpp index 878462bad7a..700d96431f6 100644 --- a/projects/rocthrust/test/test_set_symmetric_difference_by_key.cpp +++ b/projects/rocthrust/test/test_set_symmetric_difference_by_key.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetSymmetricDifferenceByKeyTests, FullTestsParams); TESTS_DEFINE(SetSymmetricDifferenceByKeyPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_symmetric_difference_by_key_descending.cpp b/projects/rocthrust/test/test_set_symmetric_difference_by_key_descending.cpp index a48e073ef58..e8a4cea4725 100644 --- a/projects/rocthrust/test/test_set_symmetric_difference_by_key_descending.cpp +++ b/projects/rocthrust/test/test_set_symmetric_difference_by_key_descending.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetSymmetricDifferenceByKeyDescendingTests, FullTestsParams); TESTS_DEFINE(SetSymmetricDifferenceByKeyDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_symmetric_difference_descending.cpp b/projects/rocthrust/test/test_set_symmetric_difference_descending.cpp index fdab3789a36..5be891ff9ba 100644 --- a/projects/rocthrust/test/test_set_symmetric_difference_descending.cpp +++ b/projects/rocthrust/test/test_set_symmetric_difference_descending.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetSymmetricDifferenceDescendingTests, FullTestsParams); TESTS_DEFINE(SetSymmetricDifferenceDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_union.cpp b/projects/rocthrust/test/test_set_union.cpp index bf5168c89af..4333ef293f4 100644 --- a/projects/rocthrust/test/test_set_union.cpp +++ b/projects/rocthrust/test/test_set_union.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetUnionTests, FullTestsParams); TESTS_DEFINE(SetUnionPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_union_by_key.cpp b/projects/rocthrust/test/test_set_union_by_key.cpp index bb8fd0e1446..4499a6c3f8b 100644 --- a/projects/rocthrust/test/test_set_union_by_key.cpp +++ b/projects/rocthrust/test/test_set_union_by_key.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetUnionByKeyTests, FullTestsParams); TESTS_DEFINE(SetUnionByKeyPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_union_by_key_descending.cpp b/projects/rocthrust/test/test_set_union_by_key_descending.cpp index 4033680c942..f6e520a6875 100644 --- a/projects/rocthrust/test/test_set_union_by_key_descending.cpp +++ b/projects/rocthrust/test/test_set_union_by_key_descending.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetUnionByKeyDescendingTests, FullTestsParams); TESTS_DEFINE(SetUnionByKeyDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_union_descending.cpp b/projects/rocthrust/test/test_set_union_descending.cpp index 4cc4fa23190..d4d0430aaa1 100644 --- a/projects/rocthrust/test/test_set_union_descending.cpp +++ b/projects/rocthrust/test/test_set_union_descending.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetUnionDescendingTests, FullTestsParams); TESTS_DEFINE(SetUnionDescendingPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_set_union_key_value.cpp b/projects/rocthrust/test/test_set_union_key_value.cpp index 1173e9a4d36..f8399df2d61 100644 --- a/projects/rocthrust/test/test_set_union_key_value.cpp +++ b/projects/rocthrust/test/test_set_union_key_value.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SetUnionKeyValueTests, FullTestsParams); TESTS_DEFINE(SetUnionKeyValuePrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_shuffle.cpp b/projects/rocthrust/test/test_shuffle.cpp index 3f6e7bb9278..e961b666f15 100644 --- a/projects/rocthrust/test/test_shuffle.cpp +++ b/projects/rocthrust/test/test_shuffle.cpp @@ -25,7 +25,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ShuffleTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_sort.cpp b/projects/rocthrust/test/test_sort.cpp index 79173d291a5..ba25697ad31 100644 --- a/projects/rocthrust/test/test_sort.cpp +++ b/projects/rocthrust/test/test_sort.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template > struct ParamsSort diff --git a/projects/rocthrust/test/test_sort_by_key.cpp b/projects/rocthrust/test/test_sort_by_key.cpp index 4c3ee94ce6a..75edc70e416 100644 --- a/projects/rocthrust/test/test_sort_by_key.cpp +++ b/projects/rocthrust/test/test_sort_by_key.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SortByKeyTests, FullTestsParams); TESTS_DEFINE(SortByKeyPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_sort_by_key_variable_bits.cpp b/projects/rocthrust/test/test_sort_by_key_variable_bits.cpp index 53abb66cc28..c3d9d4cc50e 100644 --- a/projects/rocthrust/test/test_sort_by_key_variable_bits.cpp +++ b/projects/rocthrust/test/test_sort_by_key_variable_bits.cpp @@ -17,7 +17,9 @@ #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SortByKeyVariableTests, UnsignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_sort_permutation_iterator.cpp b/projects/rocthrust/test/test_sort_permutation_iterator.cpp index 83777ae54d4..4085a8fff0d 100644 --- a/projects/rocthrust/test/test_sort_permutation_iterator.cpp +++ b/projects/rocthrust/test/test_sort_permutation_iterator.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SortPermutationIteratorsTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_sort_variables.cpp b/projects/rocthrust/test/test_sort_variables.cpp index 2d2356e9690..7f4a706a665 100644 --- a/projects/rocthrust/test/test_sort_variables.cpp +++ b/projects/rocthrust/test/test_sort_variables.cpp @@ -17,7 +17,9 @@ #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SortByKeyVariableTests, AllIntegerTestsParams); diff --git a/projects/rocthrust/test/test_stable_sort.cpp b/projects/rocthrust/test/test_stable_sort.cpp index c6120942382..24b80f33837 100644 --- a/projects/rocthrust/test/test_stable_sort.cpp +++ b/projects/rocthrust/test/test_stable_sort.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(StableSortTests, UnsignedIntegerTestsParams); TESTS_DEFINE(StableSortVectorTests, VectorIntegerTestsParams); diff --git a/projects/rocthrust/test/test_stable_sort_by_key.cpp b/projects/rocthrust/test/test_stable_sort_by_key.cpp index 7fff9b0b3ed..48ffec6e0db 100644 --- a/projects/rocthrust/test/test_stable_sort_by_key.cpp +++ b/projects/rocthrust/test/test_stable_sort_by_key.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(StableSortByKeyTests, UnsignedIntegerTestsParams); TESTS_DEFINE(StableSortByKeyVectorTests, VectorIntegerTestsParams); diff --git a/projects/rocthrust/test/test_stable_sort_by_key_large_keys.cpp b/projects/rocthrust/test/test_stable_sort_by_key_large_keys.cpp index ee3530fdc99..554f0983bc4 100644 --- a/projects/rocthrust/test/test_stable_sort_by_key_large_keys.cpp +++ b/projects/rocthrust/test/test_stable_sort_by_key_large_keys.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template void _TestStableSortByKeyWithLargeKeys(void) diff --git a/projects/rocthrust/test/test_stable_sort_by_key_large_keys_and_values.cpp b/projects/rocthrust/test/test_stable_sort_by_key_large_keys_and_values.cpp index 68bb81f97b7..86d966bc560 100644 --- a/projects/rocthrust/test/test_stable_sort_by_key_large_keys_and_values.cpp +++ b/projects/rocthrust/test/test_stable_sort_by_key_large_keys_and_values.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template void _TestStableSortByKeyWithLargeKeysAndValues() diff --git a/projects/rocthrust/test/test_stable_sort_by_key_large_values.cpp b/projects/rocthrust/test/test_stable_sort_by_key_large_values.cpp index 87479aab6ed..2222a3da77f 100644 --- a/projects/rocthrust/test/test_stable_sort_by_key_large_values.cpp +++ b/projects/rocthrust/test/test_stable_sort_by_key_large_values.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template struct greater_div_10 diff --git a/projects/rocthrust/test/test_stable_sort_large.cpp b/projects/rocthrust/test/test_stable_sort_large.cpp index 1e460507375..445b8ba5a16 100644 --- a/projects/rocthrust/test/test_stable_sort_large.cpp +++ b/projects/rocthrust/test/test_stable_sort_large.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template void _TestStableSortWithLargeKeys(void) diff --git a/projects/rocthrust/test/test_swap_ranges.cpp b/projects/rocthrust/test/test_swap_ranges.cpp index 04bc99c2a6f..7b9f9c0ef57 100644 --- a/projects/rocthrust/test/test_swap_ranges.cpp +++ b/projects/rocthrust/test/test_swap_ranges.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(SwapRangesTests, FullTestsParams); TESTS_DEFINE(PrimitiveSwapRangesTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_tabulate.cpp b/projects/rocthrust/test/test_tabulate.cpp index af877051d72..3a1c7a0c5dd 100644 --- a/projects/rocthrust/test/test_tabulate.cpp +++ b/projects/rocthrust/test/test_tabulate.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TabulateTests, FullTestsParams); TESTS_DEFINE(TabulatePrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_transform.cpp b/projects/rocthrust/test/test_transform.cpp index 04af3c149ff..215976f8348 100644 --- a/projects/rocthrust/test/test_transform.cpp +++ b/projects/rocthrust/test/test_transform.cpp @@ -26,7 +26,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_INOUT_DEFINE(TransformTests, AllInOutTestsParams); diff --git a/projects/rocthrust/test/test_transform_iterator.cpp b/projects/rocthrust/test/test_transform_iterator.cpp index 3a9c68c2742..35d291b9bcf 100644 --- a/projects/rocthrust/test/test_transform_iterator.cpp +++ b/projects/rocthrust/test/test_transform_iterator.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TransformIteratorTests, FullTestsParams); TESTS_DEFINE(PrimitiveTransformIteratorTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_transform_reduce.cpp b/projects/rocthrust/test/test_transform_reduce.cpp index 6710959d0e1..ef3b76125dc 100644 --- a/projects/rocthrust/test/test_transform_reduce.cpp +++ b/projects/rocthrust/test/test_transform_reduce.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TransformReduceTests, FullTestsParams); TESTS_DEFINE(TransformReduceIntegerTests, VectorSignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_transform_scan.cpp b/projects/rocthrust/test/test_transform_scan.cpp index bb3dd6158de..e0f998ee41d 100644 --- a/projects/rocthrust/test/test_transform_scan.cpp +++ b/projects/rocthrust/test/test_transform_scan.cpp @@ -22,7 +22,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TransformScanTests, FullTestsParams); TESTS_DEFINE(TransformScanVariablesTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_tuple.cpp b/projects/rocthrust/test/test_tuple.cpp index 73aff5ad9e2..5f70dafb927 100644 --- a/projects/rocthrust/test/test_tuple.cpp +++ b/projects/rocthrust/test/test_tuple.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TupleTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_tuple_reduce.cpp b/projects/rocthrust/test/test_tuple_reduce.cpp index 2e84309396c..1109e885dda 100644 --- a/projects/rocthrust/test/test_tuple_reduce.cpp +++ b/projects/rocthrust/test/test_tuple_reduce.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TupleReduceTests, IntegerTestsParams); diff --git a/projects/rocthrust/test/test_tuple_sort.cpp b/projects/rocthrust/test/test_tuple_sort.cpp index a754463c422..8b4c2eb8283 100644 --- a/projects/rocthrust/test/test_tuple_sort.cpp +++ b/projects/rocthrust/test/test_tuple_sort.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TupleSortTests, IntegerTestsParams); diff --git a/projects/rocthrust/test/test_tuple_transform.cpp b/projects/rocthrust/test/test_tuple_transform.cpp index f58bc358028..8f90995a0e8 100644 --- a/projects/rocthrust/test/test_tuple_transform.cpp +++ b/projects/rocthrust/test/test_tuple_transform.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(TupleTransformTests, SignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_uninitialized_copy.cpp b/projects/rocthrust/test/test_uninitialized_copy.cpp index 4cbb9a54f75..ccdfce6a888 100644 --- a/projects/rocthrust/test/test_uninitialized_copy.cpp +++ b/projects/rocthrust/test/test_uninitialized_copy.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(UninitializedCopyTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_uninitialized_fill.cpp b/projects/rocthrust/test/test_uninitialized_fill.cpp index d62ec744bd8..99495ddca51 100644 --- a/projects/rocthrust/test/test_uninitialized_fill.cpp +++ b/projects/rocthrust/test/test_uninitialized_fill.cpp @@ -23,7 +23,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(UninitializedFillTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_unique.cpp b/projects/rocthrust/test/test_unique.cpp index 214a8d1aa40..8865e95edd0 100644 --- a/projects/rocthrust/test/test_unique.cpp +++ b/projects/rocthrust/test/test_unique.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(UniqueTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_unique_by_key.cpp b/projects/rocthrust/test/test_unique_by_key.cpp index b5976aa164d..8997708b783 100644 --- a/projects/rocthrust/test/test_unique_by_key.cpp +++ b/projects/rocthrust/test/test_unique_by_key.cpp @@ -20,7 +20,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(UniqueByKeyTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_universal_memory.cpp b/projects/rocthrust/test/test_universal_memory.cpp index ce650cda75d..e3151968f47 100644 --- a/projects/rocthrust/test/test_universal_memory.cpp +++ b/projects/rocthrust/test/test_universal_memory.cpp @@ -23,7 +23,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(UniversalTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_utils.hpp b/projects/rocthrust/test/test_utils.hpp index 0135f621da2..2970eaa43b4 100644 --- a/projects/rocthrust/test/test_utils.hpp +++ b/projects/rocthrust/test/test_utils.hpp @@ -46,6 +46,140 @@ #include #include +// HIP API +#if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP +# include +# include + +// 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. +# ifndef HIP_CHECK +# define HIP_CHECK(condition) \ + do \ + { \ + hipError_t error = condition; \ + if (error != hipSuccess) \ + { \ + [error]() { \ + FAIL() << "HIP error " << error << ": " << hipGetErrorString(error); \ + }(); \ + exit(error); \ + } \ + } while (0) +# endif + +#endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP + + +namespace test +{ + +inline char* get_env(const char* name) +{ + char* env; +#ifdef _MSC_VER + size_t len; + errno_t err = _dupenv_s(&env, &len, name); + if (err) + { + return nullptr; + } +#else + env = std::getenv(name); +#endif + return env; +} + +inline void clean_env(char* name) +{ +#ifdef _MSC_VER + if (name != nullptr) + { + free(name); + } +#endif + (void) name; +} + +inline int set_device_from_ctest() +{ + static const std::string rg0 = "CTEST_RESOURCE_GROUP_0"; + char* env = get_env(rg0.c_str()); + int device = 0; + if (env != nullptr) + { + std::string amdgpu_target(env); + std::transform( + amdgpu_target.cbegin(), + amdgpu_target.cend(), + amdgpu_target.begin(), + // Feeding std::toupper plainly results in implicitly truncating conversions between int and char triggering + // warnings. + [](unsigned char c) { + return static_cast(std::toupper(c)); + }); + char* env_reqs = get_env((rg0 + "_" + amdgpu_target).c_str()); + std::string reqs(env_reqs); + device = std::atoi(reqs.substr(reqs.find(':') + 1, reqs.find(',') - (reqs.find(':') + 1)).c_str()); + clean_env(env_reqs); + HIP_CHECK(hipSetDevice(device)); + } + clean_env(env); + return device; +} +} // namespace test + +// If enabled, set up the database for inter-run bitwise reproducibility testing. +// Inter-run testing is enabled through the following environment variables: +// ROCTHRUST_BWR_PATH - path to the database (or where it should be created) +// ROCTHRUST_BWR_GENERATE - if set to 1, info about any function calls not +// found in the database will be inserted. No errors will be reported in this mode. +namespace inter_run_bwr +{ +// Disable this testing by default. +bool enabled = false; + +// This code doesn't need to be visible outside this file. +namespace +{ +const static std::string path_env = "ROCTHRUST_BWR_PATH"; +const static std::string generate_env = "ROCTHRUST_BWR_GENERATE"; + +// Check the environment variables to see if the database should be +// instantiated, and if so, what mode it should be in. +std::unique_ptr create_db() +{ + // Get the path to the database from an environment variable. + const char* db_path = std::getenv(path_env.c_str()); + const char* db_mode = std::getenv(generate_env.c_str()); + if (db_path) + { + // Check if we are allowed to insert rows into the database if + // we encounter calls that aren't already recorded. + BitwiseReproDB::Mode mode = BitwiseReproDB::Mode::test_mode; + if (db_mode && std::stoi(db_mode) > 0) + { + mode = BitwiseReproDB::Mode::generate_mode; + } + + enabled = true; + return std::make_unique(db_path, mode); + } + else if (db_mode) + { + throw std::runtime_error("ROCTHRUST_BWR_GENERATE is defined, but no database path was given.\n" + "Please set ROCTHRUST_BWR_PATH to the database path."); + } + + return nullptr; +} +} // namespace + +// Create/open the run-to-run bitwise reproducibility database. +std::unique_ptr db = create_db(); +} // namespace inter_run_bwr + #ifdef __GNUC__ inline std::string demangle(const char* name) { diff --git a/projects/rocthrust/test/test_vector.cpp b/projects/rocthrust/test/test_vector.cpp index 46c4d2d2471..c25cdbfffe8 100644 --- a/projects/rocthrust/test/test_vector.cpp +++ b/projects/rocthrust/test/test_vector.cpp @@ -19,7 +19,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(VectorTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_vector_allocators.cpp b/projects/rocthrust/test/test_vector_allocators.cpp index 4800caf25b2..b7cd8216b2f 100644 --- a/projects/rocthrust/test/test_vector_allocators.cpp +++ b/projects/rocthrust/test/test_vector_allocators.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" template class stateful_allocator : public BaseAlloc diff --git a/projects/rocthrust/test/test_vector_insert.cpp b/projects/rocthrust/test/test_vector_insert.cpp index 9ecde7e1942..3afd34436ed 100644 --- a/projects/rocthrust/test/test_vector_insert.cpp +++ b/projects/rocthrust/test/test_vector_insert.cpp @@ -17,7 +17,9 @@ #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(VectorInsertTests, FullTestsParams); TESTS_DEFINE(VectorInsertPrimitiveTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_vector_manipulation.cpp b/projects/rocthrust/test/test_vector_manipulation.cpp index fc406b67ae5..6f766bcae05 100644 --- a/projects/rocthrust/test/test_vector_manipulation.cpp +++ b/projects/rocthrust/test/test_vector_manipulation.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(VectorManipulationTests, FullTestsParams); diff --git a/projects/rocthrust/test/test_zip_iterator.cpp b/projects/rocthrust/test/test_zip_iterator.cpp index 27d3244d7e2..bb6494b98c4 100644 --- a/projects/rocthrust/test/test_zip_iterator.cpp +++ b/projects/rocthrust/test/test_zip_iterator.cpp @@ -21,7 +21,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" using ZipIteratorTests32BitParams = ::testing::Types, Params, Params>; diff --git a/projects/rocthrust/test/test_zip_iterator_reduce.cpp b/projects/rocthrust/test/test_zip_iterator_reduce.cpp index c02125ac8fa..49e950ebea9 100644 --- a/projects/rocthrust/test/test_zip_iterator_reduce.cpp +++ b/projects/rocthrust/test/test_zip_iterator_reduce.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ZipIteratorReduceTests, IntegerTestsParams); diff --git a/projects/rocthrust/test/test_zip_iterator_reduce_by_key.cpp b/projects/rocthrust/test/test_zip_iterator_reduce_by_key.cpp index 6677a849b29..3479a4ecd32 100644 --- a/projects/rocthrust/test/test_zip_iterator_reduce_by_key.cpp +++ b/projects/rocthrust/test/test_zip_iterator_reduce_by_key.cpp @@ -22,7 +22,9 @@ # include #endif -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" using UnsignedIntegralTypesParams = ::testing::Types, Params, Params, Params>; diff --git a/projects/rocthrust/test/test_zip_iterator_scan.cpp b/projects/rocthrust/test/test_zip_iterator_scan.cpp index 7fd1b8dce39..299b8ee8381 100644 --- a/projects/rocthrust/test/test_zip_iterator_scan.cpp +++ b/projects/rocthrust/test/test_zip_iterator_scan.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ZipIteratorScanVariablesTests, NumericalTestsParams); diff --git a/projects/rocthrust/test/test_zip_iterator_sort.cpp b/projects/rocthrust/test/test_zip_iterator_sort.cpp index 0dda5d97b47..48ad662a796 100644 --- a/projects/rocthrust/test/test_zip_iterator_sort.cpp +++ b/projects/rocthrust/test/test_zip_iterator_sort.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" TESTS_DEFINE(ZipIteratorStableSortTests, UnsignedIntegerTestsParams); diff --git a/projects/rocthrust/test/test_zip_iterator_sort_by_key.cpp b/projects/rocthrust/test/test_zip_iterator_sort_by_key.cpp index 9be23aa2e59..1dd6ae632d8 100644 --- a/projects/rocthrust/test/test_zip_iterator_sort_by_key.cpp +++ b/projects/rocthrust/test/test_zip_iterator_sort_by_key.cpp @@ -18,7 +18,9 @@ #include #include -#include "test_header.hpp" +#include "test_real_assertions.hpp" +#include "test_param_fixtures.hpp" +#include "test_utils.hpp" #if defined(_WIN32) && defined(__HIP__) using TestParams = ::testing::Types, Params>;