diff --git a/CMakeLists.txt b/CMakeLists.txt index eb9e57b..7c13746 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -121,8 +121,15 @@ if (USE_ONEDPL) endmacro() endif () add_subdirectory(${onedpl_SOURCE_DIR} ${onedpl_BINARY_DIR} EXCLUDE_FROM_ALL) + + # Fixup oneDPL's omission on setting DPCPP definitions. + # We do this after the creation of the oneDPL target. + if (ONEDPL_BACKEND MATCHES "^(dpcpp|dpcpp_only)$") + target_compile_definitions(oneDPL INTERFACE ONEDPL_USE_DPCPP_BACKEND=1) + endif () + endif () -endif() +endif () # include our macros diff --git a/src/dpl_shim.h b/src/dpl_shim.h index e47ae99..d341a59 100644 --- a/src/dpl_shim.h +++ b/src/dpl_shim.h @@ -34,6 +34,8 @@ T *alloc_raw(size_t size) { return sycl::malloc_shared(size, exe_policy.queue template void dealloc_raw(T *ptr) { sycl::free(ptr, exe_policy.queue()); } +static void sync_device(){exe_policy.queue().wait_and_throw(); } + #else // auto exe_policy = dpl::execution::seq; @@ -72,4 +74,6 @@ T *alloc_raw(size_t size) { return (T *) aligned_alloc(ALIGNMENT, sizeof(T) * si template void dealloc_raw(T *ptr) { free(ptr); } +static void sync_device(){ /*no-op*/ } + #endif diff --git a/src/std-data/STDDataStream.cpp b/src/std-data/STDDataStream.cpp index b6641de..d4dc17f 100644 --- a/src/std-data/STDDataStream.cpp +++ b/src/std-data/STDDataStream.cpp @@ -69,6 +69,7 @@ void STDDataStream::copy() { // c[i] = a[i] std::copy(exe_policy, BEGIN(a), END(a), BEGIN(c)); + sync_device(); } template @@ -76,6 +77,7 @@ void STDDataStream::mul() { // b[i] = scalar * c[i]; std::transform(exe_policy, BEGIN(c), END(c), BEGIN(b), [scalar = startScalar](T ci){ return scalar*ci; }); + sync_device(); } template @@ -83,6 +85,7 @@ void STDDataStream::add() { // c[i] = a[i] + b[i]; std::transform(exe_policy, BEGIN(a), END(a), BEGIN(b), BEGIN(c), std::plus()); + sync_device(); } template @@ -90,6 +93,7 @@ void STDDataStream::triad() { // a[i] = b[i] + scalar * c[i]; std::transform(exe_policy, BEGIN(b), END(b), BEGIN(c), BEGIN(a), [scalar = startScalar](T bi, T ci){ return bi+scalar*ci; }); + sync_device(); } template @@ -101,6 +105,7 @@ void STDDataStream::nstream() // 2: a[i] += scalar * c[i]; std::transform(exe_policy, BEGIN(a), END(a), BEGIN(b), BEGIN(a), [](T ai, T bi){ return ai + bi; }); std::transform(exe_policy, BEGIN(a), END(a), BEGIN(c), BEGIN(a), [scalar = startScalar](T ai, T ci){ return ai + scalar*ci; }); + sync_device(); } diff --git a/src/std-data/model.cmake b/src/std-data/model.cmake index 3d2399d..e1697b6 100644 --- a/src/std-data/model.cmake +++ b/src/std-data/model.cmake @@ -35,7 +35,7 @@ register_flag_optional(USE_ONEDPL CMake will handle any flags needed to enable OpenMP if the compiler supports it. TBB - Implements policies using TBB. TBB must be linked via USE_TBB or be available in LD_LIBRARY_PATH. - SYCL - Implements policies through SYCL2020. + DPCPP - Implements policies through SYCL2020. This requires the DPC++ compiler (other SYCL compilers are untested), required SYCL flags are added automatically." "OFF") diff --git a/src/std-indices/STDIndicesStream.cpp b/src/std-indices/STDIndicesStream.cpp index 9d98a1b..04b7829 100644 --- a/src/std-indices/STDIndicesStream.cpp +++ b/src/std-indices/STDIndicesStream.cpp @@ -77,6 +77,7 @@ void STDIndicesStream::copy() { // c[i] = a[i] std::copy(exe_policy, BEGIN(a), END(a), BEGIN(c)); + sync_device(); } template @@ -86,6 +87,7 @@ void STDIndicesStream::mul() std::transform(exe_policy, range.begin(), range.end(), BEGIN(b), [this, scalar = startScalar](int i) { return scalar * c[i]; }); + sync_device(); } template @@ -95,6 +97,7 @@ void STDIndicesStream::add() std::transform(exe_policy, range.begin(), range.end(), BEGIN(c), [this](int i) { return a[i] + b[i]; }); + sync_device(); } template @@ -104,6 +107,7 @@ void STDIndicesStream::triad() std::transform(exe_policy, range.begin(), range.end(), BEGIN(a), [this, scalar = startScalar](int i) { return b[i] + scalar * c[i]; }); + sync_device(); } template @@ -116,6 +120,7 @@ void STDIndicesStream::nstream() std::transform(exe_policy, range.begin(), range.end(), BEGIN(a), [this, scalar = startScalar](int i) { return a[i] + b[i] + scalar * c[i]; }); + sync_device(); } diff --git a/src/std-indices/model.cmake b/src/std-indices/model.cmake index befa933..c2fef28 100644 --- a/src/std-indices/model.cmake +++ b/src/std-indices/model.cmake @@ -35,7 +35,7 @@ register_flag_optional(USE_ONEDPL CMake will handle any flags needed to enable OpenMP if the compiler supports it. TBB - Implements policies using TBB. TBB must be linked via USE_TBB or be available in LD_LIBRARY_PATH. - SYCL - Implements policies through SYCL2020. + DPCPP - Implements policies through SYCL2020. This requires the DPC++ compiler (other SYCL compilers are untested), required SYCL flags are added automatically." "OFF") diff --git a/src/std-ranges/STDRangesStream.cpp b/src/std-ranges/STDRangesStream.cpp index 3ea32e4..8a77a68 100644 --- a/src/std-ranges/STDRangesStream.cpp +++ b/src/std-ranges/STDRangesStream.cpp @@ -89,6 +89,7 @@ void STDRangesStream::copy() c[i] = a[i]; } ); + sync_device(); } template @@ -103,6 +104,7 @@ void STDRangesStream::mul() b[i] = scalar * c[i]; } ); + sync_device(); } template @@ -115,6 +117,7 @@ void STDRangesStream::add() c[i] = a[i] + b[i]; } ); + sync_device(); } template @@ -129,6 +132,7 @@ void STDRangesStream::triad() a[i] = b[i] + scalar * c[i]; } ); + sync_device(); } template @@ -143,6 +147,7 @@ void STDRangesStream::nstream() a[i] += b[i] + scalar * c[i]; } ); + sync_device(); } template diff --git a/src/std-ranges/model.cmake b/src/std-ranges/model.cmake index 268cc14..35554c7 100644 --- a/src/std-ranges/model.cmake +++ b/src/std-ranges/model.cmake @@ -19,7 +19,7 @@ register_flag_optional(USE_ONEDPL CMake will handle any flags needed to enable OpenMP if the compiler supports it. TBB - Implements policies using TBB. TBB must be linked via USE_TBB or be available in LD_LIBRARY_PATH. - SYCL - Implements policies through SYCL2020. + DPCPP - Implements policies through SYCL2020. This requires the DPC++ compiler (other SYCL compilers are untested), required SYCL flags are added automatically." "OFF")