Enable compiler warnings

benchmarks/CMakeLists.txt

@@ -1,5 +1,5 @@
 #=============================================================================
-# Copyright (c) 2018-2021, NVIDIA CORPORATION.
+# Copyright (c) 2018-2022, NVIDIA CORPORATION.
 #
 # Licensed under the Apache License, Version 2.0 (the "License");
 # you may not use this file except in compliance with the License.
@@ -46,7 +46,8 @@ function(ConfigureBench BENCH_NAME BENCH_SRC)
                                         RUNTIME_OUTPUT_DIRECTORY "${CMAKE_BINARY_DIR}/gbenchmarks")
     target_include_directories(${BENCH_NAME} PRIVATE
                                              "${CMAKE_CURRENT_SOURCE_DIR}")
-    target_compile_options(${BENCH_NAME} PRIVATE --expt-extended-lambda --expt-relaxed-constexpr -Xcompiler -Wno-subobject-linkage)
+    target_compile_options(${BENCH_NAME} PRIVATE --compiler-options=-Wall --compiler-options=-Wextra
+      --expt-extended-lambda --expt-relaxed-constexpr -Xcompiler -Wno-subobject-linkage)
     target_link_libraries(${BENCH_NAME} PRIVATE
                                         benchmark benchmark_main
                                         pthread

benchmarks/hash_table/dynamic_map_bench.cu

@@ -75,7 +75,7 @@ static void BM_dynamic_insert(::benchmark::State& state)
 
   generate_keys<Dist, Key>(h_keys.begin(), h_keys.end());
 
-  for (auto i = 0; i < num_keys; ++i) {
+  for (std::size_t i = 0; i < num_keys; ++i) {
     Key key           = h_keys[i];
     Value val         = h_keys[i];
     h_pairs[i].first  = key;
@@ -90,7 +90,7 @@ static void BM_dynamic_insert(::benchmark::State& state)
       initial_size, cuco::sentinel::empty_key<Key>{-1}, cuco::sentinel::empty_value<Value>{-1}};
     {
       cuda_event_timer raii{state};
-      for (auto i = 0; i < num_keys; i += batch_size) {
+      for (std::size_t i = 0; i < num_keys; i += batch_size) {
         map.insert(d_pairs.begin() + i, d_pairs.begin() + i + batch_size);
       }
     }
@@ -113,7 +113,7 @@ static void BM_dynamic_search_all(::benchmark::State& state)
 
   generate_keys<Dist, Key>(h_keys.begin(), h_keys.end());
 
-  for (auto i = 0; i < num_keys; ++i) {
+  for (std::size_t i = 0; i < num_keys; ++i) {
     Key key           = h_keys[i];
     Value val         = h_keys[i];
     h_pairs[i].first  = key;

benchmarks/hash_table/static_map_bench.cu

@@ -22,7 +22,6 @@
 #include <benchmark/benchmark.h>
 
 #include <fstream>
-#include <iostream>
 #include <random>
 
 enum class dist_type { UNIQUE, UNIFORM, GAUSSIAN };
@@ -82,7 +81,7 @@ static void BM_static_map_insert(::benchmark::State& state)
 
   generate_keys<Dist, Key>(h_keys.begin(), h_keys.end());
 
-  for (auto i = 0; i < num_keys; ++i) {
+  for (std::size_t i = 0; i < num_keys; ++i) {
     Key key           = h_keys[i];
     Value val         = h_keys[i];
     h_pairs[i].first  = key;
@@ -124,7 +123,6 @@ static void BM_static_map_search_all(::benchmark::State& state)
   std::size_t size     = num_keys / occupancy;
 
   map_type map{size, cuco::sentinel::empty_key<Key>{-1}, cuco::sentinel::empty_value<Value>{-1}};
-  auto view = map.get_device_mutable_view();
 
   std::vector<Key> h_keys(num_keys);
   std::vector<Value> h_values(num_keys);
@@ -133,7 +131,7 @@ static void BM_static_map_search_all(::benchmark::State& state)
 
   generate_keys<Dist, Key>(h_keys.begin(), h_keys.end());
 
-  for (auto i = 0; i < num_keys; ++i) {
+  for (std::size_t i = 0; i < num_keys; ++i) {
     Key key           = h_keys[i];
     Value val         = h_keys[i];
     h_pairs[i].first  = key;
@@ -171,7 +169,6 @@ static void BM_static_map_erase_all(::benchmark::State& state)
                cuco::sentinel::empty_key<Key>{-1},
                cuco::sentinel::empty_value<Value>{-1},
                cuco::sentinel::erased_key<Key>{-2}};
-  auto view = map.get_device_mutable_view();
 
   std::vector<Key> h_keys(num_keys);
   std::vector<Value> h_values(num_keys);
@@ -180,7 +177,7 @@ static void BM_static_map_erase_all(::benchmark::State& state)
 
   generate_keys<Dist, Key>(h_keys.begin(), h_keys.end());
 
-  for (auto i = 0; i < num_keys; ++i) {
+  for (std::size_t i = 0; i < num_keys; ++i) {
     Key key           = h_keys[i];
     Value val         = h_keys[i];
     h_pairs[i].first  = key;

examples/CMakeLists.txt

@@ -1,4 +1,19 @@
-cmake_minimum_required(VERSION 3.18 FATAL_ERROR)
+#=============================================================================
+# Copyright (c) 2018-2022, NVIDIA CORPORATION.
+#
+# 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.
+#=============================================================================
+cmake_minimum_required(VERSION 3.18 FATAL_ERROR)
 
 ###################################################################################################
 # - compiler function -----------------------------------------------------------------------------
@@ -9,12 +24,13 @@ function(ConfigureExample EXAMPLE_NAME EXAMPLE_SRC)
                                           RUNTIME_OUTPUT_DIRECTORY "${CMAKE_BINARY_DIR}/examples")
     target_include_directories(${EXAMPLE_NAME} PRIVATE
                                              "${CMAKE_CURRENT_SOURCE_DIR}")
-    target_compile_options(${EXAMPLE_NAME} PRIVATE --expt-extended-lambda --expt-relaxed-constexpr)
+    target_compile_options(${EXAMPLE_NAME} PRIVATE --compiler-options=-Wall --compiler-options=-Wextra
+      --expt-extended-lambda --expt-relaxed-constexpr -Xcompiler -Wno-subobject-linkage)
     target_link_libraries(${EXAMPLE_NAME} PRIVATE cuco CUDA::cudart)
 endfunction(ConfigureExample)
 
 ###################################################################################################
-### Example sources ##################################################################################
+### Example sources ###############################################################################
 ###################################################################################################
 
 ConfigureExample(STATIC_MAP_EXAMPLE "${CMAKE_CURRENT_SOURCE_DIR}/static_map/static_map_example.cu")

include/cuco/detail/dynamic_map.inl

@@ -51,7 +51,7 @@ template <typename Key, typename Value, cuda::thread_scope Scope, typename Alloc
 void dynamic_map<Key, Value, Scope, Allocator>::reserve(std::size_t n)
 {
   int64_t num_elements_remaining = n;
-  auto submap_idx                = 0;
+  uint32_t submap_idx            = 0;
   while (num_elements_remaining > 0) {
     std::size_t submap_capacity;
 
@@ -160,4 +160,4 @@ void dynamic_map<Key, Value, Scope, Allocator>::contains(
   CUCO_CUDA_TRY(cudaDeviceSynchronize());
 }
 
-}  // namespace cuco
+}  // namespace cuco

include/cuco/detail/static_map.inl

@@ -216,7 +216,6 @@ std::pair<KeyOut, ValueOut> static_map<Key, Value, Scope, Allocator>::retrieve_a
   auto slots_begin = reinterpret_cast<value_type*>(slots_);
 
   auto begin  = thrust::make_transform_iterator(slots_begin, detail::slot_to_tuple<Key, Value>{});
-  auto end    = begin + get_capacity();
   auto filled = detail::slot_is_filled<Key>{get_empty_key_sentinel()};
   auto zipped_out_begin = thrust::make_zip_iterator(thrust::make_tuple(keys_out, values_out));
 

tests/CMakeLists.txt

@@ -43,7 +43,8 @@ function(ConfigureTest TEST_NAME)
     target_include_directories(${TEST_NAME} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR})
     set_target_properties(${TEST_NAME} PROPERTIES
                                        RUNTIME_OUTPUT_DIRECTORY "${CMAKE_BINARY_DIR}/tests")
-    target_compile_options(${TEST_NAME} PRIVATE --expt-extended-lambda --expt-relaxed-constexpr -Xcompiler -Wno-subobject-linkage)
+    target_compile_options(${TEST_NAME} PRIVATE --compiler-options=-Wall --compiler-options=-Wextra
+      --expt-extended-lambda --expt-relaxed-constexpr -Xcompiler -Wno-subobject-linkage)
     catch_discover_tests(${TEST_NAME})
 endfunction(ConfigureTest)
 

tests/static_map/duplicate_keys_test.cu

@@ -41,9 +41,6 @@ TEMPLATE_TEST_CASE_SIG("Duplicate keys",
   cuco::static_map<Key, Value> map{
     num_keys * 2, cuco::sentinel::empty_key<Key>{-1}, cuco::sentinel::empty_value<Value>{-1}};
 
-  auto m_view = map.get_device_mutable_view();
-  auto view   = map.get_device_view();
-
   thrust::device_vector<Key> d_keys(num_keys);
   thrust::device_vector<Value> d_values(num_keys);
 

tests/static_map/erase_test.cu

@@ -38,9 +38,6 @@ TEMPLATE_TEST_CASE_SIG("erase key", "", ((typename T), T), (int32_t), (int64_t))
                                    cuco::sentinel::empty_value<Value>{-1},
                                    cuco::sentinel::erased_key<Key>{-2}};
 
-  auto m_view = map.get_device_mutable_view();
-  auto view   = map.get_device_view();
-
   thrust::device_vector<Key> d_keys(num_keys);
   thrust::device_vector<Value> d_values(num_keys);
   thrust::device_vector<bool> d_keys_exist(num_keys);

tests/static_map/stream_test.cu

@@ -47,9 +47,6 @@ TEMPLATE_TEST_CASE_SIG("Unique sequence of keys on given stream",
                                    cuco::cuda_allocator<char>{},
                                    stream};
 
-  auto m_view = map.get_device_mutable_view();
-  auto view   = map.get_device_view();
-
   thrust::device_vector<Key> d_keys(num_keys);
   thrust::device_vector<Value> d_values(num_keys);
 

tests/static_multimap/custom_type_test.cu

@@ -54,7 +54,7 @@ struct value_pair {
 };
 
 template <typename Map>
-__inline__ void test_custom_key_value_type(Map& map, size_t num_pairs)
+__inline__ void test_custom_key_value_type(Map& map, std::size_t num_pairs)
 {
   using Key   = key_pair;
   using Value = value_pair;
@@ -97,7 +97,7 @@ __inline__ void test_custom_key_value_type(Map& map, size_t num_pairs)
     thrust::device_vector<cuco::pair_type<Key, Value>> found_pairs(num_pairs);
     auto output_end = map.retrieve(
       key_begin, key_begin + num_pairs, found_pairs.begin(), stream, key_pair_equals{});
-    auto size = output_end - found_pairs.begin();
+    std::size_t const size = std::distance(found_pairs.begin(), output_end);
 
     REQUIRE(size == num_pairs);
 
@@ -140,7 +140,7 @@ __inline__ void test_custom_key_value_type(Map& map, size_t num_pairs)
     thrust::device_vector<cuco::pair_type<Key, Value>> found_pairs(num_pairs);
     auto output_end = map.retrieve(
       query_key_begin, query_key_begin + num, found_pairs.begin(), stream, key_pair_equals{});
-    auto size = output_end - found_pairs.begin();
+    std::size_t const size = std::distance(found_pairs.begin(), output_end);
 
     REQUIRE(size == num_pairs);
 
@@ -182,7 +182,7 @@ __inline__ void test_custom_key_value_type(Map& map, size_t num_pairs)
     thrust::device_vector<cuco::pair_type<Key, Value>> found_pairs(num);
     auto output_end = map.retrieve_outer(
       query_key_begin, query_key_begin + num, found_pairs.begin(), stream, key_pair_equals{});
-    auto size_outer = output_end - found_pairs.begin();
+    std::size_t const size_outer = std::distance(found_pairs.begin(), output_end);
 
     REQUIRE(size_outer == num);
   }

tests/static_multimap/multiplicity_test.cu

@@ -85,9 +85,9 @@ __inline__ void test_multiplicity_two(Map& map, std::size_t num_items)
 
     REQUIRE(num == num_items);
 
-    auto output_begin = result_begin;
-    auto output_end   = map.retrieve(key_begin, key_begin + num_keys, output_begin);
-    auto size         = thrust::distance(output_begin, output_end);
+    auto output_begin      = result_begin;
+    auto output_end        = map.retrieve(key_begin, key_begin + num_keys, output_begin);
+    std::size_t const size = thrust::distance(output_begin, output_end);
 
     REQUIRE(size == num_items);
 

tests/static_multimap/non_match_test.cu

@@ -43,9 +43,9 @@ __inline__ void test_non_matches(Map& map, PairIt pair_begin, KeyIt key_begin, s
 
     REQUIRE(num == num_keys);
 
-    auto output_begin = d_results.data().get();
-    auto output_end   = map.retrieve(key_begin, key_begin + num_keys, output_begin);
-    auto size         = thrust::distance(output_begin, output_end);
+    auto output_begin      = d_results.data().get();
+    auto output_end        = map.retrieve(key_begin, key_begin + num_keys, output_begin);
+    std::size_t const size = thrust::distance(output_begin, output_end);
 
     REQUIRE(size == num_keys);
 
@@ -75,9 +75,9 @@ __inline__ void test_non_matches(Map& map, PairIt pair_begin, KeyIt key_begin, s
 
     REQUIRE(num == (num_keys + num_keys / 2));
 
-    auto output_begin = d_results.data().get();
-    auto output_end   = map.retrieve_outer(key_begin, key_begin + num_keys, output_begin);
-    auto size         = thrust::distance(output_begin, output_end);
+    auto output_begin      = d_results.data().get();
+    auto output_end        = map.retrieve_outer(key_begin, key_begin + num_keys, output_begin);
+    std::size_t const size = thrust::distance(output_begin, output_end);
 
     REQUIRE(size == (num_keys + num_keys / 2));
 

tests/static_multimap/pair_function_test.cu

@@ -69,8 +69,9 @@ __inline__ void test_pair_functions(Map& map, PairIt pair_begin, std::size_t num
 
     auto [out1_end, out2_end] = map.pair_retrieve(
       pair_begin, pair_begin + num_pairs, out1_begin, out2_begin, pair_equal<Key, Value>{});
+    std::size_t const size = std::distance(out2_begin, out1_end);
 
-    REQUIRE((out1_end - out1_begin) == num_pairs);
+    REQUIRE(size == num_pairs);
   }
 
   SECTION("Output of pair_count_outer and pair_retrieve_outer should be coherent.")
@@ -86,8 +87,9 @@ __inline__ void test_pair_functions(Map& map, PairIt pair_begin, std::size_t num
 
     auto [out1_end, out2_end] = map.pair_retrieve_outer(
       pair_begin, pair_begin + num_pairs, out1_begin, out2_begin, pair_equal<Key, Value>{});
+    std::size_t const size = std::distance(out1_begin, out1_end);
 
-    REQUIRE((out1_end - out1_begin) == (num_pairs + num_pairs / 2));
+    REQUIRE(size == (num_pairs + num_pairs / 2));
   }
 }