LLNL · adrienbernede · Nov 26, 2024 · Nov 28, 2024 · Dec 2, 2024 · Dec 3, 2024
diff --git a/.gitlab-ci.yml b/.gitlab-ci.yml
@@ -72,7 +72,7 @@ stages:
     include:
       - local: '.gitlab/custom-jobs-and-variables.yml'
       - project: 'radiuss/radiuss-shared-ci'
-        ref: 'v2024.06.0'
+        ref: 'v2024.12.0'
         file: 'pipelines/${CI_MACHINE}.yml'
       - artifact: '${CI_MACHINE}-jobs.yml'
         job: 'generate-job-lists'
@@ -85,7 +85,7 @@ include:
     file: 'id_tokens.yml'
   # [Optional] checks preliminary to running the actual CI test
   #- project: 'radiuss/radiuss-shared-ci'
-  #  ref: 'v2024.06.0'
+  #  ref: 'v2024.12.0'
   #  file: 'utilities/preliminary-ignore-draft-pr.yml'
   # pipelines subscribed by the project
   - local: '.gitlab/subscribed-pipelines.yml'
diff --git a/.gitlab/custom-jobs-and-variables.yml b/.gitlab/custom-jobs-and-variables.yml
@@ -71,6 +71,9 @@ variables:
   artifacts:
     reports:
       junit: junit.xml
+    name: "${CI_PROJECT_NAME}-${CI_MACHINE}-${CI_JOB_NAME}-${CI_PIPELINE_ID}"
+    paths:
+      - ./*.cmake
 
 .reproducer_vars:
   script:

diff --git a/.gitlab/jobs/corona.yml b/.gitlab/jobs/corona.yml
@@ -27,8 +27,10 @@
 # ${PROJECT_<MACHINE>_DEPS} in the extra jobs. There is no reason not to fully
 # describe the spec here.
 
-clang_19_0_0_sycl_gcc_10_3_1_rocmcc_5_7_1_hip:
+clang_20_0_0_sycl_gcc_10_3_1_rocmcc_6_0_2_hip:
   variables:
-    SPEC: " ~shared +sycl ~openmp tests=basic %clang@=19.0.0 cxxflags==\"-w -fsycl -fsycl-unnamed-lambda -fsycl-targets=amdgcn-amd-amdhsa -Xsycl-target-backend --offload-arch=gfx906\" ^blt@develop"
-    MODULE_LIST: "rocm/5.7.1"
+    SPEC: " ~shared +sycl ~openmp tests=basic %clang@=20.0.0 cxxflags==\"-w -fsycl -fsycl-unnamed-lambda -fsycl-targets=amdgcn-amd-amdhsa -Xsycl-target-backend --offload-arch=gfx906\" ^blt@develop"
+    MODULE_LIST: "rocm/6.0.2"
+    SYCL_PATH: "/usr/WS2/raja-dev/clang_sycl_730cd3a5275f_hip_gcc10.3.1_rocm6.0.2/"
+    LD_LIBRARY_PATH: "${SYCL_PATH}/lib:${SYCL_PATH}/lib64:${LD_LIBRARY_PATH}"
   extends: .job_on_corona
diff --git a/.gitlab/jobs/poodle.yml b/.gitlab/jobs/poodle.yml
@@ -19,9 +19,9 @@
 # the comparison with the original job is easier.
 
 # custom variant
-intel_2023_2_1:
+oneapi_2023_2_1:
   variables:
-    SPEC: "${PROJECT_POODLE_VARIANTS} +lowopttest cxxflags==-fp-model=precise %intel@=2023.2.1 ${PROJECT_POODLE_DEPS}"
+    SPEC: "${PROJECT_POODLE_VARIANTS} +lowopttest cxxflags==-fp-model=precise %oneapi@=2023.2.1 ${PROJECT_POODLE_DEPS}"
   extends: .job_on_poodle
 
 # omptask variant
@@ -43,7 +43,7 @@ gcc_10_3_1:
 # ${PROJECT_<MACHINE>_DEPS} in the extra jobs. There is no reason not to fully
 # describe the spec here.
 
-intel_2023_2_1_mpi:
+oneapi_2023_2_1_mpi:
   variables:
-    SPEC: "~shared +openmp +mpi +lowopttest cxxflags==-fp-model=precise %intel@=2023.2.1 ^mvapich2 ^blt@develop"
+    SPEC: "~shared +openmp +mpi +lowopttest cxxflags==-fp-model=precise %oneapi@=2023.2.1 ^mvapich2 ^blt@develop"
   extends: .job_on_poodle
diff --git a/.gitlab/jobs/ruby.yml b/.gitlab/jobs/ruby.yml
@@ -19,9 +19,9 @@
 # the comparison with the original job is easier.
 
 # custom variant
-intel_2023_2_1:
+oneapi_2023_2_1:
   variables:
-    SPEC: "${PROJECT_RUBY_VARIANTS} +lowopttest cxxflags==-fp-model=precise %intel@=2023.2.1 ${PROJECT_RUBY_DEPS}"
+    SPEC: "${PROJECT_RUBY_VARIANTS} +lowopttest cxxflags==-fp-model=precise %oneapi@=2023.2.1 ${PROJECT_RUBY_DEPS}"
   extends: .job_on_ruby
 
 # omptask variant
@@ -44,7 +44,7 @@ gcc_10_3_1:
 # ${PROJECT_<MACHINE>_DEPS} in the extra jobs. There is no reason not to fully
 # describe the spec here.
 
-intel_2023_2_1_mpi:
+oneapi_2023_2_1_mpi:
   variables:
-    SPEC: "~shared +openmp +mpi +lowopttest cxxflags==-fp-model=precise %intel@=2023.2.1 ^mvapich2  ^blt@develop"
+    SPEC: "~shared +openmp +mpi +lowopttest cxxflags==-fp-model=precise %oneapi@=2023.2.1 ^mvapich2  ^blt@develop"
   extends: .job_on_ruby
diff --git a/.gitlab/jobs/tioga.yml b/.gitlab/jobs/tioga.yml
@@ -32,12 +32,12 @@ cce_17_0_1:
     SPEC: "~shared +openmp %cce@=17.0.1 ^blt@develop"
   extends: .job_on_tioga
 
-rocmcc_6_2_0_hip_openmp:
+rocmcc_6_3_0_hip_openmp:
   variables:
-    SPEC: "~shared +rocm +openmp amdgpu_target=gfx90a %rocmcc@=6.2.0 ^hip@6.2.0 ^blt@develop"
+    SPEC: "~shared +rocm +openmp amdgpu_target=gfx90a %rocmcc@=6.3.0 ^hip@6.3.0 ^blt@develop"
   extends: .job_on_tioga
 
-rocmcc_6_2_0_hip_openmp_mpi:
+rocmcc_6_3_0_hip_openmp_mpi:
   variables:
-    SPEC: "~shared +rocm +openmp +mpi amdgpu_target=gfx90a %rocmcc@=6.2.0 ^hip@6.2.0 ^blt@develop"
+    SPEC: "~shared +rocm +openmp +mpi amdgpu_target=gfx90a %rocmcc@=6.3.0 ^hip@6.3.0 ^blt@develop"
   extends: .job_on_tioga
diff --git a/.uberenv_config.json b/.uberenv_config.json
@@ -4,7 +4,7 @@
 "package_final_phase" : "initconfig",
 "package_source_dir" : "../..",
 "spack_url": "https://github.com/spack/spack.git",
-"spack_branch": "develop-2024-10-06",
+"spack_branch": "develop-2025-01-12",
 "spack_activate" : {},
 "spack_configs_path": "tpl/RAJA/scripts/radiuss-spack-configs",
 "spack_packages_path": "tpl/RAJA/scripts/radiuss-spack-configs/packages",

diff --git a/scripts/gitlab/build_and_test.sh b/scripts/gitlab/build_and_test.sh
@@ -241,10 +241,6 @@ then
     fi
 
     date
-    if [[ "${truehostname}" == "tioga" ]]
-    then
-        module unload rocm
-    fi
     $cmake_exe \
       -C ${hostconfig_path} \
       ${cmake_options} \

diff --git a/scripts/lc-builds/corona_sycl.sh b/scripts/lc-builds/corona_sycl.sh
@@ -1,8 +1,8 @@
 #!/usr/bin/env bash
 
 ###############################################################################
-# Copyright (c) 2016-24, Lawrence Livermore National Security, LLC
-# and RAJA project contributors. See the RAJA/LICENSE file for details.
+# Copyright (c) 2017-25, Lawrence Livermore National Security, LLC
+# and RAJA project contributors. See the RAJAPerf/LICENSE file for details.
 #
 # SPDX-License-Identifier: (BSD-3-Clause)
 ###############################################################################
@@ -13,7 +13,7 @@ if [[ $# -lt 1 ]]; then
   echo "   1) SYCL compiler installation path"
   echo
   echo "For example: "
-  echo "    corona_sycl.sh /usr/workspace/raja-dev/clang_sycl_2f03ef85fee5_hip_gcc10.3.1_rocm5.7.1"
+  echo "    corona_sycl.sh /usr/workspace/raja-dev/clang_sycl_730cd3a5275f_hip_gcc10.3.1_rocm6.0.2"
   exit
 fi
 

diff --git a/src/apps/CONVECTION3DPA-Sycl.cpp b/src/apps/CONVECTION3DPA-Sycl.cpp
@@ -40,7 +40,7 @@ void CONVECTION3DPA::runSyclVariantImpl(VariantID vid) {
     startTimer();
     for (RepIndex_type irep = 0; irep < run_reps; ++irep) {
 
-      qu->submit([&](cl::sycl::handler& h) {
+      qu->submit([&](::sycl::handler& h) {
 
         constexpr int max_D1D = CPA_D1D;
         constexpr int max_Q1D = CPA_Q1D;
@@ -54,8 +54,8 @@ void CONVECTION3DPA::runSyclVariantImpl(VariantID vid) {
         auto sm5_vec = ::sycl::local_accessor<double, 1>(::sycl::range<1>(max_DQ*max_DQ*max_DQ), h);
 
         h.parallel_for
-          (cl::sycl::nd_range<3>(gridSize, workGroupSize),
-           [=] (cl::sycl::nd_item<3> itm) {
+          (::sycl::nd_range<3>(gridSize, workGroupSize),
+           [=] (::sycl::nd_item<3> itm) {
 
              const Index_type e = itm.get_group(2);
 

diff --git a/src/apps/DIFFUSION3DPA-Sycl.cpp b/src/apps/DIFFUSION3DPA-Sycl.cpp
@@ -41,7 +41,7 @@ void DIFFUSION3DPA::runSyclVariantImpl(VariantID vid) {
     startTimer();
     for (RepIndex_type irep = 0; irep < run_reps; ++irep) {
 
-      qu->submit([&](cl::sycl::handler& h) {
+      qu->submit([&](::sycl::handler& h) {
 
         constexpr int MQ1 = DPA_Q1D;
         constexpr int MD1 = DPA_D1D;
@@ -57,8 +57,8 @@ void DIFFUSION3DPA::runSyclVariantImpl(VariantID vid) {
         auto sm1_2_vec = ::sycl::local_accessor<double, 1>(::sycl::range<1>(MDQ*MDQ*MDQ), h);
 
         h.parallel_for
-          (cl::sycl::nd_range<3>(gridSize, workGroupSize),
-           [=] (cl::sycl::nd_item<3> itm) {
+          (::sycl::nd_range<3>(gridSize, workGroupSize),
+           [=] (::sycl::nd_item<3> itm) {
 
              const Index_type e = itm.get_group(2);
 

diff --git a/src/apps/MASS3DEA-Sycl.cpp b/src/apps/MASS3DEA-Sycl.cpp
@@ -39,14 +39,14 @@ void MASS3DEA::runSyclVariantImpl(VariantID vid) {
     for (RepIndex_type irep = 0; irep < run_reps; ++irep) {
 
       constexpr size_t shmem = 0;
-      qu->submit([&](cl::sycl::handler& h) {
+      qu->submit([&](::sycl::handler& h) {
 
       ::sycl::local_accessor<double, 2> s_B(::sycl::range<2>(MEA_Q1D,MEA_D1D),h);
       ::sycl::local_accessor<double, 3> s_D(::sycl::range<3>(MEA_Q1D,MEA_Q1D,MEA_Q1D),h);
 
       h.parallel_for
-        (cl::sycl::nd_range<3>(gridSize, workGroupSize),
-         [=] (cl::sycl::nd_item<3> itm) {
+        (::sycl::nd_range<3>(gridSize, workGroupSize),
+         [=] (::sycl::nd_item<3> itm) {
 
            const Index_type e = itm.get_group(2);
 

diff --git a/src/apps/MASS3DPA-Sycl.cpp b/src/apps/MASS3DPA-Sycl.cpp
@@ -41,7 +41,7 @@ void MASS3DPA::runSyclVariantImpl(VariantID vid) {
     startTimer();
     for (RepIndex_type irep = 0; irep < run_reps; ++irep) {
 
-      qu->submit([&](cl::sycl::handler& h) {
+      qu->submit([&](::sycl::handler& h) {
 
         constexpr int MQ1 = MPA_Q1D;
         constexpr int MD1 = MPA_D1D;
@@ -52,8 +52,8 @@ void MASS3DPA::runSyclVariantImpl(VariantID vid) {
         auto sm1_vec = ::sycl::local_accessor<double, 1>(::sycl::range<1>(MDQ * MDQ * MDQ), h);
 
         h.parallel_for
-          (cl::sycl::nd_range<3>(gridSize, workGroupSize),
-           [=] (cl::sycl::nd_item<3> itm) {
+          (::sycl::nd_range<3>(gridSize, workGroupSize),
+           [=] (::sycl::nd_item<3> itm) {
 
              const Index_type e = itm.get_group(2);
 

diff --git a/src/basic/MAT_MAT_SHARED-Sycl.cpp b/src/basic/MAT_MAT_SHARED-Sycl.cpp
@@ -47,15 +47,15 @@ void MAT_MAT_SHARED::runSyclVariantImpl(VariantID vid)
     startTimer();
     for (RepIndex_type irep = 0; irep < run_reps; ++irep) {
 
-      qu->submit([&](cl::sycl::handler& h) {
+      qu->submit([&](::sycl::handler& h) {
 
        ::sycl::local_accessor<double, 2> As(::sycl::range<2>(tile_size, tile_size), h);
        ::sycl::local_accessor<double, 2> Bs(::sycl::range<2>(tile_size, tile_size), h);
        ::sycl::local_accessor<double, 2> Cs(::sycl::range<2>(tile_size, tile_size), h);
 
         h.parallel_for
-          (cl::sycl::nd_range<3>(gridSize, workGroupSize),
-           [=] (cl::sycl::nd_item<3> itm) {
+          (::sycl::nd_range<3>(gridSize, workGroupSize),
+           [=] (::sycl::nd_item<3> itm) {
 
              Index_type tx = itm.get_local_id(2);
              Index_type ty = itm.get_local_id(1);

diff --git a/src/basic/NESTED_INIT-Sycl.cpp b/src/basic/NESTED_INIT-Sycl.cpp
@@ -48,7 +48,7 @@ void NESTED_INIT::runSyclVariantImpl(VariantID vid)
     startTimer();
     for (RepIndex_type irep = 0; irep < run_reps; ++irep) {
 
-      qu->submit([&] (cl::sycl::handler& h) {
+      qu->submit([&] (::sycl::handler& h) {
         h.parallel_for(sycl::nd_range<3> ( global_dim, wkgroup_dim),
                        [=] (sycl::nd_item<3> item) { 
 

diff --git a/src/common/KernelBase.hpp b/src/common/KernelBase.hpp
@@ -26,7 +26,7 @@
 #include "RAJA/policy/hip/raja_hiperrchk.hpp"
 #endif
 #if defined(RAJA_ENABLE_SYCL)
-#include <sycl.hpp>
+#include "RAJA/util/sycl_compat.hpp"
 #endif
 
 #include "camp/resource.hpp"

diff --git a/src/common/SyclDataUtils.hpp b/src/common/SyclDataUtils.hpp
@@ -20,7 +20,7 @@
 
 #include "common/GPUUtils.hpp"
 
-#include <sycl.hpp>
+#include "RAJA/util/sycl_compat.hpp"
 
 
 namespace rajaperf

diff --git a/tpl/RAJA b/tpl/RAJA
+2 −2		.gitlab-ci.yml
+3 −0		.gitlab/custom-jobs-and-variables.yml
+5 −3		.gitlab/jobs/corona.yml
+3 −6		.gitlab/jobs/lassen.yml
+2 −2		.gitlab/jobs/poodle.yml
+2 −2		.gitlab/jobs/ruby.yml
+6 −6		.gitlab/jobs/tioga.yml
+1 −1		.uberenv_config.json
+1 −0		CMakeLists.txt
+1 −1		cmake/SetupRajaOptions.cmake
+15 −18		examples/dynamic-forall.cpp
+1 −1		examples/memoryManager.hpp
+1 −1		examples/resource-dynamic-forall.cpp
+1 −1		exercises/memoryManager.hpp
+2 −0		include/RAJA/config.hpp.in
+69 −74		include/RAJA/pattern/forall.hpp
+1 −1		include/RAJA/pattern/launch/launch_core.hpp
+3 −2		include/RAJA/policy/hip/policy.hpp
+2 −2		include/RAJA/policy/sycl/MemUtils_SYCL.hpp
+2 −2		include/RAJA/policy/sycl/forall.hpp
+1 −1		include/RAJA/policy/sycl/kernel/Conditional.hpp
+7 −7		include/RAJA/policy/sycl/kernel/For.hpp
+7 −7		include/RAJA/policy/sycl/kernel/ForICount.hpp
+1 −1		include/RAJA/policy/sycl/kernel/Lambda.hpp
+9 −9		include/RAJA/policy/sycl/kernel/SyclKernel.hpp
+5 −5		include/RAJA/policy/sycl/kernel/Tile.hpp
+5 −5		include/RAJA/policy/sycl/kernel/TileTCount.hpp
+9 −9		include/RAJA/policy/sycl/kernel/internal.hpp
+16 −16		include/RAJA/policy/sycl/launch.hpp
+1 −1		include/RAJA/policy/sycl/policy.hpp
+33 −33		include/RAJA/policy/sycl/reduce.hpp
+5 −4		include/RAJA/policy/tensor/arch/hip/hip_wave.hpp
+2 −1		include/RAJA/policy/tensor/arch/hip/traits.hpp
+0 −4		scripts/gitlab/build_and_test.sh
+1 −2		scripts/lc-builds/corona_sycl.sh
+1 −1		scripts/radiuss-spack-configs
+1 −1		scripts/uberenv
+2 −2		src/MemUtils_SYCL.cpp
+1 −1		test/functional/dynamic_forall/resource-segment/tests/test-dynamic-forall-resource-RangeSegment.hpp
+2 −2		test/functional/dynamic_forall/segment/tests/test-dynamic-forall-RangeSegment.hpp
+9 −0		test/functional/forall/CombiningAdapter/CMakeLists.txt
+2 −37		test/functional/forall/atomic-ref/CMakeLists.txt
+1 −35		test/functional/forall/atomic-view/CMakeLists.txt
+1 −6		test/functional/forall/indexset-view/CMakeLists.txt
+4 −32		test/functional/forall/multi-reduce-basic/CMakeLists.txt
+5 −44		test/functional/forall/reduce-basic/CMakeLists.txt
+4 −3		test/functional/forall/reduce-multiple-indexset/CMakeLists.txt
+5 −3		test/functional/forall/reduce-multiple-segment/CMakeLists.txt
+1 −6		test/functional/forall/resource-indexset/CMakeLists.txt
+9 −0		test/functional/kernel/basic-fission-fusion-loop/CMakeLists.txt
+9 −0		test/functional/kernel/conditional-fission-fusion-loop/CMakeLists.txt
+5 −7		test/functional/kernel/multi-reduce-nested/CMakeLists.txt
+4 −32		test/functional/kernel/nested-loop-reducesum/CMakeLists.txt
+1 −16		test/functional/kernel/nested-loop-segment-types/CMakeLists.txt
+0 −35		test/functional/kernel/nested-loop/CMakeLists.txt
+19 −12		test/functional/kernel/single-loop-tile-icount-tcount/CMakeLists.txt
+9 −6		test/functional/kernel/tile-variants/CMakeLists.txt
+4 −24		test/functional/workgroup/CMakeLists.txt
+3 −1		test/include/RAJA_test-tensor.hpp
+3 −1		test/integration/CMakeLists.txt
+3 −0		test/unit/algorithm/CMakeLists.txt
+1 −1		test/unit/indexing/CMakeLists.txt
+1 −1		test/unit/multi_reducer/CMakeLists.txt
+16 −0		test/unit/reducer/test-reducer-constructors-openmp-target.cpp
+1 −1		tpl/camp