From a3e20154507d4a17b39da386c48523c8cf387415 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Fri, 17 Dec 2021 18:21:34 -0500 Subject: [PATCH 01/11] Add .clang-tidy configuration file --- .clang-tidy | 68 +++++++++++++++++++++++++++++++++++++++++++++++++++++ 1 file changed, 68 insertions(+) create mode 100644 .clang-tidy diff --git a/.clang-tidy b/.clang-tidy new file mode 100644 index 000000000..e53a18d4b --- /dev/null +++ b/.clang-tidy @@ -0,0 +1,68 @@ +--- +Checks: 'clang-diagnostic-*, + clang-analyzer-*, + cppcoreguidelines-*, + modernize-*, + bugprone-*, + performance-*, + readability-*, + llvm-*, + -cppcoreguidelines-macro-usage, + -llvm-header-guard, + -modernize-use-trailing-return-type, + -readability-static-accessed-through-instance, + -readability-named-parameter' +WarningsAsErrors: '' +HeaderFilterRegex: '' +AnalyzeTemporaryDtors: false +FormatStyle: none +CheckOptions: + - key: cert-dcl16-c.NewSuffixes + value: 'L;LL;LU;LLU' + - key: cert-oop54-cpp.WarnOnlyIfThisHasSuspiciousField + value: '0' + - key: cert-str34-c.DiagnoseSignedUnsignedCharComparisons + value: '0' + - key: cppcoreguidelines-explicit-virtual-functions.IgnoreDestructors + value: '1' + - key: cppcoreguidelines-non-private-member-variables-in-classes.IgnoreClassesWithAllMemberVariablesBeingPublic + value: '1' + - key: google-readability-braces-around-statements.ShortStatementLines + value: '1' + - key: google-readability-function-size.StatementThreshold + value: '800' + - key: google-readability-namespace-comments.ShortNamespaceLines + value: '10' + - key: google-readability-namespace-comments.SpacesBeforeComments + value: '2' + - key: llvm-else-after-return.WarnOnConditionVariables + value: '0' + - key: llvm-else-after-return.WarnOnUnfixable + value: '0' + - key: llvm-qualified-auto.AddConstToQualified + value: '0' + - key: modernize-loop-convert.MaxCopySize + value: '16' + - key: modernize-loop-convert.MinConfidence + value: reasonable + - key: modernize-loop-convert.NamingStyle + value: CamelCase + - key: modernize-pass-by-value.IncludeStyle + value: llvm + - key: modernize-replace-auto-ptr.IncludeStyle + value: llvm + - key: modernize-use-nullptr.NullMacros + value: 'NULL' + - key: readability-identifier-length.IgnoredParameterNames + value: 'mr|os' + - key: readability-identifier-length.IgnoredVariableNames + value: 'mr|_' + #- key: readability-function-cognitive-complexity.IgnoreMacros + # value: '1' + - key: bugprone-easily-swappable-parameters.IgnoredParameterNames + value: 'alignment' + - key: cppcoreguidelines-avoid-magic-numbers.IgnorePowersOf2IntegerValues + value: '1' + - key: readability-magic-numbers.IgnorePowersOf2IntegerValues + value: '1' +... From c296f8364c897d3cb21b65c6b3a74bf9ba13bbe7 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Fri, 17 Dec 2021 18:27:55 -0500 Subject: [PATCH 02/11] Add run_clang_tidy.py script --- run_clang_tidy.py | 251 ++++++++++++++++++++++++++++++++++++++++++++++ 1 file changed, 251 insertions(+) create mode 100644 run_clang_tidy.py diff --git a/run_clang_tidy.py b/run_clang_tidy.py new file mode 100644 index 000000000..39065c223 --- /dev/null +++ b/run_clang_tidy.py @@ -0,0 +1,251 @@ +# Copyright (c) 2021, 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. +# + +from __future__ import print_function +import re +import os +import subprocess +import argparse +import json +import multiprocessing as mp +import shutil + + +EXPECTED_VERSION = "12.0.0" +VERSION_REGEX = re.compile(r" LLVM version ([0-9.]+)") +GPU_ARCH_REGEX = re.compile(r"sm_(\d+)") +SPACES = re.compile(r"\s+") +SEPARATOR = "-" * 16 + + +def parse_args(): + argparser = argparse.ArgumentParser("Runs clang-tidy on a project") + argparser.add_argument("-cdb", type=str, + # TODO This is a hack, needs to be fixed + default="build/compile_commands.json", + help="Path to cmake-generated compilation database" + " file. It is always found inside the root of the " + "cmake build folder. So make sure that `cmake` has " + "been run once before running this script!") + argparser.add_argument("-exe", type=str, default="clang-tidy-12", + help="Path to clang-tidy exe") + argparser.add_argument("-ignore", type=str, default="[.]cxx$|[.]cpp$|[.]cc$", + help="Regex used to ignore files from checking") + argparser.add_argument("-select", type=str, default=None, + help="Regex used to select files for checking") + argparser.add_argument("-j", type=int, default=-1, + help="Number of parallel jobs to launch.") + args = argparser.parse_args() + if args.j <= 0: + args.j = mp.cpu_count() + args.ignore_compiled = re.compile(args.ignore) if args.ignore else None + args.select_compiled = re.compile(args.select) if args.select else None + if not os.path.exists(args.cdb): + raise Exception("Compilation database '%s' missing" % args.cdb) + return args + + +def get_all_commands(cdb): + with open(cdb, "r") as fp: + return json.load(fp) + + +def get_gpu_archs(command): + archs = [] + for info in command: + if "--generate-code" not in info: + continue + arch_flag = [] + if "compute_" in info: + arch_flag = info.split("compute_")[1][:2] + elif "sm_" in info: + arch_flag = info.split("sm_")[1][:2] + if arch_flag.isdigit(): + archs.append("--cuda-gpu-arch=sm_%s" % arch_flag) + return archs + + +def get_index(arr, item): + try: + return arr.index(item) + except: + return -1 + + +def remove_item(arr, item): + loc = get_index(arr, item) + if loc >= 0: + del arr[loc] + return loc + + +def remove_item_plus_one(arr, item): + loc = get_index(arr, item) + if loc >= 0: + del arr[loc + 1] + del arr[loc] + return loc + + +def get_clang_includes(exe): + dir = os.getenv("CONDA_PREFIX") + if dir is None: + ret = subprocess.check_output("which %s 2>&1" % exe, shell=True) + ret = ret.decode("utf-8") + dir = os.path.dirname(os.path.dirname(ret)) + header = os.path.join(dir, "include", "ClangHeaders") + return ["-I", header] + + +def get_tidy_args(cmd, exe): + command, file = cmd["command"], cmd["file"] + is_cuda = file.endswith(".cu") + command = re.split(SPACES, command) + # compiler is always clang++! + command[0] = "clang++" + # remove compilation and output targets from the original command + remove_item_plus_one(command, "-c") + remove_item_plus_one(command, "-o") + if is_cuda: + # replace nvcc's "-gencode ..." with clang's "--cuda-gpu-arch ..." + archs = get_gpu_archs(command) + command.extend(archs) + for info in command: + if "--generate-code" in info: + remove_item(command, info) + # "-x cuda" is the right usage in clang + loc = get_index(command, "-x") + if loc >= 0: + command[loc + 1] = "cuda" + remove_item_plus_one(command, "-ccbin") + remove_item(command, "--expt-extended-lambda") + remove_item(command, "--expt-relaxed-constexpr") + remove_item(command, "--diag_suppress=unrecognized_gcc_pragma") + remove_item(command, "-forward-unknown-to-host-compiler") + command.extend(get_clang_includes(exe)) + return command, is_cuda + + +def run_clang_tidy_command(tidy_cmd): + cmd = " ".join(tidy_cmd) + result = subprocess.run(cmd, check=False, shell=True, + stdout=subprocess.PIPE, stderr=subprocess.STDOUT) + status = result.returncode == 0 + if status: + out = "" + else: + out = "CMD: " + cmd + out += result.stdout.decode("utf-8").rstrip() + return status, out + + +def run_clang_tidy(cmd, args): + command, is_cuda = get_tidy_args(cmd, args.exe) + tidy_cmd = [args.exe, + "-header-filter=include/cuco", + cmd["file"], "--", ] + tidy_cmd.extend(command) + status = True + out = "" + if is_cuda: + tidy_cmd.append("--cuda-device-only") + tidy_cmd.append(cmd["file"]) + ret, out1 = run_clang_tidy_command(tidy_cmd) + out += out1 + out += "%s" % SEPARATOR + if not ret: + status = ret + tidy_cmd[-2] = "--cuda-host-only" + ret, out1 = run_clang_tidy_command(tidy_cmd) + if not ret: + status = ret + out += out1 + else: + tidy_cmd.append(cmd["file"]) + ret, out1 = run_clang_tidy_command(tidy_cmd) + if not ret: + status = ret + out += out1 + return status, out, cmd["file"] + + +# yikes! global var :( +results = [] +def collect_result(result): + global results + results.append(result) + + +def print_result(passed, stdout, file): + status_str = "PASSED" if passed else "FAILED" + print("%s File:%s %s %s" % (SEPARATOR, file, status_str, SEPARATOR)) + if stdout: + print(stdout) + print("%s File:%s ENDS %s" % (SEPARATOR, file, SEPARATOR)) + + +def print_results(): + global results + status = True + for passed, stdout, file in results: + # skip third-party files + if "_deps" not in file: + print_result(passed, stdout, file) + if not passed: + status = False + return status + + +def run_tidy_for_all_files(args, all_files): + pool = None if args.j == 1 else mp.Pool(args.j) + # actual tidy checker + for cmd in all_files: + # skip files that we don't want to look at + if args.ignore_compiled is not None and \ + re.search(args.ignore_compiled, cmd["file"]) is not None: + continue + if args.select_compiled is not None and \ + re.search(args.select_compiled, cmd["file"]) is None: + continue + if pool is not None: + pool.apply_async(run_clang_tidy, args=(cmd, args), + callback=collect_result) + else: + passed, stdout, file = run_clang_tidy(cmd, args) + collect_result((passed, stdout, file)) + if pool is not None: + pool.close() + pool.join() + return print_results() + + +def main(): + args = parse_args() + # Attempt to making sure that we run this script from root of repo always + if not os.path.exists(".git"): + raise Exception("This needs to always be run from the root of repo") + # Check whether clang-tidy exists + # print(args) + if "exe" not in args and shutil.which("clang-tidy") is not None: + print("clang-tidy not found. Exiting...") + return + all_files = get_all_commands(args.cdb) + status = run_tidy_for_all_files(args, all_files) + if not status: + raise Exception("clang-tidy failed! Refer to the errors above.") + + +if __name__ == "__main__": + main() From 706c9ff7844a33a2515f93d898fadd6626507ac9 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Fri, 17 Dec 2021 18:28:09 -0500 Subject: [PATCH 03/11] Update cmake --- CMakeLists.txt | 3 +++ 1 file changed, 3 insertions(+) diff --git a/CMakeLists.txt b/CMakeLists.txt index 956241ea6..43db6a214 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -29,6 +29,9 @@ include(rapids-find) # * Enable the CMake CUDA language rapids_cuda_init_architectures(CUCO) +# this is needed for clang-tidy runs +set(CMAKE_EXPORT_COMPILE_COMMANDS ON) + project(CUCO VERSION 0.0.1 LANGUAGES CXX CUDA) # Write the version header From 08e2d5bc02ddb271be1a67907e0cc847ee600282 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Fri, 17 Dec 2021 18:42:46 -0500 Subject: [PATCH 04/11] Get rid of modernize-use-nullptr warnings --- benchmarks/synchronization.hpp | 6 ++-- include/cuco/static_map.cuh | 10 +++---- include/cuco/static_multimap.cuh | 28 +++++++++---------- tests/static_map/static_map_test.cu | 2 +- tests/static_multimap/static_multimap_test.cu | 4 +-- 5 files changed, 26 insertions(+), 24 deletions(-) diff --git a/benchmarks/synchronization.hpp b/benchmarks/synchronization.hpp index e91511281..09c373da7 100644 --- a/benchmarks/synchronization.hpp +++ b/benchmarks/synchronization.hpp @@ -81,7 +81,9 @@ class cuda_event_timer { * every iteration. * @param[in] stream_ The CUDA stream we are measuring time on. */ - cuda_event_timer(benchmark::State &state, bool flush_l2_cache = false, cudaStream_t stream = 0) + cuda_event_timer(benchmark::State &state, + bool flush_l2_cache = false, + cudaStream_t stream = nullptr) : p_state(&state), stream_(stream) { // flush all of L2$ @@ -129,4 +131,4 @@ class cuda_event_timer { cudaEvent_t stop_; cudaStream_t stream_; benchmark::State *p_state; -}; \ No newline at end of file +}; diff --git a/include/cuco/static_map.cuh b/include/cuco/static_map.cuh index 321b1f3da..af780376c 100644 --- a/include/cuco/static_map.cuh +++ b/include/cuco/static_map.cuh @@ -195,7 +195,7 @@ class static_map { Key empty_key_sentinel, Value empty_value_sentinel, Allocator const& alloc = Allocator{}, - cudaStream_t stream = 0); + cudaStream_t stream = nullptr); /** * @brief Destroys the map and frees its contents. @@ -228,7 +228,7 @@ class static_map { InputIt last, Hash hash = Hash{}, KeyEqual key_equal = KeyEqual{}, - cudaStream_t stream = 0); + cudaStream_t stream = nullptr); /** * @brief Inserts key/value pairs in the range `[first, last)` if `pred` @@ -264,7 +264,7 @@ class static_map { Predicate pred, Hash hash = Hash{}, KeyEqual key_equal = KeyEqual{}, - cudaStream_t stream = 0); + cudaStream_t stream = nullptr); /** * @brief Finds the values corresponding to all keys in the range `[first, last)`. @@ -294,7 +294,7 @@ class static_map { OutputIt output_begin, Hash hash = Hash{}, KeyEqual key_equal = KeyEqual{}, - cudaStream_t stream = 0); + cudaStream_t stream = nullptr); /** * @brief Indicates whether the keys in the range @@ -324,7 +324,7 @@ class static_map { OutputIt output_begin, Hash hash = Hash{}, KeyEqual key_equal = KeyEqual{}, - cudaStream_t stream = 0); + cudaStream_t stream = nullptr); private: class device_view_base { diff --git a/include/cuco/static_multimap.cuh b/include/cuco/static_multimap.cuh index fca14ebe2..59ea1befd 100644 --- a/include/cuco/static_multimap.cuh +++ b/include/cuco/static_multimap.cuh @@ -222,7 +222,7 @@ class static_multimap { static_multimap(std::size_t capacity, Key empty_key_sentinel, Value empty_value_sentinel, - cudaStream_t stream = 0, + cudaStream_t stream = nullptr, Allocator const& alloc = Allocator{}); /** @@ -236,7 +236,7 @@ class static_multimap { * @param stream CUDA stream used for insert */ template - void insert(InputIt first, InputIt last, cudaStream_t stream = 0); + void insert(InputIt first, InputIt last, cudaStream_t stream = nullptr); /** * @brief Inserts key/value pairs in the range `[first, first + n)` if `pred` @@ -260,7 +260,7 @@ class static_multimap { */ template void insert_if( - InputIt first, InputIt last, StencilIt stencil, Predicate pred, cudaStream_t stream = 0); + InputIt first, InputIt last, StencilIt stencil, Predicate pred, cudaStream_t stream = nullptr); /** * @brief Indicates whether the keys in the range `[first, last)` are contained in the map. @@ -283,7 +283,7 @@ class static_multimap { void contains(InputIt first, InputIt last, OutputIt output_begin, - cudaStream_t stream = 0, + cudaStream_t stream = nullptr, KeyEqual key_equal = KeyEqual{}) const; /** @@ -303,7 +303,7 @@ class static_multimap { template > std::size_t count(InputIt first, InputIt last, - cudaStream_t stream = 0, + cudaStream_t stream = nullptr, KeyEqual key_equal = KeyEqual{}) const; /** @@ -325,7 +325,7 @@ class static_multimap { template > std::size_t count_outer(InputIt first, InputIt last, - cudaStream_t stream = 0, + cudaStream_t stream = nullptr, KeyEqual key_equal = KeyEqual{}) const; /** @@ -348,7 +348,7 @@ class static_multimap { std::size_t pair_count(InputIt first, InputIt last, PairEqual pair_equal, - cudaStream_t stream = 0) const; + cudaStream_t stream = nullptr) const; /** * @brief Counts the occurrences of key/value pairs in `[first, last)` contained in the multimap. @@ -372,7 +372,7 @@ class static_multimap { std::size_t pair_count_outer(InputIt first, InputIt last, PairEqual pair_equal, - cudaStream_t stream = 0) const; + cudaStream_t stream = nullptr) const; /** * @brief Retrieves all the values corresponding to all keys in the range `[first, last)`. @@ -399,7 +399,7 @@ class static_multimap { OutputIt retrieve(InputIt first, InputIt last, OutputIt output_begin, - cudaStream_t stream = 0, + cudaStream_t stream = nullptr, KeyEqual key_equal = KeyEqual{}) const; /** @@ -428,7 +428,7 @@ class static_multimap { OutputIt retrieve_outer(InputIt first, InputIt last, OutputIt output_begin, - cudaStream_t stream = 0, + cudaStream_t stream = nullptr, KeyEqual key_equal = KeyEqual{}) const; /** @@ -465,7 +465,7 @@ class static_multimap { OutputIt1 probe_output_begin, OutputIt2 contained_output_begin, PairEqual pair_equal, - cudaStream_t stream = 0) const; + cudaStream_t stream = nullptr) const; /** * @brief Retrieves all pairs matching the input probe pair in the range `[first, last)`. @@ -503,7 +503,7 @@ class static_multimap { OutputIt1 probe_output_begin, OutputIt2 contained_output_begin, PairEqual pair_equal, - cudaStream_t stream = 0) const; + cudaStream_t stream = nullptr) const; private: /** @@ -1202,7 +1202,7 @@ class static_multimap { * @param stream CUDA stream used to get the number of inserted elements * @return The number of elements in the map */ - std::size_t get_size(cudaStream_t stream = 0) const noexcept; + std::size_t get_size(cudaStream_t stream = nullptr) const noexcept; /** * @brief Gets the load factor of the hash map. @@ -1210,7 +1210,7 @@ class static_multimap { * @param stream CUDA stream used to get the load factor * @return The load factor of the hash map */ - float get_load_factor(cudaStream_t stream = 0) const noexcept; + float get_load_factor(cudaStream_t stream = nullptr) const noexcept; /** * @brief Gets the sentinel value used to represent an empty key slot. diff --git a/tests/static_map/static_map_test.cu b/tests/static_map/static_map_test.cu index 622e3a80d..31b3e980e 100644 --- a/tests/static_map/static_map_test.cu +++ b/tests/static_map/static_map_test.cu @@ -29,7 +29,7 @@ namespace cg = cooperative_groups; // User-defined logical algorithms to reduce compilation time template -bool all_of(Iterator begin, Iterator end, Predicate p, cudaStream_t stream = 0) +bool all_of(Iterator begin, Iterator end, Predicate p, cudaStream_t stream = nullptr) { auto size = thrust::distance(begin, end); auto out = thrust::count_if(thrust::cuda::par.on(stream), begin, end, p); diff --git a/tests/static_multimap/static_multimap_test.cu b/tests/static_multimap/static_multimap_test.cu index 6a48c7b18..ee25ef6d9 100644 --- a/tests/static_multimap/static_multimap_test.cu +++ b/tests/static_multimap/static_multimap_test.cu @@ -83,7 +83,7 @@ __inline__ void test_custom_key_value_type(Map& map, KeyIt key_begin, size_t num_pairs) { - constexpr cudaStream_t stream = 0; + constexpr cudaStream_t stream{nullptr}; SECTION("All inserted keys-value pairs should be correctly recovered during find") { @@ -614,7 +614,7 @@ template __inline__ void test_pair_functions(Map& map, PairIt pair_begin, std::size_t num_pairs) { map.insert(pair_begin, pair_begin + num_pairs); - cudaStreamSynchronize(0); + cudaStreamSynchronize(nullptr); auto res = map.get_size(); REQUIRE(res == num_pairs); From d41a133f08b637d319fff5b178784cc78a30e927 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Fri, 17 Dec 2021 19:08:52 -0500 Subject: [PATCH 05/11] nodiscard --- include/cuco/detail/hash_functions.cuh | 4 +- include/cuco/detail/probe_sequence_impl.cuh | 7 ++- include/cuco/detail/static_map.inl | 11 ++-- .../static_multimap/device_view_impl.inl | 8 +-- include/cuco/dynamic_map.cuh | 11 ++-- include/cuco/static_map.cuh | 57 ++++++++++++------- include/cuco/static_multimap.cuh | 21 +++---- 7 files changed, 72 insertions(+), 47 deletions(-) diff --git a/include/cuco/detail/hash_functions.cuh b/include/cuco/detail/hash_functions.cuh index d5dcd0f64..58d82e047 100644 --- a/include/cuco/detail/hash_functions.cuh +++ b/include/cuco/detail/hash_functions.cuh @@ -81,12 +81,12 @@ struct MurmurHash3_32 { } private: - constexpr __host__ __device__ uint32_t rotl32(uint32_t x, int8_t r) const noexcept + [[nodiscard]] constexpr __host__ __device__ uint32_t rotl32(uint32_t x, int8_t r) const noexcept { return (x << r) | (x >> (32 - r)); } - constexpr __host__ __device__ uint32_t fmix32(uint32_t h) const noexcept + [[nodiscard]] constexpr __host__ __device__ uint32_t fmix32(uint32_t h) const noexcept { h ^= h >> 16; h *= 0x85ebca6b; diff --git a/include/cuco/detail/probe_sequence_impl.cuh b/include/cuco/detail/probe_sequence_impl.cuh index 549d59df3..e7adfbec9 100644 --- a/include/cuco/detail/probe_sequence_impl.cuh +++ b/include/cuco/detail/probe_sequence_impl.cuh @@ -112,7 +112,7 @@ class probe_sequence_impl_base { /** * @brief Returns the capacity of the hash map. */ - __host__ __device__ __forceinline__ std::size_t get_capacity() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ std::size_t get_capacity() const noexcept { return capacity_; } @@ -125,7 +125,10 @@ class probe_sequence_impl_base { /** * @brief Returns slots array. */ - __device__ __forceinline__ const_iterator get_slots() const noexcept { return slots_; } + [[nodiscard]] __device__ __forceinline__ const_iterator get_slots() const noexcept + { + return slots_; + } protected: iterator slots_; ///< Pointer to beginning of the hash map slots diff --git a/include/cuco/detail/static_map.inl b/include/cuco/detail/static_map.inl index 1719970a7..00184e397 100644 --- a/include/cuco/detail/static_map.inl +++ b/include/cuco/detail/static_map.inl @@ -383,10 +383,11 @@ static_map::device_view::find(Key const& k, template template -__device__ typename static_map::device_view::const_iterator -static_map::device_view::find(Key const& k, - Hash hash, - KeyEqual key_equal) const noexcept +[[nodiscard]] __device__ + typename static_map::device_view::const_iterator + static_map::device_view::find(Key const& k, + Hash hash, + KeyEqual key_equal) const noexcept { auto current_slot = initial_slot(k, hash); @@ -502,7 +503,7 @@ __device__ bool static_map::device_view::contains( template template -__device__ bool static_map::device_view::contains( +[[nodiscard]] __device__ bool static_map::device_view::contains( CG g, Key const& k, Hash hash, KeyEqual key_equal) const noexcept { auto current_slot = initial_slot(g, k, hash); diff --git a/include/cuco/detail/static_multimap/device_view_impl.inl b/include/cuco/detail/static_multimap/device_view_impl.inl index 2af06ca7b..49c76ad1f 100644 --- a/include/cuco/detail/static_multimap/device_view_impl.inl +++ b/include/cuco/detail/static_multimap/device_view_impl.inl @@ -145,7 +145,7 @@ class static_multimap::device_view_ * * @return The sentinel value used to represent an empty key slot */ - __host__ __device__ __forceinline__ Key get_empty_key_sentinel() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ Key get_empty_key_sentinel() const noexcept { return empty_key_sentinel_; } @@ -155,7 +155,7 @@ class static_multimap::device_view_ * * @return The sentinel value used to represent an empty value slot */ - __host__ __device__ __forceinline__ Value get_empty_value_sentinel() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ Value get_empty_value_sentinel() const noexcept { return empty_value_sentinel_; } @@ -175,7 +175,7 @@ class static_multimap::device_view_ * * @return Slots array */ - __device__ __forceinline__ pair_atomic_type const* get_slots() const noexcept + [[nodiscard]] __device__ __forceinline__ pair_atomic_type const* get_slots() const noexcept { return probe_sequence_.get_slots(); } @@ -185,7 +185,7 @@ class static_multimap::device_view_ * * @return The maximum number of elements the hash map can hold */ - __host__ __device__ __forceinline__ std::size_t get_capacity() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ std::size_t get_capacity() const noexcept { return probe_sequence_.get_capacity(); } diff --git a/include/cuco/dynamic_map.cuh b/include/cuco/dynamic_map.cuh index 2e57ac6d2..25c26f4f5 100644 --- a/include/cuco/dynamic_map.cuh +++ b/include/cuco/dynamic_map.cuh @@ -224,21 +224,24 @@ class dynamic_map { * * @return The current number of elements in the map */ - std::size_t get_size() const noexcept { return size_; } + [[nodiscard]] std::size_t get_size() const noexcept { return size_; } /** * @brief Gets the maximum number of elements the hash map can hold. * * @return The maximum number of elements the hash map can hold */ - std::size_t get_capacity() const noexcept { return capacity_; } + [[nodiscard]] std::size_t get_capacity() const noexcept { return capacity_; } /** * @brief Gets the load factor of the hash map. * * @return The load factor of the hash map */ - float get_load_factor() const noexcept { return static_cast(size_) / capacity_; } + [[nodiscard]] float get_load_factor() const noexcept + { + return static_cast(size_) / capacity_; + } private: key_type empty_key_sentinel_{}; ///< Key value that represents an empty slot @@ -258,4 +261,4 @@ class dynamic_map { }; } // namespace cuco -#include \ No newline at end of file +#include diff --git a/include/cuco/static_map.cuh b/include/cuco/static_map.cuh index af780376c..c27fc64dc 100644 --- a/include/cuco/static_map.cuh +++ b/include/cuco/static_map.cuh @@ -376,7 +376,7 @@ class static_map { * @return Pointer to the initial slot for `k` */ template - __device__ const_iterator initial_slot(Key const& k, Hash hash) const noexcept + [[nodiscard]] __device__ const_iterator initial_slot(Key const& k, Hash hash) const noexcept { return &slots_[hash(k) % capacity_]; } @@ -412,7 +412,9 @@ class static_map { * @return Pointer to the initial slot for `k` */ template - __device__ const_iterator initial_slot(CG g, Key const& k, Hash hash) const noexcept + [[nodiscard]] __device__ const_iterator initial_slot(CG g, + Key const& k, + Hash hash) const noexcept { return &slots_[(hash(k) + g.thread_rank()) % capacity_]; } @@ -516,28 +518,37 @@ class static_map { * * @return Slots array */ - __host__ __device__ pair_atomic_type const* get_slots() const noexcept { return slots_; } + [[nodiscard]] __host__ __device__ pair_atomic_type const* get_slots() const noexcept + { + return slots_; + } /** * @brief Gets the maximum number of elements the hash map can hold. * * @return The maximum number of elements the hash map can hold */ - __host__ __device__ std::size_t get_capacity() const noexcept { return capacity_; } + [[nodiscard]] __host__ __device__ std::size_t get_capacity() const noexcept + { + return capacity_; + } /** * @brief Gets the sentinel value used to represent an empty key slot. * * @return The sentinel value used to represent an empty key slot */ - __host__ __device__ Key get_empty_key_sentinel() const noexcept { return empty_key_sentinel_; } + [[nodiscard]] __host__ __device__ Key get_empty_key_sentinel() const noexcept + { + return empty_key_sentinel_; + } /** * @brief Gets the sentinel value used to represent an empty value slot. * * @return The sentinel value used to represent an empty value slot */ - __host__ __device__ Value get_empty_value_sentinel() const noexcept + [[nodiscard]] __host__ __device__ Value get_empty_value_sentinel() const noexcept { return empty_value_sentinel_; } @@ -570,14 +581,17 @@ class static_map { * * @return Iterator to the first slot */ - __device__ const_iterator begin_slot() const noexcept { return slots_; } + [[nodiscard]] __device__ const_iterator begin_slot() const noexcept { return slots_; } /** * @brief Returns a const_iterator to one past the last slot. * * @return A const_iterator to one past the last slot */ - __host__ __device__ const_iterator end_slot() const noexcept { return slots_ + capacity_; } + [[nodiscard]] __host__ __device__ const_iterator end_slot() const noexcept + { + return slots_ + capacity_; + } /** * @brief Returns an iterator to one past the last slot. @@ -594,7 +608,7 @@ class static_map { * * @return A const_iterator to one past the last slot */ - __host__ __device__ const_iterator end() const noexcept { return end_slot(); } + [[nodiscard]] __host__ __device__ const_iterator end() const noexcept { return end_slot(); } /** * @brief Returns an iterator to one past the last slot. @@ -934,9 +948,9 @@ class static_map { */ template , typename KeyEqual = thrust::equal_to> - __device__ const_iterator find(Key const& k, - Hash hash = Hash{}, - KeyEqual key_equal = KeyEqual{}) const noexcept; + [[nodiscard]] __device__ const_iterator find(Key const& k, + Hash hash = Hash{}, + KeyEqual key_equal = KeyEqual{}) const noexcept; /** * @brief Finds the value corresponding to the key `k`. @@ -987,7 +1001,7 @@ class static_map { template , typename KeyEqual = thrust::equal_to> - __device__ const_iterator + [[nodiscard]] __device__ const_iterator find(CG g, Key const& k, Hash hash = Hash{}, KeyEqual key_equal = KeyEqual{}) const noexcept; /** @@ -1045,42 +1059,45 @@ class static_map { * * @return The maximum number of elements the hash map can hold */ - std::size_t get_capacity() const noexcept { return capacity_; } + [[nodiscard]] std::size_t get_capacity() const noexcept { return capacity_; } /** * @brief Gets the number of elements in the hash map. * * @return The number of elements in the map */ - std::size_t get_size() const noexcept { return size_; } + [[nodiscard]] std::size_t get_size() const noexcept { return size_; } /** * @brief Gets the load factor of the hash map. * * @return The load factor of the hash map */ - float get_load_factor() const noexcept { return static_cast(size_) / capacity_; } + [[nodiscard]] float get_load_factor() const noexcept + { + return static_cast(size_) / capacity_; + } /** * @brief Gets the sentinel value used to represent an empty key slot. * * @return The sentinel value used to represent an empty key slot */ - Key get_empty_key_sentinel() const noexcept { return empty_key_sentinel_; } + [[nodiscard]] Key get_empty_key_sentinel() const noexcept { return empty_key_sentinel_; } /** * @brief Gets the sentinel value used to represent an empty value slot. * * @return The sentinel value used to represent an empty value slot */ - Value get_empty_value_sentinel() const noexcept { return empty_value_sentinel_; } + [[nodiscard]] Value get_empty_value_sentinel() const noexcept { return empty_value_sentinel_; } /** * @brief Constructs a device_view object based on the members of the `static_map` object. * * @return A device_view object based on the members of the `static_map` object */ - device_view get_device_view() const noexcept + [[nodiscard]] device_view get_device_view() const noexcept { return device_view(slots_, capacity_, empty_key_sentinel_, empty_value_sentinel_); } @@ -1090,7 +1107,7 @@ class static_map { * * @return A device_mutable_view object based on the members of the `static_map` object */ - device_mutable_view get_device_mutable_view() const noexcept + [[nodiscard]] device_mutable_view get_device_mutable_view() const noexcept { return device_mutable_view(slots_, capacity_, empty_key_sentinel_, empty_value_sentinel_); } diff --git a/include/cuco/static_multimap.cuh b/include/cuco/static_multimap.cuh index 59ea1befd..505c81920 100644 --- a/include/cuco/static_multimap.cuh +++ b/include/cuco/static_multimap.cuh @@ -592,7 +592,7 @@ class static_multimap { * * @return Slots array */ - __device__ __forceinline__ pair_atomic_type const* get_slots() const noexcept + [[nodiscard]] __device__ __forceinline__ pair_atomic_type const* get_slots() const noexcept { return impl_.get_slots(); } @@ -602,7 +602,7 @@ class static_multimap { * * @return The maximum number of elements the hash map can hold */ - __host__ __device__ __forceinline__ std::size_t get_capacity() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ std::size_t get_capacity() const noexcept { return impl_.get_capacity(); } @@ -612,7 +612,7 @@ class static_multimap { * * @return The sentinel value used to represent an empty key slot */ - __host__ __device__ __forceinline__ Key get_empty_key_sentinel() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ Key get_empty_key_sentinel() const noexcept { return impl_.get_empty_key_sentinel(); } @@ -622,7 +622,8 @@ class static_multimap { * * @return The sentinel value used to represent an empty value slot */ - __host__ __device__ __forceinline__ Value get_empty_value_sentinel() const noexcept + [[nodiscard]] __host__ __device__ __forceinline__ Value + get_empty_value_sentinel() const noexcept { return impl_.get_empty_value_sentinel(); } @@ -1182,7 +1183,7 @@ class static_multimap { /** * @brief Return the raw pointer of the hash map slots. */ - value_type const* raw_slots() const noexcept + [[nodiscard]] value_type const* raw_slots() const noexcept { // Unsafe access to the slots stripping away their atomic-ness to allow non-atomic access. // TODO: to be replace by atomic_ref when it's ready @@ -1194,7 +1195,7 @@ class static_multimap { * * @return The maximum number of elements the hash map can hold */ - std::size_t get_capacity() const noexcept { return capacity_; } + [[nodiscard]] std::size_t get_capacity() const noexcept { return capacity_; } /** * @brief Gets the number of elements in the hash map. @@ -1217,14 +1218,14 @@ class static_multimap { * * @return The sentinel value used to represent an empty key slot */ - Key get_empty_key_sentinel() const noexcept { return empty_key_sentinel_; } + [[nodiscard]] Key get_empty_key_sentinel() const noexcept { return empty_key_sentinel_; } /** * @brief Gets the sentinel value used to represent an empty value slot. * * @return The sentinel value used to represent an empty value slot */ - Value get_empty_value_sentinel() const noexcept { return empty_value_sentinel_; } + [[nodiscard]] Value get_empty_value_sentinel() const noexcept { return empty_value_sentinel_; } /** * @brief Constructs a device_view object based on the members of the `static_multimap` @@ -1232,7 +1233,7 @@ class static_multimap { * * @return A device_view object based on the members of the `static_multimap` object */ - device_view get_device_view() const noexcept + [[nodiscard]] device_view get_device_view() const noexcept { return device_view(slots_.get(), capacity_, empty_key_sentinel_, empty_value_sentinel_); } @@ -1243,7 +1244,7 @@ class static_multimap { * * @return A device_mutable_view object based on the members of the `static_multimap` object */ - device_mutable_view get_device_mutable_view() const noexcept + [[nodiscard]] device_mutable_view get_device_mutable_view() const noexcept { return device_mutable_view(slots_.get(), capacity_, empty_key_sentinel_, empty_value_sentinel_); } From 47b6b4d261313a7f400bbf01936acbfa56beb93b Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Fri, 17 Dec 2021 19:39:12 -0500 Subject: [PATCH 06/11] Move python script to scripts --- include/cuco/static_multimap.cuh | 6 +++--- run_clang_tidy.py => scripts/run_clang_tidy.py | 0 2 files changed, 3 insertions(+), 3 deletions(-) rename run_clang_tidy.py => scripts/run_clang_tidy.py (100%) diff --git a/include/cuco/static_multimap.cuh b/include/cuco/static_multimap.cuh index 505c81920..56b9cc962 100644 --- a/include/cuco/static_multimap.cuh +++ b/include/cuco/static_multimap.cuh @@ -747,7 +747,7 @@ class static_multimap { */ template __device__ __forceinline__ static device_view make_copy( - CG g, pair_atomic_type* const memory_to_use, device_view source_device_view) noexcept; + CG g, pair_atomic_type* memory_to_use, device_view source_device_view) noexcept; /** * @brief Flushes per-CG buffer into the output sequence. @@ -770,7 +770,7 @@ class static_multimap { */ template __device__ __forceinline__ void flush_output_buffer(CG const& g, - uint32_t const num_outputs, + uint32_t num_outputs, value_type* output_buffer, atomicT* num_matches, OutputIt output_begin) noexcept; @@ -800,7 +800,7 @@ class static_multimap { */ template __device__ __forceinline__ void flush_output_buffer(CG const& g, - uint32_t const num_outputs, + uint32_t num_outputs, value_type* probe_output_buffer, value_type* contained_output_buffer, atomicT* num_matches, diff --git a/run_clang_tidy.py b/scripts/run_clang_tidy.py similarity index 100% rename from run_clang_tidy.py rename to scripts/run_clang_tidy.py From 877a47dc4635a2cbfd50358186f1d88bd27c6ce9 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Mon, 20 Dec 2021 10:42:05 -0500 Subject: [PATCH 07/11] Add style check bash script --- ci/checks/style.sh | 24 ++++++++++++++++++++++++ 1 file changed, 24 insertions(+) create mode 100644 ci/checks/style.sh diff --git a/ci/checks/style.sh b/ci/checks/style.sh new file mode 100644 index 000000000..a69babacc --- /dev/null +++ b/ci/checks/style.sh @@ -0,0 +1,24 @@ +#!/bin/bash +# Copyright (c) 2018-2021, NVIDIA CORPORATION. + +# Ignore errors and set path +set +e +PATH=/conda/bin:$PATH +LC_ALL=C.UTF-8 +LANG=C.UTF-8 + +# Activate common conda env +. /opt/conda/etc/profile.d/conda.sh +conda activate rapids + +# Run clang-format and check for a consistent code format +CLANG_FORMAT=`python cpp/scripts/run-clang-format.py 2>&1` +CLANG_FORMAT_RETVAL=$? + +if [ "$CLANG_FORMAT_RETVAL" != "0" ]; then + echo -e "\n\n>>>> FAILED: clang format check; begin output\n\n" + echo -e "$CLANG_FORMAT" + echo -e "\n\n>>>> FAILED: clang format check; end output\n\n" +else + echo -e "\n\n>>>> PASSED: clang format check\n\n" +fi From 0c789b5a9d15b2fb2f92329e9f668593e08aa280 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Mon, 20 Dec 2021 11:54:05 -0500 Subject: [PATCH 08/11] Add run clang format python script --- scripts/run-clang-format.py | 186 ++++++++++++++++++++++++++++++++++++ 1 file changed, 186 insertions(+) create mode 100644 scripts/run-clang-format.py diff --git a/scripts/run-clang-format.py b/scripts/run-clang-format.py new file mode 100644 index 000000000..84683f478 --- /dev/null +++ b/scripts/run-clang-format.py @@ -0,0 +1,186 @@ +# Copyright (c) 2019-2021, 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. +# + +from __future__ import print_function + +import argparse +import os +import re +import subprocess +import sys +import tempfile + +EXPECTED_VERSION = "11.1.0" +VERSION_REGEX = re.compile(r"clang-format version ([0-9.]+)") +# NOTE: populate this list with more top-level dirs as we add more of them to +# the cudf repo +DEFAULT_DIRS = [ + "benchmarks", + "examples", + "include", + "tests", +] + + +def parse_args(): + argparser = argparse.ArgumentParser("Runs clang-format on a project") + argparser.add_argument( + "-dstdir", + type=str, + default=None, + help="Directory to store the temporary outputs of" + " clang-format. If nothing is passed for this, then" + " a temporary dir will be created using `mkdtemp`", + ) + argparser.add_argument( + "-exe", + type=str, + default="clang-format", + help="Path to clang-format exe", + ) + argparser.add_argument( + "-inplace", + default=False, + action="store_true", + help="Replace the source files itself.", + ) + argparser.add_argument( + "-regex", + type=str, + default=r"[.](cu|cuh|h|hpp|cpp|inl)$", + help="Regex string to filter in sources", + ) + argparser.add_argument( + "-ignore", + type=str, + default=r"cannylab/bh[.]cu$", + help="Regex used to ignore files from matched list", + ) + argparser.add_argument( + "-v", + dest="verbose", + action="store_true", + help="Print verbose messages", + ) + argparser.add_argument( + "dirs", type=str, nargs="*", help="List of dirs where to find sources" + ) + args = argparser.parse_args() + args.regex_compiled = re.compile(args.regex) + args.ignore_compiled = re.compile(args.ignore) + if args.dstdir is None: + args.dstdir = tempfile.mkdtemp() + ret = subprocess.check_output("%s --version" % args.exe, shell=True) + ret = ret.decode("utf-8") + version = VERSION_REGEX.match(ret) + if version is None: + raise Exception("Failed to figure out clang-format version!") + version = version.group(1) + if version != EXPECTED_VERSION: + raise Exception( + "clang-format exe must be v%s found '%s'" + % (EXPECTED_VERSION, version) + ) + if len(args.dirs) == 0: + args.dirs = DEFAULT_DIRS + return args + + +def list_all_src_files(file_regex, ignore_regex, srcdirs, dstdir, inplace): + allFiles = [] + for srcdir in srcdirs: + for root, dirs, files in os.walk(srcdir): + for f in files: + if re.search(file_regex, f): + src = os.path.join(root, f) + if re.search(ignore_regex, src): + continue + if inplace: + _dir = root + else: + _dir = os.path.join(dstdir, root) + dst = os.path.join(_dir, f) + allFiles.append((src, dst)) + return allFiles + + +def run_clang_format(src, dst, exe, verbose, inplace): + dstdir = os.path.dirname(dst) + if not os.path.exists(dstdir): + os.makedirs(dstdir) + # run the clang format command itself + if src == dst: + cmd = "%s -i %s" % (exe, src) + else: + cmd = "%s %s > %s" % (exe, src, dst) + try: + subprocess.check_call(cmd, shell=True) + except subprocess.CalledProcessError: + print("Failed to run clang-format! Maybe your env is not proper?") + raise + # run the diff to check if there are any formatting issues + if inplace: + cmd = "diff -q %s %s >/dev/null" % (src, dst) + else: + cmd = "diff %s %s" % (src, dst) + + try: + subprocess.check_call(cmd, shell=True) + if verbose: + print("%s passed" % os.path.basename(src)) + except subprocess.CalledProcessError: + print( + "%s failed! 'diff %s %s' will show formatting violations!" + % (os.path.basename(src), src, dst) + ) + return False + return True + + +def main(): + args = parse_args() + # Attempt to making sure that we run this script from root of repo always + if not os.path.exists(".git"): + print("Error!! This needs to always be run from the root of repo") + sys.exit(-1) + all_files = list_all_src_files( + args.regex_compiled, + args.ignore_compiled, + args.dirs, + args.dstdir, + args.inplace, + ) + # actual format checker + status = True + for src, dst in all_files: + if not run_clang_format( + src, dst, args.exe, args.verbose, args.inplace + ): + status = False + if not status: + print("clang-format failed! You have 2 options:") + print(" 1. Look at formatting differences above and fix them manually") + print(" 2. Or run the below command to bulk-fix all these at once") + print("Bulk-fix command: ") + print( + " python scripts/run-clang-format.py %s -inplace" + % " ".join(sys.argv[1:]) + ) + sys.exit(-1) + return + + +if __name__ == "__main__": + main() From 31fc0b22c5e6b1a4f678b8bd38f1e79d3159e5f7 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Mon, 20 Dec 2021 11:54:26 -0500 Subject: [PATCH 09/11] Update clang format configuration file --- .clang-format | 26 +++++++++++++------------- 1 file changed, 13 insertions(+), 13 deletions(-) diff --git a/.clang-format b/.clang-format index 2f0f549cf..6019a6f3d 100644 --- a/.clang-format +++ b/.clang-format @@ -6,15 +6,21 @@ Language: Cpp AccessModifierOffset: -1 AlignAfterOpenBracket: Align AlignConsecutiveAssignments: true +AlignConsecutiveBitFields: true AlignConsecutiveDeclarations: false +AlignConsecutiveMacros: true AlignEscapedNewlines: Left AlignOperands: true AlignTrailingComments: true +AllowAllArgumentsOnNextLine: true +AllowAllConstructorInitializersOnNextLine: true AllowAllParametersOfDeclarationOnNextLine: true AllowShortBlocksOnASingleLine: true AllowShortCaseLabelsOnASingleLine: true +AllowShortEnumsOnASingleLine: true AllowShortFunctionsOnASingleLine: All AllowShortIfStatementsOnASingleLine: true +AllowShortLambdasOnASingleLine: true AllowShortLoopsOnASingleLine: false # This is deprecated AlwaysBreakAfterDefinitionReturnType: None @@ -40,14 +46,14 @@ BraceWrapping: SplitEmptyFunction: false SplitEmptyRecord: false SplitEmptyNamespace: false +BreakAfterJavaFieldAnnotations: false BreakBeforeBinaryOperators: None BreakBeforeBraces: WebKit BreakBeforeInheritanceComma: false -BreakInheritanceList: BeforeColon BreakBeforeTernaryOperators: true BreakConstructorInitializersBeforeComma: false BreakConstructorInitializers: BeforeColon -BreakAfterJavaFieldAnnotations: false +BreakInheritanceList: BeforeColon BreakStringLiterals: true ColumnLimit: 100 CommentPragmas: '^ IWYU pragma:' @@ -57,7 +63,7 @@ ConstructorInitializerAllOnOneLineOrOnePerLine: true ConstructorInitializerIndentWidth: 2 ContinuationIndentWidth: 2 Cpp11BracedListStyle: true -DerivePointerAlignment: true +DerivePointerAlignment: false DisableFormat: false ExperimentalAutoDetectBinPacking: false FixNamespaceComments: true @@ -66,15 +72,6 @@ ForEachMacros: - Q_FOREACH - BOOST_FOREACH IncludeBlocks: Preserve -IncludeCategories: - - Regex: '^' - Priority: 2 - - Regex: '^<.*\.h>' - Priority: 1 - - Regex: '^<.*' - Priority: 2 - - Regex: '.*' - Priority: 3 IncludeIsMainRegex: '([-_](test|unittest))?$' IndentCaseLabels: true IndentPPDirectives: None @@ -139,14 +136,17 @@ SpaceBeforeCtorInitializerColon: true SpaceBeforeInheritanceColon: true SpaceBeforeParens: ControlStatements SpaceBeforeRangeBasedForLoopColon: true +SpaceBeforeSquareBrackets: false +SpaceInEmptyBlock: false SpaceInEmptyParentheses: false SpacesBeforeTrailingComments: 2 SpacesInAngles: false +SpacesInConditionalStatement: false SpacesInContainerLiterals: true SpacesInCStyleCastParentheses: false SpacesInParentheses: false SpacesInSquareBrackets: false -Standard: Cpp11 +Standard: c++17 StatementMacros: - Q_UNUSED - QT_REQUIRE_VERSION From 0969a3962519982b281223054257911356cfcda3 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Mon, 20 Dec 2021 11:55:28 -0500 Subject: [PATCH 10/11] Code formatting --- benchmarks/hash_table/static_map_bench.cu | 6 +++--- .../hash_table/static_multimap/retrieve_bench.cu | 2 +- .../static_multimap/static_multimap_bench.cu | 2 +- benchmarks/reduce_by_key/reduce_by_key.cu | 12 +++++------- benchmarks/synchronization.hpp | 6 +++--- include/cuco/detail/error.hpp | 2 +- include/cuco/dynamic_map.cuh | 2 +- include/cuco/static_map.cuh | 4 ++-- include/cuco/static_multimap.cuh | 4 ++-- tests/dynamic_map/dynamic_map_test.cu | 6 +++--- 10 files changed, 22 insertions(+), 24 deletions(-) diff --git a/benchmarks/hash_table/static_map_bench.cu b/benchmarks/hash_table/static_map_bench.cu index af8e09de1..59c325d5d 100644 --- a/benchmarks/hash_table/static_map_bench.cu +++ b/benchmarks/hash_table/static_map_bench.cu @@ -14,13 +14,13 @@ * limitations under the License. */ +#include "cuco/static_map.cuh" #include -#include -#include #include #include #include -#include "cuco/static_map.cuh" +#include +#include enum class dist_type { UNIQUE, UNIFORM, GAUSSIAN }; diff --git a/benchmarks/hash_table/static_multimap/retrieve_bench.cu b/benchmarks/hash_table/static_multimap/retrieve_bench.cu index 5856ed7c5..265988444 100644 --- a/benchmarks/hash_table/static_multimap/retrieve_bench.cu +++ b/benchmarks/hash_table/static_multimap/retrieve_bench.cu @@ -16,8 +16,8 @@ #include -#include #include +#include #include "cuco/static_multimap.cuh" diff --git a/benchmarks/hash_table/static_multimap/static_multimap_bench.cu b/benchmarks/hash_table/static_multimap/static_multimap_bench.cu index 4ba873147..d26e04a13 100644 --- a/benchmarks/hash_table/static_multimap/static_multimap_bench.cu +++ b/benchmarks/hash_table/static_multimap/static_multimap_bench.cu @@ -16,9 +16,9 @@ #include +#include #include #include -#include #include #include diff --git a/benchmarks/reduce_by_key/reduce_by_key.cu b/benchmarks/reduce_by_key/reduce_by_key.cu index 0ca08144f..3e4776c6b 100644 --- a/benchmarks/reduce_by_key/reduce_by_key.cu +++ b/benchmarks/reduce_by_key/reduce_by_key.cu @@ -16,13 +16,13 @@ #include +#include #include #include #include #include #include #include -#include /** * @brief Generates input sizes and number of unique keys @@ -76,13 +76,11 @@ static void BM_thrust(::benchmark::State& state) } } BENCHMARK_TEMPLATE(BM_thrust, int32_t, int32_t) - ->Unit(benchmark::kMillisecond) - ->Apply(generate_size_and_num_unique); + ->Unit(benchmark::kMillisecond) + ->Apply(generate_size_and_num_unique); BENCHMARK_TEMPLATE(BM_thrust, int64_t, int64_t) - ->Unit(benchmark::kMillisecond) - ->Apply(generate_size_and_num_unique); + ->Unit(benchmark::kMillisecond) + ->Apply(generate_size_and_num_unique); // TODO: Hash based reduce by key benchmark - - diff --git a/benchmarks/synchronization.hpp b/benchmarks/synchronization.hpp index e91511281..d9d6019ca 100644 --- a/benchmarks/synchronization.hpp +++ b/benchmarks/synchronization.hpp @@ -81,7 +81,7 @@ class cuda_event_timer { * every iteration. * @param[in] stream_ The CUDA stream we are measuring time on. */ - cuda_event_timer(benchmark::State &state, bool flush_l2_cache = false, cudaStream_t stream = 0) + cuda_event_timer(benchmark::State& state, bool flush_l2_cache = false, cudaStream_t stream = 0) : p_state(&state), stream_(stream) { // flush all of L2$ @@ -95,7 +95,7 @@ class cuda_event_timer { if (l2_cache_bytes > 0) { const int memset_value = 0; - int *l2_cache_buffer = nullptr; + int* l2_cache_buffer = nullptr; BENCH_CUDA_TRY(cudaMalloc(&l2_cache_buffer, l2_cache_bytes)); BENCH_CUDA_TRY(cudaMemsetAsync(l2_cache_buffer, memset_value, l2_cache_bytes, stream_)); BENCH_CUDA_TRY(cudaFree(l2_cache_buffer)); @@ -128,5 +128,5 @@ class cuda_event_timer { cudaEvent_t start_; cudaEvent_t stop_; cudaStream_t stream_; - benchmark::State *p_state; + benchmark::State* p_state; }; \ No newline at end of file diff --git a/include/cuco/detail/error.hpp b/include/cuco/detail/error.hpp index ff2f106a6..f5f331222 100644 --- a/include/cuco/detail/error.hpp +++ b/include/cuco/detail/error.hpp @@ -32,7 +32,7 @@ struct cuda_error : public std::runtime_error { } // namespace cuco #define STRINGIFY_DETAIL(x) #x -#define CUCO_STRINGIFY(x) STRINGIFY_DETAIL(x) +#define CUCO_STRINGIFY(x) STRINGIFY_DETAIL(x) /** * @brief Error checking macro for CUDA runtime API functions. diff --git a/include/cuco/dynamic_map.cuh b/include/cuco/dynamic_map.cuh index 2e57ac6d2..d762de41e 100644 --- a/include/cuco/dynamic_map.cuh +++ b/include/cuco/dynamic_map.cuh @@ -17,12 +17,12 @@ #pragma once #include -#include #include #include #include #include #include +#include namespace cuco { diff --git a/include/cuco/static_map.cuh b/include/cuco/static_map.cuh index 321b1f3da..199dcd838 100644 --- a/include/cuco/static_map.cuh +++ b/include/cuco/static_map.cuh @@ -17,11 +17,11 @@ #pragma once #include -#include -#include #include #include #include +#include +#include #include #include diff --git a/include/cuco/static_multimap.cuh b/include/cuco/static_multimap.cuh index fca14ebe2..fea5b0565 100644 --- a/include/cuco/static_multimap.cuh +++ b/include/cuco/static_multimap.cuh @@ -17,10 +17,10 @@ #pragma once #include -#include -#include #include #include +#include +#include #include #include diff --git a/tests/dynamic_map/dynamic_map_test.cu b/tests/dynamic_map/dynamic_map_test.cu index fd38ea642..7bf25a594 100644 --- a/tests/dynamic_map/dynamic_map_test.cu +++ b/tests/dynamic_map/dynamic_map_test.cu @@ -14,13 +14,13 @@ * limitations under the License. */ -#include -#include -#include #include #include #include #include +#include +#include +#include enum class dist_type { UNIQUE, UNIFORM, GAUSSIAN }; From b95b5e9de8b6fafbccb3d306c4a45aa0ddb97139 Mon Sep 17 00:00:00 2001 From: Yunsong Wang Date: Mon, 20 Dec 2021 14:56:48 -0500 Subject: [PATCH 11/11] Add execute permission for scripts --- ci/checks/style.sh | 2 +- scripts/run-clang-format.py | 0 2 files changed, 1 insertion(+), 1 deletion(-) mode change 100644 => 100755 ci/checks/style.sh mode change 100644 => 100755 scripts/run-clang-format.py diff --git a/ci/checks/style.sh b/ci/checks/style.sh old mode 100644 new mode 100755 index a69babacc..96272b675 --- a/ci/checks/style.sh +++ b/ci/checks/style.sh @@ -12,7 +12,7 @@ LANG=C.UTF-8 conda activate rapids # Run clang-format and check for a consistent code format -CLANG_FORMAT=`python cpp/scripts/run-clang-format.py 2>&1` +CLANG_FORMAT=`python scripts/run-clang-format.py 2>&1` CLANG_FORMAT_RETVAL=$? if [ "$CLANG_FORMAT_RETVAL" != "0" ]; then diff --git a/scripts/run-clang-format.py b/scripts/run-clang-format.py old mode 100644 new mode 100755