diff --git a/sycl/include/sycl/backend.hpp b/sycl/include/sycl/backend.hpp
index 668735af5faa5..611c794c9a8f6 100644
--- a/sycl/include/sycl/backend.hpp
+++ b/sycl/include/sycl/backend.hpp
@@ -127,7 +127,19 @@ auto get_native(const SyclObjectT &Obj)
     throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
                               PI_ERROR_INVALID_OPERATION);
   }
-  return Obj.template get_native<BackendName>();
+  return reinterpret_cast<backend_return_t<BackendName, SyclObjectT>>(
+      Obj.getNative());
+}
+
+template <backend BackendName, bundle_state State>
+auto get_native(const kernel_bundle<State> &Obj)
+    -> backend_return_t<BackendName, kernel_bundle<State>> {
+  // TODO use SYCL 2020 exception when implemented
+  if (Obj.get_backend() != BackendName) {
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_ERROR_INVALID_OPERATION);
+  }
+  return Obj.template getNative<BackendName>();
 }
 
 template <backend BackendName, typename DataT, int Dimensions,
diff --git a/sycl/include/sycl/context.hpp b/sycl/include/sycl/context.hpp
index 10c60ca1626e1..b38a820e0b65f 100644
--- a/sycl/include/sycl/context.hpp
+++ b/sycl/include/sycl/context.hpp
@@ -217,15 +217,6 @@ class __SYCL_EXPORT context {
   /// \return a vector of valid SYCL device instances.
   std::vector<device> get_devices() const;
 
-  /// Gets the native handle of the SYCL context.
-  ///
-  /// \return a native handle, the type of which defined by the backend.
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, context> get_native() const {
-    return reinterpret_cast<backend_return_t<Backend, context>>(getNative());
-  }
-
 private:
   /// Constructs a SYCL context object from a valid context_impl instance.
   context(std::shared_ptr<detail::context_impl> Impl);
diff --git a/sycl/include/sycl/device.hpp b/sycl/include/sycl/device.hpp
index 5e969ca71ecc6..6e7f794af5c98 100644
--- a/sycl/include/sycl/device.hpp
+++ b/sycl/include/sycl/device.hpp
@@ -24,6 +24,9 @@ __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 // Forward declarations
 class device_selector;
+template <backend BackendName, class SyclObjectT>
+auto get_native(const SyclObjectT &Obj)
+    -> backend_return_t<BackendName, SyclObjectT>;
 namespace detail {
 class device_impl;
 auto getDeviceComparisonLambda();
@@ -184,19 +187,6 @@ class __SYCL_EXPORT device {
   /// \return the backend associated with this device.
   backend get_backend() const noexcept;
 
-  /// Gets the native handle of the SYCL device.
-  ///
-  /// \return a native handle, the type of which defined by the backend.
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, device> get_native() const {
-    // In CUDA CUdevice isn't an opaque pointer, unlike a lot of the others,
-    // but instead a 32-bit int (on all relevant systems). Different
-    // backends use the same function for this purpose so static_cast is
-    // needed in some cases but not others, so a C-style cast was chosen.
-    return (backend_return_t<Backend, device>)getNative();
-  }
-
   /// Indicates if the SYCL device has the given feature.
   ///
   /// \param Aspect is one of the values in Table 4.20 of the SYCL 2020
@@ -223,6 +213,10 @@ class __SYCL_EXPORT device {
   friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
 
   friend auto detail::getDeviceComparisonLambda();
+
+  template <backend BackendName, class SyclObjectT>
+  friend auto get_native(const SyclObjectT &Obj)
+      -> backend_return_t<BackendName, SyclObjectT>;
 };
 
 } // namespace sycl
diff --git a/sycl/include/sycl/event.hpp b/sycl/include/sycl/event.hpp
index 396df38fa675e..d677f2f07d9ce 100644
--- a/sycl/include/sycl/event.hpp
+++ b/sycl/include/sycl/event.hpp
@@ -21,6 +21,11 @@ __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 // Forward declaration
 class context;
+
+template <backend BackendName, class SyclObjectT>
+auto get_native(const SyclObjectT &Obj)
+    -> backend_return_t<BackendName, SyclObjectT>;
+
 namespace detail {
 class event_impl;
 }
@@ -128,15 +133,6 @@ class __SYCL_EXPORT event {
   /// \return the backend associated with this platform
   backend get_backend() const noexcept;
 
-  /// Gets the native handle of the SYCL event.
-  ///
-  /// \return a native handle, the type of which defined by the backend.
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, event> get_native() const {
-    return reinterpret_cast<backend_return_t<Backend, event>>(getNative());
-  }
-
 private:
   event(std::shared_ptr<detail::event_impl> EventImpl);
 
diff --git a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
index 6deb8fb0630a7..8b3c5e8792b00 100644
--- a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
+++ b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
@@ -221,21 +221,6 @@ make_buffer(
       !(BackendObject.Ownership == ext::oneapi::level_zero::ownership::keep));
 }
 
-// TODO: remove this specialization when generic is changed to call
-// .GetNative() instead of .get_native() member of kernel_bundle.
-template <>
-inline auto get_native<backend::ext_oneapi_level_zero>(
-    const kernel_bundle<bundle_state::executable> &Obj)
-    -> backend_return_t<backend::ext_oneapi_level_zero,
-                        kernel_bundle<bundle_state::executable>> {
-  // TODO use SYCL 2020 exception when implemented
-  if (Obj.get_backend() != backend::ext_oneapi_level_zero)
-    throw runtime_error(errc::backend_mismatch, "Backends mismatch",
-                        PI_ERROR_INVALID_OPERATION);
-
-  return Obj.template getNative<backend::ext_oneapi_level_zero>();
-}
-
 namespace __SYCL2020_DEPRECATED("use 'ext::oneapi::level_zero' instead")
     level_zero {
   using namespace ext::oneapi::level_zero;
diff --git a/sycl/include/sycl/ext/oneapi/experimental/backend/cuda.hpp b/sycl/include/sycl/ext/oneapi/experimental/backend/cuda.hpp
index f61a3da215ecf..1d76960cc5ad5 100644
--- a/sycl/include/sycl/ext/oneapi/experimental/backend/cuda.hpp
+++ b/sycl/include/sycl/ext/oneapi/experimental/backend/cuda.hpp
@@ -10,6 +10,7 @@
 
 #include <sycl/backend.hpp>
 #include <sycl/context.hpp>
+#include <sycl/ext/oneapi/experimental/backend/backend_traits_cuda.hpp>
 
 #include <vector>
 
@@ -52,13 +53,6 @@ inline auto get_native<backend::ext_oneapi_cuda, context>(const context &C)
   return ret;
 }
 
-// Specialisation of non-free context get_native
-template <>
-inline backend_return_t<backend::ext_oneapi_cuda, context>
-context::get_native<backend::ext_oneapi_cuda>() const {
-  return sycl::get_native<backend::ext_oneapi_cuda, context>(*this);
-}
-
 // Specialisation of interop_handles get_native_context
 template <>
 inline backend_return_t<backend::ext_oneapi_cuda, context>
@@ -79,6 +73,20 @@ inline device make_device<backend::ext_oneapi_cuda>(
   return ext::oneapi::cuda::make_device(NativeHandle);
 }
 
+template <>
+backend_return_t<backend::ext_oneapi_cuda, device>
+get_native<backend::ext_oneapi_cuda, device>(const device &Obj) {
+  // TODO use SYCL 2020 exception when implemented
+  if (Obj.get_backend() != backend::ext_oneapi_cuda) {
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_ERROR_INVALID_OPERATION);
+  }
+  // CUDA uses a 32-bit int instead of an opaque pointer like other backends,
+  // so we need a specialization with static_cast instead of reinterpret_cast.
+  return static_cast<backend_return_t<backend::ext_oneapi_cuda, device>>(
+      Obj.getNative());
+}
+
 // CUDA event specialization
 template <>
 inline event make_event<backend::ext_oneapi_cuda>(
diff --git a/sycl/include/sycl/kernel.hpp b/sycl/include/sycl/kernel.hpp
index f7d13a1b73756..bfa871bb97f7e 100644
--- a/sycl/include/sycl/kernel.hpp
+++ b/sycl/include/sycl/kernel.hpp
@@ -27,6 +27,9 @@ class program;
 class context;
 template <backend Backend> class backend_traits;
 template <bundle_state State> class kernel_bundle;
+template <backend BackendName, class SyclObjectT>
+auto get_native(const SyclObjectT &Obj)
+    -> backend_return_t<BackendName, SyclObjectT>;
 
 namespace detail {
 class kernel_impl;
@@ -190,12 +193,6 @@ class __SYCL_EXPORT kernel {
                      param>::input_type Value) const;
   // clang-format on
 
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, kernel> get_native() const {
-    return detail::pi::cast<backend_return_t<Backend, kernel>>(getNative());
-  }
-
 private:
   /// Constructs a SYCL kernel object from a valid kernel_impl instance.
   kernel(std::shared_ptr<detail::kernel_impl> Impl);
@@ -211,6 +208,9 @@ class __SYCL_EXPORT kernel {
   friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
   template <class T>
   friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
+  template <backend BackendName, class SyclObjectT>
+  friend auto get_native(const SyclObjectT &Obj)
+      -> backend_return_t<BackendName, SyclObjectT>;
 };
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/include/sycl/kernel_bundle.hpp b/sycl/include/sycl/kernel_bundle.hpp
index 47f4a8fa344ca..19f31c3969bb2 100644
--- a/sycl/include/sycl/kernel_bundle.hpp
+++ b/sycl/include/sycl/kernel_bundle.hpp
@@ -26,8 +26,9 @@ __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 // Forward declaration
 template <backend Backend> class backend_traits;
-template <backend Backend, class SyclT>
-auto get_native(const SyclT &Obj) -> backend_return_t<Backend, SyclT>;
+template <backend Backend, bundle_state State>
+auto get_native(const kernel_bundle<State> &Obj)
+    -> backend_return_t<Backend, kernel_bundle<State>>;
 
 namespace detail {
 class kernel_id_impl;
@@ -310,12 +311,6 @@ class kernel_bundle : public detail::kernel_bundle_plain {
     return reinterpret_cast<device_image_iterator>(kernel_bundle_plain::end());
   }
 
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, kernel_bundle<State>> get_native() const {
-    return getNative<Backend>();
-  }
-
 private:
   kernel_bundle(detail::KernelBundleImplPtr Impl)
       : kernel_bundle_plain(std::move(Impl)) {}
@@ -326,8 +321,9 @@ class kernel_bundle : public detail::kernel_bundle_plain {
   template <class T>
   friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
 
-  template <backend Backend, class SyclT>
-  friend auto get_native(const SyclT &Obj) -> backend_return_t<Backend, SyclT>;
+  template <backend Backend, bundle_state StateB>
+  friend auto get_native(const kernel_bundle<StateB> &Obj)
+      -> backend_return_t<Backend, kernel_bundle<StateB>>;
 
   template <backend Backend>
   backend_return_t<Backend, kernel_bundle<State>> getNative() const {
diff --git a/sycl/include/sycl/platform.hpp b/sycl/include/sycl/platform.hpp
index 535f3f49bee52..49efb14abfb7c 100644
--- a/sycl/include/sycl/platform.hpp
+++ b/sycl/include/sycl/platform.hpp
@@ -24,6 +24,9 @@ namespace sycl {
 // Forward declaration
 class device_selector;
 class device;
+template <backend BackendName, class SyclObjectT>
+auto get_native(const SyclObjectT &Obj)
+    -> backend_return_t<BackendName, SyclObjectT>;
 namespace detail {
 class platform_impl;
 }
@@ -117,15 +120,6 @@ class __SYCL_EXPORT platform {
   /// \return the backend associated with this platform
   backend get_backend() const noexcept;
 
-  /// Gets the native handle of the SYCL platform.
-  ///
-  /// \return a native handle, the type of which defined by the backend.
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, platform> get_native() const {
-    return reinterpret_cast<backend_return_t<Backend, platform>>(getNative());
-  }
-
   /// Indicates if all of the SYCL devices on this platform have the
   /// given feature.
   ///
@@ -152,6 +146,9 @@ class __SYCL_EXPORT platform {
   template <class Obj>
   friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
 
+  template <backend BackendName, class SyclObjectT>
+  friend auto get_native(const SyclObjectT &Obj)
+      -> backend_return_t<BackendName, SyclObjectT>;
 }; // class platform
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/include/sycl/program.hpp b/sycl/include/sycl/program.hpp
index 6017ce29f8b4d..0c5faf472d718 100644
--- a/sycl/include/sycl/program.hpp
+++ b/sycl/include/sycl/program.hpp
@@ -28,6 +28,10 @@ namespace sycl {
 // Forward declarations
 class context;
 class device;
+template <backend BackendName, class SyclObjectT>
+auto get_native(const SyclObjectT &Obj)
+    -> backend_return_t<BackendName, SyclObjectT>;
+
 namespace detail {
 class program_impl;
 }
@@ -365,15 +369,6 @@ class __SYCL_EXPORT __SYCL2020_DEPRECATED(
   /// \return the backend associated with this program.
   backend get_backend() const noexcept;
 
-  /// Gets the native handle of the SYCL platform.
-  ///
-  /// \return a native handle, the type of which defined by the backend.
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, program> get_native() const {
-    return reinterpret_cast<backend_return_t<Backend, program>>(getNative());
-  }
-
 private:
   pi_native_handle getNative() const;
   program(std::shared_ptr<detail::program_impl> impl);
@@ -419,6 +414,9 @@ class __SYCL_EXPORT __SYCL2020_DEPRECATED(
   friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
   template <class T>
   friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
+  template <backend BackendName, class SyclObjectT>
+  friend auto get_native(const SyclObjectT &Obj)
+      -> backend_return_t<BackendName, SyclObjectT>;
 };
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/include/sycl/queue.hpp b/sycl/include/sycl/queue.hpp
index bb9deaa90820b..96e74d564dae4 100644
--- a/sycl/include/sycl/queue.hpp
+++ b/sycl/include/sycl/queue.hpp
@@ -81,8 +81,13 @@ class context;
 class device;
 class queue;
 
+template <backend BackendName, class SyclObjectT>
+auto get_native(const SyclObjectT &Obj)
+    -> backend_return_t<BackendName, SyclObjectT>;
+
 namespace detail {
 class queue_impl;
+
 #if __SYCL_USE_FALLBACK_ASSERT
 static event submitAssertCapture(queue &, event &, queue *,
                                  const detail::code_location &);
@@ -1033,15 +1038,6 @@ class __SYCL_EXPORT queue {
   /// \return the backend associated with this queue.
   backend get_backend() const noexcept;
 
-  /// Gets the native handle of the SYCL queue.
-  ///
-  /// \return a native handle, the type of which defined by the backend.
-  template <backend Backend>
-  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
-  backend_return_t<Backend, queue> get_native() const {
-    return reinterpret_cast<backend_return_t<Backend, queue>>(getNative());
-  }
-
 private:
   pi_native_handle getNative() const;
 
@@ -1053,6 +1049,10 @@ class __SYCL_EXPORT queue {
   template <class T>
   friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
 
+  template <backend BackendName, class SyclObjectT>
+  friend auto get_native(const SyclObjectT &Obj)
+      -> backend_return_t<BackendName, SyclObjectT>;
+
 #if __SYCL_USE_FALLBACK_ASSERT
   friend event detail::submitAssertCapture(queue &, event &, queue *,
                                            const detail::code_location &);
diff --git a/sycl/test/basic_tests/interop-cuda-experimental.cpp b/sycl/test/basic_tests/interop-cuda-experimental.cpp
index 40363384b022b..fc48f37d3fd9e 100644
--- a/sycl/test/basic_tests/interop-cuda-experimental.cpp
+++ b/sycl/test/basic_tests/interop-cuda-experimental.cpp
@@ -56,20 +56,6 @@ int main() {
   cu_event = get_native<backend::ext_oneapi_cuda>(Event);
   cu_queue = get_native<backend::ext_oneapi_cuda>(Queue);
 
-  // Check deprecated
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_cuda>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  cu_device = Device.get_native<backend::ext_oneapi_cuda>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_cuda>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  cu_context = Context.get_native<backend::ext_oneapi_cuda>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_cuda>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  cu_event = Event.get_native<backend::ext_oneapi_cuda>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_cuda>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  cu_queue = Queue.get_native<backend::ext_oneapi_cuda>();
-
   // 4.5.1.1 For each SYCL runtime class T which supports SYCL application
   // interoperability with the SYCL backend, a specialization of input_type must
   // be defined as the type of SYCL application interoperability native backend
diff --git a/sycl/test/basic_tests/interop-level-zero-2020.cpp b/sycl/test/basic_tests/interop-level-zero-2020.cpp
index 770626294ae53..5035eef62f4d1 100644
--- a/sycl/test/basic_tests/interop-level-zero-2020.cpp
+++ b/sycl/test/basic_tests/interop-level-zero-2020.cpp
@@ -75,30 +75,6 @@ int main() {
   ZeKernelBundle = get_native<backend::ext_oneapi_level_zero>(KernelBundle);
   ZeKernel = get_native<backend::ext_oneapi_level_zero>(Kernel);
 
-  // Check deprecated
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  ZeDriver = Platform.get_native<backend::ext_oneapi_level_zero>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  ZeDevice = Device.get_native<backend::ext_oneapi_level_zero>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  ZeContext = Context.get_native<backend::ext_oneapi_level_zero>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  ZeQueue = Queue.get_native<backend::ext_oneapi_level_zero>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  ZeEvent = Event.get_native<backend::ext_oneapi_level_zero>();
-  // expected-warning@+3 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+2 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  /*ZeKernelBundle*/ (
-      void)KernelBundle.get_native<backend::ext_oneapi_level_zero>();
-  // expected-warning@+2 {{'get_native' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  // expected-warning@+1 {{'get_native<sycl::backend::ext_oneapi_level_zero>' is deprecated: Use SYCL 2020 sycl::get_native free function}}
-  ZeKernel = Kernel.get_native<backend::ext_oneapi_level_zero>();
-
   // 4.5.1.1 For each SYCL runtime class T which supports SYCL application
   // interoperability with the SYCL backend, a specialization of input_type must
   // be defined as the type of SYCL application interoperability native backend