From 1afd55bb24dfe96d33073f1f461ce18f249b04a0 Mon Sep 17 00:00:00 2001 From: Stewart Martin-Haugh Date: Thu, 6 Feb 2025 14:47:00 +0000 Subject: [PATCH] Make traccc::alpaka executables work in pure C++ (#832) * Make traccc::alpaka executables work in pure C++: move device code to utils. Built on work by Ryan Cross * Fix formatting * Fix choice of CPU backend, move to alpaka device_copy in track_params_estimation * Fix formatting * Remove variable only used for one file * Move to CDNA warp size for AMD * Fix ifdef logic for SYCL * Formatting * Move back to vecmem::copy * Move back to vecmem::copy * Move tag logic to device_tag.hpp, make get_device_info return a string * Remove position-independent executable logic: not needed now that this is pure C++ * Fix copy-paste error in alpaka cuda device_memory_resource --------- Co-authored-by: Stewart Martin-Haugh stewart.martin-haugh@stfc.ac.uk --- device/alpaka/CMakeLists.txt | 1 + .../seeding/track_params_estimation.hpp | 4 +- .../traccc/alpaka/utils/device_tag.hpp | 31 ++++++++++++ .../traccc/alpaka/utils/get_device_info.hpp | 19 +++++++ .../traccc/alpaka/utils/vecmem_types.hpp | 13 ++++- device/alpaka/src/seeding/seed_finding.cpp | 4 +- .../src/seeding/track_params_estimation.cpp | 2 +- device/alpaka/src/utils/get_device_info.cpp | 23 +++++++++ device/alpaka/src/utils/utils.hpp | 49 ++++++++++++------- examples/run/alpaka/CMakeLists.txt | 12 ----- .../run/alpaka/seeding_example_alpaka.cpp | 31 +++--------- examples/run/alpaka/seq_example_alpaka.cpp | 24 +++------ tests/alpaka/CMakeLists.txt | 15 ++---- tests/alpaka/alpaka_basic.cpp | 12 ++--- tests/alpaka/test_cca.cpp | 28 +++-------- 15 files changed, 152 insertions(+), 116 deletions(-) create mode 100644 device/alpaka/include/traccc/alpaka/utils/device_tag.hpp create mode 100644 device/alpaka/include/traccc/alpaka/utils/get_device_info.hpp create mode 100644 device/alpaka/src/utils/get_device_info.cpp diff --git a/device/alpaka/CMakeLists.txt b/device/alpaka/CMakeLists.txt index 5ed6f087e6..9e218b4139 100644 --- a/device/alpaka/CMakeLists.txt +++ b/device/alpaka/CMakeLists.txt @@ -27,6 +27,7 @@ traccc_add_alpaka_library( traccc_alpaka alpaka TYPE SHARED # Utility definitions. "include/traccc/alpaka/utils/make_prefix_sum_buff.hpp" "src/utils/make_prefix_sum_buff.cpp" + "src/utils/get_device_info.cpp" # Seed finding code. "include/traccc/alpaka/seeding/spacepoint_binning.hpp" "include/traccc/alpaka/seeding/seed_finding.hpp" diff --git a/device/alpaka/include/traccc/alpaka/seeding/track_params_estimation.hpp b/device/alpaka/include/traccc/alpaka/seeding/track_params_estimation.hpp index 5a04a0f976..866a4e3819 100644 --- a/device/alpaka/include/traccc/alpaka/seeding/track_params_estimation.hpp +++ b/device/alpaka/include/traccc/alpaka/seeding/track_params_estimation.hpp @@ -8,6 +8,8 @@ #pragma once // Project include(s) +#include + #include "traccc/edm/seed.hpp" #include "traccc/edm/spacepoint.hpp" #include "traccc/edm/track_parameters.hpp" @@ -58,7 +60,7 @@ struct track_params_estimation /// Memory resource used by the algorithm traccc::memory_resource m_mr; /// Copy object used by the algorithm - vecmem::copy& m_copy; + ::vecmem::copy& m_copy; }; } // namespace traccc::alpaka diff --git a/device/alpaka/include/traccc/alpaka/utils/device_tag.hpp b/device/alpaka/include/traccc/alpaka/utils/device_tag.hpp new file mode 100644 index 0000000000..2c3a21eb2b --- /dev/null +++ b/device/alpaka/include/traccc/alpaka/utils/device_tag.hpp @@ -0,0 +1,31 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2024 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +#include + +namespace traccc::alpaka { + +// Get alpaka tag for current device +#if defined(ALPAKA_ACC_GPU_CUDA_ENABLED) +using AccTag = ::alpaka::TagGpuCudaRt; +#elif defined(ALPAKA_ACC_GPU_HIP_ENABLED) +using AccTag = ::alpaka::TagGpuHipRt; +#elif defined(ALPAKA_ACC_SYCL_ENABLED) +#if defined(ALPAKA_SYCL_ONEAPI_CPU) +using AccTag = ::alpaka::TagCpuSycl; +#elif defined(ALPAKA_SYCL_ONEAPI_FPGA) +using AccTag = ::alpaka::TagFpgaSyclIntel; +#elif defined(ALPAKA_SYCL_ONEAPI_GPU) +using AccTag = ::alpaka::TagGpuSyclIntel; +#endif +#elif defined(ALPAKA_ACC_CPU_B_SEQ_T_THREADS_ENABLED) +using AccTag = ::alpaka::TagCpuThreads; +#endif + +} // namespace traccc::alpaka diff --git a/device/alpaka/include/traccc/alpaka/utils/get_device_info.hpp b/device/alpaka/include/traccc/alpaka/utils/get_device_info.hpp new file mode 100644 index 0000000000..24987a908d --- /dev/null +++ b/device/alpaka/include/traccc/alpaka/utils/get_device_info.hpp @@ -0,0 +1,19 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2024 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +#include + +namespace traccc::alpaka { + +/// Function that prints the current device information to the console. +/// Included as part of the traccc::alpaka namespace, to avoid having to include +/// alpaka headers in any users of the library. +std::string get_device_info(); + +} // namespace traccc::alpaka diff --git a/device/alpaka/include/traccc/alpaka/utils/vecmem_types.hpp b/device/alpaka/include/traccc/alpaka/utils/vecmem_types.hpp index 9b26f609f1..95d152a0ba 100644 --- a/device/alpaka/include/traccc/alpaka/utils/vecmem_types.hpp +++ b/device/alpaka/include/traccc/alpaka/utils/vecmem_types.hpp @@ -32,7 +32,7 @@ #include #endif -#include +#include "traccc/alpaka/utils/device_tag.hpp" // Forward declarations so we can compile the types below namespace vecmem { @@ -69,7 +69,7 @@ struct host_device_types { }; template <> struct host_device_types<::alpaka::TagGpuCudaRt> { - using device_memory_resource = ::vecmem::cuda::host_memory_resource; + using device_memory_resource = ::vecmem::cuda::device_memory_resource; using host_memory_resource = ::vecmem::cuda::host_memory_resource; using managed_memory_resource = ::vecmem::cuda::managed_memory_resource; using device_copy = ::vecmem::cuda::copy; @@ -102,4 +102,13 @@ struct host_device_types<::alpaka::TagGpuSyclIntel> { using managed_memory_resource = ::vecmem::sycl::shared_memory_resource; using device_copy = ::vecmem::sycl::copy; }; + +using device_memory_resource = + typename host_device_types::device_memory_resource; +using host_memory_resource = + typename host_device_types::host_memory_resource; +using managed_memory_resource = + typename host_device_types::managed_memory_resource; +using device_copy = typename host_device_types::device_copy; + } // namespace traccc::alpaka::vecmem diff --git a/device/alpaka/src/seeding/seed_finding.cpp b/device/alpaka/src/seeding/seed_finding.cpp index 903a987f51..f246624fee 100644 --- a/device/alpaka/src/seeding/seed_finding.cpp +++ b/device/alpaka/src/seeding/seed_finding.cpp @@ -343,7 +343,9 @@ seed_finding::output_type seed_finding::operator()( // Calculate the number of threads and thread blocks to run the weight // updating kernel for. - threadsPerBlock = warpSize * 2 < maxThreads ? warpSize * 2 : maxThreads; + threadsPerBlock = getWarpSize() * 2 < maxThreads + ? getWarpSize() * 2 + : maxThreads; blocksPerGrid = (pBufHost_counter->m_nTriplets + threadsPerBlock - 1) / threadsPerBlock; workDiv = makeWorkDiv(blocksPerGrid, threadsPerBlock); diff --git a/device/alpaka/src/seeding/track_params_estimation.cpp b/device/alpaka/src/seeding/track_params_estimation.cpp index d518b71ed8..a0079161dc 100644 --- a/device/alpaka/src/seeding/track_params_estimation.cpp +++ b/device/alpaka/src/seeding/track_params_estimation.cpp @@ -63,7 +63,7 @@ track_params_estimation::output_type track_params_estimation::operator()( // Run the kernel ::alpaka::exec(queue, workDiv, EstimateTrackParamsKernel{}, spacepoints_view, seeds_view, bfield, stddev, - vecmem::get_data(params_buffer)); + ::vecmem::get_data(params_buffer)); ::alpaka::wait(queue); return params_buffer; diff --git a/device/alpaka/src/utils/get_device_info.cpp b/device/alpaka/src/utils/get_device_info.cpp new file mode 100644 index 0000000000..b0ec52af11 --- /dev/null +++ b/device/alpaka/src/utils/get_device_info.cpp @@ -0,0 +1,23 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2025 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +// Local include(s). +#include "utils.hpp" + +// Project include(s). +#include "traccc/alpaka/utils/get_device_info.hpp" + +namespace traccc::alpaka { + +std::string get_device_info() { + int device = 0; + auto devAcc = ::alpaka::getDevByIdx(::alpaka::Platform{}, 0u); + return std::string("Using Alpaka device: " + ::alpaka::getName(devAcc) + + " [id: " + std::to_string(device) + "] "); +} + +} // namespace traccc::alpaka diff --git a/device/alpaka/src/utils/utils.hpp b/device/alpaka/src/utils/utils.hpp index 0ed40d6e2c..19330a7c31 100644 --- a/device/alpaka/src/utils/utils.hpp +++ b/device/alpaka/src/utils/utils.hpp @@ -8,17 +8,6 @@ #pragma once #include -#include - -#ifdef ALPAKA_ACC_GPU_CUDA_ENABLED -#include -#endif - -#ifdef ALPAKA_ACC_GPU_HIP_ENABLED -#include -#endif - -#include namespace traccc::alpaka { @@ -26,16 +15,40 @@ using Dim = ::alpaka::DimInt<1>; using Idx = uint32_t; using WorkDiv = ::alpaka::WorkDivMembers; -using Acc = ::alpaka::ExampleDefaultAcc; +// Get alpaka accelerator - based on alpaka/examples/ExampleDefaultAcc.hpp +#if defined(ALPAKA_ACC_GPU_CUDA_ENABLED) +using Acc = ::alpaka::AccGpuCudaRt; +#elif defined(ALPAKA_ACC_GPU_HIP_ENABLED) +using Acc = ::alpaka::AccGpuHipRt; +#elif defined(ALPAKA_ACC_SYCL_ENABLED) +#if defined(ALPAKA_SYCL_ONEAPI_CPU) +using Acc = ::alpaka::AccCpuSycl; +#elif defined(ALPAKA_SYCL_ONEAPI_FPGA) +using Acc = ::alpaka::AccFpgaSyclIntel; +#elif defined(ALPAKA_SYCL_ONEAPI_GPU) +using Acc = ::alpaka::AccGpuSyclIntel; +#endif +#elif defined(ALPAKA_ACC_CPU_B_SEQ_T_THREADS_ENABLED) +using Acc = ::alpaka::AccCpuThreads; +#else +#error "No supported backend selected." //we definitely want to fail the build if no matching accelerator is found +#endif + using Host = ::alpaka::DevCpu; using Queue = ::alpaka::Queue; -static constexpr std::size_t warpSize = -#if defined(ALPAKA_ACC_GPU_CUDA_ENABLED) || defined(ALPAKA_ACC_GPU_HIP_ENABLED) - 32; -#else - 4; -#endif +template +consteval std::size_t getWarpSize() { + if constexpr (::alpaka::accMatchesTags) { + return 32; + } + if constexpr (::alpaka::accMatchesTags) { + return 64; + } else { + return 4; + } +} template inline WorkDiv makeWorkDiv(Idx blocks, Idx threadsOrElements) { diff --git a/examples/run/alpaka/CMakeLists.txt b/examples/run/alpaka/CMakeLists.txt index 5696f0259b..23a3e02152 100644 --- a/examples/run/alpaka/CMakeLists.txt +++ b/examples/run/alpaka/CMakeLists.txt @@ -15,16 +15,12 @@ include(traccc-alpaka-functions) traccc_enable_language_alpaka() if(alpaka_ACC_GPU_CUDA_ENABLE) - set_source_files_properties(${TRACCC_ALPAKA_EXAMPLE_SOURCES} PROPERTIES LANGUAGE CUDA) - list (APPEND EXTRA_LIBS vecmem::cuda) elseif(alpaka_ACC_GPU_HIP_ENABLE) find_package( HIPToolkit REQUIRED ) - set_source_files_properties(${TRACCC_ALPAKA_EXAMPLE_SOURCES} PROPERTIES LANGUAGE HIP) list(APPEND EXTRA_LIBS HIP::hiprt vecmem::hip) elseif(alpaka_ACC_SYCL_ENABLE) list(APPEND EXTRA_LIBS vecmem::sycl) - set_source_files_properties(${TRACCC_ALPAKA_EXAMPLE_SOURCES} PROPERTIES LANGUAGE SYCL) endif() set(LIBRARIES vecmem::core traccc::io traccc::performance @@ -35,11 +31,3 @@ traccc_add_executable( seq_example_alpaka "seq_example_alpaka.cpp" LINK_LIBRARIES ${LIBRARIES} ) traccc_add_executable( seeding_example_alpaka "seeding_example_alpaka.cpp" LINK_LIBRARIES ${LIBRARIES} ) - -#Can only do this once target is defined, so need another if here -if(alpaka_ACC_GPU_HIP_ENABLE) - set_target_properties( traccc_seq_example_alpaka PROPERTIES - POSITION_INDEPENDENT_CODE TRUE ) - set_target_properties( traccc_seeding_example_alpaka PROPERTIES - POSITION_INDEPENDENT_CODE TRUE ) -endif() diff --git a/examples/run/alpaka/seeding_example_alpaka.cpp b/examples/run/alpaka/seeding_example_alpaka.cpp index 5574b8adb2..2514a35558 100644 --- a/examples/run/alpaka/seeding_example_alpaka.cpp +++ b/examples/run/alpaka/seeding_example_alpaka.cpp @@ -35,7 +35,6 @@ #include "traccc/seeding/track_params_estimation.hpp" // Detray include(s). -#include "alpaka/example/ExampleDefaultAcc.hpp" #include "detray/core/detector.hpp" #include "detray/detectors/bfield.hpp" #include "detray/io/frontend/detector_reader.hpp" @@ -63,33 +62,19 @@ int seq_run(const traccc::opts::track_seeding& seeding_opts, const traccc::opts::performance& performance_opts, const traccc::opts::accelerator& accelerator_opts) { - using Dim = ::alpaka::DimInt<1>; - using Idx = uint32_t; - - using Acc = ::alpaka::ExampleDefaultAcc; #ifdef ALPAKA_ACC_SYCL_ENABLED ::sycl::queue q; vecmem::sycl::queue_wrapper qw{&q}; - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::device_copy copy(qw); - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::host_memory_resource host_mr(qw); - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::device_memory_resource - device_mr(qw); - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::managed_memory_resource - mng_mr(qw); + traccc::alpaka::vecmem::device_copy copy(qw); + traccc::alpaka::vecmem::host_memory_resource host_mr(qw); + traccc::alpaka::vecmem::device_memory_resource device_mr(qw); + traccc::alpaka::vecmem::managed_memory_resource mng_mr(qw); traccc::memory_resource mr{device_mr, &host_mr}; #else - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::device_copy copy; - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::host_memory_resource host_mr; - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::device_memory_resource device_mr; - traccc::alpaka::vecmem::host_device_types< - ::alpaka::trait::AccToTag::type>::managed_memory_resource mng_mr; + traccc::alpaka::vecmem::device_copy copy; + traccc::alpaka::vecmem::host_memory_resource host_mr; + traccc::alpaka::vecmem::device_memory_resource device_mr; + traccc::alpaka::vecmem::managed_memory_resource mng_mr; traccc::memory_resource mr{device_mr, &host_mr}; #endif diff --git a/examples/run/alpaka/seq_example_alpaka.cpp b/examples/run/alpaka/seq_example_alpaka.cpp index 9c0a8291d2..ab1f2973a7 100644 --- a/examples/run/alpaka/seq_example_alpaka.cpp +++ b/examples/run/alpaka/seq_example_alpaka.cpp @@ -6,7 +6,6 @@ */ // Project include(s). -#include "alpaka/example/ExampleDefaultAcc.hpp" #include "traccc/alpaka/clusterization/clusterization_algorithm.hpp" #include "traccc/alpaka/clusterization/measurement_sorting_algorithm.hpp" #include "traccc/alpaka/seeding/seeding_algorithm.hpp" @@ -66,28 +65,17 @@ int seq_run(const traccc::opts::detector& detector_opts, const traccc::vector3 field_vec = {0.f, 0.f, seeding_opts.seedfinder.bFieldInZ}; - using Dim = ::alpaka::DimInt<1>; - using Idx = uint32_t; - - using Acc = ::alpaka::ExampleDefaultAcc; // Memory resources used by the application. #ifdef ALPAKA_ACC_SYCL_ENABLED ::sycl::queue q; vecmem::sycl::queue_wrapper qw{&q}; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_copy copy(qw); - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::host_memory_resource host_mr(qw); - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_memory_resource - device_mr(qw); + traccc::alpaka::vecmem::device_copy copy(qw); + traccc::alpaka::vecmem::host_memory_resource host_mr(qw); + traccc::alpaka::vecmem::device_memory_resource device_mr(qw); #else - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_copy copy; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::host_memory_resource host_mr; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_memory_resource device_mr; + traccc::alpaka::vecmem::device_copy copy; + traccc::alpaka::vecmem::host_memory_resource host_mr; + traccc::alpaka::vecmem::device_memory_resource device_mr; #endif traccc::memory_resource mr{device_mr, &host_mr}; diff --git a/tests/alpaka/CMakeLists.txt b/tests/alpaka/CMakeLists.txt index b319741e32..be4e4f3e97 100644 --- a/tests/alpaka/CMakeLists.txt +++ b/tests/alpaka/CMakeLists.txt @@ -4,28 +4,23 @@ # # Mozilla Public License Version 2.0 -set(TRACCC_ALPAKA_TEST_SOURCES - alpaka_basic.cpp - test_cca.cpp -) - - include(traccc-alpaka-functions) traccc_enable_language_alpaka() if(alpaka_ACC_GPU_CUDA_ENABLE) - set_source_files_properties(${TRACCC_ALPAKA_TEST_SOURCES} PROPERTIES LANGUAGE CUDA) + set_source_files_properties(alpaka_basic.cpp PROPERTIES LANGUAGE CUDA) list(APPEND DEVICE_LIBRARIES vecmem::cuda) elseif(alpaka_ACC_GPU_HIP_ENABLE) - set_source_files_properties(${TRACCC_ALPAKA_TEST_SOURCES} PROPERTIES LANGUAGE HIP) + set_source_files_properties(alpaka_basic.cpp PROPERTIES LANGUAGE HIP) list(APPEND DEVICE_LIBRARIES vecmem::hip) elseif(alpaka_ACC_SYCL_ENABLE) list(APPEND DEVICE_LIBRARIES vecmem::sycl) - set_source_files_properties(${TRACCC_ALPAKA_TEST_SOURCES} PROPERTIES LANGUAGE SYCL) + set_source_files_properties(alpaka_basic.cpp PROPERTIES LANGUAGE SYCL) endif() traccc_add_test( alpaka - ${TRACCC_ALPAKA_TEST_SOURCES} + alpaka_basic.cpp + test_cca.cpp LINK_LIBRARIES GTest::gtest_main traccc_tests_common diff --git a/tests/alpaka/alpaka_basic.cpp b/tests/alpaka/alpaka_basic.cpp index ad96423094..d3f5cd6eae 100644 --- a/tests/alpaka/alpaka_basic.cpp +++ b/tests/alpaka/alpaka_basic.cpp @@ -139,17 +139,13 @@ GTEST_TEST(AlpakaBasic, VecMemOp) { #ifdef ALPAKA_ACC_SYCL_ENABLED ::sycl::queue q; vecmem::sycl::queue_wrapper qw{&q}; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_copy vm_copy(qw); + traccc::alpaka::vecmem::device_copy vm_copy(qw); #else - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_copy vm_copy; + traccc::alpaka::vecmem::device_copy vm_copy; #endif - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::host_memory_resource host_mr; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_memory_resource device_mr; + traccc::alpaka::vecmem::host_memory_resource host_mr; + traccc::alpaka::vecmem::device_memory_resource device_mr; vecmem::vector host_vector{n, &host_mr}; diff --git a/tests/alpaka/test_cca.cpp b/tests/alpaka/test_cca.cpp index bab2952289..9c19b61233 100644 --- a/tests/alpaka/test_cca.cpp +++ b/tests/alpaka/test_cca.cpp @@ -7,8 +7,6 @@ #include -#include -#include #include #include @@ -31,30 +29,16 @@ cca_function_t get_f_with(traccc::clustering_config cfg) { std::map> result; - using namespace alpaka; - using Dim = DimInt<1>; - using Idx = uint32_t; - - using Acc = ExampleDefaultAcc; #ifdef ALPAKA_ACC_SYCL_ENABLED ::sycl::queue q; vecmem::sycl::queue_wrapper qw{&q}; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::host_memory_resource - host_mr(qw); - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_copy copy(qw); - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_memory_resource - device_mr; + traccc::alpaka::vecmem::host_memory_resource host_mr(qw); + traccc::alpaka::vecmem::device_copy copy(qw); + traccc::alpaka::vecmem::device_memory_resource device_mr; #else - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::host_memory_resource host_mr; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_copy copy; - traccc::alpaka::vecmem::host_device_types< - alpaka::trait::AccToTag::type>::device_memory_resource - device_mr; + traccc::alpaka::vecmem::host_memory_resource host_mr; + traccc::alpaka::vecmem::device_copy copy; + traccc::alpaka::vecmem::device_memory_resource device_mr; #endif traccc::alpaka::clusterization_algorithm cc({device_mr}, copy, cfg);