Regards #505, #506: Make the use of the cooperative_groups header in …

…examples explicitly controllable - with MSVC never using them.
eyalroz · May 1, 2023 · 167af65 · 167af65
1 parent ae070e7
commit 167af65
Show file tree

Hide file tree

Showing 2 changed files with 16 additions and 7 deletions.
diff --git a/examples/CMakeLists.txt b/examples/CMakeLists.txt
@@ -1,5 +1,7 @@
 include(CheckIncludeFiles)
 
+option(CAW_WITH_COOPERATIVE_GROUPS "Enable cooperative-groups-related code in example programs" ON)
+
 enable_language(CXX)
 enable_language(CUDA)
 
@@ -23,7 +25,9 @@ set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} ${CUDA_ARCH_FLAGS_STR}")
 include(CompileWithWarnings)
 
 if(MSVC)
-	# MSVC is very problematic!
+	# MSVC is very problematic! ... lets go easy on it
+
+	set(CAW_WITH_COOPERATIVE_GROUPS OFF)
 	set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /Zc:__cplusplus")
 	# set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /NODEFAULTLIB:nvptxcompiler_static.lib")
 	set(CMAKE_CXX_STANDARD 17)
@@ -40,6 +44,11 @@ set(CMAKE_CUDA_STANDARD 11)
 set(CMAKE_CUDA_STANDARD_REQUIRED ON)
 set(CMAKE_CUDA_EXTENSIONS OFF)
 
+if(CAW_WITH_COOPERATIVE_GROUPS AND CUDAToolkit_VERSION VERSION_GREATER_EQUAL "9.0")
+	add_compile_definitions(USE_COOPERATIVE_GROUPS)
+	message(STATUS "Using cooperative groups in example programs")
+endif()
+
 link_libraries(cuda-api-wrappers::runtime-and-driver)
 
 set(CMAKE_RUNTIME_OUTPUT_DIRECTORY "bin")
@@ -123,7 +132,7 @@ IF((NOT ${CMAKE_SYSTEM_PROCESSOR} MATCHES "arm")
 endif()
 add_executable(p2pBandwidthLatencyTest modified_cuda_samples/p2pBandwidthLatencyTest/p2pBandwidthLatencyTest.cu)
 add_executable(asyncAPI modified_cuda_samples/asyncAPI/asyncAPI.cu)
-if(CUDAToolkit_VERSION VERSION_GREATER_EQUAL "11.0")
+if(USE_COOPERATIVE_GROUPS AND CUDAToolkit_VERSION VERSION_GREATER_EQUAL "11.0")
 	# This one needs cooperative_groups/reduce.h
 	add_executable(binaryPartitionCG modified_cuda_samples/binaryPartitionCG/binaryPartitionCG.cu)
 endif()

diff --git a/examples/by_api_module/execution_control.cu b/examples/by_api_module/execution_control.cu
@@ -14,11 +14,11 @@
  */
 #include "../common.hpp"
 
-#if __CUDACC_VER_MAJOR__ >= 9
-#define TEST_COOPERATIVE_GROUPS 1
+#ifdef USE_COOPERATIVE_GROUPS
+#if __CUDACC_VER_MAJOR__ < 9
+#error "Can't use the cooperative groups header with CUDA versions before 9.x!"
+#endif
 #include <cooperative_groups.h>
-#else
-#define TEST_COOPERATIVE_GROUPS 0
 #endif
 
 __global__ void foo(int bar)
@@ -28,7 +28,7 @@ __global__ void foo(int bar)
 	}
 }
 
-#if TEST_COOPERATIVE_GROUPS
+#if USE_COOPERATIVE_GROUPS
 __global__ void grid_cooperating_foo(int bar)
 {
 #ifdef _CG_HAS_GRID_GROUP