diff --git a/.github/workflows/ci.yml b/.github/workflows/ci.yml index e8aadae25b45..efb14d78be46 100644 --- a/.github/workflows/ci.yml +++ b/.github/workflows/ci.yml @@ -16,7 +16,7 @@ concurrency: # Therefore we aim to have each value been set in at lest one job. # ALPAKA_CI_CXX : {g++, clang++} # [g++] ALPAKA_CI_GCC_VER : {10, 11, 12, 13} -# [clang++] ALPAKA_CI_CLANG_VER : {10, 11, 12, 13, 14} +# [clang++] ALPAKA_CI_CLANG_VER : {13, 14, 15, 16, 17, 18, 19} # [cl.exe] ALPAKA_CI_CL_VER : {2022} # ALPAKA_CI_STDLIB : {libstdc++, [ALPAKA_CI_CXX==clang++]:libc++} # CMAKE_BUILD_TYPE : {Debug, Release} @@ -173,18 +173,18 @@ jobs: # - libc++ # - sanitizer jobs # clang++ - - name: linux_clang-10_release + - name: linux_clang-13_release os: ubuntu-20.04 - env: {ALPAKA_CI_CXX: clang++, ALPAKA_CI_CLANG_VER: 10, ALPAKA_CI_STDLIB: libc++, CMAKE_BUILD_TYPE: Release, ALPAKA_BOOST_VERSION: 1.75.0, ALPAKA_CI_CMAKE_VER: 3.25.3, OMP_NUM_THREADS: 4, alpaka_ACC_CPU_B_TBB_T_SEQ_ENABLE: OFF, CMAKE_CXX_EXTENSIONS: OFF} + env: {ALPAKA_CI_CXX: clang++, ALPAKA_CI_CLANG_VER: 13, ALPAKA_CI_STDLIB: libc++, CMAKE_BUILD_TYPE: Release, ALPAKA_BOOST_VERSION: 1.75.0, ALPAKA_CI_CMAKE_VER: 3.25.3, OMP_NUM_THREADS: 4, alpaka_ACC_CPU_B_TBB_T_SEQ_ENABLE: OFF, CMAKE_CXX_EXTENSIONS: OFF} container: ubuntu:20.04 # clang-11 tested in GitLab CI - - name: linux_clang-12_release + - name: linux_clang-14_release os: ubuntu-20.04 - env: {ALPAKA_CI_CXX: clang++, ALPAKA_CI_CLANG_VER: 12, ALPAKA_CI_STDLIB: libc++, CMAKE_BUILD_TYPE: Release, ALPAKA_BOOST_VERSION: 1.77.0, ALPAKA_CI_CMAKE_VER: 3.25.3, OMP_NUM_THREADS: 4, alpaka_ACC_CPU_B_TBB_T_SEQ_ENABLE: OFF, CMAKE_CXX_EXTENSIONS: OFF} + env: {ALPAKA_CI_CXX: clang++, ALPAKA_CI_CLANG_VER: 14, ALPAKA_CI_STDLIB: libc++, CMAKE_BUILD_TYPE: Release, ALPAKA_BOOST_VERSION: 1.77.0, ALPAKA_CI_CMAKE_VER: 3.25.3, OMP_NUM_THREADS: 4, alpaka_ACC_CPU_B_TBB_T_SEQ_ENABLE: OFF, CMAKE_CXX_EXTENSIONS: OFF} container: ubuntu:20.04 - - name: linux_clang-13_debug + - name: linux_clang-15_debug os: ubuntu-22.04 - env: {ALPAKA_CI_CXX: clang++, ALPAKA_CI_CLANG_VER: 13, ALPAKA_CI_STDLIB: libstdc++, CMAKE_BUILD_TYPE: Debug, ALPAKA_BOOST_VERSION: 1.79.0, ALPAKA_CI_CMAKE_VER: 3.25.3, OMP_NUM_THREADS: 3, CMAKE_CXX_EXTENSIONS: OFF} + env: {ALPAKA_CI_CXX: clang++, ALPAKA_CI_CLANG_VER: 15, ALPAKA_CI_STDLIB: libstdc++, CMAKE_BUILD_TYPE: Debug, ALPAKA_BOOST_VERSION: 1.79.0, ALPAKA_CI_CMAKE_VER: 3.25.3, OMP_NUM_THREADS: 3, CMAKE_CXX_EXTENSIONS: OFF} container: ubuntu:22.04 - name: linux_clang-16_debug_ubsan os: ubuntu-22.04 diff --git a/example/bufferCopy/src/bufferCopy.cpp b/example/bufferCopy/src/bufferCopy.cpp index 7c2f9e164d6a..06e4200af6af 100644 --- a/example/bufferCopy/src/bufferCopy.cpp +++ b/example/bufferCopy/src/bufferCopy.cpp @@ -59,7 +59,7 @@ struct FillBufferKernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain diff --git a/example/complex/src/complex.cpp b/example/complex/src/complex.cpp index d9da52a0718a..6c507f78b041 100644 --- a/example/complex/src/complex.cpp +++ b/example/complex/src/complex.cpp @@ -32,7 +32,7 @@ struct ComplexKernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { using Idx = std::size_t; diff --git a/example/conv2DWithMdspan/src/conv2DWithMdspan.cpp b/example/conv2DWithMdspan/src/conv2DWithMdspan.cpp index 698a8d12fa75..4a3f58f7c0e4 100644 --- a/example/conv2DWithMdspan/src/conv2DWithMdspan.cpp +++ b/example/conv2DWithMdspan/src/conv2DWithMdspan.cpp @@ -77,7 +77,7 @@ auto FuzzyEqual(float a, float b) -> bool // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain diff --git a/example/convolution1D/src/convolution1D.cpp b/example/convolution1D/src/convolution1D.cpp index 1e3aec673af7..c39b4c07944e 100644 --- a/example/convolution1D/src/convolution1D.cpp +++ b/example/convolution1D/src/convolution1D.cpp @@ -68,7 +68,7 @@ auto FuzzyEqual(float a, float b) -> bool // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Size of 1D arrays to be used in convolution integral diff --git a/example/convolution2D/src/convolution2D.cpp b/example/convolution2D/src/convolution2D.cpp index 87f618c7380e..ed7c311956e3 100644 --- a/example/convolution2D/src/convolution2D.cpp +++ b/example/convolution2D/src/convolution2D.cpp @@ -212,7 +212,7 @@ auto FuzzyEqual(float a, float b) -> bool // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain diff --git a/example/counterBasedRng/src/counterBasedRng.cpp b/example/counterBasedRng/src/counterBasedRng.cpp index d96ab2b775a2..d2f1fc8dc95b 100644 --- a/example/counterBasedRng/src/counterBasedRng.cpp +++ b/example/counterBasedRng/src/counterBasedRng.cpp @@ -96,7 +96,7 @@ class CounterBasedRngKernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain diff --git a/example/heatEquation/src/heatEquation.cpp b/example/heatEquation/src/heatEquation.cpp index a13b3f00bc26..3ebcad172b6f 100644 --- a/example/heatEquation/src/heatEquation.cpp +++ b/example/heatEquation/src/heatEquation.cpp @@ -67,7 +67,7 @@ auto exactSolution(double const x, double const t) -> double //! Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the //! selected accelerator only. If you use the example as the starting point for your project, you can rename the //! example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Parameters (a user is supposed to change numNodesX, numTimeSteps) diff --git a/example/heatEquation2D/src/heatEquation2D.cpp b/example/heatEquation2D/src/heatEquation2D.cpp index b186e9ec6dba..04af0cdc3fa1 100644 --- a/example/heatEquation2D/src/heatEquation2D.cpp +++ b/example/heatEquation2D/src/heatEquation2D.cpp @@ -31,7 +31,7 @@ //! Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the //! selected accelerator only. If you use the example as the starting point for your project, you can rename the //! example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Set Dim and Idx type diff --git a/example/helloWorld/src/helloWorld.cpp b/example/helloWorld/src/helloWorld.cpp index e18df95a5a90..40ef657571d4 100644 --- a/example/helloWorld/src/helloWorld.cpp +++ b/example/helloWorld/src/helloWorld.cpp @@ -51,7 +51,7 @@ struct HelloWorldKernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain diff --git a/example/helloWorldLambda/src/helloWorldLambda.cpp b/example/helloWorldLambda/src/helloWorldLambda.cpp index 143b9e7c8cbc..4c951b2a9522 100644 --- a/example/helloWorldLambda/src/helloWorldLambda.cpp +++ b/example/helloWorldLambda/src/helloWorldLambda.cpp @@ -43,7 +43,7 @@ void ALPAKA_FN_ACC hiWorldFunction(TAcc const& acc, size_t const nExclamationMar // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // It requires support for extended lambdas when using nvcc as CUDA compiler. diff --git a/example/kernelSpecialization/src/kernelSpecialization.cpp b/example/kernelSpecialization/src/kernelSpecialization.cpp index f33306ec5100..9c640df42340 100644 --- a/example/kernelSpecialization/src/kernelSpecialization.cpp +++ b/example/kernelSpecialization/src/kernelSpecialization.cpp @@ -57,7 +57,7 @@ struct Kernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the accelerator diff --git a/example/matrixMulWithMdspan/src/matrixMulMdSpan.cpp b/example/matrixMulWithMdspan/src/matrixMulMdSpan.cpp index e34dcb2d60fe..c55fe078a476 100644 --- a/example/matrixMulWithMdspan/src/matrixMulMdSpan.cpp +++ b/example/matrixMulWithMdspan/src/matrixMulMdSpan.cpp @@ -90,7 +90,7 @@ inline void initializeMatrx(TMdSpan& span) // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Set number of dimensions (i.e 2) as a type diff --git a/example/monteCarloIntegration/src/monteCarloIntegration.cpp b/example/monteCarloIntegration/src/monteCarloIntegration.cpp index b26cd2af10fa..cd8d9d9a01db 100644 --- a/example/monteCarloIntegration/src/monteCarloIntegration.cpp +++ b/example/monteCarloIntegration/src/monteCarloIntegration.cpp @@ -77,7 +77,7 @@ struct Kernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Defines and setup. diff --git a/example/parallelLoopPatterns/src/parallelLoopPatterns.cpp b/example/parallelLoopPatterns/src/parallelLoopPatterns.cpp index a7e05b362b98..b4d06d25899c 100644 --- a/example/parallelLoopPatterns/src/parallelLoopPatterns.cpp +++ b/example/parallelLoopPatterns/src/parallelLoopPatterns.cpp @@ -404,7 +404,7 @@ void openMPSimdStyle(TDev& dev, TQueue& queue, TBufAcc& bufAcc) // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain, this example is only for 1d diff --git a/example/randomCells2D/src/randomCells2D.cpp b/example/randomCells2D/src/randomCells2D.cpp index 36bc1258d3b0..18b992bbcd35 100644 --- a/example/randomCells2D/src/randomCells2D.cpp +++ b/example/randomCells2D/src/randomCells2D.cpp @@ -145,7 +145,7 @@ struct RunTimestepKernelVector // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { using Dim = alpaka::DimInt<2>; diff --git a/example/randomStrategies/src/randomStrategies.cpp b/example/randomStrategies/src/randomStrategies.cpp index 6a1940c8b244..b8b5ac214c09 100644 --- a/example/randomStrategies/src/randomStrategies.cpp +++ b/example/randomStrategies/src/randomStrategies.cpp @@ -25,7 +25,7 @@ using RandomEngine = alpaka::rand::Philox4x32x10; /// Parameters to set up the default accelerator, queue, and buffers -template +template struct Box { // accelerator, queue, and work division typedefs @@ -184,7 +184,7 @@ struct FillKernel * * File is in TSV format. One line for each "point"; line length is the number of "rolls". */ -template +template void saveDataAndShowAverage(std::string filename, float const* buffer, Box const& box) { std::ofstream output(filename); @@ -207,7 +207,7 @@ struct Writer; template<> struct Writer { - template + template static void save(float const* buffer, Box const& box) { saveDataAndShowAverage("out_seed.csv", buffer, box); @@ -217,7 +217,7 @@ struct Writer template<> struct Writer { - template + template static void save(float const* buffer, Box const& box) { saveDataAndShowAverage("out_subsequence.csv", buffer, box); @@ -227,14 +227,14 @@ struct Writer template<> struct Writer { - template + template static void save(float const* buffer, Box const& box) { saveDataAndShowAverage("out_offset.csv", buffer, box); } }; -template +template void runStrategy(Box& box) { // Set up the pointer to the PRNG states buffer @@ -326,7 +326,7 @@ void runStrategy(Box& box) // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { Box box; // Initialize the box diff --git a/example/tagSpecialization/src/tagSpecialization.cpp b/example/tagSpecialization/src/tagSpecialization.cpp index dac94bbaa32b..735d0678220f 100644 --- a/example/tagSpecialization/src/tagSpecialization.cpp +++ b/example/tagSpecialization/src/tagSpecialization.cpp @@ -28,7 +28,7 @@ std::string host_function_ver1() } //! Function specialization via overloading -template +template std::string host_function_ver2(TTag) { return "generic host function v2"; @@ -80,7 +80,7 @@ struct WrapperKernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the accelerator diff --git a/example/vectorAdd/src/vectorAdd.cpp b/example/vectorAdd/src/vectorAdd.cpp index 91d7bc7baea4..944001b607b4 100644 --- a/example/vectorAdd/src/vectorAdd.cpp +++ b/example/vectorAdd/src/vectorAdd.cpp @@ -48,7 +48,7 @@ class VectorAddKernel // Instead, a single accelerator is selected once from the active accelerators and the kernels are executed with the // selected accelerator only. If you use the example as the starting point for your project, you can rename the // example() function to main() and move the accelerator tag to the function body. -template +template auto example(TAccTag const&) -> int { // Define the index domain diff --git a/include/alpaka/acc/AccGenericSycl.hpp b/include/alpaka/acc/AccGenericSycl.hpp index 404bd6386de1..107ee351f80d 100644 --- a/include/alpaka/acc/AccGenericSycl.hpp +++ b/include/alpaka/acc/AccGenericSycl.hpp @@ -46,13 +46,19 @@ namespace alpaka { - template + template< + alpaka::concepts::Tag TTag, + typename TAcc, + typename TDim, + typename TIdx, + typename TKernelFnObj, + typename... TArgs> class TaskKernelGenericSycl; //! The SYCL accelerator. //! //! This accelerator allows parallel kernel execution on SYCL devices. - template + template class AccGenericSycl : public WorkDivGenericSycl , public gb::IdxGbGenericSycl @@ -103,26 +109,26 @@ namespace alpaka namespace alpaka::trait { //! The SYCL accelerator type trait specialization. - template + template struct AccType> { using type = AccGenericSycl; }; //! The SYCL single thread accelerator type trait specialization. - template + template struct IsSingleThreadAcc> : std::false_type { }; //! The SYCL multi thread accelerator type trait specialization. - template + template struct IsMultiThreadAcc> : std::true_type { }; //! The SYCL accelerator device properties get trait specialization. - template + template struct GetAccDevProps> { static auto getAccDevProps(DevGenericSycl const& dev) -> AccDevProps @@ -159,7 +165,7 @@ namespace alpaka::trait }; //! The SYCL accelerator name trait specialization. - template + template struct GetAccName> { static auto getAccName() -> std::string @@ -170,21 +176,27 @@ namespace alpaka::trait }; //! The SYCL accelerator device type trait specialization. - template + template struct DevType> { using type = DevGenericSycl; }; //! The SYCL accelerator dimension getter trait specialization. - template + template struct DimType> { using type = TDim; }; //! The SYCL accelerator execution task type trait specialization. - template + template< + alpaka::concepts::Tag TTag, + typename TDim, + typename TIdx, + typename TWorkDiv, + typename TKernelFnObj, + typename... TArgs> struct CreateTaskKernel, TWorkDiv, TKernelFnObj, TArgs...> { static auto createTaskKernel(TWorkDiv const& workDiv, TKernelFnObj const& kernelFnObj, TArgs&&... args) @@ -197,14 +209,14 @@ namespace alpaka::trait }; //! The SYCL execution task platform type trait specialization. - template + template struct PlatformType> { using type = PlatformGenericSycl; }; //! The SYCL accelerator idx type trait specialization. - template + template struct IdxType> { using type = TIdx; diff --git a/include/alpaka/acc/Tag.hpp b/include/alpaka/acc/Tag.hpp index f7880afd6f15..828cd6539086 100644 --- a/include/alpaka/acc/Tag.hpp +++ b/include/alpaka/acc/Tag.hpp @@ -9,8 +9,15 @@ #include #include +namespace alpaka +{ + struct InterfaceTag + { + }; +} // namespace alpaka + #define CREATE_ACC_TAG(tag_name) \ - struct tag_name \ + struct tag_name : public alpaka::InterfaceTag \ { \ static std::string get_name() \ { \ @@ -32,12 +39,24 @@ namespace alpaka CREATE_ACC_TAG(TagGpuHipRt); CREATE_ACC_TAG(TagGpuSyclIntel); + namespace concepts + { + template + concept Tag = requires { + { + T::get_name() + } -> std::same_as; + requires std::default_initializable; + requires std::derived_from; + }; + } // namespace concepts + namespace trait { template struct AccToTag; - template + template struct TagToAcc; } // namespace trait @@ -50,10 +69,10 @@ namespace alpaka //! \tparam TTag alpaka tag type //! \tparam TDim dimension of the mapped acc type //! \tparam TIdx index type of the mapped acc type - template + template using TagToAcc = typename trait::TagToAcc::type; - template + template inline constexpr bool accMatchesTags = (std::is_same_v, TTag> || ...); //! list of all available tags diff --git a/include/alpaka/acc/TagAccIsEnabled.hpp b/include/alpaka/acc/TagAccIsEnabled.hpp index c21fd2b149aa..212cf2ebab26 100644 --- a/include/alpaka/acc/TagAccIsEnabled.hpp +++ b/include/alpaka/acc/TagAccIsEnabled.hpp @@ -11,6 +11,7 @@ #include "alpaka/acc/AccFpgaSyclIntel.hpp" #include "alpaka/acc/AccGpuCudaRt.hpp" #include "alpaka/acc/AccGpuHipRt.hpp" +#include "alpaka/acc/Tag.hpp" #include "alpaka/dim/DimIntegralConst.hpp" #include "alpaka/meta/Filter.hpp" @@ -20,12 +21,12 @@ namespace alpaka { //! \brief check if the accelerator is enabled for a given tag //! \tparam TTag alpaka tag type - template + template struct AccIsEnabled : std::false_type { }; - template + template struct AccIsEnabled, int>>> : std::true_type { }; diff --git a/include/alpaka/dev/DevGenericSycl.hpp b/include/alpaka/dev/DevGenericSycl.hpp index 76da1385c7a8..3a16fb90ef2f 100644 --- a/include/alpaka/dev/DevGenericSycl.hpp +++ b/include/alpaka/dev/DevGenericSycl.hpp @@ -4,6 +4,7 @@ #pragma once +#include "alpaka/acc/Tag.hpp" #include "alpaka/acc/Traits.hpp" #include "alpaka/core/Common.hpp" #include "alpaka/core/Sycl.hpp" @@ -38,16 +39,16 @@ namespace alpaka struct GetDevByIdx; } // namespace trait - template + template using QueueGenericSyclBlocking = detail::QueueGenericSyclBase; - template + template using QueueGenericSyclNonBlocking = detail::QueueGenericSyclBase; - template + template struct PlatformGenericSycl; - template + template class BufGenericSycl; namespace detail @@ -120,7 +121,7 @@ namespace alpaka } // namespace detail //! The SYCL device handle. - template + template class DevGenericSycl : public interface::Implements> , public interface::Implements> @@ -154,7 +155,7 @@ namespace alpaka namespace trait { //! The SYCL device name get trait specialization. - template + template struct GetName> { static auto getName(DevGenericSycl const& dev) -> std::string @@ -165,7 +166,7 @@ namespace alpaka }; //! The SYCL device available memory get trait specialization. - template + template struct GetMemBytes> { static auto getMemBytes(DevGenericSycl const& dev) -> std::size_t @@ -176,7 +177,7 @@ namespace alpaka }; //! The SYCL device free memory get trait specialization. - template + template struct GetFreeMemBytes> { static auto getFreeMemBytes(DevGenericSycl const& /* dev */) -> std::size_t @@ -189,7 +190,7 @@ namespace alpaka }; //! The SYCL device warp size get trait specialization. - template + template struct GetWarpSizes> { static auto getWarpSizes(DevGenericSycl const& dev) -> std::vector @@ -207,7 +208,7 @@ namespace alpaka }; //! The SYCL device preferred warp size get trait specialization. - template + template struct GetPreferredWarpSize> { static auto getPreferredWarpSize(DevGenericSycl const& dev) -> std::size_t @@ -217,7 +218,7 @@ namespace alpaka }; //! The SYCL device reset trait specialization. - template + template struct Reset> { static auto reset(DevGenericSycl const&) -> void @@ -229,7 +230,7 @@ namespace alpaka }; //! The SYCL device native handle trait specialization. - template + template struct NativeHandle> { [[nodiscard]] static auto getNativeHandle(DevGenericSycl const& dev) @@ -239,21 +240,21 @@ namespace alpaka }; //! The SYCL device memory buffer type trait specialization. - template + template struct BufType, TElem, TDim, TIdx> { using type = BufGenericSycl; }; //! The SYCL device platform type trait specialization. - template + template struct PlatformType> { using type = PlatformGenericSycl; }; //! The thread SYCL device wait specialization. - template + template struct CurrentThreadWaitFor> { static auto currentThreadWaitFor(DevGenericSycl const& dev) -> void @@ -263,14 +264,14 @@ namespace alpaka }; //! The SYCL blocking queue trait specialization. - template + template struct QueueType, Blocking> { using type = QueueGenericSyclBlocking; }; //! The SYCL non-blocking queue trait specialization. - template + template struct QueueType, NonBlocking> { using type = QueueGenericSyclNonBlocking; diff --git a/include/alpaka/event/EventGenericSycl.hpp b/include/alpaka/event/EventGenericSycl.hpp index 7ea8538400f5..8cdf79ddcf07 100644 --- a/include/alpaka/event/EventGenericSycl.hpp +++ b/include/alpaka/event/EventGenericSycl.hpp @@ -22,7 +22,7 @@ namespace alpaka { //! The SYCL device event. - template + template class EventGenericSycl final { public: @@ -60,7 +60,7 @@ namespace alpaka namespace alpaka::trait { //! The SYCL device event device get trait specialization. - template + template struct GetDev> { static auto getDev(EventGenericSycl const& event) -> DevGenericSycl @@ -70,7 +70,7 @@ namespace alpaka::trait }; //! The SYCL device event test trait specialization. - template + template struct IsComplete> { static auto isComplete(EventGenericSycl const& event) @@ -82,7 +82,7 @@ namespace alpaka::trait }; //! The SYCL queue enqueue trait specialization. - template + template struct Enqueue, EventGenericSycl> { static auto enqueue(QueueGenericSyclNonBlocking& queue, EventGenericSycl& event) @@ -92,7 +92,7 @@ namespace alpaka::trait }; //! The SYCL queue enqueue trait specialization. - template + template struct Enqueue, EventGenericSycl> { static auto enqueue(QueueGenericSyclBlocking& queue, EventGenericSycl& event) @@ -105,7 +105,7 @@ namespace alpaka::trait //! //! Waits until the event itself and therefore all tasks preceding it in the queue it is enqueued to have been //! completed. If the event is not enqueued to a queue the method returns immediately. - template + template struct CurrentThreadWaitFor> { static auto currentThreadWaitFor(EventGenericSycl const& event) @@ -115,7 +115,7 @@ namespace alpaka::trait }; //! The SYCL queue event wait trait specialization. - template + template struct WaiterWaitFor, EventGenericSycl> { static auto waiterWaitFor(QueueGenericSyclNonBlocking& queue, EventGenericSycl const& event) @@ -125,7 +125,7 @@ namespace alpaka::trait }; //! The SYCL queue event wait trait specialization. - template + template struct WaiterWaitFor, EventGenericSycl> { static auto waiterWaitFor(QueueGenericSyclBlocking& queue, EventGenericSycl const& event) @@ -138,7 +138,7 @@ namespace alpaka::trait //! //! Any future work submitted in any queue of this device will wait for event to complete before beginning //! execution. - template + template struct WaiterWaitFor, EventGenericSycl> { static auto waiterWaitFor(DevGenericSycl& dev, EventGenericSycl const& event) @@ -148,7 +148,7 @@ namespace alpaka::trait }; //! The SYCL device event native handle trait specialization. - template + template struct NativeHandle> { [[nodiscard]] static auto getNativeHandle(EventGenericSycl const& event) diff --git a/include/alpaka/kernel/TaskKernelGenericSycl.hpp b/include/alpaka/kernel/TaskKernelGenericSycl.hpp index 11cc2cae4590..119a3d5be901 100644 --- a/include/alpaka/kernel/TaskKernelGenericSycl.hpp +++ b/include/alpaka/kernel/TaskKernelGenericSycl.hpp @@ -71,7 +71,13 @@ namespace alpaka { //! The SYCL accelerator execution task. - template + template< + alpaka::concepts::Tag TTag, + typename TAcc, + typename TDim, + typename TIdx, + typename TKernelFnObj, + typename... TArgs> class TaskKernelGenericSycl final : public WorkDivMembers { public: @@ -286,7 +292,13 @@ namespace alpaka::trait //! \tparam TIdx The idx type of the accelerator device properties. //! \tparam TKernelFn Kernel function object type. //! \tparam TArgs Kernel function object argument types as a parameter pack. - template + template< + alpaka::concepts::Tag TTag, + typename TDev, + typename TDim, + typename TIdx, + typename TKernelFn, + typename... TArgs> struct FunctionAttributes, TDev, TKernelFn, TArgs...> { //! \param dev The device instance diff --git a/include/alpaka/mem/buf/BufGenericSycl.hpp b/include/alpaka/mem/buf/BufGenericSycl.hpp index 9beb16c77890..4ff0151c4624 100644 --- a/include/alpaka/mem/buf/BufGenericSycl.hpp +++ b/include/alpaka/mem/buf/BufGenericSycl.hpp @@ -24,7 +24,7 @@ namespace alpaka { //! The SYCL memory buffer. - template + template class BufGenericSycl : public internal::ViewAccessOps> { public: @@ -62,14 +62,14 @@ namespace alpaka namespace alpaka::trait { //! The BufGenericSycl device type trait specialization. - template + template struct DevType> { using type = DevGenericSycl; }; //! The BufGenericSycl device get trait specialization. - template + template struct GetDev> { static auto getDev(BufGenericSycl const& buf) @@ -79,21 +79,21 @@ namespace alpaka::trait }; //! The BufGenericSycl dimension getter trait specialization. - template + template struct DimType> { using type = TDim; }; //! The BufGenericSycl memory element type get trait specialization. - template + template struct ElemType> { using type = TElem; }; //! The BufGenericSycl extent get trait specialization. - template + template struct GetExtents> { auto operator()(BufGenericSycl const& buf) const @@ -103,7 +103,7 @@ namespace alpaka::trait }; //! The BufGenericSycl native pointer get trait specialization. - template + template struct GetPtrNative> { static auto getPtrNative(BufGenericSycl const& buf) -> TElem const* @@ -118,7 +118,7 @@ namespace alpaka::trait }; //! The BufGenericSycl pointer on device get trait specialization. - template + template struct GetPtrDev, DevGenericSycl> { static auto getPtrDev(BufGenericSycl const& buf, DevGenericSycl const& dev) @@ -148,7 +148,7 @@ namespace alpaka::trait }; //! The SYCL memory allocation trait specialization. - template + template struct BufAlloc> { template @@ -200,13 +200,13 @@ namespace alpaka::trait }; //! The BufGenericSycl stream-ordered memory allocation capability trait specialization. - template + template struct HasAsyncBufSupport> : std::false_type { }; //! The BufGenericSycl offset get trait specialization. - template + template struct GetOffsets> { auto operator()(BufGenericSycl const&) const -> Vec @@ -216,7 +216,7 @@ namespace alpaka::trait }; //! The pinned/mapped memory allocation trait specialization for the SYCL devices. - template + template struct BufAllocMapped, TElem, TDim, TIdx> { template @@ -238,20 +238,20 @@ namespace alpaka::trait }; //! The pinned/mapped memory allocation capability trait specialization. - template + template struct HasMappedBufSupport> : public std::true_type { }; //! The BufGenericSycl idx type trait specialization. - template + template struct IdxType> { using type = TIdx; }; //! The BufCpu pointer on SYCL device get trait specialization. - template + template struct GetPtrDev, DevGenericSycl> { static auto getPtrDev(BufCpu const& buf, DevGenericSycl const&) -> TElem const* diff --git a/include/alpaka/mem/buf/sycl/Copy.hpp b/include/alpaka/mem/buf/sycl/Copy.hpp index 44098f1ba38f..e929ef77782d 100644 --- a/include/alpaka/mem/buf/sycl/Copy.hpp +++ b/include/alpaka/mem/buf/sycl/Copy.hpp @@ -195,7 +195,7 @@ namespace alpaka::detail namespace alpaka::trait { //! The SYCL host-to-device memory copy trait specialization. - template + template struct CreateTaskMemcpy, DevCpu> { template @@ -209,7 +209,7 @@ namespace alpaka::trait }; //! The SYCL device-to-host memory copy trait specialization. - template + template struct CreateTaskMemcpy> { template @@ -223,7 +223,7 @@ namespace alpaka::trait }; //! The SYCL device-to-device memory copy trait specialization. - template + template struct CreateTaskMemcpy, DevGenericSycl> { template diff --git a/include/alpaka/mem/global/DeviceGlobalCpu.hpp b/include/alpaka/mem/global/DeviceGlobalCpu.hpp index aafcb06f5c00..a5c4f071e6d4 100644 --- a/include/alpaka/mem/global/DeviceGlobalCpu.hpp +++ b/include/alpaka/mem/global/DeviceGlobalCpu.hpp @@ -48,7 +48,7 @@ namespace alpaka } // namespace detail template< - typename TTag, + alpaka::concepts::Tag TTag, typename TViewSrc, typename TTypeDst, typename TQueue, @@ -73,7 +73,7 @@ namespace alpaka } template< - typename TTag, + alpaka::concepts::Tag TTag, typename TTypeSrc, typename TViewDstFwd, typename TQueue, @@ -98,7 +98,7 @@ namespace alpaka } template< - typename TTag, + alpaka::concepts::Tag TTag, typename TExtent, typename TViewSrc, typename TTypeDst, @@ -124,7 +124,7 @@ namespace alpaka } template< - typename TTag, + alpaka::concepts::Tag TTag, typename TExtent, typename TTypeSrc, typename TViewDstFwd, diff --git a/include/alpaka/mem/global/DeviceGlobalUniformCudaHipBuiltIn.hpp b/include/alpaka/mem/global/DeviceGlobalUniformCudaHipBuiltIn.hpp index 6b802fc9bd32..5f868815cbf6 100644 --- a/include/alpaka/mem/global/DeviceGlobalUniformCudaHipBuiltIn.hpp +++ b/include/alpaka/mem/global/DeviceGlobalUniformCudaHipBuiltIn.hpp @@ -41,7 +41,7 @@ namespace alpaka // from device to host template< - typename TTag, + alpaka::concepts::Tag TTag, typename TApi, bool TBlocking, typename TViewDst, @@ -77,7 +77,7 @@ namespace alpaka // from host to device template< - typename TTag, + alpaka::concepts::Tag TTag, typename TApi, bool TBlocking, typename TTypeDst, @@ -113,7 +113,7 @@ namespace alpaka // from device to host template< - typename TTag, + alpaka::concepts::Tag TTag, typename TApi, bool TBlocking, typename TViewDst, @@ -149,7 +149,7 @@ namespace alpaka // from host to device template< - typename TTag, + alpaka::concepts::Tag TTag, typename TApi, bool TBlocking, typename TTypeDst, diff --git a/include/alpaka/mem/global/Traits.hpp b/include/alpaka/mem/global/Traits.hpp index 7b3c3d1ccd3e..63914121f7c8 100644 --- a/include/alpaka/mem/global/Traits.hpp +++ b/include/alpaka/mem/global/Traits.hpp @@ -13,7 +13,7 @@ namespace alpaka namespace detail { - template + template struct DevGlobalImplGeneric { // does not make use of TTag @@ -31,7 +31,7 @@ namespace alpaka } }; - template + template struct DevGlobalTrait { static constexpr bool const IsImplementedFor = alpaka::meta::DependentFalseType::value; diff --git a/include/alpaka/mem/view/ViewPlainPtr.hpp b/include/alpaka/mem/view/ViewPlainPtr.hpp index dda4a179fdae..039d6762a63b 100644 --- a/include/alpaka/mem/view/ViewPlainPtr.hpp +++ b/include/alpaka/mem/view/ViewPlainPtr.hpp @@ -154,7 +154,7 @@ namespace alpaka #if defined(ALPAKA_ACC_SYCL_ENABLED) //! The SYCL device CreateViewPlainPtr trait specialization. - template + template struct CreateViewPlainPtr> { template diff --git a/include/alpaka/platform/PlatformGenericSycl.hpp b/include/alpaka/platform/PlatformGenericSycl.hpp index 34dc1894c35d..16268d8715fb 100644 --- a/include/alpaka/platform/PlatformGenericSycl.hpp +++ b/include/alpaka/platform/PlatformGenericSycl.hpp @@ -32,12 +32,12 @@ namespace alpaka { namespace detail { - template + template struct SYCLDeviceSelector; } // namespace detail //! The SYCL device manager. - template + template struct PlatformGenericSycl : interface::Implements> { PlatformGenericSycl() @@ -104,14 +104,14 @@ namespace alpaka namespace trait { //! The SYCL platform device type trait specialization. - template + template struct DevType> { using type = DevGenericSycl; }; //! The SYCL platform device count get trait specialization. - template + template struct GetDevCount> { static auto getDevCount(PlatformGenericSycl const& platform) -> std::size_t @@ -123,7 +123,7 @@ namespace alpaka }; //! The SYCL platform device get trait specialization. - template + template struct GetDevByIdx> { static auto getDevByIdx(PlatformGenericSycl const& platform, std::size_t const& devIdx) diff --git a/include/alpaka/queue/QueueGenericSyclBlocking.hpp b/include/alpaka/queue/QueueGenericSyclBlocking.hpp index 44dfb149c45c..d6e99d549fa7 100644 --- a/include/alpaka/queue/QueueGenericSyclBlocking.hpp +++ b/include/alpaka/queue/QueueGenericSyclBlocking.hpp @@ -10,7 +10,7 @@ namespace alpaka { - template + template using QueueGenericSyclBlocking = detail::QueueGenericSyclBase; } // namespace alpaka diff --git a/include/alpaka/queue/QueueGenericSyclNonBlocking.hpp b/include/alpaka/queue/QueueGenericSyclNonBlocking.hpp index 22615cae3c23..806c62c95c5e 100644 --- a/include/alpaka/queue/QueueGenericSyclNonBlocking.hpp +++ b/include/alpaka/queue/QueueGenericSyclNonBlocking.hpp @@ -10,7 +10,7 @@ namespace alpaka { - template + template using QueueGenericSyclNonBlocking = detail::QueueGenericSyclBase; } // namespace alpaka diff --git a/include/alpaka/queue/sycl/QueueGenericSyclBase.hpp b/include/alpaka/queue/sycl/QueueGenericSyclBase.hpp index 159b47e165a7..8ba6914db827 100644 --- a/include/alpaka/queue/sycl/QueueGenericSyclBase.hpp +++ b/include/alpaka/queue/sycl/QueueGenericSyclBase.hpp @@ -4,6 +4,7 @@ #pragma once +#include "alpaka/acc/Tag.hpp" #include "alpaka/dev/Traits.hpp" #include "alpaka/event/Traits.hpp" #include "alpaka/queue/Traits.hpp" @@ -25,10 +26,10 @@ namespace alpaka { - template + template class DevGenericSycl; - template + template class EventGenericSycl; namespace detail @@ -173,7 +174,7 @@ namespace alpaka sycl::queue m_queue; }; - template + template class QueueGenericSyclBase : public interface::Implements> , public interface::Implements> @@ -212,14 +213,14 @@ namespace alpaka namespace trait { //! The SYCL blocking queue device type trait specialization. - template + template struct DevType> { using type = DevGenericSycl; }; //! The SYCL blocking queue device get trait specialization. - template + template struct GetDev> { static auto getDev(alpaka::detail::QueueGenericSyclBase const& queue) @@ -230,14 +231,14 @@ namespace alpaka }; //! The SYCL blocking queue event type trait specialization. - template + template struct EventType> { using type = EventGenericSycl; }; //! The SYCL blocking queue enqueue trait specialization. - template + template struct Enqueue, TTask> { static auto enqueue(alpaka::detail::QueueGenericSyclBase& queue, TTask const& task) @@ -249,7 +250,7 @@ namespace alpaka }; //! The SYCL blocking queue test trait specialization. - template + template struct Empty> { static auto empty(alpaka::detail::QueueGenericSyclBase const& queue) -> bool @@ -263,7 +264,7 @@ namespace alpaka //! //! Blocks execution of the calling thread until the queue has finished processing all previously requested //! tasks (kernels, data copies, ...) - template + template struct CurrentThreadWaitFor> { static auto currentThreadWaitFor(alpaka::detail::QueueGenericSyclBase const& queue) @@ -275,7 +276,7 @@ namespace alpaka }; //! The SYCL queue native handle trait specialization. - template + template struct NativeHandle> { [[nodiscard]] static auto getNativeHandle( diff --git a/include/alpaka/test/event/EventHostManualTrigger.hpp b/include/alpaka/test/event/EventHostManualTrigger.hpp index 653dbbb641f3..4232a42cc8fc 100644 --- a/include/alpaka/test/event/EventHostManualTrigger.hpp +++ b/include/alpaka/test/event/EventHostManualTrigger.hpp @@ -712,7 +712,7 @@ namespace alpaka { namespace test { - template + template class EventHostManualTriggerSycl { public: @@ -727,13 +727,13 @@ namespace alpaka namespace trait { - template + template struct EventHostManualTriggerType> { using type = alpaka::test::EventHostManualTriggerSycl; }; - template + template struct IsEventHostManualTriggerSupported> { ALPAKA_FN_HOST static auto isSupported(DevGenericSycl const&) -> bool @@ -746,7 +746,7 @@ namespace alpaka namespace trait { - template + template struct Enqueue, test::EventHostManualTriggerSycl> { ALPAKA_FN_HOST static auto enqueue( @@ -756,7 +756,7 @@ namespace alpaka } }; - template + template struct Enqueue, test::EventHostManualTriggerSycl> { ALPAKA_FN_HOST static auto enqueue( @@ -766,7 +766,7 @@ namespace alpaka } }; - template + template struct IsComplete> { ALPAKA_FN_HOST static auto isComplete(test::EventHostManualTriggerSycl const& /* event */) -> bool diff --git a/include/alpaka/test/queue/Queue.hpp b/include/alpaka/test/queue/Queue.hpp index 0518e6d41e41..93c5c132d115 100644 --- a/include/alpaka/test/queue/Queue.hpp +++ b/include/alpaka/test/queue/Queue.hpp @@ -42,7 +42,7 @@ namespace alpaka::test #ifdef ALPAKA_ACC_SYCL_ENABLED //! The default queue type trait specialization for the SYCL device. - template + template struct DefaultQueueType> { # if(ALPAKA_DEBUG >= ALPAKA_DEBUG_FULL) @@ -89,13 +89,13 @@ namespace alpaka::test #endif #ifdef ALPAKA_ACC_SYCL_ENABLED - template + template struct IsBlockingQueue> { static constexpr auto value = true; }; - template + template struct IsBlockingQueue> { static constexpr auto value = false; diff --git a/script/gitlabci/job_clang.yml b/script/gitlabci/job_clang.yml index e58c22ecddfa..fce6301965e0 100644 --- a/script/gitlabci/job_clang.yml +++ b/script/gitlabci/job_clang.yml @@ -1,9 +1,9 @@ # SPDX-License-Identifier: MPL-2.0 -linux_clang-11_release: +linux_clang-13_release: extends: .base_clang variables: - ALPAKA_CI_CLANG_VER: 11 + ALPAKA_CI_CLANG_VER: 13 ALPAKA_CI_STDLIB: libstdc++ CMAKE_BUILD_TYPE: Release ALPAKA_BOOST_VERSION: 1.75.0 diff --git a/script/job_generator/versions.py b/script/job_generator/versions.py index 573767fffd4b..a4b67f9c325d 100644 --- a/script/job_generator/versions.py +++ b/script/job_generator/versions.py @@ -12,7 +12,7 @@ sw_versions: Dict[str, List[str]] = { GCC: ["10", "11", "12", "13"], - CLANG: ["10", "11", "12", "13", "14", "15", "16", "17", "18", "19"], + CLANG: ["13", "14", "15", "16", "17", "18", "19"], NVCC: [ "12.0", "12.1", diff --git a/test/common/devCompileOptions.cmake b/test/common/devCompileOptions.cmake index 957354d862a3..de5e54df7896 100644 --- a/test/common/devCompileOptions.cmake +++ b/test/common/devCompileOptions.cmake @@ -136,6 +136,8 @@ if(CMAKE_CXX_COMPILER_ID MATCHES "Clang" OR CMAKE_CXX_COMPILER_ID STREQUAL "Inte # We are not C++98 compatible list(APPEND alpaka_DEV_COMPILE_OPTIONS "-Wno-c++98-compat") list(APPEND alpaka_DEV_COMPILE_OPTIONS "-Wno-c++98-compat-pedantic") + # disable warnings, that C++20 features are not backwards compatible to C++17 and older + list(APPEND alpaka_DEV_COMPILE_OPTIONS "-Wno-c++20-compat") # Triggered by inline constants list(APPEND alpaka_DEV_COMPILE_OPTIONS "-Wno-global-constructors") # This padding warning is generated by the execution tasks depending on the argument types diff --git a/test/unit/acc/src/AccTagTest.cpp b/test/unit/acc/src/AccTagTest.cpp index ac798bc48f61..896aca65e724 100644 --- a/test/unit/acc/src/AccTagTest.cpp +++ b/test/unit/acc/src/AccTagTest.cpp @@ -14,7 +14,6 @@ #include - using Dim = alpaka::DimInt<1>; using Idx = int; using TestAccs = alpaka::test::EnabledAccs; @@ -275,3 +274,31 @@ TEST_CASE("test EnabledAccTags", "[acc][tag]") using AllAccs = alpaka::test::EnabledAccs, int>; STATIC_REQUIRE(std::tuple_size::value == std::tuple_size::value); } + +struct NoTag1 +{ + static std::string get_name() + { + return "foo"; + } +}; + +struct NoTag2 : public alpaka::InterfaceTag +{ +}; + +template +consteval bool specialize_tag() +{ + return true; +} + +TEMPLATE_LIST_TEST_CASE("test concept tag", "[tag][concept]", TagList) +{ + using TagToTest = TestType; + STATIC_REQUIRE(alpaka::concepts::Tag); + STATIC_REQUIRE_FALSE(alpaka::concepts::Tag); + STATIC_REQUIRE_FALSE(alpaka::concepts::Tag); + + STATIC_REQUIRE(specialize_tag()); +} diff --git a/test/unit/idx/src/MapIdxPitchBytes.cpp b/test/unit/idx/src/MapIdxPitchBytes.cpp index 6a341cac27cd..d05ef60544f1 100644 --- a/test/unit/idx/src/MapIdxPitchBytes.cpp +++ b/test/unit/idx/src/MapIdxPitchBytes.cpp @@ -13,7 +13,7 @@ #include #include -template +template auto mapIdxPitchBytes(TAccTag const&) { using Dim = TDim;