diff --git a/cub/benchmarks/bench/merge_sort/pairs.cu b/cub/benchmarks/bench/merge_sort/pairs.cu index 05a14ad5aa6..cf4b23b69b7 100644 --- a/cub/benchmarks/bench/merge_sort/pairs.cu +++ b/cub/benchmarks/bench/merge_sort/pairs.cu @@ -156,7 +156,15 @@ using key_types = all_types; #ifdef TUNE_ValueT using value_types = nvbench::type_list; #else // !defined(TUNE_ValueT) -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; #endif // TUNE_ValueT NVBENCH_BENCH_TYPES(pairs, NVBENCH_TYPE_AXES(key_types, value_types, offset_types)) diff --git a/cub/benchmarks/bench/radix_sort/pairs.cu b/cub/benchmarks/bench/radix_sort/pairs.cu index 1b99b41100c..006e33b98ce 100644 --- a/cub/benchmarks/bench/radix_sort/pairs.cu +++ b/cub/benchmarks/bench/radix_sort/pairs.cu @@ -230,7 +230,15 @@ using key_types = fundamental_types; #ifdef TUNE_ValueT using value_types = nvbench::type_list; #else // !defined(Tune_ValueT) -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; #endif // TUNE_ValueT NVBENCH_BENCH_TYPES(radix_sort_values, NVBENCH_TYPE_AXES(key_types, value_types, offset_types)) diff --git a/cub/benchmarks/bench/reduce/by_key.cu b/cub/benchmarks/bench/reduce/by_key.cu index 611b22b7e7e..d11667b7a5b 100644 --- a/cub/benchmarks/bench/reduce/by_key.cu +++ b/cub/benchmarks/bench/reduce/by_key.cu @@ -178,7 +178,15 @@ using some_offset_types = nvbench::type_list; #ifdef TUNE_KeyT using key_types = nvbench::type_list; #else // !defined(TUNE_KeyT) -using key_types = nvbench::type_list; +using key_types = nvbench::type_list; #endif // TUNE_KeyT #ifdef TUNE_ValueT diff --git a/cub/benchmarks/bench/scan/exclusive/by_key.cu b/cub/benchmarks/bench/scan/exclusive/by_key.cu index 66aeabca274..93e515e02cb 100644 --- a/cub/benchmarks/bench/scan/exclusive/by_key.cu +++ b/cub/benchmarks/bench/scan/exclusive/by_key.cu @@ -159,7 +159,15 @@ using key_types = all_types; #ifdef TUNE_ValueT using value_types = nvbench::type_list; #else // !defined(TUNE_ValueT) -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; #endif // TUNE_ValueT NVBENCH_BENCH_TYPES(scan, NVBENCH_TYPE_AXES(key_types, value_types, some_offset_types)) diff --git a/cub/benchmarks/bench/select/unique_by_key.cu b/cub/benchmarks/bench/select/unique_by_key.cu index c42de02f5c7..1d610e2b823 100644 --- a/cub/benchmarks/bench/select/unique_by_key.cu +++ b/cub/benchmarks/bench/select/unique_by_key.cu @@ -169,7 +169,15 @@ using some_offset_types = nvbench::type_list; #ifdef TUNE_KeyT using key_types = nvbench::type_list; #else // !defined(TUNE_KeyT) -using key_types = nvbench::type_list; +using key_types = nvbench::type_list; #endif // TUNE_KeyT #ifdef TUNE_ValueT diff --git a/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cu b/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cu index 668b18b868e..56e0d89b8f0 100644 --- a/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cu +++ b/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cu @@ -831,7 +831,10 @@ void gen_power_law_segment_offsets_device(seed_t seed, generator_t{}.power_law_segment_offsets(executor::device, seed, segment_offsets, elements); } -void do_not_optimize(const void *ptr) { asm volatile("" : : "r"(ptr) : "memory"); } +void do_not_optimize(const void *ptr) +{ + (void)ptr; +} } // namespace detail @@ -882,13 +885,16 @@ INSTANTIATE(uint8_t); INSTANTIATE(uint16_t); INSTANTIATE(uint32_t); INSTANTIATE(uint64_t); -INSTANTIATE(uint128_t); INSTANTIATE(int8_t); INSTANTIATE(int16_t); INSTANTIATE(int32_t); INSTANTIATE(int64_t); + +#if NVBENCH_HELPER_HAS_I128 INSTANTIATE(int128_t); +INSTANTIATE(uint128_t); +#endif INSTANTIATE(float); INSTANTIATE(double); diff --git a/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cuh b/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cuh index ad2c9049196..c5ae5f6c508 100644 --- a/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cuh +++ b/cub/benchmarks/nvbench_helper/nvbench_helper/nvbench_helper.cuh @@ -10,12 +10,22 @@ #include #include -using complex = cuda::std::complex; +#if defined(_MSC_VER) +#define NVBENCH_HELPER_HAS_I128 0 +#else +#define NVBENCH_HELPER_HAS_I128 1 +#endif + +#if NVBENCH_HELPER_HAS_I128 using int128_t = __int128_t; using uint128_t = __uint128_t; NVBENCH_DECLARE_TYPE_STRINGS(int128_t, "I128", "int128_t"); NVBENCH_DECLARE_TYPE_STRINGS(uint128_t, "U128", "uint128_t"); +#endif + +using complex = cuda::std::complex; + NVBENCH_DECLARE_TYPE_STRINGS(complex, "C64", "complex"); namespace detail @@ -50,7 +60,9 @@ using fundamental_types = nvbench::type_list; @@ -58,7 +70,9 @@ using all_types = nvbench::type_list; diff --git a/cub/benchmarks/nvbench_helper/test/gen_range.cu b/cub/benchmarks/nvbench_helper/test/gen_range.cu index 326565614d1..52ec72d647d 100644 --- a/cub/benchmarks/nvbench_helper/test/gen_range.cu +++ b/cub/benchmarks/nvbench_helper/test/gen_range.cu @@ -33,7 +33,15 @@ #include #include -using types = nvbench::type_list; +using types = nvbench::type_list; TEMPLATE_LIST_TEST_CASE("Generators produce data within specified range", "[gen]", types) { diff --git a/cub/benchmarks/nvbench_helper/test/gen_seed.cu b/cub/benchmarks/nvbench_helper/test/gen_seed.cu index acedc116834..63448dc1e13 100644 --- a/cub/benchmarks/nvbench_helper/test/gen_seed.cu +++ b/cub/benchmarks/nvbench_helper/test/gen_seed.cu @@ -31,8 +31,17 @@ #include #include -using types = - nvbench::type_list; +using types = nvbench::type_list; TEMPLATE_LIST_TEST_CASE("Generator seeds the data", "[gen]", types) { diff --git a/cub/benchmarks/nvbench_helper/test/gen_uniform_distribution.cu b/cub/benchmarks/nvbench_helper/test/gen_uniform_distribution.cu index 2a5005c8389..30192009973 100644 --- a/cub/benchmarks/nvbench_helper/test/gen_uniform_distribution.cu +++ b/cub/benchmarks/nvbench_helper/test/gen_uniform_distribution.cu @@ -73,7 +73,15 @@ bool is_uniform(thrust::host_vector data, T min, T max) return chi_square <= critical_value; } -using types = nvbench::type_list; +using types = nvbench::type_list; TEMPLATE_LIST_TEST_CASE("Generators produce uniformly distributed data", "[gen][uniform]", types) { diff --git a/thrust/benchmarks/bench/reduce/by_key.cu b/thrust/benchmarks/bench/reduce/by_key.cu index 724b8b5ea5b..4eaaed194e5 100644 --- a/thrust/benchmarks/bench/reduce/by_key.cu +++ b/thrust/benchmarks/bench/reduce/by_key.cu @@ -66,7 +66,16 @@ static void basic(nvbench::state &state, nvbench::type_list) }); } -using key_types = nvbench::type_list; +using key_types = nvbench::type_list; + using value_types = all_types; NVBENCH_BENCH_TYPES(basic, NVBENCH_TYPE_AXES(key_types, value_types)) diff --git a/thrust/benchmarks/bench/scan/exclusive/by_key.cu b/thrust/benchmarks/bench/scan/exclusive/by_key.cu index b49458c770b..76a5a0f9921 100644 --- a/thrust/benchmarks/bench/scan/exclusive/by_key.cu +++ b/thrust/benchmarks/bench/scan/exclusive/by_key.cu @@ -51,7 +51,15 @@ static void scan(nvbench::state &state, nvbench::type_list) } using key_types = all_types; -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; NVBENCH_BENCH_TYPES(scan, NVBENCH_TYPE_AXES(key_types, value_types)) .set_name("base") diff --git a/thrust/benchmarks/bench/scan/inclusive/by_key.cu b/thrust/benchmarks/bench/scan/inclusive/by_key.cu index 67e8f3ca410..bb468ff57dd 100644 --- a/thrust/benchmarks/bench/scan/inclusive/by_key.cu +++ b/thrust/benchmarks/bench/scan/inclusive/by_key.cu @@ -51,7 +51,15 @@ static void scan(nvbench::state &state, nvbench::type_list) } using key_types = all_types; -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; NVBENCH_BENCH_TYPES(scan, NVBENCH_TYPE_AXES(key_types, value_types)) .set_name("base") diff --git a/thrust/benchmarks/bench/shuffle/basic.cu b/thrust/benchmarks/bench/shuffle/basic.cu index b039fed7ead..f70629f2a4c 100644 --- a/thrust/benchmarks/bench/shuffle/basic.cu +++ b/thrust/benchmarks/bench/shuffle/basic.cu @@ -68,7 +68,15 @@ static void basic(nvbench::state &state, nvbench::type_list) } } -using types = nvbench::type_list; +using types = nvbench::type_list; NVBENCH_BENCH_TYPES(basic, NVBENCH_TYPE_AXES(types)) .set_name("base") diff --git a/thrust/benchmarks/bench/sort/pairs.cu b/thrust/benchmarks/bench/sort/pairs.cu index 757bac93e59..a6d45e33ed5 100644 --- a/thrust/benchmarks/bench/sort/pairs.cu +++ b/thrust/benchmarks/bench/sort/pairs.cu @@ -60,7 +60,15 @@ static void basic(nvbench::state &state, nvbench::type_list) } using key_types = fundamental_types; -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; NVBENCH_BENCH_TYPES(basic, NVBENCH_TYPE_AXES(key_types, value_types)) .set_name("base") diff --git a/thrust/benchmarks/bench/sort/pairs_custom.cu b/thrust/benchmarks/bench/sort/pairs_custom.cu index 07a2427f680..4cd73139895 100644 --- a/thrust/benchmarks/bench/sort/pairs_custom.cu +++ b/thrust/benchmarks/bench/sort/pairs_custom.cu @@ -60,7 +60,15 @@ static void basic(nvbench::state &state, nvbench::type_list) } using key_types = fundamental_types; -using value_types = nvbench::type_list; +using value_types = nvbench::type_list; NVBENCH_BENCH_TYPES(basic, NVBENCH_TYPE_AXES(key_types, value_types)) .set_name("base") diff --git a/thrust/benchmarks/bench/unique/by_key.cu b/thrust/benchmarks/bench/unique/by_key.cu index 8419c4aa148..e6961bc4d95 100644 --- a/thrust/benchmarks/bench/unique/by_key.cu +++ b/thrust/benchmarks/bench/unique/by_key.cu @@ -65,7 +65,15 @@ static void basic(nvbench::state &state, nvbench::type_list) }); } -using key_types = nvbench::type_list; +using key_types = nvbench::type_list; using value_types = all_types; NVBENCH_BENCH_TYPES(basic, NVBENCH_TYPE_AXES(key_types, value_types))