diff --git a/testing/CMakeLists.txt b/testing/CMakeLists.txt index 67e25af45..fdfc04e97 100644 --- a/testing/CMakeLists.txt +++ b/testing/CMakeLists.txt @@ -111,6 +111,14 @@ function(thrust_add_test target_name_var test_name test_src thrust_target) if (("OMP" IN_LIST config_systems) OR ("TBB" IN_LIST config_systems)) set_tests_properties(${test_target} PROPERTIES RUN_SERIAL ON) endif() + + # Check for per-test script. Script will be included in the current scope + # to allow custom property modifications. + get_filename_component(test_cmake_script "${test_src}" NAME_WLE) + set(test_cmake_script "${CMAKE_CURRENT_LIST_DIR}/${test_cmake_script}.cmake") + if (EXISTS "${test_cmake_script}") + include("${test_cmake_script}") + endif() endfunction() file(GLOB test_srcs diff --git a/testing/cuda/stream_legacy.cu b/testing/cuda/stream_legacy.cu new file mode 100644 index 000000000..51c82a096 --- /dev/null +++ b/testing/cuda/stream_legacy.cu @@ -0,0 +1,21 @@ +#include +#include +#include + +#include + +void verify_stream() +{ + auto exec = thrust::device; + auto stream = thrust::cuda_cub::stream(exec); + ASSERT_EQUAL(stream, cudaStreamLegacy); +} + +void TestLegacyDefaultStream() +{ + verify_stream(); + + std::thread t(verify_stream); + t.join(); +} +DECLARE_UNITTEST(TestLegacyDefaultStream); diff --git a/testing/cuda/stream_per_thread.cmake b/testing/cuda/stream_per_thread.cmake new file mode 100644 index 000000000..265f4fdc3 --- /dev/null +++ b/testing/cuda/stream_per_thread.cmake @@ -0,0 +1,11 @@ +# This test should always use per-thread streams on NVCC. +set_target_properties(${test_target} PROPERTIES + COMPILE_OPTIONS + $<$,$>:--default-stream=per-thread> +) + +# NVC++ does not have an equivalent option, and will always +# use the global stream by default. +if (CMAKE_CUDA_COMPILER_ID STREQUAL "Feta") + set_tests_properties(${test_target} PROPERTIES WILL_FAIL ON) +endif() diff --git a/testing/cuda/stream_per_thread.cu b/testing/cuda/stream_per_thread.cu new file mode 100644 index 000000000..ef126e78a --- /dev/null +++ b/testing/cuda/stream_per_thread.cu @@ -0,0 +1,21 @@ +#include +#include +#include + +#include + +void verify_stream() +{ + auto exec = thrust::device; + auto stream = thrust::cuda_cub::stream(exec); + ASSERT_EQUAL(stream, cudaStreamPerThread); +} + +void TestPerThreadDefaultStream() +{ + verify_stream(); + + std::thread t(verify_stream); + t.join(); +} +DECLARE_UNITTEST(TestPerThreadDefaultStream); diff --git a/testing/cuda/stream_per_thread.mk b/testing/cuda/stream_per_thread.mk new file mode 100644 index 000000000..da9adfe1b --- /dev/null +++ b/testing/cuda/stream_per_thread.mk @@ -0,0 +1 @@ +CUDACC_FLAGS += --default-stream per-thread