[SYCL] Rename detail::dim_loop -> detail::loop and make it constexpr-friendly (#9108)

Can be used like this now:

```
  loop<2>([](auto i) {
    if constexpr (i == 0)
      std::cout << "constexpr 0\n";
    std::cout << i << std::endl;
  });
```

Author: aelovikov-intel

sycl/include/sycl/accessor.hpp

@@ -1084,7 +1084,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
   template <int Dims = Dimensions> size_t getLinearIndex(id<Dims> Id) const {
 
     size_t Result = 0;
-    detail::dim_loop<Dims>([&, this](size_t I) {
+    detail::loop<Dims>([&, this](size_t I) {
       Result = Result * getMemoryRange()[I] + Id[I];
       // We've already adjusted for the accessor's offset in the __init, so
       // don't include it here in case of device.
@@ -1147,7 +1147,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
   void __init(ConcreteASPtrType Ptr, range<AdjustedDim> AccessRange,
               range<AdjustedDim> MemRange, id<AdjustedDim> Offset) {
     MData = Ptr;
-    detail::dim_loop<AdjustedDim>([&, this](size_t I) {
+    detail::loop<AdjustedDim>([&, this](size_t I) {
       if constexpr (!(PropertyListT::template has_property<
                         sycl::ext::oneapi::property::no_offset>())) {
         getOffset()[I] = Offset[I];
@@ -2252,7 +2252,7 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
 #ifdef __SYCL_DEVICE_ONLY__
   size_t getTotalOffset() const noexcept {
     size_t TotalOffset = 0;
-    detail::dim_loop<Dimensions>([&, this](size_t I) {
+    detail::loop<Dimensions>([&, this](size_t I) {
       TotalOffset = TotalOffset * impl.MemRange[I];
       if constexpr (!(PropertyListT::template has_property<
                         sycl::ext::oneapi::property::no_offset>())) {
@@ -2515,7 +2515,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
   void __init(ConcreteASPtrType Ptr, range<AdjustedDim> AccessRange,
               range<AdjustedDim>, id<AdjustedDim>) {
     MData = Ptr;
-    detail::dim_loop<AdjustedDim>(
+    detail::loop<AdjustedDim>(
         [&, this](size_t I) { getSize()[I] = AccessRange[I]; });
   }
 
@@ -2570,7 +2570,7 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
   // Method which calculates linear offset for the ID using Range and Offset.
   template <int Dims = AdjustedDim> size_t getLinearIndex(id<Dims> Id) const {
     size_t Result = 0;
-    detail::dim_loop<Dims>(
+    detail::loop<Dims>(
         [&, this](size_t I) { Result = Result * getSize()[I] + Id[I]; });
     return Result;
   }

sycl/include/sycl/detail/helpers.hpp

@@ -241,12 +241,12 @@ getSPIRVMemorySemanticsMask(const access::fence_space AccessSpace,
 
 // To ensure loop unrolling is done when processing dimensions.
 template <size_t... Inds, class F>
-void dim_loop_impl(std::integer_sequence<size_t, Inds...>, F &&f) {
-  (f(Inds), ...);
+void loop_impl(std::integer_sequence<size_t, Inds...>, F &&f) {
+  (f(std::integral_constant<size_t, Inds>{}), ...);
 }
 
-template <size_t count, class F> void dim_loop(F &&f) {
-  dim_loop_impl(std::make_index_sequence<count>{}, std::forward<F>(f));
+template <size_t count, class F> void loop(F &&f) {
+  loop_impl(std::make_index_sequence<count>{}, std::forward<F>(f));
 }
 
 } // namespace detail

sycl/include/sycl/group_algorithm.hpp

@@ -253,7 +253,7 @@ reduce_over_group(Group g, sycl::vec<T, N> x, BinaryOperation binary_op) {
       "Result type of binary_op must match reduction accumulation type.");
   sycl::vec<T, N> result;
 
-  detail::dim_loop<N>(
+  detail::loop<N>(
       [&](size_t s) { result[s] = reduce_over_group(g, x[s], binary_op); });
   return result;
 }