diff --git a/c2h/CMakeLists.txt b/c2h/CMakeLists.txt index 6e0367b1e36..0f9a437a6fd 100644 --- a/c2h/CMakeLists.txt +++ b/c2h/CMakeLists.txt @@ -36,6 +36,7 @@ add_library( generators_uniform_offsets.cu generators_vector.cu ) +target_compile_definitions(cccl.c2h PUBLIC CATCH_CONFIG_PREFIX_ALL) target_include_directories(cccl.c2h PUBLIC "${C2H_SOURCE_DIR}/include") target_link_libraries( cccl.c2h diff --git a/c2h/include/c2h/catch2_test_helper.h b/c2h/include/c2h/catch2_test_helper.h index ff62c885970..a32a54106d4 100644 --- a/c2h/include/c2h/catch2_test_helper.h +++ b/c2h/include/c2h/catch2_test_helper.h @@ -19,6 +19,7 @@ #include #include +#include #include #include #include @@ -595,26 +596,26 @@ class nvtx_fixture #define C2H_TEST_IMPL(ID, NAME, TAG, ...) \ using C2H_TEST_CONCAT(types_, ID) = c2h::cartesian_product<__VA_ARGS__>; \ - TEMPLATE_LIST_TEST_CASE_METHOD(::detail::nvtx_fixture, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) + CATCH_TEMPLATE_LIST_TEST_CASE_METHOD(::detail::nvtx_fixture, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) #define C2H_TEST(NAME, TAG, ...) C2H_TEST_IMPL(__LINE__, NAME, TAG, __VA_ARGS__) #define C2H_TEST_WITH_FIXTURE_IMPL(ID, FIXTURE, NAME, TAG, ...) \ using C2H_TEST_CONCAT(types_, ID) = c2h::cartesian_product<__VA_ARGS__>; \ - TEMPLATE_LIST_TEST_CASE_METHOD(FIXTURE, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) + CATCH_TEMPLATE_LIST_TEST_CASE_METHOD(FIXTURE, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) #define C2H_TEST_WITH_FIXTURE(FIXTURE, NAME, TAG, ...) \ C2H_TEST_WITH_FIXTURE_IMPL(__LINE__, FIXTURE, NAME, TAG, __VA_ARGS__) #define C2H_TEST_LIST_IMPL(ID, NAME, TAG, ...) \ using C2H_TEST_CONCAT(types_, ID) = c2h::type_list<__VA_ARGS__>; \ - TEMPLATE_LIST_TEST_CASE_METHOD(::detail::nvtx_fixture, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) + CATCH_TEMPLATE_LIST_TEST_CASE_METHOD(::detail::nvtx_fixture, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) #define C2H_TEST_LIST(NAME, TAG, ...) C2H_TEST_LIST_IMPL(__LINE__, NAME, TAG, __VA_ARGS__) #define C2H_TEST_LIST_WITH_FIXTURE_IMPL(ID, FIXTURE, NAME, TAG, ...) \ using C2H_TEST_CONCAT(types_, ID) = c2h::type_list<__VA_ARGS__>; \ - TEMPLATE_LIST_TEST_CASE_METHOD(FIXTURE, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) + CATCH_TEMPLATE_LIST_TEST_CASE_METHOD(FIXTURE, C2H_TEST_NAME(NAME), TAG, C2H_TEST_CONCAT(types_, ID)) #define C2H_TEST_LIST_WITH_FIXTURE(FIXTURE, NAME, TAG, ...) \ C2H_TEST_LIST_WITH_FIXTURE_IMPL(__LINE__, FIXTURE, NAME, TAG, __VA_ARGS__) diff --git a/c2h/include/c2h/catch2_test_macros.h b/c2h/include/c2h/catch2_test_macros.h new file mode 100644 index 00000000000..8de20d3330e --- /dev/null +++ b/c2h/include/c2h/catch2_test_macros.h @@ -0,0 +1,171 @@ +// SPDX-FileCopyrightText: Copyright (c) 2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception + +#pragma once + +#include + +#include + +#include +#include +#include +#include + +// This file implements Catch2's test macros that work both in host and device code. We globally define the +// CATCH_CONFIG_PREFIX_ALL macro to force Catch2 to prepend it's macros with CATCH_ prefix. That allows us to implement +// the non-prefixed versions ourselves. +// +// In host code, we just use the CATCH_-prefixed variant, in device code we implement the functionality, so it +// corresponds the desired functionality. +// +// Only a subset of the Catch2's macro are provided. If needed, feel free to extend the support. Host-only macros can +// be determined by missing NV_IF_ELSE_TARGET wrapper and immediate dispatch to CATCH_-prefixed variant. + +// We must pass the COND as a cstring parameter, because it might contain the '%' character that would break the printf +// formatting. +#define C2H_INTERNAL_DEVICE_TEST_PRINT(KIND, COND) \ + ::printf( \ + __FILE__ \ + ":" _CCCL_TO_STRING(__LINE__) ":\n " KIND "(%s) failed\n block [%u, %u, %u], thread [%u, %u, %u]\n\n", \ + COND, \ + blockIdx.x, \ + blockIdx.y, \ + blockIdx.z, \ + threadIdx.x, \ + threadIdx.y, \ + threadIdx.z) + +// + +#define REQUIRE(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_REQUIRE(__VA_ARGS__);), ({ \ + if (!(__VA_ARGS__)) \ + { \ + C2H_INTERNAL_DEVICE_TEST_PRINT("REQUIRE", #__VA_ARGS__); \ + ::__trap(); \ + } \ + })) +#define REQUIRE_FALSE(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_REQUIRE_FALSE(__VA_ARGS__);), ({ \ + if (__VA_ARGS__) \ + { \ + C2H_INTERNAL_DEVICE_TEST_PRINT("REQUIRE_FALSE", #__VA_ARGS__); \ + ::__trap(); \ + } \ + })) + +#define REQUIRE_THROWS(...) CATCH_REQUIRE_THROWS(__VA_ARGS__) +#define REQUIRE_THROWS_AS(...) CATCH_REQUIRE_THROWS_AS(__VA_ARGS__) +#define REQUIRE_NOTHROW(...) NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_REQUIRE_NOTHROW(__VA_ARGS__);), (__VA_ARGS__;)) + +#define CHECK(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_CHECK(__VA_ARGS__);), ({ \ + if (!(__VA_ARGS__)) \ + { \ + C2H_INTERNAL_DEVICE_TEST_PRINT("CHECK", #__VA_ARGS__); \ + ::__trap(); \ + } \ + })) +#define CHECK_FALSE(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_CHECK_FALSE(__VA_ARGS__);), ({ \ + if (__VA_ARGS__) \ + { \ + C2H_INTERNAL_DEVICE_TEST_PRINT("CHECK_FALSE", #__VA_ARGS__); \ + ::__trap(); \ + } \ + })) +#define CHECKED_IF(...) CATCH_CHECKED_IF(__VA_ARGS__) +#define CHECKED_ELSE(...) CATCH_CHECKED_ELSE(__VA_ARGS__) +#define CHECK_NOFAIL(...) CATCH_CHECK_NOFAIL(__VA_ARGS__) + +#define CHECK_THROWS(...) CATCH_CHECK_THROWS(__VA_ARGS__) +#define CHECK_THROWS_AS(...) CATCH_CHECK_THROWS_AS(__VA_ARGS__) +#define CHECK_NOTHROW(...) NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_CHECK_NOTHROW(__VA_ARGS__);), (__VA_ARGS__;)) + +#define TEST_CASE(...) CATCH_TEST_CASE(__VA_ARGS__) +#define TEST_CASE_METHOD(...) CATCH_TEST_CASE_METHOD(__VA_ARGS__) +#define METHOD_AS_TEST_CASE(...) CATCH_METHOD_AS_TEST_CASE(__VA_ARGS__) +#define REGISTER_TEST_CASE(...) CATCH_REGISTER_TEST_CASE(__VA_ARGS__) +#define SECTION(...) CATCH_SECTION(__VA_ARGS__) +#define DYNAMIC_SECTION(...) CATCH_DYNAMIC_SECTION(__VA_ARGS__) +#define FAIL(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_FAIL(__VA_ARGS__);), ({ \ + C2H_INTERNAL_DEVICE_TEST_PRINT("FAIL", #__VA_ARGS__); \ + ::__trap(); \ + })) +#define FAIL_CHECK(...) CATCH_FAIL_CHECK(__VA_ARGS__) +#define SUCCEED(...) CATCH_SUCCEED(__VA_ARGS__) +#define SKIP(...) CATCH_SKIP(__VA_ARGS__) + +#define STATIC_REQUIRE(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_STATIC_REQUIRE(__VA_ARGS__);), (static_assert(__VA_ARGS__, #__VA_ARGS__);)) +#define STATIC_REQUIRE_FALSE(...) \ + NV_IF_ELSE_TARGET( \ + NV_IS_HOST, (CATCH_STATIC_REQUIRE_FALSE(__VA_ARGS__);), (static_assert(!(__VA_ARGS__), "!(" #__VA_ARGS__ ")");)) +#define STATIC_CHECK(...) \ + NV_IF_ELSE_TARGET(NV_IS_HOST, (CATCH_STATIC_CHECK(__VA_ARGS__);), (static_assert(__VA_ARGS__, #__VA_ARGS__);)) +#define STATIC_CHECK_FALSE(...) \ + NV_IF_ELSE_TARGET( \ + NV_IS_HOST, (CATCH_STATIC_CHECK_FALSE(__VA_ARGS__);), (static_assert(!(__VA_ARGS__), "!(" #__VA_ARGS__ ")");)) + +#define SCENARIO(...) CATCH_SCENARIO(__VA_ARGS__) +#define SCENARIO_METHOD(...) CATCH_SCENARIO_METHOD(__VA_ARGS__) +#define GIVEN(...) CATCH_GIVEN(__VA_ARGS__) +#define AND_GIVEN(...) CATCH_AND_GIVEN(__VA_ARGS__) +#define WHEN(...) CATCH_WHEN(__VA_ARGS__) +#define AND_WHEN(...) CATCH_AND_WHEN(__VA_ARGS__) +#define THEN(...) CATCH_THEN(__VA_ARGS__) +#define AND_THEN(...) CATCH_AND_THEN(__VA_ARGS__) + +// + +#define INFO(...) CATCH_INFO(__VA_ARGS__) +#define UNSCOPED_INFO(...) CATCH_UNSCOPED_INFO(__VA_ARGS__) +#define WARN(...) CATCH_WARN(__VA_ARGS__) +#define CAPTURE(...) CATCH_CAPTURE(__VA_ARGS__) + +// + +#define TEMPLATE_TEST_CASE(...) CATCH_TEMPLATE_TEST_CASE(__VA_ARGS__) +#define TEMPLATE_TEST_CASE_SIG(...) CATCH_TEMPLATE_TEST_CASE_SIG(__VA_ARGS__) +#define TEMPLATE_TEST_CASE_METHOD(...) CATCH_TEMPLATE_TEST_CASE_METHOD(__VA_ARGS__) +#define TEMPLATE_TEST_CASE_METHOD_SIG(...) CATCH_TEMPLATE_TEST_CASE_METHOD_SIG(__VA_ARGS__) +#define TEMPLATE_PRODUCT_TEST_CASE(...) CATCH_TEMPLATE_PRODUCT_TEST_CASE(__VA_ARGS__) +#define TEMPLATE_PRODUCT_TEST_CASE_SIG(...) CATCH_TEMPLATE_PRODUCT_TEST_CASE_SIG(__VA_ARGS__) +#define TEMPLATE_PRODUCT_TEST_CASE_METHOD(...) CATCH_TEMPLATE_PRODUCT_TEST_CASE_METHOD(__VA_ARGS__) +#define TEMPLATE_PRODUCT_TEST_CASE_METHOD_SIG(...) CATCH_TEMPLATE_PRODUCT_TEST_CASE_METHOD_SIG(__VA_ARGS__) +#define TEMPLATE_LIST_TEST_CASE(...) CATCH_TEMPLATE_LIST_TEST_CASE(__VA_ARGS__) +#define TEMPLATE_LIST_TEST_CASE_METHOD(...) CATCH_TEMPLATE_LIST_TEST_CASE_METHOD(__VA_ARGS__) + +// + +#define REQUIRE_THROWS_WITH(...) CATCH_REQUIRE_THROWS_WITH(__VA_ARGS__) +#define REQUIRE_THROWS_MATCHES(...) CATCH_REQUIRE_THROWS_MATCHES(__VA_ARGS__) +#define CHECK_THROWS_WITH(...) CATCH_CHECK_THROWS_WITH(__VA_ARGS__) +#define CHECK_THROWS_MATCHES(...) CATCH_CHECK_THROWS_MATCHES(__VA_ARGS__) +#define CHECK_THAT(...) CATCH_CHECK_THAT(__VA_ARGS__) +#define REQUIRE_THAT(...) CATCH_REQUIRE_THAT(__VA_ARGS__) + +// extensions + +// Sometimes clang-cuda has problems with REQUIRE(...) when used in __device__ function - it tries to instantiate the +// host path. This is related to clang-cuda's compilation trajectory. For these cases, we provide REQUIRE_DEVICE(...) as +// a fallback. +#define REQUIRE_DEVICE(...) \ + do \ + { \ + if (!(__VA_ARGS__)) \ + { \ + C2H_INTERNAL_DEVICE_TEST_PRINT("REQUIRE", #__VA_ARGS__); \ + ::__trap(); \ + } \ + } while (false) + +// Macros to require/check success of a CUDA Driver call. +#define REQUIRE_CUDA(...) REQUIRE((__VA_ARGS__) == CUDA_SUCCESS) +#define CHECK_CUDA(...) CHECK((__VA_ARGS__) == CUDA_SUCCESS) + +// Macros to require/check success of a CUDA Runtime call. +#define REQUIRE_CUDART(...) REQUIRE((__VA_ARGS__) == cudaSuccess) +#define CHECK_CUDART(...) CHECK((__VA_ARGS__) == cudaSuccess) diff --git a/cub/test/catch2_test_device_radix_sort_custom.cu b/cub/test/catch2_test_device_radix_sort_custom.cu index 9a2e1cb2f78..5e2b21c4f08 100644 --- a/cub/test/catch2_test_device_radix_sort_custom.cu +++ b/cub/test/catch2_test_device_radix_sort_custom.cu @@ -499,12 +499,12 @@ struct decomposer_t }; // example-end custom-type -static __host__ std::ostream& operator<<(std::ostream& os, const custom_t& self) +__host__ std::ostream& operator<<(std::ostream& os, const custom_t& self) { return os << "{ " << self.f << ", " << self.lli << " }"; } -static __host__ __device__ bool operator==(const custom_t& lhs, const custom_t& rhs) +__host__ __device__ bool operator==(const custom_t& lhs, const custom_t& rhs) { return lhs.f == rhs.f && lhs.lli == rhs.lli; } diff --git a/cub/test/catch2_test_device_topk_api.cu b/cub/test/catch2_test_device_topk_api.cu index 8b7d7a52a70..8e8b229172b 100644 --- a/cub/test/catch2_test_device_topk_api.cu +++ b/cub/test/catch2_test_device_topk_api.cu @@ -226,22 +226,22 @@ struct decomposer_t }; // example-end topk-custom-type -static __host__ std::ostream& operator<<(std::ostream& os, const custom_t& self) +__host__ std::ostream& operator<<(std::ostream& os, const custom_t& self) { return os << "{ " << self.f << ", " << self.lli << " }"; } -static __host__ __device__ bool operator==(const custom_t& lhs, const custom_t& rhs) +__host__ __device__ bool operator==(const custom_t& lhs, const custom_t& rhs) { return lhs.f == rhs.f && lhs.lli == rhs.lli; } -static __host__ __device__ bool operator<(const custom_t& lhs, const custom_t& rhs) +__host__ __device__ bool operator<(const custom_t& lhs, const custom_t& rhs) { return lhs.lli == rhs.lli ? lhs.f < rhs.f : lhs.lli < rhs.lli; } -static __host__ __device__ bool operator>(const custom_t& lhs, const custom_t& rhs) +__host__ __device__ bool operator>(const custom_t& lhs, const custom_t& rhs) { return rhs < lhs; } diff --git a/cub/test/insert_nested_NVTX_range_guard.h b/cub/test/insert_nested_NVTX_range_guard.h index 1337666815b..ff1738610a1 100644 --- a/cub/test/insert_nested_NVTX_range_guard.h +++ b/cub/test/insert_nested_NVTX_range_guard.h @@ -8,7 +8,7 @@ #include #include -#include +#include inline thread_local const char* current_nvtx_range_name = nullptr; diff --git a/cudax/test/common/testing.cuh b/cudax/test/common/testing.cuh index 3d834cf7fee..7e550050fd9 100644 --- a/cudax/test/common/testing.cuh +++ b/cudax/test/common/testing.cuh @@ -32,56 +32,13 @@ namespace cudax_async = cuda::experimental::execution; // NOLINT: misc-unused-al #define CUDART(call) REQUIRE((call) == cudaSuccess) -// Unlike nvcc, when clang parses CUDA, both the host and device sections are present (see -// https://llvm.org/docs/CompileCudaWithLLVM.html#compilation-models). The upshot here is that -// certain calls of CUDAX_REQUIRE() from host code will try to use device calls because -// NV_IS_DEVICE is true. -#if _CCCL_CUDA_COMPILER(CLANG) -__host__ -#endif - __device__ inline void - cudax_require_impl(bool condition, - [[maybe_unused]] const char* condition_text, - [[maybe_unused]] const char* filename, - [[maybe_unused]] unsigned int linenum, - [[maybe_unused]] const char* funcname) -{ - if (!condition) - { -#if !_CCCL_CUDA_COMPILER(CLANG) - // TODO do warp aggregate prints for easier readability? - printf("%s:%u: %s: block: [%d,%d,%d], thread: [%d,%d,%d] Condition `%s` failed.\n", - filename, - linenum, - funcname, - blockIdx.x, - blockIdx.y, - blockIdx.z, - threadIdx.x, - threadIdx.y, - threadIdx.z, - condition_text); -#endif - ::cuda::std::terminate(); - } -} - -#define CUDAX_REQUIRE(condition) \ - NV_IF_ELSE_TARGET(NV_IS_DEVICE, \ - (cudax_require_impl(condition, #condition, __FILE__, __LINE__, __PRETTY_FUNCTION__);), \ - (REQUIRE(condition);)) +#define CUDAX_REQUIRE(condition) REQUIRE(condition) -#define CUDAX_CHECK(condition) \ - NV_IF_ELSE_TARGET(NV_IS_DEVICE, \ - (cudax_require_impl(condition, #condition, __FILE__, __LINE__, __PRETTY_FUNCTION__);), \ - (CHECK(condition);)) +#define CUDAX_CHECK(condition) CHECK(condition) -#define CUDAX_FAIL(message) /* */ \ - NV_IF_ELSE_TARGET(NV_IS_DEVICE, /* */ \ - (cudax_require_impl(false, message, __FILE__, __LINE__, __PRETTY_FUNCTION__);), \ - (FAIL(message);)) +#define CUDAX_FAIL(message) FAIL(message) -#define CUDAX_CHECK_FALSE(condition) CUDAX_CHECK(!(condition)) +#define CUDAX_CHECK_FALSE(condition) CHECK_FALSE(condition) __host__ __device__ constexpr bool operator==(const dim3& lhs, const dim3& rhs) noexcept { diff --git a/cudax/test/cufile/driver_register.cu b/cudax/test/cufile/driver_register.cu index 5cd30afdcfa..d943df8f026 100644 --- a/cudax/test/cufile/driver_register.cu +++ b/cudax/test/cufile/driver_register.cu @@ -47,7 +47,7 @@ void test_register_native_handle() CUDAX_REQUIRE(file.get() != nullptr); // 4. Reregistering the same file handle should result in an cufile_error. - CHECK_THROWS_AS(cudax::cufile_driver.register_native_handle(fd), cudax::cufile_error); + CHECK_THROWS_AS((void) cudax::cufile_driver.register_native_handle(fd), cudax::cufile_error); // 5. Deregister the cuFile handles. Can be called multiple times. cudax::cufile_driver.deregister_native_handle(file); diff --git a/cudax/test/utility/optionally_static.cu b/cudax/test/utility/optionally_static.cu index 4a8ea7b789e..e25b33d206b 100644 --- a/cudax/test/utility/optionally_static.cu +++ b/cudax/test/utility/optionally_static.cu @@ -12,7 +12,7 @@ #include -#include +#include namespace cudax = cuda::experimental; diff --git a/libcudacxx/test/libcudacxx/cuda/ccclrt/common/testing.cuh b/libcudacxx/test/libcudacxx/cuda/ccclrt/common/testing.cuh index 1f7f54ab84c..52468fa155d 100644 --- a/libcudacxx/test/libcudacxx/cuda/ccclrt/common/testing.cuh +++ b/libcudacxx/test/libcudacxx/cuda/ccclrt/common/testing.cuh @@ -26,40 +26,19 @@ #define CUDART(call) REQUIRE((call) == cudaSuccess) -// There is a problem with clang-cuda and nv/target, but we don't need the device side macros yet, -// disable them for now -#if _CCCL_CUDA_COMPILER(CLANG) -# define CCCLRT_REQUIRE(condition) REQUIRE(condition) -# define CCCLRT_CHECK(condition) CHECK(condition) -# define CCCLRT_FAIL(message) FAIL(message) -# define CCCLRT_CHECK_FALSE(condition) CCCLRT_CHECK(!(condition)) - -#else // _CCCL_CUDA_COMPILER(CLANG) -# define CCCLRT_REQUIRE(condition) \ - NV_IF_ELSE_TARGET(NV_IS_DEVICE, \ - (ccclrt_require_impl(condition, #condition, __FILE__, __LINE__, __PRETTY_FUNCTION__);), \ - (REQUIRE(condition);)) - -# define CCCLRT_CHECK(condition) \ - NV_IF_ELSE_TARGET(NV_IS_DEVICE, \ - (ccclrt_require_impl(condition, #condition, __FILE__, __LINE__, __PRETTY_FUNCTION__);), \ - (CHECK(condition);)) - -# define CCCLRT_FAIL(message) /* */ \ - NV_IF_ELSE_TARGET(NV_IS_DEVICE, /* */ \ - (ccclrt_require_impl(false, message, __FILE__, __LINE__, __PRETTY_FUNCTION__);), \ - (FAIL(message);)) - -# define CCCLRT_CHECK_FALSE(condition) CCCLRT_CHECK(!(condition)) -#endif // _CCCL_CUDA_COMPILER(CLANG) +#define CCCLRT_REQUIRE(condition) REQUIRE(condition) + +#define CCCLRT_CHECK(condition) CHECK(condition) + +#define CCCLRT_FAIL(message) FAIL(message) + +#define CCCLRT_CHECK_FALSE(condition) CHECK_FALSE(condition) // Explicit device side require macros for clang-cuda -#define CCCLRT_REQUIRE_DEVICE(condition) \ - ccclrt_require_impl(condition, #condition, __FILE__, __LINE__, __PRETTY_FUNCTION__); -#define CCCLRT_CHECK_DEVICE(condition) \ - ccclrt_require_impl(condition, #condition, __FILE__, __LINE__, __PRETTY_FUNCTION__); -#define CCCLRT_FAIL_DEVICE(message) ccclrt_require_impl(false, message, __FILE__, __LINE__, __PRETTY_FUNCTION__); -#define CCCLRT_CHECK_FALSE_DEVICE(condition) CCCLRT_CHECK_DEVICE(!(condition)) +#define CCCLRT_REQUIRE_DEVICE(condition) REQUIRE_DEVICE(condition) +#define CCCLRT_CHECK_DEVICE(condition) CHECK(condition) +#define CCCLRT_FAIL_DEVICE(message) FAIL(message) +#define CCCLRT_CHECK_FALSE_DEVICE(condition) CHECK_FALSE(condition) TEST_FUNC constexpr bool operator==(const dim3& lhs, const dim3& rhs) noexcept { diff --git a/libcudacxx/test/libcudacxx/cuda/containers/buffer/access.cu b/libcudacxx/test/libcudacxx/cuda/containers/buffer/access.cu index 36db24bb697..a5e6bcbd063 100644 --- a/libcudacxx/test/libcudacxx/cuda/containers/buffer/access.cu +++ b/libcudacxx/test/libcudacxx/cuda/containers/buffer/access.cu @@ -121,7 +121,7 @@ C2H_CCCLRT_TEST("cuda::buffer access and stream", "[container][buffer]", test_ty static_assert(cuda::std::is_same_v); CCCLRT_CHECK(v2 == T(4)); - CHECK_THROWS_AS((buf.at(4)), std::out_of_range); + CHECK_THROWS_AS(((void) buf.at(4)), std::out_of_range); } } diff --git a/libcudacxx/test/libcudacxx/cuda/memory_resource/resources/shared_memory_pools.cu b/libcudacxx/test/libcudacxx/cuda/memory_resource/resources/shared_memory_pools.cu index 6d9afa5cab8..a83df8c77cd 100644 --- a/libcudacxx/test/libcudacxx/cuda/memory_resource/resources/shared_memory_pools.cu +++ b/libcudacxx/test/libcudacxx/cuda/memory_resource/resources/shared_memory_pools.cu @@ -168,7 +168,7 @@ C2H_CCCLRT_TEST_LIST("shared_memory_pool comparison", "[memory_resource]", SHARE SECTION("Copies are equal") { - shared_pool copy(pool1); + shared_pool copy(pool1); // NOLINT(performance-unnecessary-copy-initialization) CHECK(pool1 == copy); }