Support ComputeCpp 2.6.0 - 2.8.0, set ComputeCpp SYCL version to 2020

celerity · Jan 18, 2022 · 2831b2a · 2831b2a
1 parent 008a868
commit 2831b2a
Show file tree

Hide file tree

Showing 10 changed files with 49 additions and 25 deletions.
diff --git a/.github/workflows/celerity_ci.yml b/.github/workflows/celerity_ci.yml
@@ -34,6 +34,30 @@ jobs:
           - sycl-version: "computecpp:2.6.0"
             platform: "intel"
             build-type: "Release"
+          - sycl-version: "computecpp:2.7.0"
+            platform: "intel"
+            build-type: "Debug"
+          - sycl-version: "computecpp:2.7.0"
+            platform: "intel"
+            build-type: "Release"
+          - sycl-version: "computecpp:2.7.0-experimental"
+            platform: "intel"
+            build-type: "Debug"
+          - sycl-version: "computecpp:2.7.0-experimental"
+            platform: "intel"
+            build-type: "Release"
+          - sycl-version: "computecpp:2.8.0"
+            platform: "intel"
+            build-type: "Debug"
+          - sycl-version: "computecpp:2.8.0"
+            platform: "intel"
+            build-type: "Release"
+          - sycl-version: "computecpp:2.8.0-experimental"
+            platform: "intel"
+            build-type: "Debug"
+          - sycl-version: "computecpp:2.8.0-experimental"
+            platform: "intel"
+            build-type: "Release"
           - sycl-version: "hipsycl:HEAD"
             platform: "nvidia"
             build-type: "Debug"

diff --git a/CMakeLists.txt b/CMakeLists.txt
@@ -42,7 +42,7 @@ else()
 
   find_package(ComputeCpp QUIET)
   if(ComputeCpp_FOUND)
-    message(STATUS "Found ComputeCpp: ${ComputeCpp_DIR}")
+    message(STATUS "Found ComputeCpp ${ComputeCpp_VERSION}: ${ComputeCpp_DIR}")
     list(APPEND AVAILABLE_SYCL_IMPLS ComputeCpp)
   endif()
 
@@ -65,9 +65,12 @@ if(CELERITY_SYCL_IMPL STREQUAL "DPC++")
   message(STATUS "DPC++ will target ${CELERITY_DPCPP_TARGETS}")
 endif()
 
-if(CELERITY_SYCL_IMPL STREQUAL "ComputeCpp" AND NOT COMPUTECPP_USER_FLAGS MATCHES "-D_CRT_SECURE_NO_WARNINGS")
-  set(COMPUTECPP_USER_FLAGS "${COMPUTECPP_USER_FLAGS};-D_CRT_SECURE_NO_WARNINGS=1"
-    CACHE STRING "" FORCE)
+if(CELERITY_SYCL_IMPL STREQUAL "ComputeCpp")
+  list(APPEND CELERITY_CXX_FLAGS -DSYCL_LANGUAGE_VERSION=COMPUTECPP_SYCL_VERSION_2020)
+  list(APPEND COMPUTECPP_DEVICE_COMPILER_FLAGS -DSYCL_LANGUAGE_VERSION=COMPUTECPP_SYCL_VERSION_2020)
+endif()
+if(MSVC)
+  list(APPEND CELERITY_CXX_FLAGS -D_CRT_SECURE_NO_WARNINGS)
 endif()
 
 if(CELERITY_SYCL_IMPL STREQUAL hipSYCL)
@@ -219,6 +222,8 @@ elseif(CMAKE_CXX_COMPILER_ID MATCHES "GNU|Clang|AppleClang")
   target_compile_options(celerity_runtime PRIVATE -Wall -Wextra -Wno-unused-parameter -Werror=return-type -Werror=init-self)
 endif()
 
+target_compile_options(celerity_runtime PRIVATE "${CELERITY_CXX_FLAGS}")
+
 # Examples
 
 option(CELERITY_BUILD_EXAMPLES "Build various example applications" ON)

diff --git a/cmake/AddToTarget.cmake b/cmake/AddToTarget.cmake
@@ -24,7 +24,9 @@ function(add_celerity_to_target)
     "${one_value_args}"
     "${multi_value_args}"
     ${ARGN}
-    )
+  )
+
+  target_compile_options(${ADD_CELERITY_TARGET} PRIVATE "${CELERITY_CXX_FLAGS}")
 
   set_property(
     TARGET ${ADD_CELERITY_TARGET}

diff --git a/cmake/celerity-config.cmake.in b/cmake/celerity-config.cmake.in
@@ -42,6 +42,11 @@ elseif(CELERITY_SYCL_IMPL STREQUAL "DPC++")
   endif()
 endif()
 
+set(CELERITY_CXX_FLAGS "@CELERITY_CXX_FLAGS@")
+if(CELERITY_SYCL_IMPL STREQUAL "ComputeCpp")
+  set(COMPUTECPP_DEVICE_COMPILER_FLAGS "@COMPUTECPP_DEVICE_COMPILER_FLAGS@")
+endif()
+
 set(CELERITY_FEATURE_SCALAR_REDUCTIONS "@CELERITY_FEATURE_SCALAR_REDUCTIONS@")
 set(CELERITY_FEATURE_SIMPLE_SCALAR_REDUCTIONS "@CELERITY_FEATURE_SIMPLE_SCALAR_REDUCTIONS@")
 set(CELERITY_FEATURE_LOCAL_ACCESSOR "@CELERITY_FEATURE_LOCAL_ACCESSOR@")

diff --git a/include/accessor.h b/include/accessor.h
@@ -252,11 +252,13 @@ class accessor<DataT, Dims, Mode, target::device> : public detail::accessor_base
 	friend bool operator!=(const accessor& lhs, const accessor& rhs) { return !(lhs == rhs); }
 
   private:
-#if WORKAROUND_COMPUTECPP || WORKAROUND_DPCPP
+#if WORKAROUND_DPCPP || WORKAROUND(COMPUTECPP, 2, 6)
 #pragma GCC diagnostic push
 #pragma GCC diagnostic ignored "-Wdeprecated-declarations" // target::gobal_buffer is now target::device, but only for very recent versions of DPC++
 	using sycl_accessor_t = cl::sycl::accessor<DataT, Dims, Mode, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::true_t>;
 #pragma GCC diagnostic pop
+#elif WORKAROUND_COMPUTECPP
+	using sycl_accessor_t = cl::sycl::accessor<DataT, Dims, Mode, cl::sycl::access::target::device, cl::sycl::access::placeholder::true_t>;
 #else
 	using sycl_accessor_t = cl::sycl::accessor<DataT, Dims, Mode, cl::sycl::target::device>;
 #endif

diff --git a/include/item.h b/include/item.h
@@ -225,7 +225,7 @@ class nd_item {
 
 	size_t get_group_range(int dimension) const { return group_range[dimension]; }
 
-#if !WORKAROUND_COMPUTECPP && !WORKAROUND_DPCPP // no sub_group support
+#if !WORKAROUND_DPCPP // no sub_group support
 	cl::sycl::sub_group get_sub_group() const { return sycl_item.get_sub_group(); }
 #endif
 
@@ -290,18 +290,11 @@ class nd_item {
 };
 
 
-#if !WORKAROUND_COMPUTECPP
 using cl::sycl::group_barrier;
-#endif
 
 template <int Dims>
 void group_barrier(const group<Dims>& g, memory_scope scope = memory_scope_work_group) {
-#if WORKAROUND_COMPUTECPP
-	auto space = scope > memory_scope_work_group ? cl::sycl::access::fence_space::global_and_local : cl::sycl::access::fence_space::local_space;
-	detail::get_sycl_item(g).barrier(space);
-#else
 	cl::sycl::group_barrier(detail::get_sycl_item(g).get_group(), static_cast<cl::sycl::memory_scope>(scope)); // identical representation
-#endif
 }
 
 

diff --git a/include/sycl_wrappers.h b/include/sycl_wrappers.h
@@ -15,9 +15,7 @@ namespace detail {
 	struct read_write_host_task_tag_t {};
 	struct write_only_host_task_tag_t {};
 
-#if !WORKAROUND_COMPUTECPP // no memory_scope
 	using memory_scope_ut = std::underlying_type_t<cl::sycl::memory_scope>;
-#endif
 
 } // namespace detail
 
@@ -56,16 +54,12 @@ namespace property {
 inline const property::no_init no_init;
 
 // We do not expose memory_scope::system because Celerity does not expose USM or multi-GPU kernel launches
-#if WORKAROUND_COMPUTECPP // no memory_scope
-enum class memory_scope { work_item, sub_group, work_group, device };
-#else
 enum class memory_scope : detail::memory_scope_ut {
 	work_item = static_cast<detail::memory_scope_ut>(cl::sycl::memory_scope::work_item),
 	sub_group = static_cast<detail::memory_scope_ut>(cl::sycl::memory_scope::sub_group),
 	work_group = static_cast<detail::memory_scope_ut>(cl::sycl::memory_scope::work_group),
 	device = static_cast<detail::memory_scope_ut>(cl::sycl::memory_scope::device)
 };
-#endif
 
 inline constexpr auto memory_scope_work_item = memory_scope::work_item;
 inline constexpr auto memory_scope_sub_group = memory_scope::sub_group;

diff --git a/src/device_queue.cc b/src/device_queue.cc
@@ -14,7 +14,7 @@ namespace detail {
 		if(device_profiling_enabled) { queue_logger.info("Device profiling enabled."); }
 
 		const auto props = device_profiling_enabled ? cl::sycl::property_list{cl::sycl::property::queue::enable_profiling()} : cl::sycl::property_list{};
-		const auto handle_exceptions = [this](cl::sycl::exception_list el) { this->handle_async_exceptions(el); };
+		const auto handle_exceptions = cl::sycl::async_handler{[this](cl::sycl::exception_list el) { this->handle_async_exceptions(el); }};
 		auto device = pick_device(cfg, user_device);
 		sycl_queue = std::make_unique<cl::sycl::queue>(device, handle_exceptions, props);
 	}

diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt
@@ -21,13 +21,12 @@ add_library(unit_test_suite unit_test_suite.cc unit_test_suite_celerity.cc)
 target_link_libraries(
   unit_test_suite
   PUBLIC
-  celerity_runtime
   Catch2::Catch2
 )
 
 set_property(TARGET unit_test_suite PROPERTY CXX_STANDARD 17)
 
-add_sycl_to_target(TARGET unit_test_suite SOURCES unit_test_suite_celerity.cc)
+add_celerity_to_target(TARGET unit_test_suite SOURCES unit_test_suite_celerity.cc)
 
 foreach(TEST_TARGET ${TEST_TARGETS})
 
@@ -40,7 +39,7 @@ foreach(TEST_TARGET ${TEST_TARGETS})
   set_property(TARGET ${TEST_TARGET} PROPERTY CXX_STANDARD 17)
   set_property(TARGET ${TEST_TARGET} PROPERTY FOLDER "tests")
 
-  add_sycl_to_target(TARGET ${TEST_TARGET} SOURCES ${TEST_SOURCE})
+  add_celerity_to_target(TARGET ${TEST_TARGET} SOURCES ${TEST_SOURCE})
 
   if(MSVC)
     target_compile_options(${TEST_TARGET} PRIVATE /D_CRT_SECURE_NO_WARNINGS /MP /W3 /bigobj)

diff --git a/test/system/CMakeLists.txt b/test/system/CMakeLists.txt
@@ -5,7 +5,7 @@ target_link_libraries(distr_tests PRIVATE unit_test_suite)
 set_property(TARGET distr_tests PROPERTY CXX_STANDARD 17)
 set_property(TARGET distr_tests PROPERTY FOLDER "tests/system")
 
-add_sycl_to_target(TARGET distr_tests SOURCES distr_tests.cc)
+add_celerity_to_target(TARGET distr_tests SOURCES distr_tests.cc)
 
 if(MSVC)
   target_compile_options(distr_tests PRIVATE /D_CRT_SECURE_NO_WARNINGS /MP /W3 /bigobj)