Remove sycl::nd_item<1> from TagParallelFor interface (#4646)

This PR removes the `sycl::nd_item<1>` that `detail::ParallelFor_doit`
passes into the function it was called with. As far as I can tell #4515
removed the last usage of the nd item.

Author: AlexanderSinn

Src/Base/AMReX_FBI.H

@@ -57,9 +57,6 @@ fab_to_fab (Vector<Array4CopyTag<T0, T1> > const& copy_tags, int scomp, int dcom
 
     detail::ParallelFor_doit(tv,
         [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-            sycl::nd_item<1> const& /*item*/,
-#endif
             int icell, int ncells, int i, int j, int k, Array4CopyTag<T0, T1> const& tag) noexcept
     {
         if (icell < ncells) {
@@ -331,9 +328,6 @@ FabArray<FAB>::FB_local_copy_gpu (const FB& TheFB, int scomp, int ncomp)
 
         detail::ParallelFor_doit(*tv,
             [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-                sycl::nd_item<1> const& /*item*/,
-#endif
                 int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
         {
             if (icell < ncells) {
@@ -868,9 +862,6 @@ FabArray<FAB>::pack_send_buffer_gpu (FabArray<FAB> const& src, int scomp, int nc
 
     detail::ParallelFor_doit(*tv,
         [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-            sycl::nd_item<1> const& /*item*/,
-#endif
             int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
     {
         if (icell < ncells) {
@@ -982,9 +973,6 @@ FabArray<FAB>::unpack_recv_buffer_gpu (FabArray<FAB>& dst, int dcomp, int ncomp,
         {
             detail::ParallelFor_doit(*tv,
                 [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-                    sycl::nd_item<1> const& /*item*/,
-#endif
                     int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
             {
                 if (icell < ncells) {
@@ -1001,9 +989,6 @@ FabArray<FAB>::unpack_recv_buffer_gpu (FabArray<FAB>& dst, int dcomp, int ncomp,
             if (is_thread_safe) {
                 detail::ParallelFor_doit(*tv,
                     [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-                        sycl::nd_item<1> const& /*item*/,
-#endif
                         int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
                 {
                     if (icell < ncells) {
@@ -1018,9 +1003,6 @@ FabArray<FAB>::unpack_recv_buffer_gpu (FabArray<FAB>& dst, int dcomp, int ncomp,
                 if constexpr (amrex::HasAtomicAdd<value_type>::value) {
                     detail::ParallelFor_doit(*tv,
                         [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-                            sycl::nd_item<1> const& /*item*/,
-#endif
                             int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
                     {
                         if (icell < ncells) {

Src/Base/AMReX_TagParallelFor.H

@@ -308,11 +308,7 @@ namespace detail {
 template <typename T, typename F>
 AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
 std::enable_if_t<std::is_same_v<std::decay_t<decltype(std::declval<T>().box())>, Box>, void>
-tagparfor_call_f (
-#ifdef AMREX_USE_SYCL
-    sycl::nd_item<1> const& item,
-#endif
-    int icell, T const& tag, F&& f) noexcept
+tagparfor_call_f (int icell, T const& tag, F&& f) noexcept
 {
     int ncells = tag.box().numPts();
     const auto len = amrex::length(tag.box());
@@ -323,28 +319,16 @@ tagparfor_call_f (
     i += lo.x;
     j += lo.y;
     k += lo.z;
-#ifdef AMREX_USE_SYCL
-    f(item, icell, ncells, i, j, k, tag);
-#else
-    f(      icell, ncells, i, j, k, tag);
-#endif
+    f(icell, ncells, i, j, k, tag);
 }
 
 template <typename T, typename F>
 AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
 std::enable_if_t<std::is_integral_v<std::decay_t<decltype(std::declval<T>().size())> >, void>
-tagparfor_call_f (
-#ifdef AMREX_USE_SYCL
-    sycl::nd_item<1> const& item,
-#endif
-    int i, T const& tag, F&& f) noexcept
+tagparfor_call_f (int i, T const& tag, F&& f) noexcept
 {
     int N = tag.size();
-#ifdef AMREX_USE_SYCL
-    f(item, i, N, tag);
-#else
-    f(      i, N, tag);
-#endif
+    f(i, N, tag);
 }
 
 template <class TagType, class F>
@@ -388,11 +372,7 @@ ParallelFor_doit (TagVector<TagType> const& tv, F const& f)
 #endif
         int icell = b_wid*Gpu::Device::warp_size + lane;
 
-#ifdef AMREX_USE_SYCL
-        tagparfor_call_f(item, icell, d_tags[tag_id], f);
-#else
-        tagparfor_call_f(      icell, d_tags[tag_id], f);
-#endif
+        tagparfor_call_f(icell, d_tags[tag_id], f);
     });
 }
 
@@ -454,9 +434,6 @@ ParallelFor (TagVector<TagType> const& tv, int ncomp, F const& f)
 {
     detail::ParallelFor_doit(tv,
         [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-            sycl::nd_item<1> const& /*item*/,
-#endif
             int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
         {
             if (icell < ncells) {
@@ -473,9 +450,6 @@ ParallelFor (TagVector<TagType> const& tv, F const& f)
 {
     detail::ParallelFor_doit(tv,
         [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-            sycl::nd_item<1> const& /*item*/,
-#endif
             int icell, int ncells, int i, int j, int k, TagType const& tag) noexcept
         {
             if (icell < ncells) {
@@ -490,9 +464,6 @@ ParallelFor (TagVector<TagType> const& tv, F const& f)
 {
     detail::ParallelFor_doit(tv,
         [=] AMREX_GPU_DEVICE (
-#ifdef AMREX_USE_SYCL
-            sycl::nd_item<1> const& /*item*/,
-#endif
             int icell, int ncells, TagType const& tag) noexcept
         {
             if (icell < ncells) {