diff --git a/clang-tools-extra/clang-tidy/CMakeLists.txt b/clang-tools-extra/clang-tidy/CMakeLists.txt --- a/clang-tools-extra/clang-tidy/CMakeLists.txt +++ b/clang-tools-extra/clang-tidy/CMakeLists.txt @@ -58,6 +58,7 @@ add_subdirectory(cert) add_subdirectory(concurrency) add_subdirectory(cppcoreguidelines) +add_subdirectory(cuda) add_subdirectory(darwin) add_subdirectory(fuchsia) add_subdirectory(google) @@ -85,6 +86,7 @@ clangTidyCERTModule clangTidyConcurrencyModule clangTidyCppCoreGuidelinesModule + clangTidyCudaModule clangTidyDarwinModule clangTidyFuchsiaModule clangTidyGoogleModule diff --git a/clang-tools-extra/clang-tidy/ClangTidyForceLinker.h b/clang-tools-extra/clang-tidy/ClangTidyForceLinker.h --- a/clang-tools-extra/clang-tidy/ClangTidyForceLinker.h +++ b/clang-tools-extra/clang-tidy/ClangTidyForceLinker.h @@ -55,6 +55,11 @@ static int LLVM_ATTRIBUTE_UNUSED CppCoreGuidelinesModuleAnchorDestination = CppCoreGuidelinesModuleAnchorSource; +// This anchor is used to force the linker to link the CudaModule. +extern volatile int CudaModuleAnchorSource; +static int LLVM_ATTRIBUTE_UNUSED CudaModuleAnchorDestination = + CudaModuleAnchorSource; + // This anchor is used to force the linker to link the DarwinModule. extern volatile int DarwinModuleAnchorSource; static int LLVM_ATTRIBUTE_UNUSED DarwinModuleAnchorDestination = diff --git a/clang-tools-extra/clang-tidy/cuda/CMakeLists.txt b/clang-tools-extra/clang-tidy/cuda/CMakeLists.txt new file mode 100644 --- /dev/null +++ b/clang-tools-extra/clang-tidy/cuda/CMakeLists.txt @@ -0,0 +1,15 @@ +add_clang_library(clangTidyCudaModule + CudaTidyModule.cpp + LINK_LIBS + clangTidy + clangTidyUtils + ) + +clang_target_link_libraries(clangTidyAlteraModule + PRIVATE + clangAnalysis + clangAST + clangASTMatchers + clangBasic + clangLex + ) diff --git a/clang-tools-extra/clang-tidy/cuda/CudaTidyModule.cpp b/clang-tools-extra/clang-tidy/cuda/CudaTidyModule.cpp new file mode 100644 --- /dev/null +++ b/clang-tools-extra/clang-tidy/cuda/CudaTidyModule.cpp @@ -0,0 +1,36 @@ +//===--- GoogleTidyModule.cpp - clang-tidy --------------------------------===// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +#include "../ClangTidy.h" +#include "../ClangTidyCheck.h" +#include "../ClangTidyModule.h" +#include "../ClangTidyModuleRegistry.h" + +using namespace clang::ast_matchers; + +namespace clang { +namespace tidy { +namespace cuda { + +class CudaModule : public ClangTidyModule { +public: + void addCheckFactories(ClangTidyCheckFactories &CheckFactories) override {} +}; + +// Register the CudaTidyModule using this statically initialized variable. +static ClangTidyModuleRegistry::Add + X("cuda-module", "Adds Cuda-related lint checks."); + +} // namespace cuda + +// This anchor is used to force the linker to link in the generated object file +// and thus register the CudaModule. +volatile int CudaModuleAnchorSource = 0; + +} // namespace tidy +} // namespace clang diff --git a/clang-tools-extra/test/clang-tidy/check_clang_tidy.py b/clang-tools-extra/test/clang-tidy/check_clang_tidy.py --- a/clang-tools-extra/test/clang-tidy/check_clang_tidy.py +++ b/clang-tools-extra/test/clang-tidy/check_clang_tidy.py @@ -93,7 +93,7 @@ file_name_with_extension = self.assume_file_name or self.input_file_name _, extension = os.path.splitext(file_name_with_extension) - if extension not in ['.c', '.hpp', '.m', '.mm']: + if extension not in ['.c', '.cu', '.hpp', '.m', '.mm']: extension = '.cpp' self.temp_file_name = self.temp_file_name + extension @@ -115,9 +115,14 @@ self.clang_extra_args = ['-fobjc-abi-version=2', '-fobjc-arc', '-fblocks'] + \ self.clang_extra_args - if extension in ['.cpp', '.hpp', '.mm']: + if extension in ['.cpp', '.cu', '.hpp', '.mm']: self.clang_extra_args.append('-std=' + self.std) + # Tests should not rely on a certain cuda device being available on the machine, + # or a certain version of it + if extension == '.cu': + self.clang_extra_args.extend(["--no-cuda-version-check", "-nocudalib", "-nocudainc"]) + # Tests should not rely on STL being available, and instead provide mock # implementations of relevant APIs. self.clang_extra_args.append('-nostdinc++') diff --git a/clang-tools-extra/test/clang-tidy/checkers/Inputs/Headers/cuda/cuda-initializers.h b/clang-tools-extra/test/clang-tidy/checkers/Inputs/Headers/cuda/cuda-initializers.h new file mode 100644 --- /dev/null +++ b/clang-tools-extra/test/clang-tidy/checkers/Inputs/Headers/cuda/cuda-initializers.h @@ -0,0 +1,145 @@ +// CUDA struct types with interesting initialization properties. +// Keep in sync with clang/test/SemaCUDA/Inputs/cuda-initializers.h. + +// Base classes with different initializer variants. + +// trivial constructor -- allowed +struct T { + int t; +}; + +// empty constructor +struct EC { + int ec; + __device__ EC() {} // -- allowed + __device__ EC(int) {} // -- not allowed +}; + +// empty destructor +struct ED { + __device__ ~ED() {} // -- allowed +}; + +struct ECD { + __device__ ECD() {} // -- allowed + __device__ ~ECD() {} // -- allowed +}; + +// empty templated constructor -- allowed with no arguments +struct ETC { + template __device__ ETC(T...) {} +}; + +// undefined constructor -- not allowed +struct UC { + int uc; + __device__ UC(); +}; + +// undefined destructor -- not allowed +struct UD { + int ud; + __device__ ~UD(); +}; + +// empty constructor w/ initializer list -- not allowed +struct ECI { + int eci; + __device__ ECI() : eci(1) {} +}; + +// non-empty constructor -- not allowed +struct NEC { + int nec; + __device__ NEC() { nec = 1; } +}; + +// non-empty destructor -- not allowed +struct NED { + int ned; + __device__ ~NED() { ned = 1; } +}; + +// no-constructor, virtual method -- not allowed +struct NCV { + int ncv; + __device__ virtual void vm() {} +}; + +// virtual destructor -- not allowed. +struct VD { + __device__ virtual ~VD() {} +}; + +// dynamic in-class field initializer -- not allowed +__device__ int f(); +struct NCF { + int ncf = f(); +}; + +// static in-class field initializer. NVCC does not allow it, but +// clang generates static initializer for this, so we'll accept it. +// We still can't use it on __shared__ vars as they don't allow *any* +// initializers. +struct NCFS { + int ncfs = 3; +}; + +// undefined templated constructor -- not allowed +struct UTC { + template __device__ UTC(T...); +}; + +// non-empty templated constructor -- not allowed +struct NETC { + int netc; + template __device__ NETC(T...) { netc = 1; } +}; + +// Regular base class -- allowed +struct T_B_T : T {}; + +// Incapsulated object of allowed class -- allowed +struct T_F_T { + T t; +}; + +// array of allowed objects -- allowed +struct T_FA_T { + T t[2]; +}; + + +// Calling empty base class initializer is OK +struct EC_I_EC : EC { + __device__ EC_I_EC() : EC() {} +}; + +// .. though passing arguments is not allowed. +struct EC_I_EC1 : EC { + __device__ EC_I_EC1() : EC(1) {} +}; + +// Virtual base class -- not allowed +struct T_V_T : virtual T {}; + +// Inherited from or incapsulated class with non-empty constructor -- +// not allowed +struct T_B_NEC : NEC {}; +struct T_F_NEC { + NEC nec; +}; +struct T_FA_NEC { + NEC nec[2]; +}; + + +// Inherited from or incapsulated class with non-empty desstructor -- +// not allowed +struct T_B_NED : NED {}; +struct T_F_NED { + NED ned; +}; +struct T_FA_NED { + NED ned[2]; +}; diff --git a/clang-tools-extra/test/clang-tidy/checkers/Inputs/Headers/cuda/cuda.h b/clang-tools-extra/test/clang-tidy/checkers/Inputs/Headers/cuda/cuda.h new file mode 100644 --- /dev/null +++ b/clang-tools-extra/test/clang-tidy/checkers/Inputs/Headers/cuda/cuda.h @@ -0,0 +1,28 @@ +/* Minimal declarations for CUDA support. Testing purposes only. */ + +#include + +#define __constant__ __attribute__((constant)) +#define __device__ __attribute__((device)) +#define __global__ __attribute__((global)) +#define __host__ __attribute__((host)) +#define __shared__ __attribute__((shared)) + +struct dim3 { + unsigned x, y, z; + __host__ __device__ dim3(unsigned x, unsigned y = 1, unsigned z = 1) : x(x), y(y), z(z) {} +}; + +typedef struct cudaStream *cudaStream_t; +typedef enum cudaError {} cudaError_t; +extern "C" int cudaConfigureCall(dim3 gridSize, dim3 blockSize, + size_t sharedSize = 0, + cudaStream_t stream = 0); +extern "C" int __cudaPushCallConfiguration(dim3 gridSize, dim3 blockSize, + size_t sharedSize = 0, + cudaStream_t stream = 0); +extern "C" cudaError_t cudaLaunchKernel(const void *func, dim3 gridDim, + dim3 blockDim, void **args, + size_t sharedMem, cudaStream_t stream); + +extern "C" __device__ int printf(const char*, ...); diff --git a/clang-tools-extra/test/clang-tidy/checkers/cuda/.keep b/clang-tools-extra/test/clang-tidy/checkers/cuda/.keep new file mode 100644