diff --git a/sycl/include/CL/sycl/backend/cuda.hpp b/sycl/include/CL/sycl/backend/cuda.hpp
index 08a2b80c8a073..7e6353503ab2b 100644
--- a/sycl/include/CL/sycl/backend/cuda.hpp
+++ b/sycl/include/CL/sycl/backend/cuda.hpp
@@ -41,7 +41,9 @@ template <> struct interop<backend::cuda, queue> { using type = CUstream; };
 
 template <> struct interop<backend::cuda, event> { using type = CUevent; };
 
+#ifdef __SYCL_INTERNAL_API
 template <> struct interop<backend::cuda, program> { using type = CUmodule; };
+#endif
 
 template <typename DataT, int Dimensions, access::mode AccessMode>
 struct interop<backend::cuda,
diff --git a/sycl/include/CL/sycl/backend/opencl.hpp b/sycl/include/CL/sycl/backend/opencl.hpp
index 7820f0e398102..83f3c65501f9a 100644
--- a/sycl/include/CL/sycl/backend/opencl.hpp
+++ b/sycl/include/CL/sycl/backend/opencl.hpp
@@ -37,9 +37,11 @@ template <> struct interop<backend::opencl, queue> {
   using type = cl_command_queue;
 };
 
+#ifdef __SYCL_INTERNAL_API
 template <> struct interop<backend::opencl, program> {
   using type = cl_program;
 };
+#endif
 
 template <> struct interop<backend::opencl, event> { using type = cl_event; };
 
@@ -110,8 +112,10 @@ namespace opencl {
 __SYCL_EXPORT platform make_platform(pi_native_handle NativeHandle);
 __SYCL_EXPORT device make_device(pi_native_handle NativeHandle);
 __SYCL_EXPORT context make_context(pi_native_handle NativeHandle);
+#ifdef __SYCL_INTERNAL_API
 __SYCL_EXPORT program make_program(const context &Context,
                                    pi_native_handle NativeHandle);
+#endif
 __SYCL_EXPORT queue make_queue(const context &Context,
                                pi_native_handle InteropHandle);
 
@@ -137,12 +141,14 @@ T make(typename interop<backend::opencl, T>::type Interop) {
 }
 
 // Construction of SYCL program.
+#ifdef __SYCL_INTERNAL_API
 template <typename T, typename detail::enable_if_t<
                           std::is_same<T, program>::value> * = nullptr>
 T make(const context &Context,
        typename interop<backend::opencl, T>::type Interop) {
   return make_program(Context, detail::pi::cast<pi_native_handle>(Interop));
 }
+#endif
 
 // Construction of SYCL queue.
 template <typename T, typename detail::enable_if_t<
diff --git a/sycl/include/CL/sycl/info/info_desc.hpp b/sycl/include/CL/sycl/info/info_desc.hpp
index cc74ed463dd3e..d749f1125e054 100644
--- a/sycl/include/CL/sycl/info/info_desc.hpp
+++ b/sycl/include/CL/sycl/info/info_desc.hpp
@@ -15,7 +15,9 @@
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 
+#ifdef __SYCL_INTERNAL_API
 class program;
+#endif
 class device;
 class platform;
 
@@ -220,7 +222,9 @@ enum class kernel : cl_kernel_info {
   function_name = CL_KERNEL_FUNCTION_NAME,
   num_args = CL_KERNEL_NUM_ARGS,
   context = CL_KERNEL_CONTEXT,
+#ifdef __SYCL_INTERNAL_API
   program = CL_KERNEL_PROGRAM,
+#endif
   reference_count = CL_KERNEL_REFERENCE_COUNT,
   attributes = CL_KERNEL_ATTRIBUTES
 };
@@ -257,11 +261,13 @@ enum class kernel_device_specific : cl_kernel_work_group_info {
 };
 
 // A.6 Program information desctiptors
+#ifdef __SYCL_INTERNAL_API
 enum class program : cl_program_info {
   context = CL_PROGRAM_CONTEXT,
   devices = CL_PROGRAM_DEVICES,
   reference_count = CL_PROGRAM_REFERENCE_COUNT
 };
+#endif
 
 // A.7 Event information desctiptors
 enum class event : cl_event_info {
@@ -315,7 +321,9 @@ template <typename T, T param> struct compatibility_param_traits {};
 
 #include <CL/sycl/info/platform_traits.def>
 
+#ifdef __SYCL_INTERNAL_API
 #include <CL/sycl/info/program_traits.def>
+#endif
 
 #include <CL/sycl/info/queue_traits.def>
 
diff --git a/sycl/include/CL/sycl/info/kernel_traits.def b/sycl/include/CL/sycl/info/kernel_traits.def
index 789f33ce577fc..00126cf257390 100644
--- a/sycl/include/CL/sycl/info/kernel_traits.def
+++ b/sycl/include/CL/sycl/info/kernel_traits.def
@@ -3,5 +3,7 @@ __SYCL_PARAM_TRAITS_SPEC(kernel, num_args, cl_uint)
 __SYCL_PARAM_TRAITS_SPEC(kernel, reference_count, cl_uint)
 __SYCL_PARAM_TRAITS_SPEC(kernel, attributes, std::string)
 __SYCL_PARAM_TRAITS_SPEC(kernel, context, cl::sycl::context)
+#ifdef __SYCL_INTERNAL_API
 __SYCL_PARAM_TRAITS_SPEC(kernel, program, cl::sycl::program)
+#endif
 
diff --git a/sycl/include/CL/sycl/kernel.hpp b/sycl/include/CL/sycl/kernel.hpp
index a73e8b9948a12..272c9080766d8 100644
--- a/sycl/include/CL/sycl/kernel.hpp
+++ b/sycl/include/CL/sycl/kernel.hpp
@@ -20,7 +20,9 @@
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 // Forward declaration
+#ifdef __SYCL_INTERNAL_API
 class program;
+#endif
 class context;
 template <backend Backend> class backend_traits;
 template <bundle_state State> class kernel_bundle;
@@ -127,7 +129,9 @@ class __SYCL_EXPORT kernel {
   /// get_info<info::kernel::program>().
   ///
   /// \return a valid SYCL program
+#ifdef __SYCL_INTERNAL_API
   program get_program() const;
+#endif
 
   /// Query information from the kernel object using the info::kernel_info
   /// descriptor.
diff --git a/sycl/include/CL/sycl/program.hpp b/sycl/include/CL/sycl/program.hpp
index 7973ff03cf273..6ecc35492c62b 100644
--- a/sycl/include/CL/sycl/program.hpp
+++ b/sycl/include/CL/sycl/program.hpp
@@ -8,6 +8,8 @@
 
 #pragma once
 
+#ifdef __SYCL_INTERNAL_API
+
 #include <CL/sycl/context.hpp>
 #include <CL/sycl/detail/export.hpp>
 #include <CL/sycl/detail/kernel_desc.hpp>
@@ -428,3 +430,5 @@ template <> struct hash<cl::sycl::program> {
   }
 };
 } // namespace std
+
+#endif
diff --git a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
index d12e649e2b4c1..654767b27d387 100644
--- a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
+++ b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
@@ -35,9 +35,11 @@ template <> struct interop<backend::level_zero, event> {
   using type = ze_event_handle_t;
 };
 
+#ifdef __SYCL_INTERNAL_API
 template <> struct interop<backend::level_zero, program> {
   using type = ze_module_handle_t;
 };
+#endif
 
 template <> struct interop<backend::level_zero, kernel> {
   using type = ze_kernel_handle_t;
@@ -154,8 +156,10 @@ __SYCL_EXPORT device make_device(const platform &Platform,
 __SYCL_EXPORT context make_context(const std::vector<device> &DeviceList,
                                    pi_native_handle NativeHandle,
                                    bool keep_ownership = false);
+#ifdef __SYCL_INTERNAL_API
 __SYCL_EXPORT program make_program(const context &Context,
                                    pi_native_handle NativeHandle);
+#endif
 __SYCL_EXPORT queue make_queue(const context &Context,
                                pi_native_handle InteropHandle,
                                bool keep_ownership = false);
@@ -200,6 +204,7 @@ T make(const std::vector<device> &DeviceList,
 }
 
 // Construction of SYCL program.
+#ifdef __SYCL_INTERNAL_API
 template <typename T, typename detail::enable_if_t<
                           std::is_same<T, program>::value> * = nullptr>
 __SYCL_DEPRECATED("Use SYCL 2020 sycl::make_kernel_bundle free function")
@@ -207,6 +212,7 @@ T make(const context &Context,
        typename interop<backend::level_zero, T>::type Interop) {
   return make_program(Context, reinterpret_cast<pi_native_handle>(Interop));
 }
+#endif
 
 // Construction of SYCL queue.
 template <typename T, typename detail::enable_if_t<
diff --git a/sycl/source/detail/kernel_impl.hpp b/sycl/source/detail/kernel_impl.hpp
index f4b734d37ebe8..857efd7bf577d 100644
--- a/sycl/source/detail/kernel_impl.hpp
+++ b/sycl/source/detail/kernel_impl.hpp
@@ -226,10 +226,12 @@ inline context kernel_impl::get_info<info::kernel::context>() const {
   return createSyclObjFromImpl<context>(MContext);
 }
 
+#ifdef __SYCL_INTERNAL_API
 template <>
 inline program kernel_impl::get_info<info::kernel::program>() const {
   return createSyclObjFromImpl<program>(MProgramImpl);
 }
+#endif
 
 template <info::kernel_device_specific param>
 inline typename info::param_traits<info::kernel_device_specific,
diff --git a/sycl/test/abi/symbol_size_alignment.cpp b/sycl/test/abi/symbol_size_alignment.cpp
index 94582e9e595aa..68cd05e158745 100644
--- a/sycl/test/abi/symbol_size_alignment.cpp
+++ b/sycl/test/abi/symbol_size_alignment.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t
+// RUN: %clangxx -fsycl %s -D__SYCL_INTERNAL_API -o %t
 // UNSUPPORTED: libcxx
 
 // Changing symbol size or alignment is a breaking change. If it happens, refer
diff --git a/sycl/test/abi/user_mangling.cpp b/sycl/test/abi/user_mangling.cpp
index 3d41f5d8f9bd7..44fa3a866900b 100644
--- a/sycl/test/abi/user_mangling.cpp
+++ b/sycl/test/abi/user_mangling.cpp
@@ -1,6 +1,6 @@
 // clang-format off
-// RUN: %clangxx -fsycl -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-HOST
-// RUN: %clangxx -fsycl -fsycl-device-only -O0 -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-DEVICE
+// RUN: %clangxx -fsycl -c -emit-llvm -D__SYCL_INTERNAL_API -S -o - %s | FileCheck %s --check-prefix CHK-HOST
+// RUN: %clangxx -fsycl -fsycl-device-only -D__SYCL_INTERNAL_API -O0 -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-DEVICE
 // REQUIRES: linux
 // UNSUPPORTED: libcxx
 
diff --git a/sycl/test/basic_tests/interop-level-zero-2020.cpp b/sycl/test/basic_tests/interop-level-zero-2020.cpp
index 1f9893c3079a8..4bc313fbf76c4 100644
--- a/sycl/test/basic_tests/interop-level-zero-2020.cpp
+++ b/sycl/test/basic_tests/interop-level-zero-2020.cpp
@@ -33,8 +33,6 @@ int main() {
   context Context(Device);
   queue Queue(Device);
   event Event;
-  // expected-warning@+1 {{'program' is deprecated: program class is deprecated, use kernel_bundle instead}}
-  program Program(Context);
   kernel_bundle<bundle_state::executable> KernelBundle =
       get_kernel_bundle<bundle_state::executable>(Context);
   kernel Kernel = KernelBundle.get_kernel(get_kernel_ids().front());
@@ -55,9 +53,6 @@ int main() {
       ZeContext;
   backend_traits<backend::ext_oneapi_level_zero>::return_type<queue> ZeQueue;
   backend_traits<backend::ext_oneapi_level_zero>::return_type<event> ZeEvent;
-  // expected-warning@+1 {{'program' is deprecated: program class is deprecated, use kernel_bundle instead}}
-  backend_traits<backend::ext_oneapi_level_zero>::return_type<program>
-      ZeProgram;
   backend_traits<backend::ext_oneapi_level_zero>::return_type<
       kernel_bundle<bundle_state::executable>>
       ZeKernelBundle;
@@ -94,9 +89,6 @@ int main() {
   // 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@+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}}
-  ZeProgram = Program.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*/ (
@@ -165,9 +157,6 @@ int main() {
   // expected-warning@+1 {{'make<sycl::event, nullptr>' is deprecated: Use SYCL 2020 sycl::make_event free function}}
   auto E = ext::oneapi::level_zero::make<event>(
       Context, ZeEvent, ext::oneapi::level_zero::ownership::keep);
-  // expected-warning@+2 {{'program' is deprecated: program class is deprecated, use kernel_bundle instead}}
-  // expected-warning@+1 {{'make<sycl::program, nullptr>' is deprecated: Use SYCL 2020 sycl::make_kernel_bundle free function}}
-  auto PR = ext::oneapi::level_zero::make<program>(Context, ZeProgram);
 
   return 0;
 }
diff --git a/sycl/test/basic_tests/spec_const_types.cpp b/sycl/test/basic_tests/spec_const_types.cpp
index 8e43cd66feaee..494d04e3a3cf3 100644
--- a/sycl/test/basic_tests/spec_const_types.cpp
+++ b/sycl/test/basic_tests/spec_const_types.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl -fsycl-device-only -c -o %t.bc %s
+// RUN: %clangxx -fsycl -fsycl-device-only -D__SYCL_INTERNAL_API -c -o %t.bc %s
 // RUN: sycl-post-link %t.bc -spec-const=rt -o %t-split.txt
 // RUN: cat %t-split_0.prop | FileCheck %s
 // RUN: llvm-spirv -o %t-split_0.spv -spirv-max-version=1.1 -spirv-ext=+all %t-split_0.bc
diff --git a/sycl/test/warnings/spec_consts_ext_deprecation.cpp b/sycl/test/warnings/spec_consts_ext_deprecation.cpp
deleted file mode 100644
index f5ff97a04dfd9..0000000000000
--- a/sycl/test/warnings/spec_consts_ext_deprecation.cpp
+++ /dev/null
@@ -1,13 +0,0 @@
-// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out
-
-#include <CL/sycl.hpp>
-
-using namespace sycl;
-
-int main() {
-  // expected-warning@+1 {{Specialization constats extension is deprecated, use SYCL 2020 specialization constants instead}}
-  ext::oneapi::experimental::spec_constant<int> SC;
-  (void)SC;
-  return 0;
-}
-
diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp
index 81c4792149c33..34c9a9fbe40c3 100644
--- a/sycl/test/warnings/sycl_2020_deprecations.cpp
+++ b/sycl/test/warnings/sycl_2020_deprecations.cpp
@@ -61,7 +61,7 @@ int main() {
   // expected-warning@+1 {{'get' is deprecated: OpenCL interop get() functions are deprecated, use get_native() instead}}
   (void)Kernel.get();
 
-  // expected-warning@+1 {{'program' is deprecated: program class is deprecated, use kernel_bundle instead}}
+  // expected-error@+1 {{no type named 'program' in namespace 'sycl'}}
   sycl::program Prog{Ctx};
 
   sycl::buffer<int, 1> Buffer(4);
diff --git a/sycl/unittests/SYCL2020/GetNativeOpenCL.cpp b/sycl/unittests/SYCL2020/GetNativeOpenCL.cpp
index 77835923dbd0f..2076113ebb5b1 100644
--- a/sycl/unittests/SYCL2020/GetNativeOpenCL.cpp
+++ b/sycl/unittests/SYCL2020/GetNativeOpenCL.cpp
@@ -7,6 +7,7 @@
 //===----------------------------------------------------------------------===//
 
 #define SYCL2020_DISABLE_DEPRECATION_WARNINGS
+#define __SYCL_INTERNAL_API
 
 #include <CL/sycl.hpp>
 #include <CL/sycl/backend/opencl.hpp>
diff --git a/sycl/unittests/kernel-and-program/CMakeLists.txt b/sycl/unittests/kernel-and-program/CMakeLists.txt
index 6f8f6e5417653..0527bb1be0eef 100644
--- a/sycl/unittests/kernel-and-program/CMakeLists.txt
+++ b/sycl/unittests/kernel-and-program/CMakeLists.txt
@@ -5,3 +5,4 @@ add_sycl_unittest(KernelAndProgramTests OBJECT
   DeviceInfo.cpp
   PersistentDeviceCodeCache.cpp
 )
+target_compile_definitions(KernelAndProgramTests PRIVATE -D__SYCL_INTERNAL_API)
diff --git a/sycl/unittests/misc/KernelBuildOptions.cpp b/sycl/unittests/misc/KernelBuildOptions.cpp
index 90f3c3235aab6..7a83d6f689375 100644
--- a/sycl/unittests/misc/KernelBuildOptions.cpp
+++ b/sycl/unittests/misc/KernelBuildOptions.cpp
@@ -7,6 +7,9 @@
 //===---------------------------------------------------------------------===//
 
 #define SYCL2020_DISABLE_DEPRECATION_WARNINGS
+#ifndef __SYCL_INTERNAL_API
+#define __SYCL_INTERNAL_API
+#endif
 
 #include <CL/sycl.hpp>
 #include <helpers/PiImage.hpp>