From ef350b0a77518c9600b73de4de35578806de0323 Mon Sep 17 00:00:00 2001 From: Allison Vacanti Date: Tue, 15 Dec 2020 16:29:30 -0500 Subject: [PATCH] WIP, dropping more suppressions. --- cmake/ThrustBuildCompilerTargets.cmake | 12 +-- testing/binary_search_descending.cu | 102 +++++++++++----------- testing/constant_iterator.cu | 5 +- testing/mr_new.cu | 2 +- testing/pair_reduce.cu | 2 +- testing/unittest/assertions.h | 2 +- thrust/system/detail/generic/sequence.inl | 21 ++++- thrust/system/detail/generic/shuffle.inl | 5 +- 8 files changed, 84 insertions(+), 67 deletions(-) diff --git a/cmake/ThrustBuildCompilerTargets.cmake b/cmake/ThrustBuildCompilerTargets.cmake index 38b1d1dc74..8c8f2c391b 100644 --- a/cmake/ThrustBuildCompilerTargets.cmake +++ b/cmake/ThrustBuildCompilerTargets.cmake @@ -31,7 +31,7 @@ function(thrust_build_compiler_targets) if (THRUST_TBB_FOUND) # There's a ton of these in the TBB backend, even though the code is correct. # TODO: silence these warnings in code instead - append_option_if_available("-Wno-unused-parameter" cxx_compile_options) +# append_option_if_available("-Wno-unused-parameter" cxx_compile_options) endif() if ("MSVC" STREQUAL "${CMAKE_CXX_COMPILER_ID}") @@ -61,15 +61,15 @@ function(thrust_build_compiler_targets) # Disabled loss-of-data conversion warnings. # TODO Re-enable. - append_option_if_available("/wd4244" cxx_compile_options) - append_option_if_available("/wd4267" cxx_compile_options) +# append_option_if_available("/wd4244" cxx_compile_options) +# append_option_if_available("/wd4267" cxx_compile_options) # Suppress numeric conversion-to-bool warnings. # TODO Re-enable. - append_option_if_available("/wd4800" cxx_compile_options) +# append_option_if_available("/wd4800" cxx_compile_options) # Disable warning about applying unary operator- to unsigned type. - append_option_if_available("/wd4146" cxx_compile_options) +# append_option_if_available("/wd4146" cxx_compile_options) # MSVC STL assumes that `allocator_traits`'s allocator will use raw pointers, # and the `__DECLSPEC_ALLOCATOR` macro causes issues with thrust's universal @@ -122,7 +122,7 @@ function(thrust_build_compiler_targets) # xlC and Clang warn about unused parameters in uninstantiated templates. # This causes xlC to choke on the OMP backend, which is mostly #ifdef'd out # (and thus has unused parameters) when you aren't using it. - append_option_if_available("-Wno-unused-parameters" cxx_compile_options) +# append_option_if_available("-Wno-unused-parameters" cxx_compile_options) endif() if ("Feta" STREQUAL "${CMAKE_CUDA_COMPILER_ID}") diff --git a/testing/binary_search_descending.cu b/testing/binary_search_descending.cu index 5228c45673..08294c044f 100644 --- a/testing/binary_search_descending.cu +++ b/testing/binary_search_descending.cu @@ -22,16 +22,16 @@ void TestScalarLowerBoundDescendingSimple(void) vec[3] = 2; vec[4] = 0; - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::lower_bound(vec.begin(), vec.end(), 0, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::lower_bound(vec.begin(), vec.end(), 1, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), 2, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), 3, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), 4, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::lower_bound(vec.begin(), vec.end(), 5, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::lower_bound(vec.begin(), vec.end(), 6, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::lower_bound(vec.begin(), vec.end(), 7, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::lower_bound(vec.begin(), vec.end(), 8, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::lower_bound(vec.begin(), vec.end(), 9, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::lower_bound(vec.begin(), vec.end(), T{0}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::lower_bound(vec.begin(), vec.end(), T{1}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), T{2}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), T{3}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), T{4}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::lower_bound(vec.begin(), vec.end(), T{5}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::lower_bound(vec.begin(), vec.end(), T{6}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::lower_bound(vec.begin(), vec.end(), T{7}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::lower_bound(vec.begin(), vec.end(), T{8}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::lower_bound(vec.begin(), vec.end(), T{9}, thrust::greater())); } DECLARE_VECTOR_UNITTEST(TestScalarLowerBoundDescendingSimple); @@ -49,16 +49,16 @@ void TestScalarUpperBoundDescendingSimple(void) vec[3] = 2; vec[4] = 0; - ASSERT_EQUAL_QUIET(vec.begin() + 5, thrust::upper_bound(vec.begin(), vec.end(), 0, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::upper_bound(vec.begin(), vec.end(), 1, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::upper_bound(vec.begin(), vec.end(), 2, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), 3, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), 4, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), 5, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::upper_bound(vec.begin(), vec.end(), 6, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::upper_bound(vec.begin(), vec.end(), 7, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::upper_bound(vec.begin(), vec.end(), 8, thrust::greater())); - ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::upper_bound(vec.begin(), vec.end(), 9, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 5, thrust::upper_bound(vec.begin(), vec.end(), T{0}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::upper_bound(vec.begin(), vec.end(), T{1}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::upper_bound(vec.begin(), vec.end(), T{2}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), T{3}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), T{4}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), T{5}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::upper_bound(vec.begin(), vec.end(), T{6}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::upper_bound(vec.begin(), vec.end(), T{7}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::upper_bound(vec.begin(), vec.end(), T{8}, thrust::greater())); + ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::upper_bound(vec.begin(), vec.end(), T{9}, thrust::greater())); } DECLARE_VECTOR_UNITTEST(TestScalarUpperBoundDescendingSimple); @@ -76,16 +76,16 @@ void TestScalarBinarySearchDescendingSimple(void) vec[3] = 2; vec[4] = 0; - ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), 0, thrust::greater())); - ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), 1, thrust::greater())); - ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), 2, thrust::greater())); - ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), 3, thrust::greater())); - ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), 4, thrust::greater())); - ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), 5, thrust::greater())); - ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), 6, thrust::greater())); - ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), 7, thrust::greater())); - ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), 8, thrust::greater())); - ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), 9, thrust::greater())); + ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{0}, thrust::greater())); + ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{1}, thrust::greater())); + ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{2}, thrust::greater())); + ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{3}, thrust::greater())); + ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{4}, thrust::greater())); + ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{5}, thrust::greater())); + ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{6}, thrust::greater())); + ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{7}, thrust::greater())); + ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{8}, thrust::greater())); + ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{9}, thrust::greater())); } DECLARE_VECTOR_UNITTEST(TestScalarBinarySearchDescendingSimple); @@ -103,27 +103,27 @@ void TestScalarEqualRangeDescendingSimple(void) vec[3] = 2; vec[4] = 0; - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), 0, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), 1, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), 2, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), 3, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), 4, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), 5, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), 6, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::equal_range(vec.begin(), vec.end(), 7, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::equal_range(vec.begin(), vec.end(), 8, thrust::greater()).first); - ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::equal_range(vec.begin(), vec.end(), 9, thrust::greater()).first); - - ASSERT_EQUAL_QUIET(vec.begin() + 5, thrust::equal_range(vec.begin(), vec.end(), 0, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), 1, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), 2, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), 3, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), 4, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), 5, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), 6, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), 7, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::equal_range(vec.begin(), vec.end(), 8, thrust::greater()).second); - ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::equal_range(vec.begin(), vec.end(), 9, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), T{0}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), T{1}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), T{2}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), T{3}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), T{4}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), T{5}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), T{6}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::equal_range(vec.begin(), vec.end(), T{7}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::equal_range(vec.begin(), vec.end(), T{8}, thrust::greater()).first); + ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::equal_range(vec.begin(), vec.end(), T{9}, thrust::greater()).first); + + ASSERT_EQUAL_QUIET(vec.begin() + 5, thrust::equal_range(vec.begin(), vec.end(), T{0}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), T{1}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::equal_range(vec.begin(), vec.end(), T{2}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), T{3}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), T{4}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::equal_range(vec.begin(), vec.end(), T{5}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), T{6}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::equal_range(vec.begin(), vec.end(), T{7}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::equal_range(vec.begin(), vec.end(), T{8}, thrust::greater()).second); + ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::equal_range(vec.begin(), vec.end(), T{9}, thrust::greater()).second); } DECLARE_VECTOR_UNITTEST(TestScalarEqualRangeDescendingSimple); diff --git a/testing/constant_iterator.cu b/testing/constant_iterator.cu index cbf771c9ab..e42cfea8d3 100644 --- a/testing/constant_iterator.cu +++ b/testing/constant_iterator.cu @@ -109,11 +109,12 @@ void TestConstantIteratorCopy(void) { using namespace thrust; - typedef constant_iterator ConstIter; + using ValueType = typename Vector::value_type; + using ConstIter = constant_iterator; Vector result(4); - ConstIter first = make_constant_iterator(7); + ConstIter first = make_constant_iterator(7); ConstIter last = first + result.size(); thrust::copy(first, last, result.begin()); diff --git a/testing/mr_new.cu b/testing/mr_new.cu index df0f3fde54..02f34eccf0 100644 --- a/testing/mr_new.cu +++ b/testing/mr_new.cu @@ -9,7 +9,7 @@ void TestAlignment(MemoryResource memres, std::size_t size, std::size_t alignmen ASSERT_EQUAL(reinterpret_cast(ptr) % alignment, 0u); char * char_ptr = reinterpret_cast(ptr); - thrust::fill(char_ptr, char_ptr + size, 0); + thrust::fill(char_ptr, char_ptr + size, char{}); memres.do_deallocate(ptr, size, alignment); } diff --git a/testing/pair_reduce.cu b/testing/pair_reduce.cu index ebdab65979..120f8c1516 100644 --- a/testing/pair_reduce.cu +++ b/testing/pair_reduce.cu @@ -43,7 +43,7 @@ template thrust::device_vector d_p2 = h_p2; thrust::device_vector

d_pairs = h_pairs; - P init = thrust::make_pair(13,13); + P init = thrust::make_pair(T{13}, T{13}); // reduce on the host P h_result = thrust::reduce(h_pairs.begin(), h_pairs.end(), init, add_pairs()); diff --git a/testing/unittest/assertions.h b/testing/unittest/assertions.h index 3528e09b9c..5e33c297f9 100644 --- a/testing/unittest/assertions.h +++ b/testing/unittest/assertions.h @@ -428,7 +428,7 @@ void assert_equal(ForwardIterator1 first1, ForwardIterator1 last1, ForwardIterat if(mismatches <= MAX_OUTPUT_LINES) { - if (sizeof(InputType) == 1) + THRUST_IF_CONSTEXPR(sizeof(InputType) == 1) f << " [" << i << "] " << *first1 + InputType() << " " << *first2 + InputType() << "\n"; // unprintable chars are a problem else f << " [" << i << "] " << *first1 << " " << *first2 << "\n"; diff --git a/thrust/system/detail/generic/sequence.inl b/thrust/system/detail/generic/sequence.inl index 16631c7f43..2e7622a405 100644 --- a/thrust/system/detail/generic/sequence.inl +++ b/thrust/system/detail/generic/sequence.inl @@ -15,7 +15,6 @@ */ #include -#include #include #include #include @@ -52,6 +51,19 @@ __host__ __device__ thrust::sequence(exec, first, last, init, T(1)); } // end sequence() +namespace detail +{ +template +struct compute_sequence_value +{ + T init; + T step; + T operator()(std::size_t i) const + { + return init + step * static_cast(i); + } +}; +} template __host__ __device__ @@ -61,9 +73,12 @@ __host__ __device__ T init, T step) { - using thrust::placeholders::_1; - thrust::tabulate(exec, first, last, init + step * _1); + thrust::tabulate(exec, + first, + last, + detail::compute_sequence_value{std::move(init), + std::move(step)}); } // end sequence() diff --git a/thrust/system/detail/generic/shuffle.inl b/thrust/system/detail/generic/shuffle.inl index b3c187f823..e522e7e923 100644 --- a/thrust/system/detail/generic/shuffle.inl +++ b/thrust/system/detail/generic/shuffle.inl @@ -100,9 +100,10 @@ class feistel_bijection { // but sufficient for generating permutations. __host__ __device__ uint32_t round_function(uint64_t value, const uint64_t key_) const { - uint64_t hash0 = thrust::random::taus88(value)(); + uint64_t hash0 = thrust::random::taus88(static_cast(value))(); uint64_t hash1 = thrust::random::ranlux48(value)(); - return hash_combine(hash_combine(hash0, key_), hash1) & left_side_mask; + return static_cast( + hash_combine(hash_combine(hash0, key_), hash1) & left_side_mask); } __host__ __device__ round_state do_round(const round_state state,