[SYCL] Remove deprecated kernel launch queries and unused ABI entry points

Also restore some of the kernel launch queries which were incorrectly
removed under preview breaking macro.

Author: againull

sycl/include/sycl/kernel.hpp

@@ -245,11 +245,6 @@ class __SYCL_EXPORT kernel : public detail::OwnerLessBase<kernel> {
 
   ur_native_handle_t getNative() const;
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-  __SYCL_DEPRECATED("Use getNative() member function")
-  ur_native_handle_t getNativeImpl() const;
-#endif
-
   std::shared_ptr<detail::kernel_impl> impl;
 
   template <backend BackendName, class SyclObjectT>

sycl/source/detail/kernel_impl.hpp

@@ -115,17 +115,13 @@ class kernel_impl {
   typename Param::return_type get_info(const device &Device,
                                        const range<3> &WGSize) const;
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-  // This function is unused and should be removed in the next ABI breaking.
-
   /// Query queue/launch-specific information from a kernel using the
   /// info::kernel_queue_specific descriptor for a specific Queue.
   ///
   /// \param Queue is a valid SYCL queue.
   /// \return depends on information being queried.
   template <typename Param>
   typename Param::return_type ext_oneapi_get_info(queue Queue) const;
-#endif // __INTEL_PREVIEW_BREAKING_CHANGES
 
   /// Query queue/launch-specific information from a kernel using the
   /// info::kernel_queue_specific descriptor for a specific Queue and values.
@@ -457,9 +453,6 @@ inline typename ext::intel::info::kernel_device_specific::spill_memory_size::
       getAdapter());
 }
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-// These functions are unused and should be removed in the next ABI breaking.
-
 template <>
 inline typename syclex::info::kernel_queue_specific::max_work_group_size::
     return_type
@@ -511,8 +504,6 @@ ADD_TEMPLATE_METHOD_SPEC(3)
 
 #undef ADD_TEMPLATE_METHOD_SPEC
 
-#endif // __INTEL_PREVIEW_BREAKING_CHANGES
-
 #define ADD_TEMPLATE_METHOD_SPEC(QueueSpec, Num, Kind, Reg)                    \
   template <>                                                                  \
   inline typename syclex::info::kernel_queue_specific::QueueSpec::return_type  \

sycl/source/kernel.cpp

@@ -106,15 +106,13 @@ template __SYCL_EXPORT uint32_t
 kernel::get_info<info::kernel_device_specific::max_sub_group_size>(
     const device &, const sycl::range<3> &) const;
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
 // This function is unused and should be removed in the next ABI-breaking
 // window.
 template <typename Param>
 typename detail::is_kernel_queue_specific_info_desc<Param>::return_type
 kernel::ext_oneapi_get_info(queue Queue) const {
   return impl->ext_oneapi_get_info<Param>(std::move(Queue));
 }
-#endif // __INTEL_PREVIEW_BREAKING_CHANGES
 
 template <typename Param>
 typename detail::is_kernel_queue_specific_info_desc<Param>::return_type
@@ -158,10 +156,6 @@ kernel::ext_oneapi_get_info(queue Queue, const range<3> &WorkGroupSize,
                                           DynamicLocalMemorySize);
 }
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-// These functions are unused and should be removed in the next ABI-breaking
-// window.
-
 template __SYCL_EXPORT typename ext::oneapi::experimental::info::
     kernel_queue_specific::max_work_group_size::return_type
     kernel::ext_oneapi_get_info<ext::oneapi::experimental::info::
@@ -186,8 +180,6 @@ template __SYCL_EXPORT typename ext::oneapi::experimental::info::
         ext::oneapi::experimental::info::kernel_queue_specific::
             max_work_item_sizes<3>>(queue Queue) const;
 
-#endif // __INTEL_PREVIEW_BREAKING_CHANGES
-
 template __SYCL_EXPORT typename ext::oneapi::experimental::info::
     kernel_queue_specific::max_sub_group_size::return_type
     kernel::ext_oneapi_get_info<ext::oneapi::experimental::info::
@@ -245,55 +237,5 @@ kernel::kernel(std::shared_ptr<detail::kernel_impl> Impl) : impl(Impl) {}
 
 ur_native_handle_t kernel::getNative() const { return impl->getNative(); }
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-ur_native_handle_t kernel::getNativeImpl() const { return impl->getNative(); }
-
-// The following query was deprecated since it doesn't include a way to specify
-// the invdividual dimensions of the work group. All of the contents of this
-// #ifndef block should be removed during the next ABI breaking window.
-namespace ext::oneapi::experimental::info::kernel_queue_specific {
-struct max_num_work_group_sync {
-  using return_type = size_t;
-};
-} // namespace ext::oneapi::experimental::info::kernel_queue_specific
-template <>
-struct detail::is_kernel_queue_specific_info_desc<
-    ext::oneapi::experimental::info::kernel_queue_specific::
-        max_num_work_group_sync> : std::true_type {
-  using return_type = ext::oneapi::experimental::info::kernel_queue_specific::
-      max_num_work_group_sync::return_type;
-};
-template <>
-__SYCL2020_DEPRECATED(
-    "The 'max_num_work_group_sync' query is deprecated. See "
-    "'sycl_ext_oneapi_launch_queries' for the new 'max_num_work_groups' query.")
-__SYCL_EXPORT typename ext::oneapi::experimental::info::kernel_queue_specific::
-    max_num_work_group_sync::return_type kernel::ext_oneapi_get_info<
-        ext::oneapi::experimental::info::kernel_queue_specific::
-            max_num_work_group_sync>(queue Queue, const range<3> &WorkGroupSize,
-                                     size_t DynamicLocalMemorySize) const {
-  return ext_oneapi_get_info<ext::oneapi::experimental::info::
-                                 kernel_queue_specific::max_num_work_groups>(
-      std::move(Queue), WorkGroupSize, DynamicLocalMemorySize);
-}
-template <>
-__SYCL2020_DEPRECATED(
-    "The 'max_num_work_group_sync' query is deprecated. See "
-    "'sycl_ext_oneapi_launch_queries' for the new 'max_num_work_groups' query.")
-__SYCL_EXPORT typename ext::oneapi::experimental::info::kernel_queue_specific::
-    max_num_work_group_sync::return_type kernel::ext_oneapi_get_info<
-        ext::oneapi::experimental::info::kernel_queue_specific::
-            max_num_work_group_sync>(queue Queue) const {
-  auto Device = Queue.get_device();
-  const auto MaxWorkGroupSize =
-      get_info<info::kernel_device_specific::work_group_size>(Device);
-  const sycl::range<3> WorkGroupSize{MaxWorkGroupSize, 1, 1};
-  return ext_oneapi_get_info<ext::oneapi::experimental::info::
-                                 kernel_queue_specific::max_num_work_groups>(
-      std::move(Queue), WorkGroupSize,
-      /* DynamicLocalMemorySize */ 0);
-}
-#endif
-
 } // namespace _V1
 } // namespace sycl

sycl/test-e2e/Experimental/launch_queries/max_work_group_size.cpp

@@ -1,9 +1,6 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-// XFAIL: preview-mode
-// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18910
-
 #include <sycl/detail/core.hpp>
 #include <sycl/kernel_bundle.hpp>
 

sycl/test-e2e/Experimental/launch_queries/max_work_item_sizes.cpp

@@ -1,9 +1,6 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-// XFAIL: preview-mode
-// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18910
-
 #include <sycl/detail/core.hpp>
 #include <sycl/detail/info_desc_helpers.hpp>
 #include <sycl/kernel.hpp>

sycl/test/abi/sycl_symbols_linux.dump

@@ -3942,7 +3942,6 @@ _ZNK4sycl3_V16device6is_gpuEv
 _ZNK4sycl3_V16device9getNativeEv
 _ZNK4sycl3_V16kernel11get_backendEv
 _ZNK4sycl3_V16kernel11get_contextEv
-_ZNK4sycl3_V16kernel13getNativeImplEv
 _ZNK4sycl3_V16kernel13get_info_implINS0_4info6kernel10attributesEEENS0_6detail11ABINeutralTINS6_19is_kernel_info_descIT_E11return_typeEE4typeEv
 _ZNK4sycl3_V16kernel13get_info_implINS0_4info6kernel13function_nameEEENS0_6detail11ABINeutralTINS6_19is_kernel_info_descIT_E11return_typeEE4typeEv
 _ZNK4sycl3_V16kernel13get_info_implINS0_4info6kernel15reference_countEEENS0_6detail11ABINeutralTINS6_19is_kernel_info_descIT_E11return_typeEE4typeEv
@@ -3962,8 +3961,6 @@ _ZNK4sycl3_V16kernel19ext_oneapi_get_infoINS0_3ext6oneapi12experimental4info21ke
 _ZNK4sycl3_V16kernel19ext_oneapi_get_infoINS0_3ext6oneapi12experimental4info21kernel_queue_specific19max_work_item_sizesILi1EEEEENS0_6detail34is_kernel_queue_specific_info_descIT_E11return_typeENS0_5queueE
 _ZNK4sycl3_V16kernel19ext_oneapi_get_infoINS0_3ext6oneapi12experimental4info21kernel_queue_specific19max_work_item_sizesILi2EEEEENS0_6detail34is_kernel_queue_specific_info_descIT_E11return_typeENS0_5queueE
 _ZNK4sycl3_V16kernel19ext_oneapi_get_infoINS0_3ext6oneapi12experimental4info21kernel_queue_specific19max_work_item_sizesILi3EEEEENS0_6detail34is_kernel_queue_specific_info_descIT_E11return_typeENS0_5queueE
-_ZNK4sycl3_V16kernel19ext_oneapi_get_infoINS0_3ext6oneapi12experimental4info21kernel_queue_specific23max_num_work_group_syncEEENS0_6detail34is_kernel_queue_specific_info_descIT_E11return_typeENS0_5queueE
-_ZNK4sycl3_V16kernel19ext_oneapi_get_infoINS0_3ext6oneapi12experimental4info21kernel_queue_specific23max_num_work_group_syncEEENS0_6detail34is_kernel_queue_specific_info_descIT_E11return_typeENS0_5queueERKNS0_5rangeILi3EEEm
 _ZNK4sycl3_V16kernel3getEv
 _ZNK4sycl3_V16kernel8get_infoINS0_3ext5intel4info22kernel_device_specific17spill_memory_sizeEEENS0_6detail35is_kernel_device_specific_info_descIT_E11return_typeERKNS0_6deviceE
 _ZNK4sycl3_V16kernel8get_infoINS0_4info22kernel_device_specific15work_group_sizeEEENS0_6detail35is_kernel_device_specific_info_descIT_E11return_typeERKNS0_6deviceE

sycl/test/abi/sycl_symbols_windows.dump

@@ -14,8 +14,6 @@
 ??$ext_oneapi_get_info@U?$max_work_item_sizes@$00@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA?AV?$id@$00@12@Vqueue@12@@Z
 ??$ext_oneapi_get_info@U?$max_work_item_sizes@$01@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA?AV?$id@$01@12@Vqueue@12@@Z
 ??$ext_oneapi_get_info@U?$max_work_item_sizes@$02@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA?AV?$id@$02@12@Vqueue@12@@Z
-??$ext_oneapi_get_info@Umax_num_work_group_sync@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA_KVqueue@12@@Z
-??$ext_oneapi_get_info@Umax_num_work_group_sync@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA_KVqueue@12@AEBV?$range@$02@12@_K@Z
 ??$ext_oneapi_get_info@Umax_num_work_groups@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA_KVqueue@12@AEBV?$range@$00@12@_K@Z
 ??$ext_oneapi_get_info@Umax_num_work_groups@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA_KVqueue@12@AEBV?$range@$01@12@_K@Z
 ??$ext_oneapi_get_info@Umax_num_work_groups@kernel_queue_specific@info@experimental@oneapi@ext@_V1@sycl@@@kernel@_V1@sycl@@QEBA_KVqueue@12@AEBV?$range@$02@12@_K@Z
@@ -4003,7 +4001,6 @@
 ?getNativeContext@interop_handle@_V1@sycl@@AEBA_KXZ
 ?getNativeDevice@interop_handle@_V1@sycl@@AEBA_KXZ
 ?getNativeGraph@interop_handle@_V1@sycl@@AEBA_KXZ
-?getNativeImpl@kernel@_V1@sycl@@AEBA_KXZ
 ?getNativeMem@interop_handle@_V1@sycl@@AEBA_KPEAVAccessorImplHost@detail@23@@Z
 ?getNativeQueue@interop_handle@_V1@sycl@@AEBA_KAEAH@Z
 ?getNativeVector@buffer_plain@detail@_V1@sycl@@IEBA?AV?$vector@_KV?$allocator@_K@std@@@std@@W4backend@34@@Z

sycl/unittests/Extensions/LaunchQueries.cpp

@@ -61,9 +61,6 @@ auto getKernel(const sycl::queue &Q) {
   return KernelBundle.get_kernel(KernelID);
 }
 
-// These tests fail in preview breaking mode.
-// Failure Tracker: https://github.com/intel/llvm/issues/18910
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
 TEST(LaunchQueries, GetWorkGroupSizeSuccess) {
   sycl::unittest::UrMock<> Mock;
   mock::getCallbacks().set_replace_callback(
@@ -149,7 +146,6 @@ TEST(LaunchQueries, GetMaxWorkGroupItemSizesExceptionCode) {
           syclex::info::kernel_queue_specific::max_work_item_sizes<3>>(Queue),
       sycl::exception);
 }
-#endif // __INTEL_PREVIEW_BREAKING_CHANGES
 
 TEST(LaunchQueries, GetMaxSubGroupSize3DSuccess) {
   sycl::unittest::UrMock<> Mock;