Addressing code review.

Author: Alexandr-Konovalov

sycl/include/sycl/detail/cg_types.hpp

@@ -240,8 +240,8 @@ class HostKernelRefBase : public HostKernelBase {
 public:
   virtual std::unique_ptr<HostKernelBase> takeOrCopyOwnership() const = 0;
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-  // This function can't be called from old user code, because there is no
-  // HostKernelRef in old user code. So, make it empty.
+  // The kernels that are passed via HostKernelRefBase are instantiated along
+  // ctor call with GetInstantiateKernelOnHostPtr().
   void InstantiateKernelOnHost() override {}
 #endif
 };
@@ -288,11 +288,6 @@ class HostKernelRef<KernelType, KernelTypeUniversalRef &, KernelArgType, Dims>
   }
 
   ~HostKernelRef() noexcept override = default;
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-  // This function can't be called from old user code, because there is no
-  // HostKernelRef in old user code. So, make it empty.
-  void InstantiateKernelOnHost() override {}
-#endif
 };
 
 // This function is needed for host-side compilation to keep kernels

sycl/include/sycl/khr/free_function_commands.hpp

@@ -148,11 +148,14 @@ void launch_grouped(handler &h, range<3> r, range<3> size,
   h.parallel_for(nd_range<3>(r, size), k);
 }
 
+template <typename KernelType>
+constexpr bool enable_kernel_function_overload = !std::is_same_v<
+    typename std::decay_t<KernelType>, sycl::kernel>;
+
 template <
     typename KernelType,
-    // exclude sycl::kernel from overload resolution
-    typename = typename std::enable_if_t<!std::is_same_v<
-        typename std::decay_t<KernelType>, sycl::kernel>>>
+    typename = typename std::enable_if_t<enable_kernel_function_overload<
+        KernelType>>>
 void launch_grouped(const queue &q, range<1> r, range<1> size, KernelType &&k,
                     const sycl::detail::code_location &codeLoc =
                         sycl::detail::code_location::current()) {
@@ -168,8 +171,8 @@ void launch_grouped(const queue &q, range<1> r, range<1> size, KernelType &&k,
 }
 template <
     typename KernelType,
-    typename = typename std::enable_if_t<!std::is_same_v<
-        typename std::decay_t<KernelType>, sycl::kernel>>>
+    typename = typename std::enable_if_t<enable_kernel_function_overload<
+        KernelType>>>
 void launch_grouped(const queue &q, range<2> r, range<2> size, KernelType &&k,
                     const sycl::detail::code_location &codeLoc =
                         sycl::detail::code_location::current()) {
@@ -185,8 +188,8 @@ void launch_grouped(const queue &q, range<2> r, range<2> size, KernelType &&k,
 }
 template <
     typename KernelType,
-    typename = typename std::enable_if_t<!std::is_same_v<
-        typename std::decay_t<KernelType>, sycl::kernel>>>
+    typename = typename std::enable_if_t<enable_kernel_function_overload<
+        KernelType>>>
 void launch_grouped(const queue &q, range<3> r, range<3> size, KernelType &&k,
                     const sycl::detail::code_location &codeLoc =
                         sycl::detail::code_location::current()) {

sycl/unittests/Extensions/FreeFunctionCommands/FreeFunctionCommandsEvents.cpp

@@ -35,7 +35,7 @@ class TestMoveFunctor {
   void operator()(sycl::nd_item<3> Item) const {}
 };
 
-int TestMoveFunctor::MoveCtorCalls;
+int TestMoveFunctor::MoveCtorCalls = 0;
 
 namespace sycl {
 inline namespace _V1 {
@@ -44,12 +44,6 @@ template <>
 struct KernelInfo<TestFunctor> : public unittest::MockKernelInfoBase {
   static constexpr const char *getName() { return "TestFunctor"; }
   static constexpr int64_t getKernelSize() { return sizeof(TestFunctor); }
-  static constexpr const char *getFileName() { return "TestFunctor.hpp"; }
-  static constexpr const char *getFunctionName() {
-    return "TestFunctorFunctionName";
-  }
-  static constexpr unsigned getLineNumber() { return 13; }
-  static constexpr unsigned getColumnNumber() { return 8; }
 };
 
 template <>