[llvm] [libsycl] Implement nd_range kernel submissions (PR #206505)

[Sergey Semenov via llvm-commits]

================
@@ -332,6 +332,33 @@ class _LIBSYCL_EXPORT queue {
                                        std::forward<Rest>(rest)...);
   }
 
+  template <typename KernelName = detail::AutoName, int Dims, typename... Rest>
+  event parallel_for(nd_range<Dims> executionRange, Rest &&...rest) {
+    return parallel_for<KernelName, Dims, Rest...>(executionRange, {},
+                                                   std::forward<Rest>(rest)...);
+  }
+
+  template <typename KernelName = detail::AutoName, int Dims, typename... Rest>
+  event parallel_for(nd_range<Dims> executionRange, event depEvent,
+                     Rest &&...rest) {
+    return parallel_for<KernelName, Dims, Rest...>(executionRange, {depEvent},
+                                                   std::forward<Rest>(rest)...);
+  }
+
+  template <typename KernelName = detail::AutoName, int Dims, typename... Rest>
+  event parallel_for(nd_range<Dims> executionRange,
+                     const std::vector<event> &depEvents, Rest &&...rest) {
+    if (executionRange.get_global_range() != range<Dims>{} &&
+        (executionRange.get_local_range() == range<Dims>{} ||
+         executionRange.get_global_range() % executionRange.get_local_range() !=
----------------
sergey-semenov wrote:

Two things stand out to me here: it looks like the check won't catch the case where only some of the local range dimensions are set to 0, or the case where both ranges are set to 0.

https://github.com/llvm/llvm-project/pull/206505