Changes to work with latest Intel/SYCL compiler on corona

LLNL · Jan 23, 2025 · 7612584 · 7612584
1 parent 3ff3770
commit 7612584
Show file tree

Hide file tree

Showing 9 changed files with 21 additions and 21 deletions.
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