[SYCL] Re-allow auto& parameter in range parallel_for (#11315)

Before #10864, our header implementation allowed kernels with an auto&
parameter, e.g. `q.parallel_for(1, [=](auto &id) {})` , but strictly
speaking, this would not follow the criteria in [4.9.4.2.2. parallel_for
invoke](https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#_parallel_for_invoke).
After that PR, this behavior unintentionally became disallowed. This PR
restores the old behavior while disallowing it under
`-DSYCL2020_CONFORMANT_APIS`.

Author: jzc

sycl/include/sycl/handler.hpp

@@ -349,12 +349,12 @@ template <int Dims> class RoundedRangeIDGenerator {
   id<Dims> getId() { return Id; }
 
   template <typename KernelType> auto getItem() {
-    if constexpr (std::is_invocable_v<KernelType, item<Dims>> ||
-                  std::is_invocable_v<KernelType, item<Dims>, kernel_handler>)
+    if constexpr (std::is_invocable_v<KernelType, item<Dims> &> ||
+                  std::is_invocable_v<KernelType, item<Dims> &, kernel_handler>)
       return detail::Builder::createItem<Dims, true>(UserRange, getId(), {});
     else {
-      static_assert(std::is_invocable_v<KernelType, item<Dims, false>> ||
-                        std::is_invocable_v<KernelType, item<Dims, false>,
+      static_assert(std::is_invocable_v<KernelType, item<Dims, false> &> ||
+                        std::is_invocable_v<KernelType, item<Dims, false> &,
                                             kernel_handler>,
                     "Kernel must be invocable with an item!");
       return detail::Builder::createItem<Dims, false>(UserRange, getId());
@@ -376,8 +376,10 @@ class RoundedRangeKernel {
   void operator()(item<Dims> It) const {
     auto RoundedRange = It.get_range();
     for (RoundedRangeIDGenerator Gen(It.get_id(), UserRange, RoundedRange); Gen;
-         Gen.updateId())
-      KernelFunc(Gen.template getItem<KernelType>());
+         Gen.updateId()) {
+      auto item = Gen.template getItem<KernelType>();
+      KernelFunc(item);
+    }
   }
 };
 
@@ -389,8 +391,10 @@ class RoundedRangeKernelWithKH {
   void operator()(item<Dims> It, kernel_handler KH) const {
     auto RoundedRange = It.get_range();
     for (RoundedRangeIDGenerator Gen(It.get_id(), UserRange, RoundedRange); Gen;
-         Gen.updateId())
-      KernelFunc(Gen.template getItem<KernelType>(), KH);
+         Gen.updateId()) {
+      auto item = Gen.template getItem<KernelType>();
+      KernelFunc(item, KH);
+    }
   }
 };
 
@@ -1239,6 +1243,19 @@ class __SYCL_EXPORT handler {
                   "Kernel argument cannot have a sycl::nd_item type in "
                   "sycl::parallel_for with sycl::range");
 
+#ifdef SYCL2020_CONFORMANT_APIS
+    static_assert(
+        (std::is_invocable_v<KernelType, item<Dims>> ||
+         std::is_invocable_v<
+             KernelType, item<Dims>,
+             kernel_handler>)&&(std::is_invocable_v<KernelType,
+                                                    item<Dims, false>> ||
+                                std::is_invocable_v<KernelType,
+                                                    item<Dims, false>,
+                                                    kernel_handler>),
+        "Kernel should be invocable with a sycl::item");
+#endif
+
     // TODO: Properties may change the kernel function, so in order to avoid
     //       conflicts they should be included in the name.
     using NameT =

sycl/test/basic_tests/handler/parallel_for_arg_restrictions.cpp

@@ -71,7 +71,28 @@ int main() {
     CGH.parallel_for(sycl::nd_range{sycl::range{1, 1, 1}, sycl::range{1, 1, 1}},
                      [=](ConvertibleFromItem<3>) {});
   });
+  Q.submit([&](sycl::handler &CGH) {
+    // expected-error@sycl/handler.hpp:* {{Kernel should be invocable with a sycl::item}}
+    CGH.parallel_for(sycl::range{1}, [=](auto &) {});
+  });
+  Q.submit([&](sycl::handler &CGH) {
+    // expected-error@sycl/handler.hpp:* {{Kernel should be invocable with a sycl::item}}
+    CGH.parallel_for(sycl::range{1, 1}, [=](auto &) {});
+  });
+  Q.submit([&](sycl::handler &CGH) {
+    // expected-error@sycl/handler.hpp:* {{Kernel should be invocable with a sycl::item}}
+    CGH.parallel_for(sycl::range{1, 1, 1}, [=](auto &) {});
+  });
 #endif // SYCL2020_CONFORMANT_APIS
+  Q.submit([&](sycl::handler &CGH) {
+    CGH.parallel_for(sycl::range{1}, [=](auto &) {});
+  });
+  Q.submit([&](sycl::handler &CGH) {
+    CGH.parallel_for(sycl::range{1, 1}, [=](auto &) {});
+  });
+  Q.submit([&](sycl::handler &CGH) {
+    CGH.parallel_for(sycl::range{1, 1, 1}, [=](auto &) {});
+  });
 
   // Range parallel_for with nd_item.
   Q.submit([&](sycl::handler &CGH) {