diff --git a/sycl/CMakeLists.txt b/sycl/CMakeLists.txt
index 2fedec8470756..6c93f7dc42027 100644
--- a/sycl/CMakeLists.txt
+++ b/sycl/CMakeLists.txt
@@ -185,6 +185,8 @@ function( add_common_options LIB_NAME)
   endif()
 endfunction(add_common_options)
 
+set(SYCL_SOURCE_DIR ${CMAKE_CURRENT_SOURCE_DIR})
+
 # SYCL runtime library
 add_subdirectory( source )
 
diff --git a/sycl/include/CL/sycl/accessor.hpp b/sycl/include/CL/sycl/accessor.hpp
index ae211384f7778..74bb361bb96f2 100644
--- a/sycl/include/CL/sycl/accessor.hpp
+++ b/sycl/include/CL/sycl/accessor.hpp
@@ -16,7 +16,6 @@
 #include <CL/sycl/detail/generic_type_traits.hpp>
 #include <CL/sycl/detail/image_accessor_util.hpp>
 #include <CL/sycl/detail/image_ocl_types.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
 #include <CL/sycl/exception.hpp>
 #include <CL/sycl/handler.hpp>
 #include <CL/sycl/id.hpp>
@@ -167,6 +166,8 @@ static T<NewDim> convertToArrayOfN(T<OldDim> OldObj) {
   return NewObj;
 }
 
+device getDeviceFromHandler(handler &CommandGroupHandlerRef);
+
 template <typename DataT, int Dimensions, access::mode AccessMode,
           access::target AccessTarget, access::placeholder IsPlaceholder>
 class accessor_common {
@@ -397,10 +398,7 @@ class image_accessor
         MImageCount(ImageRef.get_count()),
         MImgChannelOrder(detail::getSyclObjImpl(ImageRef)->getChannelOrder()),
         MImgChannelType(detail::getSyclObjImpl(ImageRef)->getChannelType()) {
-    detail::EventImplPtr Event =
-        detail::Scheduler::getInstance().addHostAccessor(
-            AccessorBaseHost::impl.get());
-    Event->wait(Event);
+    addHostAccessorAndWait(AccessorBaseHost::impl.get());
   }
 #endif
 
@@ -429,7 +427,7 @@ class image_accessor
         MImgChannelOrder(detail::getSyclObjImpl(ImageRef)->getChannelOrder()),
         MImgChannelType(detail::getSyclObjImpl(ImageRef)->getChannelType()) {
     checkDeviceFeatureSupported<info::device::image_support>(
-        CommandGroupHandlerRef.MQueue->get_device());
+        getDeviceFromHandler(CommandGroupHandlerRef));
   }
 #endif
 
@@ -770,12 +768,8 @@ class accessor :
             detail::convertToArrayOfN<3, 1>(BufferRef.get_range()), AccessMode,
             detail::getSyclObjImpl(BufferRef).get(), AdjustedDim, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    if (!IsPlaceH) {
-      detail::EventImplPtr Event =
-          detail::Scheduler::getInstance().addHostAccessor(
-              AccessorBaseHost::impl.get());
-      Event->wait(Event);
-    }
+    if (!IsPlaceH)
+      addHostAccessorAndWait(AccessorBaseHost::impl.get());
 #endif
   }
 
@@ -814,12 +808,8 @@ class accessor :
             detail::convertToArrayOfN<3, 1>(BufferRef.get_range()), AccessMode,
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    if (!IsPlaceH) {
-      detail::EventImplPtr Event =
-          detail::Scheduler::getInstance().addHostAccessor(
-              AccessorBaseHost::impl.get());
-      Event->wait(Event);
-    }
+    if (!IsPlaceH)
+      addHostAccessorAndWait(AccessorBaseHost::impl.get());
   }
 #endif
 
@@ -858,12 +848,8 @@ class accessor :
                          AccessMode, detail::getSyclObjImpl(BufferRef).get(),
                          Dimensions, sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer) {
-    if (!IsPlaceH) {
-      detail::EventImplPtr Event =
-          detail::Scheduler::getInstance().addHostAccessor(
-              AccessorBaseHost::impl.get());
-      Event->wait(Event);
-    }
+    if (!IsPlaceH)
+      addHostAccessorAndWait(AccessorBaseHost::impl.get());
   }
 #endif
 
diff --git a/sycl/include/CL/sycl/buffer.hpp b/sycl/include/CL/sycl/buffer.hpp
index 8181f2628a173..b894bfeda9961 100644
--- a/sycl/include/CL/sycl/buffer.hpp
+++ b/sycl/include/CL/sycl/buffer.hpp
@@ -194,11 +194,8 @@ class buffer {
          event AvailableEvent = {})
       : Range{0} {
 
-    size_t BufSize = 0;
-    const detail::plugin &Plugin = detail::getSyclObjImpl(SyclContext)->getPlugin();
-    Plugin.call<detail::PiApiKind::piMemGetInfo>(
-        detail::pi::cast<detail::RT::PiMem>(MemObject), CL_MEM_SIZE,
-        sizeof(size_t), &BufSize, nullptr);
+    size_t BufSize = detail::SYCLMemObjT::getBufSizeForContext(
+        detail::getSyclObjImpl(SyclContext), MemObject);
 
     Range[0] = BufSize / sizeof(T);
     impl = std::make_shared<detail::buffer_impl>(
diff --git a/sycl/include/CL/sycl/detail/accessor_impl.hpp b/sycl/include/CL/sycl/detail/accessor_impl.hpp
index 23985145fdc74..83bb2fadbed88 100644
--- a/sycl/include/CL/sycl/detail/accessor_impl.hpp
+++ b/sycl/include/CL/sycl/detail/accessor_impl.hpp
@@ -185,6 +185,8 @@ class LocalAccessorBaseHost {
 
 using Requirement = AccessorImplHost;
 
+void addHostAccessorAndWait(Requirement *Req);
+
 } // namespace detail
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/include/CL/sycl/detail/image_impl.hpp b/sycl/include/CL/sycl/detail/image_impl.hpp
index 653ddcba79caf..d6a28ac26063f 100644
--- a/sycl/include/CL/sycl/detail/image_impl.hpp
+++ b/sycl/include/CL/sycl/detail/image_impl.hpp
@@ -219,15 +219,6 @@ template <int Dimensions> class image_impl final : public SYCLMemObjT {
   ~image_impl() { BaseT::updateHostMemory(); }
 
 private:
-  template <typename T>
-  void getImageInfo(const ContextImplPtr Context, RT::PiMemImageInfo Info,
-                    T &Dest) {
-    const detail::plugin &Plugin = Context->getPlugin();
-    RT::PiMem Mem = pi::cast<RT::PiMem>(BaseT::MInteropMemObject);
-    Plugin.call<PiApiKind::piMemImageGetInfo>(Mem, Info, sizeof(T), &Dest,
-                                              nullptr);
-  }
-
   vector_class<device> getDevices(const ContextImplPtr Context);
 
   template <info::device Param>
diff --git a/sycl/include/CL/sycl/detail/sycl_mem_obj_i.hpp b/sycl/include/CL/sycl/detail/sycl_mem_obj_i.hpp
index 1e1d4a0a6f059..5d9eda8c3edd5 100644
--- a/sycl/include/CL/sycl/detail/sycl_mem_obj_i.hpp
+++ b/sycl/include/CL/sycl/detail/sycl_mem_obj_i.hpp
@@ -9,7 +9,7 @@
 #pragma once
 
 #include <CL/sycl/detail/pi.hpp>
-#include <memory>
+#include <CL/sycl/stl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -20,8 +20,8 @@ class event_impl;
 class context_impl;
 struct MemObjRecord;
 
-using EventImplPtr = std::shared_ptr<detail::event_impl>;
-using ContextImplPtr = std::shared_ptr<detail::context_impl>;
+using EventImplPtr = shared_ptr_class<detail::event_impl>;
+using ContextImplPtr = shared_ptr_class<detail::context_impl>;
 
 // The class serves as an interface in the scheduler for all SYCL memory
 // objects.
@@ -64,7 +64,10 @@ class SYCLMemObjI {
 protected:
   // Pointer to the record that contains the memory commands. This is managed
   // by the scheduler.
-  std::unique_ptr<MemObjRecord> MRecord;
+  // fixme replace with unique_ptr_class once it is implemented. Standard
+  // unique_ptr requires knowlege of sizeof(MemObjRecord) at compile time
+  // which is unavailable.
+  shared_ptr_class<MemObjRecord> MRecord;
   friend class Scheduler;
 };
 
diff --git a/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp b/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp
index 6337f34eeeedc..940210a7a76cd 100644
--- a/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp
+++ b/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp
@@ -9,9 +9,7 @@
 #pragma once
 
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
 #include <CL/sycl/detail/sycl_mem_obj_allocator.hpp>
-#include <CL/sycl/detail/plugin.hpp>
 #include <CL/sycl/detail/sycl_mem_obj_i.hpp>
 #include <CL/sycl/detail/type_traits.hpp>
 #include <CL/sycl/event.hpp>
@@ -26,7 +24,9 @@ namespace sycl {
 namespace detail {
 
 // Forward declarations
+class context_impl;
 class event_impl;
+class plugin;
 
 using ContextImplPtr = shared_ptr_class<context_impl>;
 using EventImplPtr = shared_ptr_class<event_impl>;
@@ -83,11 +83,8 @@ class SYCLMemObjT : public SYCLMemObjI {
 
   virtual ~SYCLMemObjT() = default;
 
-  const plugin &getPlugin() const {
-    assert((MInteropContext != nullptr) &&
-           "Trying to get Plugin from SYCLMemObjT with nullptr ContextImpl.");
-    return (MInteropContext->getPlugin());
-  }
+  const plugin &getPlugin() const;
+
   size_t getSize() const override { return MSizeInBytes; }
   size_t get_count() const {
     size_t AllocatorValueSize = MAllocator->getValueSize();
@@ -259,6 +256,9 @@ class SYCLMemObjT : public SYCLMemObjI {
     MAllocator->setAlignment(RequiredAlign);
   }
 
+  static size_t getBufSizeForContext(const ContextImplPtr &Context,
+                                     cl_mem MemObject);
+
 protected:
   // Allocator used for allocation memory on host.
   unique_ptr_class<SYCLMemObjAllocator> MAllocator;
diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp
index 8dbc3fbe5c1e3..845645e7b594a 100644
--- a/sycl/include/CL/sycl/handler.hpp
+++ b/sycl/include/CL/sycl/handler.hpp
@@ -59,6 +59,7 @@ namespace sycl {
 
 // Forward declaration
 
+class handler;
 template <typename T, int Dimensions, typename AllocatorT> class buffer;
 namespace detail {
 
@@ -104,6 +105,7 @@ template <typename Type> struct get_kernel_name_t<detail::auto_name, Type> {
   using name = Type;
 };
 
+device getDeviceFromHandler(handler &);
 } // namespace detail
 
 /// 4.8.3 Command group handler class
@@ -1278,6 +1280,7 @@ class handler {
   template <typename DataT, int Dims, access::mode AccMode,
             access::target AccTarget, access::placeholder isPlaceholder>
   friend class accessor;
+  friend device detail::getDeviceFromHandler(handler &);
 
   template <typename DataT, int Dimensions, access::mode AccessMode,
             access::target AccessTarget, access::placeholder IsPlaceholder>
diff --git a/sycl/include/CL/sycl/intel/function_pointer.hpp b/sycl/include/CL/sycl/intel/function_pointer.hpp
index b3faf3f282c16..3a59bb027aaac 100644
--- a/sycl/include/CL/sycl/intel/function_pointer.hpp
+++ b/sycl/include/CL/sycl/intel/function_pointer.hpp
@@ -8,7 +8,6 @@
 
 #pragma once
 
-#include <CL/sycl/detail/plugin.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/program.hpp>
 #include <CL/sycl/stl.hpp>
diff --git a/sycl/include/CL/sycl/platform.hpp b/sycl/include/CL/sycl/platform.hpp
index 2bfc7bbc2cd5d..4e7088f19f6f2 100644
--- a/sycl/include/CL/sycl/platform.hpp
+++ b/sycl/include/CL/sycl/platform.hpp
@@ -8,7 +8,6 @@
 
 #pragma once
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/platform_info.hpp>
 #include <CL/sycl/stl.hpp>
 
 // 4.6.2 Platform class
diff --git a/sycl/include/CL/sycl/queue.hpp b/sycl/include/CL/sycl/queue.hpp
index 30a26ccfb30ea..bb811decac723 100644
--- a/sycl/include/CL/sycl/queue.hpp
+++ b/sycl/include/CL/sycl/queue.hpp
@@ -9,9 +9,11 @@
 #pragma once
 
 #include <CL/sycl/detail/common.hpp>
+#include <CL/sycl/device.hpp>
 #include <CL/sycl/device_selector.hpp>
 #include <CL/sycl/event.hpp>
 #include <CL/sycl/exception_list.hpp>
+#include <CL/sycl/handler.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/property_list.hpp>
 #include <CL/sycl/stl.hpp>
@@ -107,10 +109,7 @@ class queue {
   /// @param SyclDevice is an instance of SYCL device.
   /// @param PropList is a list of properties for queue construction.
   queue(const context &SyclContext, const device &SyclDevice,
-        const property_list &PropList = {})
-      : queue(SyclContext, SyclDevice,
-              detail::getSyclObjImpl(SyclContext)->get_async_handler(),
-              PropList) {};
+        const property_list &PropList = {});
 
   /// Constructs a SYCL queue associated with the given context, device,
   /// asynchronous exception handler and optional properties list.
@@ -447,9 +446,7 @@ class queue {
   /// Returns whether the queue is in order or OoO
   ///
   /// Equivalent to has_property<property::queue::in_order>()
-  bool is_in_order() const {
-    return impl->has_property<property::queue::in_order>();
-  }
+  bool is_in_order() const;
 
 private:
   shared_ptr_class<detail::queue_impl> impl;
diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt
index a41e02136d885..5327ff140c677 100644
--- a/sycl/source/CMakeLists.txt
+++ b/sycl/source/CMakeLists.txt
@@ -82,6 +82,7 @@ set(SYCL_SOURCES
     "detail/usm/usm_dispatch.cpp"
     "detail/usm/usm_impl.cpp"
     "detail/util.cpp"
+    "accessor.cpp"
     "context.cpp"
     "device.cpp"
     "device_selector.cpp"
diff --git a/sycl/source/accessor.cpp b/sycl/source/accessor.cpp
new file mode 100644
index 0000000000000..7731615ea2b74
--- /dev/null
+++ b/sycl/source/accessor.cpp
@@ -0,0 +1,20 @@
+//==------------ accessor.cpp - SYCL standard source file ------------------==//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <CL/sycl/accessor.hpp>
+#include <detail/queue_impl.hpp>
+
+__SYCL_INLINE_NAMESPACE(cl) {
+namespace sycl {
+namespace detail {
+device getDeviceFromHandler(handler &CommandGroupHandlerRef) {
+  return CommandGroupHandlerRef.MQueue->get_device();
+}
+} // namespace detail
+} // namespace sycl
+} // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/source/context.cpp b/sycl/source/context.cpp
index f8b502f81cae3..c631a70336ec8 100644
--- a/sycl/source/context.cpp
+++ b/sycl/source/context.cpp
@@ -8,13 +8,14 @@
 
 #include <CL/sycl/context.hpp>
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/device_selector.hpp>
 #include <CL/sycl/exception.hpp>
 #include <CL/sycl/exception_list.hpp>
 #include <CL/sycl/platform.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/context_impl.hpp>
+
 #include <algorithm>
 #include <memory>
 #include <utility>
diff --git a/sycl/source/detail/accessor_impl.cpp b/sycl/source/detail/accessor_impl.cpp
index 4a119f4d78ab7..bb8931392cf7e 100644
--- a/sycl/source/detail/accessor_impl.cpp
+++ b/sycl/source/detail/accessor_impl.cpp
@@ -7,7 +7,8 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/accessor_impl.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -17,6 +18,12 @@ AccessorImplHost::~AccessorImplHost() {
   if (MBlockedCmd)
     detail::Scheduler::getInstance().releaseHostAccessor(this);
 }
+
+void addHostAccessorAndWait(Requirement *Req) {
+  detail::EventImplPtr Event =
+      detail::Scheduler::getInstance().addHostAccessor(Req);
+  Event->wait(Event);
+}
 }
 }
 }
diff --git a/sycl/source/detail/buffer_impl.cpp b/sycl/source/detail/buffer_impl.cpp
index 9117eddd64af1..a9aa0e4194f9e 100644
--- a/sycl/source/detail/buffer_impl.cpp
+++ b/sycl/source/detail/buffer_impl.cpp
@@ -7,9 +7,9 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/buffer_impl.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
 #include <CL/sycl/detail/memory_manager.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/detail/config.hpp b/sycl/source/detail/config.hpp
index 903b95916dc15..90ec14f8582d8 100644
--- a/sycl/source/detail/config.hpp
+++ b/sycl/source/detail/config.hpp
@@ -8,6 +8,8 @@
 
 #pragma once
 
+#include <CL/sycl/detail/defines.hpp>
+
 #include <cstdlib>
 
 __SYCL_INLINE_NAMESPACE(cl) {
diff --git a/sycl/source/detail/context_impl.cpp b/sycl/source/detail/context_impl.cpp
index 70deee347d11a..9d4c6bf368bf1 100644
--- a/sycl/source/detail/context_impl.cpp
+++ b/sycl/source/detail/context_impl.cpp
@@ -8,14 +8,14 @@
 
 #include <CL/sycl/detail/clusm.hpp>
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/context_info.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/exception.hpp>
 #include <CL/sycl/exception_list.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/platform.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/context_info.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/include/CL/sycl/detail/context_impl.hpp b/sycl/source/detail/context_impl.hpp
similarity index 91%
rename from sycl/include/CL/sycl/detail/context_impl.hpp
rename to sycl/source/detail/context_impl.hpp
index 1cceed1837a16..631cc5061e88a 100644
--- a/sycl/include/CL/sycl/detail/context_impl.hpp
+++ b/sycl/source/detail/context_impl.hpp
@@ -8,16 +8,16 @@
 
 #pragma once
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/device_impl.hpp>
-#include <CL/sycl/detail/kernel_program_cache.hpp>
 #include <CL/sycl/detail/os_util.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/platform_impl.hpp>
-#include <CL/sycl/detail/program_manager/program_manager.hpp>
-#include <CL/sycl/detail/usm_dispatch.hpp>
 #include <CL/sycl/exception_list.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/kernel_program_cache.hpp>
+#include <detail/platform_impl.hpp>
+#include <detail/program_manager/program_manager.hpp>
+#include <detail/usm/usm_dispatch.hpp>
 
 #include <map>
 #include <memory>
@@ -32,8 +32,8 @@ class context_impl {
 public:
   /// Constructs a context_impl using a single SYCL devices.
   ///
-  /// The constructed context_impl will use the AsyncHandler parameter to handle
-  /// exceptions.
+  /// The constructed context_impl will use the AsyncHandler parameter to
+  /// handle exceptions.
   ///
   /// @param Device is an instance of SYCL device.
   /// @param AsyncHandler is an instance of async_handler.
@@ -44,8 +44,8 @@ class context_impl {
   /// Newly created instance will save each SYCL device in the list. This
   /// requres that all devices in the list are associated with the same
   /// SYCL platform.
-  /// The constructed context_impl will use the AsyncHandler parameter to handle
-  /// exceptions.
+  /// The constructed context_impl will use the AsyncHandler parameter to
+  /// handle exceptions.
   ///
   /// @param DeviceList is a list of SYCL device instances.
   /// @param AsyncHandler is an instance of async_handler.
@@ -54,13 +54,13 @@ class context_impl {
 
   /// Construct a context_impl using plug-in interoperability handle.
   ///
-  /// The constructed context_impl will use the AsyncHandler parameter to handle
-  /// exceptions.
+  /// The constructed context_impl will use the AsyncHandler parameter to
+  /// handle exceptions.
   ///
   /// @param PiContext is an instance of a valid plug-in context handle.
   /// @param AsyncHandler is an instance of async_handler.
-  /// @param &Plugin is the reference to the underlying Plugin that this context
-  /// is associated with.
+  /// @param &Plugin is the reference to the underlying Plugin that this
+  /// context is associated with.
   context_impl(RT::PiContext PiContext, async_handler AsyncHandler,
                const plugin &Plugin);
 
@@ -116,9 +116,7 @@ class context_impl {
 
   /// Unlike `get_info<info::context::devices>', this function returns a
   /// reference.
-  const vector_class<device> &getDevices() const {
-    return MDevices;
-  }
+  const vector_class<device> &getDevices() const { return MDevices; }
 
   /// In contrast to user programs, which are compiled from user code, library
   /// programs come from the SYCL runtime. They are identified by the
diff --git a/sycl/include/CL/sycl/detail/context_info.hpp b/sycl/source/detail/context_info.hpp
similarity index 96%
rename from sycl/include/CL/sycl/detail/context_info.hpp
rename to sycl/source/detail/context_info.hpp
index 032b36762cd5e..cead0d2e09c2f 100644
--- a/sycl/include/CL/sycl/detail/context_info.hpp
+++ b/sycl/source/detail/context_info.hpp
@@ -9,8 +9,8 @@
 #pragma once
 
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/context_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/detail/device_impl.cpp b/sycl/source/detail/device_impl.cpp
index 1a4c4980623bc..7198592acebb5 100644
--- a/sycl/source/detail/device_impl.cpp
+++ b/sycl/source/detail/device_impl.cpp
@@ -6,8 +6,8 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/device_impl.hpp>
 #include <CL/sycl/device.hpp>
+#include <detail/device_impl.hpp>
 
 #include <algorithm>
 
diff --git a/sycl/include/CL/sycl/detail/device_impl.hpp b/sycl/source/detail/device_impl.hpp
similarity index 88%
rename from sycl/include/CL/sycl/detail/device_impl.hpp
rename to sycl/source/detail/device_impl.hpp
index 30f3006b8790d..ad9aa993a5228 100644
--- a/sycl/include/CL/sycl/detail/device_impl.hpp
+++ b/sycl/source/detail/device_impl.hpp
@@ -8,9 +8,11 @@
 
 #pragma once
 
-#include <CL/sycl/detail/device_info.hpp>
 #include <CL/sycl/detail/pi.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/device_info.hpp>
+#include <detail/platform_impl.hpp>
+
 #include <memory>
 
 __SYCL_INLINE_NAMESPACE(cl) {
@@ -43,8 +45,8 @@ class device_impl {
 
   /// Get instance of OpenCL device
   ///
-  /// @return a valid cl_device_id instance in accordance with the requirements
-  /// described in 4.3.1.
+  /// @return a valid cl_device_id instance in accordance with the
+  /// requirements described in 4.3.1.
   cl_device_id get() const;
 
   /// Get reference to PI device
@@ -120,18 +122,21 @@ class device_impl {
 
   /// Partition device into sub devices
   ///
-  /// If this SYCL device does not support info::partition_property::partition_equally
-  /// a feature_not_supported exception must be thrown.
+  /// If this SYCL device does not support
+  /// info::partition_property::partition_equally a feature_not_supported
+  /// exception must be thrown.
   ///
-  /// @param ComputeUnits is a desired count of compute units in each sub device.
+  /// @param ComputeUnits is a desired count of compute units in each sub
+  /// device.
   /// @return A vector class of sub devices partitioned equally from this
   /// SYCL device based on the ComputeUnits parameter.
   vector_class<device> create_sub_devices(size_t ComputeUnits) const;
 
   /// Partition device into sub devices
   ///
-  /// If this SYCL device does not support info::partition_property::partition_by_counts
-  /// a feature_not_supported exception must be thrown.
+  /// If this SYCL device does not support
+  /// info::partition_property::partition_by_counts a feature_not_supported
+  /// exception must be thrown.
   ///
   /// @param Counts is a vector_class of desired compute units in sub devices.
   /// @return a vector_class of sub devices partitioned from this SYCL device
@@ -141,11 +146,13 @@ class device_impl {
 
   /// Partition device into sub devices
   ///
-  /// If this SYCL device does not support info::partition_property::partition_by_affinity_domain
-  /// or the SYCL device does not support info::affinity_domain provided
-  /// a feature_not_supported exception must be thrown.
+  /// If this SYCL device does not support
+  /// info::partition_property::partition_by_affinity_domain or the SYCL
+  /// device does not support info::affinity_domain provided a
+  /// feature_not_supported exception must be thrown.
   ///
-  /// @param AffinityDomain is one of the values described in Table 4.20 of SYCL Spec
+  /// @param AffinityDomain is one of the values described in Table 4.20 of
+  /// SYCL Spec
   /// @return a vector class of sub devices partitioned from this SYCL device
   /// by affinity domain based on the AffinityDomain parameter
   vector_class<device>
@@ -158,7 +165,8 @@ class device_impl {
   /// @return true if Prop is supported by device.
   bool is_partition_supported(info::partition_property Prop) const;
 
-  /// Queries this SYCL device for information requested by the template parameter param
+  /// Queries this SYCL device for information requested by the template
+  /// parameter param
   ///
   /// Specializations of info::param_traits must be defined in accordance
   /// with the info parameters in Table 4.20 of SYCL Spec to facilitate
@@ -176,9 +184,11 @@ class device_impl {
         param>::get(this->getHandleRef(), this->getPlugin());
   }
 
-  /// Check if affinity partitioning by specified domain is supported by device
+  /// Check if affinity partitioning by specified domain is supported by
+  /// device
   ///
-  /// @param AffinityDomain is one of the values described in Table 4.20 of SYCL Spec
+  /// @param AffinityDomain is one of the values described in Table 4.20 of
+  /// SYCL Spec
   /// @return true if AffinityDomain is supported by device.
   bool
   is_affinity_supported(info::partition_affinity_domain AffinityDomain) const;
diff --git a/sycl/source/detail/device_info.cpp b/sycl/source/detail/device_info.cpp
index 948b402585e78..1cdfee599fcc1 100644
--- a/sycl/source/detail/device_info.cpp
+++ b/sycl/source/detail/device_info.cpp
@@ -6,11 +6,12 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/device_impl.hpp>
-#include <CL/sycl/detail/device_info.hpp>
 #include <CL/sycl/detail/os_util.hpp>
-#include <CL/sycl/detail/platform_util.hpp>
 #include <CL/sycl/device.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/device_info.hpp>
+#include <detail/platform_util.hpp>
+
 #include <chrono>
 #include <thread>
 
diff --git a/sycl/include/CL/sycl/detail/device_info.hpp b/sycl/source/detail/device_info.hpp
similarity index 95%
rename from sycl/include/CL/sycl/detail/device_info.hpp
rename to sycl/source/detail/device_info.hpp
index 574a81fba69e1..06676fad25442 100644
--- a/sycl/include/CL/sycl/detail/device_info.hpp
+++ b/sycl/source/detail/device_info.hpp
@@ -10,9 +10,9 @@
 #include <CL/sycl/detail/common_info.hpp>
 #include <CL/sycl/detail/defines.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/platform_impl.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/platform.hpp>
+#include <detail/plugin.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -28,8 +28,8 @@ read_execution_bitfield(cl_device_exec_capabilities bits);
 
 // Mapping expected SYCL return types to those returned by PI calls
 template <typename T> struct sycl_to_pi { using type = T; };
-template <> struct sycl_to_pi<bool>     { using type = pi_bool; };
-template <> struct sycl_to_pi<device>   { using type = RT::PiDevice; };
+template <> struct sycl_to_pi<bool> { using type = pi_bool; };
+template <> struct sycl_to_pi<device> { using type = RT::PiDevice; };
 template <> struct sycl_to_pi<platform> { using type = RT::PiPlatform; };
 
 // Mapping fp_config device info types to the values used to check fp support
@@ -155,8 +155,8 @@ template <> struct get_device_info<bool, info::device::queue_profiling> {
 template <>
 struct get_device_info<vector_class<info::execution_capability>,
                        info::device::execution_capabilities> {
-  static vector_class<info::execution_capability>
-  get(RT::PiDevice dev, const plugin &Plugin) {
+  static vector_class<info::execution_capability> get(RT::PiDevice dev,
+                                                      const plugin &Plugin) {
     cl_device_exec_capabilities result;
     Plugin.call<PiApiKind::piDeviceGetInfo>(
         dev, pi::cast<RT::PiDeviceInfo>(info::device::execution_capabilities),
@@ -273,8 +273,7 @@ struct get_device_info<info::partition_affinity_domain,
 template <>
 struct get_device_info<info::partition_property,
                        info::device::partition_type_property> {
-  static info::partition_property get(RT::PiDevice dev,
-                                      const plugin &Plugin) {
+  static info::partition_property get(RT::PiDevice dev, const plugin &Plugin) {
     size_t resultSize;
     Plugin.call<PiApiKind::piDeviceGetInfo>(dev, PI_DEVICE_INFO_PARTITION_TYPE,
                                             0, nullptr, &resultSize);
@@ -342,7 +341,7 @@ typename info::param_traits<info::device, param>::return_type
 get_device_info_host() = delete;
 
 #define PARAM_TRAITS_SPEC(param_type, param, ret_type)                         \
-      template <> ret_type get_device_info_host<info::param_type::param>();
+  template <> ret_type get_device_info_host<info::param_type::param>();
 
 #include <CL/sycl/info/device_traits.def>
 
@@ -353,8 +352,7 @@ cl_uint get_native_vector_width(size_t idx);
 // USM
 
 // Specialization for device usm query.
-template <>
-struct get_device_info<bool, info::device::usm_device_allocations> {
+template <> struct get_device_info<bool, info::device::usm_device_allocations> {
   static bool get(RT::PiDevice dev, const plugin &Plugin) {
     pi_usm_capabilities caps;
     pi_result Err = Plugin.call_nocheck<PiApiKind::piDeviceGetInfo>(
@@ -366,8 +364,7 @@ struct get_device_info<bool, info::device::usm_device_allocations> {
 };
 
 // Specialization for host usm query.
-template <>
-struct get_device_info<bool, info::device::usm_host_allocations> {
+template <> struct get_device_info<bool, info::device::usm_host_allocations> {
   static bool get(RT::PiDevice dev, const plugin &Plugin) {
     pi_usm_capabilities caps;
     pi_result Err = Plugin.call_nocheck<PiApiKind::piDeviceGetInfo>(
@@ -379,8 +376,7 @@ struct get_device_info<bool, info::device::usm_host_allocations> {
 };
 
 // Specialization for shared usm query.
-template <>
-struct get_device_info<bool, info::device::usm_shared_allocations> {
+template <> struct get_device_info<bool, info::device::usm_shared_allocations> {
   static bool get(RT::PiDevice dev, const plugin &Plugin) {
     pi_usm_capabilities caps;
     pi_result Err = Plugin.call_nocheck<PiApiKind::piDeviceGetInfo>(
@@ -408,8 +404,7 @@ struct get_device_info<bool, info::device::usm_restricted_shared_allocations> {
 };
 
 // Specialization for system usm query
-template <>
-struct get_device_info<bool, info::device::usm_system_allocator> {
+template <> struct get_device_info<bool, info::device::usm_system_allocator> {
   static bool get(RT::PiDevice dev, const plugin &Plugin) {
     pi_usm_capabilities caps;
     pi_result Err = Plugin.call_nocheck<PiApiKind::piDeviceGetInfo>(
diff --git a/sycl/source/detail/error_handling/enqueue_kernel.cpp b/sycl/source/detail/error_handling/enqueue_kernel.cpp
index 729770f373124..7b954f114740f 100644
--- a/sycl/source/detail/error_handling/enqueue_kernel.cpp
+++ b/sycl/source/detail/error_handling/enqueue_kernel.cpp
@@ -13,7 +13,7 @@
 #include "error_handling.hpp"
 
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/plugin.hpp>
+#include <detail/plugin.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/detail/error_handling/error_handling.hpp b/sycl/source/detail/error_handling/error_handling.hpp
index 29c2df1979cba..06bfe4cec173c 100644
--- a/sycl/source/detail/error_handling/error_handling.hpp
+++ b/sycl/source/detail/error_handling/error_handling.hpp
@@ -9,8 +9,8 @@
 #pragma once
 
 #include <CL/sycl/detail/cg.hpp>
-#include <CL/sycl/detail/device_impl.hpp>
 #include <CL/sycl/detail/pi.h>
+#include <detail/device_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/detail/event_impl.cpp b/sycl/source/detail/event_impl.cpp
index 7d18dea1d4c67..a7b3b467b8988 100644
--- a/sycl/source/detail/event_impl.cpp
+++ b/sycl/source/detail/event_impl.cpp
@@ -7,11 +7,11 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/context.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
-#include <CL/sycl/detail/event_info.hpp>
-#include <CL/sycl/detail/plugin.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/event_info.hpp>
+#include <detail/plugin.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include "detail/config.hpp"
 
diff --git a/sycl/include/CL/sycl/detail/event_impl.hpp b/sycl/source/detail/event_impl.hpp
similarity index 90%
rename from sycl/include/CL/sycl/detail/event_impl.hpp
rename to sycl/source/detail/event_impl.hpp
index 7b99145c82a24..b9c8d4fa17de5 100644
--- a/sycl/include/CL/sycl/detail/event_impl.hpp
+++ b/sycl/source/detail/event_impl.hpp
@@ -9,10 +9,10 @@
 #pragma once
 
 #include <CL/sycl/detail/common.hpp>
+#include <CL/sycl/detail/host_profiling_info.hpp>
 #include <CL/sycl/detail/pi.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/stl.hpp>
-#include <CL/sycl/detail/host_profiling_info.hpp>
 
 #include <cassert>
 
@@ -34,7 +34,8 @@ class event_impl {
   event_impl() = default;
   /// Constructs an event instance from a plug-in event handle.
   ///
-  /// The SyclContext must match the plug-in context associated with the ClEvent.
+  /// The SyclContext must match the plug-in context associated with the
+  /// ClEvent.
   ///
   /// @param Event is a valid instance of plug-in event.
   /// @param SyclContext is an instance of SYCL context.
@@ -63,21 +64,21 @@ class event_impl {
 
   /// Waits for the event.
   ///
-  /// If any uncaught asynchronous errors occurred on the context that the event
-  /// is waiting on executions from, then call that context's asynchronous error
-  /// handler with those errors.
-  /// Self is needed in order to pass shared_ptr to Scheduler.
+  /// If any uncaught asynchronous errors occurred on the context that the
+  /// event is waiting on executions from, then call that context's
+  /// asynchronous error handler with those errors. Self is needed in order to
+  /// pass shared_ptr to Scheduler.
   ///
   /// @param Self is a pointer to this event.
   void wait_and_throw(std::shared_ptr<cl::sycl::detail::event_impl> Self);
 
   /// Queries this event for profiling information.
   ///
-  /// If the requested info is not available when this member function is called
-  /// due to incompletion of command groups associated with the event, then the
-  /// call to this member function will block until the requested info is
-  /// available. If the queue which submitted the command group this event is
-  /// associated with was not constructed with the
+  /// If the requested info is not available when this member function is
+  /// called due to incompletion of command groups associated with the event,
+  /// then the call to this member function will block until the requested
+  /// info is available. If the queue which submitted the command group this
+  /// event is associated with was not constructed with the
   /// property::queue::enable_profiling property, an invalid_object_error SYCL
   /// exception is thrown.
   ///
diff --git a/sycl/include/CL/sycl/detail/event_info.hpp b/sycl/source/detail/event_info.hpp
similarity index 95%
rename from sycl/include/CL/sycl/detail/event_info.hpp
rename to sycl/source/detail/event_info.hpp
index 8963d849b1144..e2ba912cd6d70 100644
--- a/sycl/include/CL/sycl/detail/event_info.hpp
+++ b/sycl/source/detail/event_info.hpp
@@ -9,9 +9,9 @@
 #pragma once
 
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
-#include <CL/sycl/detail/plugin.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/plugin.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/detail/helpers.cpp b/sycl/source/detail/helpers.cpp
index 8dcd3871787cf..73ed626761904 100644
--- a/sycl/source/detail/helpers.cpp
+++ b/sycl/source/detail/helpers.cpp
@@ -8,9 +8,9 @@
 
 #include <CL/sycl/detail/helpers.hpp>
 
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
 #include <CL/sycl/event.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/event_impl.hpp>
 
 #include <memory>
 
diff --git a/sycl/source/detail/image_impl.cpp b/sycl/source/detail/image_impl.cpp
index 3c4fc11917f6f..8102fec48ee3a 100644
--- a/sycl/source/detail/image_impl.cpp
+++ b/sycl/source/detail/image_impl.cpp
@@ -6,11 +6,10 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/context_impl.hpp>
 #include <CL/sycl/detail/image_impl.hpp>
 #include <CL/sycl/detail/memory_manager.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/image.hpp>
+#include <detail/context_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -232,6 +231,15 @@ image_channel_type convertChannelType(RT::PiMemImageChannelType Type) {
   }
 }
 
+template <typename T>
+static void getImageInfo(const ContextImplPtr Context, RT::PiMemImageInfo Info,
+                         T &Dest, RT::PiMem InteropMemObject) {
+  const detail::plugin &Plugin = Context->getPlugin();
+  RT::PiMem Mem = pi::cast<RT::PiMem>(InteropMemObject);
+  Plugin.call<PiApiKind::piMemImageGetInfo>(Mem, Info, sizeof(T), &Dest,
+                                            nullptr);
+}
+
 template <int Dimensions>
 image_impl<Dimensions>::image_impl(
     cl_mem MemObject, const context &SyclContext, event AvailableEvent,
@@ -246,26 +254,26 @@ image_impl<Dimensions>::image_impl(
                                        &(BaseT::MSizeInBytes), nullptr);
 
   RT::PiMemImageFormat Format;
-  getImageInfo(Context, PI_IMAGE_INFO_FORMAT, Format);
+  getImageInfo(Context, PI_IMAGE_INFO_FORMAT, Format, Mem);
   MOrder = detail::convertChannelOrder(Format.image_channel_order);
   MType = detail::convertChannelType(Format.image_channel_data_type);
   MNumChannels = getImageNumberChannels(MOrder);
 
-  getImageInfo(Context, PI_IMAGE_INFO_ELEMENT_SIZE, MElementSize);
+  getImageInfo(Context, PI_IMAGE_INFO_ELEMENT_SIZE, MElementSize, Mem);
   assert(getImageElementSize(MNumChannels, MType) == MElementSize);
 
-  getImageInfo(Context, PI_IMAGE_INFO_ROW_PITCH, MRowPitch);
-  getImageInfo(Context, PI_IMAGE_INFO_SLICE_PITCH, MSlicePitch);
+  getImageInfo(Context, PI_IMAGE_INFO_ROW_PITCH, MRowPitch, Mem);
+  getImageInfo(Context, PI_IMAGE_INFO_SLICE_PITCH, MSlicePitch, Mem);
 
   switch (Dimensions) {
   case 3:
-    getImageInfo(Context, PI_IMAGE_INFO_DEPTH, MRange[2]);
+    getImageInfo(Context, PI_IMAGE_INFO_DEPTH, MRange[2], Mem);
     // fall through
   case 2:
-    getImageInfo(Context, PI_IMAGE_INFO_HEIGHT, MRange[1]);
+    getImageInfo(Context, PI_IMAGE_INFO_HEIGHT, MRange[1], Mem);
     // fall through
   case 1:
-    getImageInfo(Context, PI_IMAGE_INFO_WIDTH, MRange[0]);
+    getImageInfo(Context, PI_IMAGE_INFO_WIDTH, MRange[0], Mem);
   }
 }
 
diff --git a/sycl/source/detail/kernel_impl.cpp b/sycl/source/detail/kernel_impl.cpp
index 2b50fccc67a09..948f772e6da96 100644
--- a/sycl/source/detail/kernel_impl.cpp
+++ b/sycl/source/detail/kernel_impl.cpp
@@ -7,11 +7,11 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/kernel_impl.hpp>
-#include <CL/sycl/detail/kernel_info.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/program.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/kernel_impl.hpp>
+#include <detail/kernel_info.hpp>
 
 #include <memory>
 
diff --git a/sycl/include/CL/sycl/detail/kernel_impl.hpp b/sycl/source/detail/kernel_impl.hpp
similarity index 94%
rename from sycl/include/CL/sycl/detail/kernel_impl.hpp
rename to sycl/source/detail/kernel_impl.hpp
index f30450949415a..50e375e4fa831 100644
--- a/sycl/include/CL/sycl/detail/kernel_impl.hpp
+++ b/sycl/source/detail/kernel_impl.hpp
@@ -9,12 +9,12 @@
 #pragma once
 
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/device_impl.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/program_impl.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/program_impl.hpp>
 
 #include <cassert>
 #include <memory>
@@ -54,15 +54,13 @@ class kernel_impl {
   /// @param IsCreatedFromSource is a flag that indicates whether program
   /// is created from source code
   kernel_impl(RT::PiKernel Kernel, ContextImplPtr ContextImpl,
-              ProgramImplPtr ProgramImpl,
-              bool IsCreatedFromSource);
+              ProgramImplPtr ProgramImpl, bool IsCreatedFromSource);
 
   /// Constructs a SYCL kernel for host device
   ///
   /// @param SyclContext is a valid SYCL context
   /// @param ProgramImpl is a valid instance of program_impl
-  kernel_impl(ContextImplPtr Context,
-              ProgramImplPtr ProgramImpl);
+  kernel_impl(ContextImplPtr Context, ProgramImplPtr ProgramImpl);
 
   ~kernel_impl();
 
@@ -131,8 +129,8 @@ class kernel_impl {
   RT::PiKernel &getHandleRef() { return MKernel; }
   /// Get a constant reference to a raw kernel object.
   ///
-  /// @return a constant reference to a valid PiKernel instance with raw kernel
-  /// object.
+  /// @return a constant reference to a valid PiKernel instance with raw
+  /// kernel object.
   const RT::PiKernel &getHandleRef() const { return MKernel; }
 
   /// Check if kernel was created from a program that had been created from
diff --git a/sycl/source/detail/kernel_info.cpp b/sycl/source/detail/kernel_info.cpp
index 555835296cb07..ff6fd5a64a3cb 100644
--- a/sycl/source/detail/kernel_info.cpp
+++ b/sycl/source/detail/kernel_info.cpp
@@ -6,8 +6,8 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/kernel_info.hpp>
 #include <CL/sycl/device.hpp>
+#include <detail/kernel_info.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/include/CL/sycl/detail/kernel_info.hpp b/sycl/source/detail/kernel_info.hpp
similarity index 97%
rename from sycl/include/CL/sycl/detail/kernel_info.hpp
rename to sycl/source/detail/kernel_info.hpp
index 0b2e04b300646..fae537341b133 100644
--- a/sycl/include/CL/sycl/detail/kernel_info.hpp
+++ b/sycl/source/detail/kernel_info.hpp
@@ -10,9 +10,9 @@
 
 #include <CL/sycl/detail/common.hpp>
 #include <CL/sycl/detail/common_info.hpp>
-#include <CL/sycl/detail/kernel_impl.hpp>
 #include <CL/sycl/detail/pi.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/kernel_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -54,8 +54,7 @@ template <info::kernel Param> struct get_kernel_info<cl_uint, Param> {
 
 template <typename T, info::kernel_work_group Param>
 struct get_kernel_work_group_info {
-  static T get(RT::PiKernel Kernel, RT::PiDevice Device,
-               const plugin &Plugin) {
+  static T get(RT::PiKernel Kernel, RT::PiDevice Device, const plugin &Plugin) {
     T Result;
     // TODO catch an exception and put it to list of asynchronous exceptions
     Plugin.call<PiApiKind::piKernelGetGroupInfo>(
diff --git a/sycl/source/detail/kernel_program_cache.cpp b/sycl/source/detail/kernel_program_cache.cpp
index de8e03159dafb..18b13f3fd589c 100644
--- a/sycl/source/detail/kernel_program_cache.cpp
+++ b/sycl/source/detail/kernel_program_cache.cpp
@@ -6,9 +6,9 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/kernel_program_cache.hpp>
-#include <CL/sycl/detail/plugin.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/kernel_program_cache.hpp>
+#include <detail/plugin.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/include/CL/sycl/detail/kernel_program_cache.hpp b/sycl/source/detail/kernel_program_cache.hpp
similarity index 88%
rename from sycl/include/CL/sycl/detail/kernel_program_cache.hpp
rename to sycl/source/detail/kernel_program_cache.hpp
index ebb970bf2643c..3c85ba1f2d5fa 100644
--- a/sycl/include/CL/sycl/detail/kernel_program_cache.hpp
+++ b/sycl/source/detail/kernel_program_cache.hpp
@@ -11,7 +11,7 @@
 #include <CL/sycl/detail/locked.hpp>
 #include <CL/sycl/detail/os_util.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/platform_impl.hpp>
+#include <detail/platform_impl.hpp>
 
 #include <atomic>
 #include <condition_variable>
@@ -41,15 +41,12 @@ class KernelProgramCache {
   /// The pointer is not null if and only if the entity is usable.
   /// State of the entity is provided by the user of cache instance.
   /// Currently there is only a single user - ProgramManager class.
-  template<typename T>
-  struct BuildResult {
+  template<typename T> struct BuildResult {
     std::atomic<T *> Ptr;
     std::atomic<int> State;
     BuildError Error;
 
-    BuildResult(T* P, int S)
-      : Ptr{P}, State{S}, Error{"", 0, false}
-    {}
+    BuildResult(T* P, int S) : Ptr{P}, State{S}, Error{"", 0, false} {}
   };
 
   using PiProgramT = std::remove_pointer<RT::PiProgram>::type;
@@ -76,16 +73,13 @@ class KernelProgramCache {
     return {MKernelsPerProgramCache, MKernelsPerProgramCacheMutex};
   }
 
-  template<class Predicate>
-  void waitUntilBuilt(Predicate Pred) const {
+  template <class Predicate> void waitUntilBuilt(Predicate Pred) const {
     std::unique_lock<std::mutex> Lock(MBuildCVMutex);
 
     MBuildCV.wait(Lock, Pred);
   }
 
-  void notifyAllBuild() const {
-    MBuildCV.notify_all();
-  }
+  void notifyAllBuild() const { MBuildCV.notify_all(); }
 
 private:
   std::mutex MProgramCacheMutex;
@@ -98,6 +92,6 @@ class KernelProgramCache {
   KernelCacheT MKernelsPerProgramCache;
   ContextPtr MParentContext;
 };
-}
-}
-}
+} // namespace detail
+} // namespace sycl
+} // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/source/detail/memory_manager.cpp b/sycl/source/detail/memory_manager.cpp
index c44248f86a662..577045ac4aaa9 100644
--- a/sycl/source/detail/memory_manager.cpp
+++ b/sycl/source/detail/memory_manager.cpp
@@ -6,11 +6,11 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
 #include <CL/sycl/detail/memory_manager.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/usm_dispatch.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/usm/usm_dispatch.hpp>
 
 #include <algorithm>
 #include <cassert>
diff --git a/sycl/source/detail/pi.cpp b/sycl/source/detail/pi.cpp
index 991346a3c41dd..3e310db2e4e20 100644
--- a/sycl/source/detail/pi.cpp
+++ b/sycl/source/detail/pi.cpp
@@ -7,7 +7,7 @@
 //===----------------------------------------------------------------------===//
 #include <CL/sycl/detail/common.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/plugin.hpp>
+#include <detail/plugin.hpp>
 
 #include <cstdarg>
 #include <cstring>
diff --git a/sycl/source/detail/platform_impl.cpp b/sycl/source/detail/platform_impl.cpp
index 364ca7158153f..c0d360f68a473 100644
--- a/sycl/source/detail/platform_impl.cpp
+++ b/sycl/source/detail/platform_impl.cpp
@@ -6,11 +6,11 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/device_impl.hpp>
-#include <CL/sycl/detail/platform_impl.hpp>
-#include <CL/sycl/detail/platform_info.hpp>
 #include <CL/sycl/device.hpp>
 #include <detail/config.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/platform_impl.hpp>
+#include <detail/platform_info.hpp>
 
 #include <algorithm>
 #include <cstring>
diff --git a/sycl/include/CL/sycl/detail/platform_impl.hpp b/sycl/source/detail/platform_impl.hpp
similarity index 96%
rename from sycl/include/CL/sycl/detail/platform_impl.hpp
rename to sycl/source/detail/platform_impl.hpp
index ef707d5910dbc..8cb7aaab89828 100644
--- a/sycl/include/CL/sycl/detail/platform_impl.hpp
+++ b/sycl/source/detail/platform_impl.hpp
@@ -10,9 +10,9 @@
 #include <CL/sycl/detail/common.hpp>
 #include <CL/sycl/detail/force_device.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/plugin.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/plugin.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -30,13 +30,15 @@ class platform_impl {
   /// Constructs platform_impl for a SYCL host platform.
   platform_impl() : MHostPlatform(true) {}
 
-  /// Constructs platform_impl from a plug-in interoperability platform handle.
+  /// Constructs platform_impl from a plug-in interoperability platform
+  /// handle.
   ///
   /// @param Platform is a raw plug-in platform handle.
   explicit platform_impl(RT::PiPlatform APlatform, const plugin &APlugin)
       : MPlatform(APlatform), MPlugin(std::make_shared<plugin>(APlugin)) {}
 
-  explicit platform_impl(RT::PiPlatform APlatform, std::shared_ptr<plugin> APlugin)
+  explicit platform_impl(RT::PiPlatform APlatform,
+                         std::shared_ptr<plugin> APlugin)
       : MPlatform(APlatform), MPlugin(APlugin) {}
 
   ~platform_impl() = default;
diff --git a/sycl/source/detail/platform_info.cpp b/sycl/source/detail/platform_info.cpp
index 9161e00dc4d1b..8f32d50432561 100644
--- a/sycl/source/detail/platform_info.cpp
+++ b/sycl/source/detail/platform_info.cpp
@@ -6,7 +6,7 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/platform_info.hpp>
+#include <detail/platform_info.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/include/CL/sycl/detail/platform_info.hpp b/sycl/source/detail/platform_info.hpp
similarity index 95%
rename from sycl/include/CL/sycl/detail/platform_info.hpp
rename to sycl/source/detail/platform_info.hpp
index 606e9305b92d1..e9caa58db140d 100644
--- a/sycl/include/CL/sycl/detail/platform_info.hpp
+++ b/sycl/source/detail/platform_info.hpp
@@ -10,8 +10,8 @@
 #include <CL/sycl/detail/common.hpp>
 #include <CL/sycl/detail/common_info.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/plugin.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/plugin.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -20,8 +20,7 @@ namespace detail {
 // The platform information methods
 template <typename T, info::platform param> struct get_platform_info {};
 
-template <info::platform param>
-struct get_platform_info<string_class, param> {
+template <info::platform param> struct get_platform_info<string_class, param> {
   static string_class get(RT::PiPlatform plt, const plugin &Plugin) {
     size_t resultSize;
     // TODO catch an exception and put it to list of asynchronous exceptions
diff --git a/sycl/source/detail/platform_util.cpp b/sycl/source/detail/platform_util.cpp
index 5929f049e277b..5ea2b5ae06c3d 100644
--- a/sycl/source/detail/platform_util.cpp
+++ b/sycl/source/detail/platform_util.cpp
@@ -7,8 +7,8 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/os_util.hpp>
-#include <CL/sycl/detail/platform_util.hpp>
 #include <CL/sycl/exception.hpp>
+#include <detail/platform_util.hpp>
 
 #if defined(SYCL_RT_OS_LINUX)
 #include <cpuid.h>
diff --git a/sycl/include/CL/sycl/detail/platform_util.hpp b/sycl/source/detail/platform_util.hpp
similarity index 100%
rename from sycl/include/CL/sycl/detail/platform_util.hpp
rename to sycl/source/detail/platform_util.hpp
diff --git a/sycl/include/CL/sycl/detail/plugin.hpp b/sycl/source/detail/plugin.hpp
similarity index 100%
rename from sycl/include/CL/sycl/detail/plugin.hpp
rename to sycl/source/detail/plugin.hpp
diff --git a/sycl/source/detail/program_impl.cpp b/sycl/source/detail/program_impl.cpp
index a529d974082a4..adbbe010c69ef 100644
--- a/sycl/source/detail/program_impl.cpp
+++ b/sycl/source/detail/program_impl.cpp
@@ -8,8 +8,8 @@
 
 #include <CL/sycl/detail/common.hpp>
 #include <CL/sycl/detail/kernel_desc.hpp>
-#include <CL/sycl/detail/program_impl.hpp>
 #include <CL/sycl/kernel.hpp>
+#include <detail/program_impl.hpp>
 
 #include <algorithm>
 #include <fstream>
diff --git a/sycl/include/CL/sycl/detail/program_impl.hpp b/sycl/source/detail/program_impl.hpp
similarity index 93%
rename from sycl/include/CL/sycl/detail/program_impl.hpp
rename to sycl/source/detail/program_impl.hpp
index e52bbe62935c1..9580d2fb20994 100644
--- a/sycl/include/CL/sycl/detail/program_impl.hpp
+++ b/sycl/source/detail/program_impl.hpp
@@ -10,11 +10,11 @@
 #include <CL/sycl/context.hpp>
 #include <CL/sycl/detail/common_info.hpp>
 #include <CL/sycl/detail/kernel_desc.hpp>
-#include <CL/sycl/detail/kernel_impl.hpp>
-#include <CL/sycl/detail/program_manager/program_manager.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/program.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/kernel_impl.hpp>
+#include <detail/program_manager/program_manager.hpp>
 
 #include <algorithm>
 #include <cassert>
@@ -61,11 +61,10 @@ class program_impl {
   /// state and must be associated with the same SYCL context. Otherwise an
   /// invalid_object_error SYCL exception will be thrown. A
   /// feature_not_supported exception will be thrown if any device that the
-  /// program is to be linked for returns false for the device information query
-  /// info::device::is_linker_available.
-  /// Kernels caching for linked programs won't be allowed due to only compiled
-  /// state of each and every program in the list and thus unknown state of
-  /// caching resolution.
+  /// program is to be linked for returns false for the device information
+  /// query info::device::is_linker_available. Kernels caching for linked
+  /// programs won't be allowed due to only compiled state of each and every
+  /// program in the list and thus unknown state of caching resolution.
   ///
   /// @param ProgramList is a list of program_impl instances.
   /// @param LinkOptions is a string containing valid OpenCL link options.
@@ -75,11 +74,13 @@ class program_impl {
   /// Constructs a program instance from plugin interface interoperability
   /// handle.
   ///
-  /// The state of the constructed program can be either program_state::compiled
-  /// or program_state::linked, depending on the state of the ClProgram.
-  /// Otherwise an invalid_object_error SYCL exception is thrown.
+  /// The state of the constructed program can be either
+  /// program_state::compiled or program_state::linked, depending on the state
+  /// of the ClProgram. Otherwise an invalid_object_error SYCL exception is
+  /// thrown.
   ///
-  /// The instance of plugin interface program will be retained on construction.
+  /// The instance of plugin interface program will be retained on
+  /// construction.
   ///
   /// @param Context is a pointer to SYCL context impl.
   /// @param Program is an instance of plugin interface interoperability
@@ -98,17 +99,17 @@ class program_impl {
   /// Returns a valid cl_program instance.
   ///
   /// The instance of cl_program will be retained before returning.
-  /// If the program is created for a SYCL host device, an invalid_object_error
-  /// exception is thrown.
+  /// If the program is created for a SYCL host device, an
+  /// invalid_object_error exception is thrown.
   ///
   /// @return a valid OpenCL cl_program instance.
   cl_program get() const;
 
-  /// @return a reference to a raw PI program handle. PI program is not retained
-  /// before return.
-  RT::PiProgram &getHandleRef() { return MProgram; }
-  /// @return a constant reference to a raw PI program handle. PI program is not
+  /// @return a reference to a raw PI program handle. PI program is not
   /// retained before return.
+  RT::PiProgram &getHandleRef() { return MProgram; }
+  /// @return a constant reference to a raw PI program handle. PI program is
+  /// not retained before return.
   const RT::PiProgram &getHandleRef() const { return MProgram; }
 
   /// @return true if this SYCL program is a host program.
@@ -155,8 +156,8 @@ class program_impl {
   /// compilation fails, a compile_program_error SYCL exception is thrown. If
   /// any device that the program is being built for returns false for the
   /// device information queries info::device::is_compiler_available or
-  /// info::device::is_linker_available, a feature_not_supported SYCL exception
-  /// is thrown.
+  /// info::device::is_linker_available, a feature_not_supported SYCL
+  /// exception is thrown.
   ///
   /// @param KernelName is a string containing SYCL kernel name.
   /// @param BuildOptions is a string containing OpenCL compile options.
@@ -171,8 +172,8 @@ class program_impl {
   /// fails, a compile_program_error SYCL exception is thrown. If any device
   /// that the program is being built for returns false for the device
   /// information queries info::device::is_compiler_available or
-  /// info::device::is_linker_available, a feature_not_supported SYCL exception
-  /// is thrown.
+  /// info::device::is_linker_available, a feature_not_supported SYCL
+  /// exception is thrown.
   ///
   /// @param KernelSource is a string containing OpenCL C kernel source code.
   /// @param BuildOptions is a string containing OpenCL build options.
@@ -184,8 +185,8 @@ class program_impl {
   /// This member function sets the state of this SYCL program to
   /// program_state::linked. If the program was not in the
   /// program_state::compiled state, an invalid_object_error SYCL exception is
-  /// thrown. If linking fails, a compile_program_error is thrown. If any device
-  /// that the program is to be linked for returns false for the device
+  /// thrown. If linking fails, a compile_program_error is thrown. If any
+  /// device that the program is to be linked for returns false for the device
   /// information query info::device::is_linker_available, a
   /// feature_not_supported exception is thrown.
   ///
@@ -200,7 +201,8 @@ class program_impl {
   /// @return true if the SYCL kernel is available.
   bool has_kernel(string_class KernelName, bool IsCreatedFromSource) const;
 
-  /// Returns a SYCL kernel for the SYCL kernel function defined by kernel name.
+  /// Returns a SYCL kernel for the SYCL kernel function defined by kernel
+  /// name.
   ///
   /// If program is in the program_state::none state or if the SYCL kernel
   /// function is not available, an invalid_object_error exception is thrown.
@@ -249,8 +251,8 @@ class program_impl {
   /// If the program was built instead of explicitly compiled, if the program
   /// has not yet been compiled, or if the program has been compiled for only
   /// the host device, then an empty string is return, unless the underlying
-  /// cl_program was explicitly compiled, in which case the compile options used
-  /// in the explicit compile are returned.
+  /// cl_program was explicitly compiled, in which case the compile options
+  /// used in the explicit compile are returned.
   ///
   /// @return a string of valid OpenCL compile options.
   string_class get_compile_options() const { return MCompileOptions; }
@@ -264,9 +266,9 @@ class program_impl {
   /// is returned. If the program was constructed from cl_program, then an
   /// empty string is returned unless the cl_program was explicitly linked,
   /// in which case the link options used in that explicit link are returned.
-  /// If the program object was constructed using a constructor form that links
-  /// a vector of programs, then the link options passed to this constructor
-  /// are returned.
+  /// If the program object was constructed using a constructor form that
+  /// links a vector of programs, then the link options passed to this
+  /// constructor are returned.
   ///
   /// @return a string of valid OpenCL compile options.
   string_class get_link_options() const { return MLinkOptions; }
@@ -354,8 +356,8 @@ class program_impl {
   /// @param State is a program state to match against.
   void throw_if_state_is(program_state State) const;
 
-  /// Throws an invalid_object_exception if state of this program is not in the
-  /// specified state.
+  /// Throws an invalid_object_exception if state of this program is not in
+  /// the specified state.
   ///
   /// @param State is a program state to match against.
   void throw_if_state_is_not(program_state State) const;
diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp
index 720bbfe92de8b..a1369bdd0c868 100644
--- a/sycl/source/detail/program_manager/program_manager.cpp
+++ b/sycl/source/detail/program_manager/program_manager.cpp
@@ -8,15 +8,15 @@
 
 #include <CL/sycl/context.hpp>
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/device_impl.hpp>
 #include <CL/sycl/detail/os_util.hpp>
-#include <CL/sycl/detail/program_manager/program_manager.hpp>
 #include <CL/sycl/detail/type_traits.hpp>
 #include <CL/sycl/detail/util.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/exception.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/program_manager/program_manager.hpp>
 
 #include <algorithm>
 #include <cassert>
diff --git a/sycl/include/CL/sycl/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp
similarity index 88%
rename from sycl/include/CL/sycl/detail/program_manager/program_manager.hpp
rename to sycl/source/detail/program_manager/program_manager.hpp
index 7974a257ed512..826ab66d8e187 100644
--- a/sycl/include/CL/sycl/detail/program_manager/program_manager.hpp
+++ b/sycl/source/detail/program_manager/program_manager.hpp
@@ -13,6 +13,7 @@
 #include <CL/sycl/detail/pi.hpp>
 #include <CL/sycl/stl.hpp>
 
+#include <map>
 #include <unordered_map>
 #include <vector>
 
@@ -55,8 +56,8 @@ enum DeviceLibExt {
 // that is necessary for no interoperability cases with lambda.
 class ProgramManager {
 public:
-  // Returns the single instance of the program manager for the entire process.
-  // Can only be called after staticInit is done.
+  // Returns the single instance of the program manager for the entire
+  // process. Can only be called after staticInit is done.
   static ProgramManager &getInstance();
   DeviceImage &getDeviceImage(OSModuleHandle M, const string_class &KernelName,
                               const context &Context);
@@ -104,17 +105,16 @@ class ProgramManager {
   /// The three maps below are used during kernel resolution. Any kernel is
   /// identified by its name and the OS module it's coming from, allowing
   /// kernels with identical names in different OS modules. The following
-  /// assumption is made: for any two device images in a SYCL application their
-  /// kernel sets are either identical or disjoint.
-  /// Based on this assumption, m_KernelSets is used to group kernels together
-  /// into sets by assigning a set ID to them during device image registration.
-  /// This ID is then mapped to a vector of device images containing kernels
-  /// from the set (m_DeviceImages).
-  /// An exception is made for device images with no entry information: a
-  /// special kernel set ID is used for them which is assigned to just the OS
-  /// module. These kernel set ids are stored in m_OSModuleKernelSets and device
-  /// images associated with them are assumed to contain all kernels coming from
-  /// that OS module.
+  /// assumption is made: for any two device images in a SYCL application
+  /// their kernel sets are either identical or disjoint. Based on this
+  /// assumption, m_KernelSets is used to group kernels together into sets by
+  /// assigning a set ID to them during device image registration. This ID is
+  /// then mapped to a vector of device images containing kernels from the set
+  /// (m_DeviceImages). An exception is made for device images with no entry
+  /// information: a special kernel set ID is used for them which is assigned
+  /// to just the OS module. These kernel set ids are stored in
+  /// m_OSModuleKernelSets and device images associated with them are assumed
+  /// to contain all kernels coming from that OS module.
 
   /// Keeps all available device executable images added via \ref addImages.
   /// Organizes the images as a map from a kernel set id to the vector of images
diff --git a/sycl/source/detail/queue_impl.cpp b/sycl/source/detail/queue_impl.cpp
index aa639322befc0..5b7690cf5fa36 100644
--- a/sycl/source/detail/queue_impl.cpp
+++ b/sycl/source/detail/queue_impl.cpp
@@ -10,9 +10,9 @@
 #include <CL/sycl/detail/clusm.hpp>
 #include <CL/sycl/detail/memory_manager.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/usm_dispatch.hpp>
 #include <CL/sycl/device.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/usm/usm_dispatch.hpp>
 
 #include <cstring>
 
diff --git a/sycl/include/CL/sycl/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp
similarity index 93%
rename from sycl/include/CL/sycl/detail/queue_impl.hpp
rename to sycl/source/detail/queue_impl.hpp
index 52210e221817e..79d62ae2912ae 100644
--- a/sycl/include/CL/sycl/detail/queue_impl.hpp
+++ b/sycl/source/detail/queue_impl.hpp
@@ -9,11 +9,6 @@
 #pragma once
 
 #include <CL/sycl/context.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/device_impl.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
-#include <CL/sycl/detail/plugin.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/event.hpp>
 #include <CL/sycl/exception.hpp>
@@ -21,6 +16,11 @@
 #include <CL/sycl/handler.hpp>
 #include <CL/sycl/property_list.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/plugin.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -39,14 +39,14 @@ class queue_impl {
   /// Constructs a SYCL queue from a device using an async_handler and
   /// property_list provided.
   ///
-  /// @param Device is a SYCL device that is used to dispatch tasks submitted to
-  /// the queue.
+  /// @param Device is a SYCL device that is used to dispatch tasks submitted
+  /// to the queue.
   /// @param AsyncHandler is a SYCL asynchronous exception handler.
-  /// @param Order specifies whether the queue being constructed as in-order or
-  /// out-of-order.
+  /// @param Order specifies whether the queue being constructed as in-order
+  /// or out-of-order.
   /// @param PropList is a list of properties to use for queue construction.
-  queue_impl(DeviceImplPtr Device, async_handler AsyncHandler,
-             QueueOrder Order, const property_list &PropList)
+  queue_impl(DeviceImplPtr Device, async_handler AsyncHandler, QueueOrder Order,
+             const property_list &PropList)
       : queue_impl(Device,
                    detail::getSyclObjImpl(
                        context(createSyclObjFromImpl<device>(Device))),
@@ -55,13 +55,13 @@ class queue_impl {
   /// Constructs a SYCL queue with an async_handler and property_list provided
   /// form a device and a context.
   ///
-  /// @param Device is a SYCL device that is used to dispatch tasks submitted to
-  /// the queue.
+  /// @param Device is a SYCL device that is used to dispatch tasks submitted
+  /// to the queue.
   /// @param Context is a SYCL context to associate with the queue being
   /// constructed.
   /// @param AsyncHandler is a SYCL asynchronous exception handler.
-  /// @param Order specifies whether the queue being constructed as in-order or
-  /// out-of-order.
+  /// @param Order specifies whether the queue being constructed as in-order
+  /// or out-of-order.
   /// @param PropList is a list of properties to use for queue construction.
   queue_impl(DeviceImplPtr Device, ContextImplPtr Context,
              async_handler AsyncHandler, QueueOrder Order,
@@ -96,7 +96,8 @@ class queue_impl {
     // TODO catch an exception and put it to list of asynchronous exceptions
     Plugin.call<PiApiKind::piQueueGetInfo>(MCommandQueue, PI_QUEUE_INFO_DEVICE,
                                            sizeof(Device), &Device, nullptr);
-    MDevice = DeviceImplPtr(new device_impl(Device, Context->getPlatformImpl()));
+    MDevice =
+        DeviceImplPtr(new device_impl(Device, Context->getPlatformImpl()));
 
     // TODO catch an exception and put it to list of asynchronous exceptions
     Plugin.call<PiApiKind::piQueueRetain>(MCommandQueue);
@@ -197,10 +198,10 @@ class queue_impl {
   /// Performs a blocking wait for the completion of all enqueued tasks in the
   /// queue.
   ///
-  /// Synchronous errors will be reported through SYCL exceptions. Asynchronous
-  /// errors will be passed to the async_handler passed to the queue on
-  /// construction. If no async_handler was provided then asynchronous
-  /// exceptions will be lost.
+  /// Synchronous errors will be reported through SYCL exceptions.
+  /// Asynchronous errors will be passed to the async_handler passed to the
+  /// queue on construction. If no async_handler was provided then
+  /// asynchronous exceptions will be lost.
   void throw_asynchronous() {
     std::unique_lock<mutex_class> lock(MMutex);
 
@@ -220,8 +221,8 @@ class queue_impl {
 
   /// Creates PI queue.
   ///
-  /// @param Order specifies whether the queue being constructed as in-order or
-  /// out-of-order.
+  /// @param Order specifies whether the queue being constructed as in-order
+  /// or out-of-order.
   RT::PiQueue createQueue(QueueOrder Order) {
     RT::PiQueueProperties CreationFlags = 0;
 
@@ -264,8 +265,8 @@ class queue_impl {
       return MQueues.back();
     }
 
-    // If the limit of OpenCL queues is going to be exceeded - take the earliest
-    // used queue, wait until it finished and then reuse it.
+    // If the limit of OpenCL queues is going to be exceeded - take the
+    // earliest used queue, wait until it finished and then reuse it.
     MQueueNumber %= MaxNumQueues;
     size_t FreeQueueNum = MQueueNumber++;
 
@@ -273,8 +274,8 @@ class queue_impl {
     return MQueues[FreeQueueNum];
   }
 
-  /// @return a raw PI queue handle. The returned handle is not retained. It is
-  /// caller responsibility to make sure queue is still alive.
+  /// @return a raw PI queue handle. The returned handle is not retained. It
+  /// is caller responsibility to make sure queue is still alive.
   RT::PiQueue &getHandleRef() {
     if (MSupportOOO) {
       return MCommandQueue;
diff --git a/sycl/source/detail/sampler_impl.cpp b/sycl/source/detail/sampler_impl.cpp
index 56c2d345b7c1d..bad995ad752bc 100644
--- a/sycl/source/detail/sampler_impl.cpp
+++ b/sycl/source/detail/sampler_impl.cpp
@@ -6,8 +6,8 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/context_impl.hpp>
 #include <CL/sycl/detail/sampler_impl.hpp>
+#include <detail/context_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp
index 6e1b8d246364b..4f6989a445cc5 100644
--- a/sycl/source/detail/scheduler/commands.cpp
+++ b/sycl/source/detail/scheduler/commands.cpp
@@ -11,18 +11,18 @@
 #include "CL/sycl/access/access.hpp"
 #include <CL/cl.h>
 #include <CL/sycl/detail/clusm.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
 #include <CL/sycl/detail/kernel_desc.hpp>
-#include <CL/sycl/detail/kernel_impl.hpp>
-#include <CL/sycl/detail/kernel_info.hpp>
 #include <CL/sycl/detail/memory_manager.hpp>
-#include <CL/sycl/detail/program_manager/program_manager.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/scheduler/commands.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/detail/stream_impl.hpp>
 #include <CL/sycl/sampler.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/kernel_impl.hpp>
+#include <detail/kernel_info.hpp>
+#include <detail/program_manager/program_manager.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/scheduler/commands.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <string>
 #include <vector>
diff --git a/sycl/include/CL/sycl/detail/scheduler/commands.hpp b/sycl/source/detail/scheduler/commands.hpp
similarity index 95%
rename from sycl/include/CL/sycl/detail/scheduler/commands.hpp
rename to sycl/source/detail/scheduler/commands.hpp
index 016816615560e..5150b815880ee 100644
--- a/sycl/include/CL/sycl/detail/scheduler/commands.hpp
+++ b/sycl/source/detail/scheduler/commands.hpp
@@ -70,11 +70,11 @@ struct DepDesc {
   AllocaCommandBase *MAllocaCmd = nullptr;
 };
 
-// The Command represents some action that needs to be performed on one or more
-// memory objects. The command has vector of Depdesc objects that represent
-// dependencies of the command. It has vector of pointer to commands that depend
-// on the command. It has pointer to sycl::queue object. And has event that is
-// associated with the command.
+// The Command represents some action that needs to be performed on one or
+// more memory objects. The command has vector of Depdesc objects that
+// represent dependencies of the command. It has vector of pointer to commands
+// that depend on the command. It has pointer to sycl::queue object. And has
+// event that is associated with the command.
 class Command {
 public:
   enum CommandType {
@@ -104,9 +104,9 @@ class Command {
   // Return type of the command, e.g. Allocate, MemoryCopy.
   CommandType getType() const { return MType; }
 
-  // The method checks if the command is enqueued, waits for it to be unblocked
-  // if "Blocking" argument is true, then calls enqueueImp.
-  // Returns true if the command is enqueued. Sets EnqueueResult to the specific
+  // The method checks if the command is enqueued, waits for it to be
+  // unblocked if "Blocking" argument is true, then calls enqueueImp. Returns
+  // true if the command is enqueued. Sets EnqueueResult to the specific
   // status otherwise.
   bool enqueue(EnqueueResultT &EnqueueResult, BlockingT Blocking);
 
@@ -251,8 +251,8 @@ class AllocaCommand : public AllocaCommandBase {
 private:
   cl_int enqueueImp() final;
 
-  // The flag indicates that alloca should try to reuse pointer provided by the
-  // user during memory object construction
+  // The flag indicates that alloca should try to reuse pointer provided by
+  // the user during memory object construction
   bool MInitFromUserData = false;
 };
 
diff --git a/sycl/source/detail/scheduler/graph_builder.cpp b/sycl/source/detail/scheduler/graph_builder.cpp
index bbc5d13b79b19..bf9b6f76be0f1 100644
--- a/sycl/source/detail/scheduler/graph_builder.cpp
+++ b/sycl/source/detail/scheduler/graph_builder.cpp
@@ -6,14 +6,14 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/access/access.hpp>
 #include "detail/config.hpp"
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
+#include <CL/sycl/access/access.hpp>
 #include <CL/sycl/detail/memory_manager.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/exception.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <cstdlib>
 #include <fstream>
@@ -765,7 +765,7 @@ void Scheduler::GraphBuilder::removeRecordForMemObj(SYCLMemObjI *MemObject) {
       [MemObject](const SYCLMemObjI *Obj) { return Obj == MemObject; });
   if (It != MMemObjs.end())
     MMemObjs.erase(It);
-  MemObject->MRecord.reset(nullptr);
+  MemObject->MRecord.reset();
 }
 
 } // namespace detail
diff --git a/sycl/source/detail/scheduler/graph_processor.cpp b/sycl/source/detail/scheduler/graph_processor.cpp
index e9ccb51f2b57c..5681d2f5bbd31 100644
--- a/sycl/source/detail/scheduler/graph_processor.cpp
+++ b/sycl/source/detail/scheduler/graph_processor.cpp
@@ -6,9 +6,9 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/event_impl.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <memory>
 #include <vector>
diff --git a/sycl/source/detail/scheduler/scheduler.cpp b/sycl/source/detail/scheduler/scheduler.cpp
index b4f8c3af47174..b86367f5a0cfc 100644
--- a/sycl/source/detail/scheduler/scheduler.cpp
+++ b/sycl/source/detail/scheduler/scheduler.cpp
@@ -7,9 +7,9 @@
 //===----------------------------------------------------------------------===//
 
 #include "CL/sycl/detail/sycl_mem_obj_i.hpp"
-#include <CL/sycl/detail/queue_impl.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/device_selector.hpp>
+#include <detail/queue_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <memory>
 #include <mutex>
diff --git a/sycl/include/CL/sycl/detail/scheduler/scheduler.hpp b/sycl/source/detail/scheduler/scheduler.hpp
similarity index 91%
rename from sycl/include/CL/sycl/detail/scheduler/scheduler.hpp
rename to sycl/source/detail/scheduler/scheduler.hpp
index 10bfb4a30d81b..d5ede48160c19 100644
--- a/sycl/include/CL/sycl/detail/scheduler/scheduler.hpp
+++ b/sycl/source/detail/scheduler/scheduler.hpp
@@ -10,8 +10,8 @@
 
 #include <CL/sycl/detail/cg.hpp>
 #include <CL/sycl/detail/circular_buffer.hpp>
-#include <CL/sycl/detail/scheduler/commands.hpp>
 #include <CL/sycl/detail/sycl_mem_obj_i.hpp>
+#include <detail/scheduler/commands.hpp>
 
 #include <cstddef>
 #include <memory>
@@ -65,16 +65,16 @@ class Scheduler {
 
   EventImplPtr addCopyBack(Requirement *Req);
 
-  // Blocking call that waits for the event passed. For the eager execution mode
-  // this method invokes corresponding function of device API. In the lazy
-  // execution mode the method may enqueue the command associated with the event
-  // passed and its dependency before calling device API.
+  // Blocking call that waits for the event passed. For the eager execution
+  // mode this method invokes corresponding function of device API. In the
+  // lazy execution mode the method may enqueue the command associated with
+  // the event passed and its dependency before calling device API.
   void waitForEvent(EventImplPtr Event);
 
   // Removes buffer pointed by MemObj from the graph: ensures all commands
-  // accessing the memory objects are executed and triggers deallocation of all
-  // memory assigned to the memory object. It's called from the sycl::buffer and
-  // sycl::image destructors.
+  // accessing the memory objects are executed and triggers deallocation of
+  // all memory assigned to the memory object. It's called from the
+  // sycl::buffer and sycl::image destructors.
   void removeMemoryObject(detail::SYCLMemObjI *MemObj);
 
   // Removes finished non-leaf non-alloca commands from the subgraph (assuming
@@ -82,11 +82,10 @@ class Scheduler {
   void cleanupFinishedCommands(Command *FinishedCmd);
 
   // Creates nodes in the graph, that update Req with the pointer to the host
-  // memory which contains the latest data of the memory object. New operations
-  // with the same memory object that have side effects are blocked until
-  // releaseHostAccessor is called.
-  // Returns an event which indicates when these nodes are completed and host
-  // accessor is ready for using.
+  // memory which contains the latest data of the memory object. New
+  // operations with the same memory object that have side effects are blocked
+  // until releaseHostAccessor is called. Returns an event which indicates
+  // when these nodes are completed and host accessor is ready for using.
   EventImplPtr addHostAccessor(Requirement *Req);
 
   // Unblocks operations with the memory object.
@@ -125,8 +124,8 @@ class Scheduler {
     // [Provisional] Optimizes the whole graph.
     void optimize();
 
-    // [Provisional] Optimizes subgraph that consists of command associated with
-    // Event passed and its dependencies.
+    // [Provisional] Optimizes subgraph that consists of command associated
+    // with Event passed and its dependencies.
     void optimize(EventImplPtr Event);
 
     // Removes finished non-leaf non-alloca commands from the subgraph (assuming
@@ -191,8 +190,7 @@ class Scheduler {
                                                Requirement *Req,
                                                QueueImplPtr Queue);
 
-    void markModifiedIfWrite(MemObjRecord *Record,
-                             Requirement *Req);
+    void markModifiedIfWrite(MemObjRecord *Record, Requirement *Req);
 
     // Print contents of graph to text file in DOT format
     void printGraphAsDot(const char *ModeName);
diff --git a/sycl/source/detail/stream_impl.cpp b/sycl/source/detail/stream_impl.cpp
index a7b57e2e0fbf0..968192af6810c 100644
--- a/sycl/source/detail/stream_impl.cpp
+++ b/sycl/source/detail/stream_impl.cpp
@@ -7,6 +7,8 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/stream_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
+
 #include <cstdio>
 
 __SYCL_INLINE_NAMESPACE(cl) {
diff --git a/sycl/source/detail/sycl_mem_obj_t.cpp b/sycl/source/detail/sycl_mem_obj_t.cpp
index 0069bd384daf9..142803dd33290 100644
--- a/sycl/source/detail/sycl_mem_obj_t.cpp
+++ b/sycl/source/detail/sycl_mem_obj_t.cpp
@@ -7,10 +7,11 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/memory_manager.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/detail/sycl_mem_obj_t.hpp>
-#include <CL/sycl/detail/context_impl.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
+#include <detail/context_impl.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/plugin.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
@@ -81,6 +82,21 @@ void SYCLMemObjT::updateHostMemory() {
         pi::cast<RT::PiMem>(MInteropMemObject));
   }
 }
+const plugin &SYCLMemObjT::getPlugin() const {
+  assert((MInteropContext != nullptr) &&
+         "Trying to get Plugin from SYCLMemObjT with nullptr ContextImpl.");
+  return (MInteropContext->getPlugin());
+}
+
+size_t SYCLMemObjT::getBufSizeForContext(const ContextImplPtr &Context,
+                                         cl_mem MemObject) {
+  size_t BufSize = 0;
+  const detail::plugin &Plugin = Context->getPlugin();
+  Plugin.call<detail::PiApiKind::piMemGetInfo>(
+      detail::pi::cast<detail::RT::PiMem>(MemObject), CL_MEM_SIZE,
+      sizeof(size_t), &BufSize, nullptr);
+  return BufSize;
+}
 } // namespace detail
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/source/detail/usm/usm_dispatch.cpp b/sycl/source/detail/usm/usm_dispatch.cpp
index 7c521a227ccd8..980410f778bf4 100644
--- a/sycl/source/detail/usm/usm_dispatch.cpp
+++ b/sycl/source/detail/usm/usm_dispatch.cpp
@@ -7,8 +7,8 @@
 // ===--------------------------------------------------------------------=== //
 
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/plugin.hpp>
-#include <CL/sycl/detail/usm_dispatch.hpp>
+#include <detail/plugin.hpp>
+#include <detail/usm/usm_dispatch.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
   namespace sycl {
diff --git a/sycl/include/CL/sycl/detail/usm_dispatch.hpp b/sycl/source/detail/usm/usm_dispatch.hpp
similarity index 100%
rename from sycl/include/CL/sycl/detail/usm_dispatch.hpp
rename to sycl/source/detail/usm/usm_dispatch.hpp
diff --git a/sycl/source/detail/usm/usm_impl.cpp b/sycl/source/detail/usm/usm_impl.cpp
index ce4978e20aa0d..e6cb56c936cd6 100644
--- a/sycl/source/detail/usm/usm_impl.cpp
+++ b/sycl/source/detail/usm/usm_impl.cpp
@@ -10,9 +10,9 @@
 #include <CL/sycl/detail/aligned_allocator.hpp>
 #include <CL/sycl/detail/os_util.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/queue_impl.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/usm.hpp>
+#include <detail/queue_impl.hpp>
 
 #include <cstdlib>
 
diff --git a/sycl/source/device.cpp b/sycl/source/device.cpp
index ac6abf7a70c4a..6da02b72eeeda 100644
--- a/sycl/source/device.cpp
+++ b/sycl/source/device.cpp
@@ -6,11 +6,11 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/device_impl.hpp>
 #include <CL/sycl/detail/force_device.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/device_selector.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/device_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/event.cpp b/sycl/source/event.cpp
index a0b4b2c1ee07d..bf0b0780edca9 100644
--- a/sycl/source/event.cpp
+++ b/sycl/source/event.cpp
@@ -8,13 +8,12 @@
 
 #include <CL/sycl/context.hpp>
 #include <CL/sycl/detail/common.hpp>
-#include <CL/sycl/detail/event_impl.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
-#include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/event.hpp>
-
+#include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/event_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <memory>
 #include <unordered_set>
diff --git a/sycl/source/function_pointer.cpp b/sycl/source/function_pointer.cpp
index 92e19664c3a0c..c273ae817c8bf 100644
--- a/sycl/source/function_pointer.cpp
+++ b/sycl/source/function_pointer.cpp
@@ -7,8 +7,8 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/intel/function_pointer.hpp>
-#include <CL/sycl/detail/device_impl.hpp>
-#include <CL/sycl/detail/program_impl.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/program_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/half_type.cpp b/sycl/source/half_type.cpp
index bab3009a8ed15..b7b06f767cd34 100644
--- a/sycl/source/half_type.cpp
+++ b/sycl/source/half_type.cpp
@@ -8,7 +8,8 @@
 
 #include <CL/sycl/half_type.hpp>
 // This is included to enable __builtin_expect()
-#include <CL/sycl/detail/platform_util.hpp>
+#include <detail/platform_util.hpp>
+
 #include <cstring>
 
 __SYCL_INLINE_NAMESPACE(cl) {
diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp
index a271509dd64ab..f8ff54c1b9c91 100644
--- a/sycl/source/handler.cpp
+++ b/sycl/source/handler.cpp
@@ -11,11 +11,11 @@
 #include <CL/sycl/detail/common.hpp>
 #include <CL/sycl/detail/helpers.hpp>
 #include <CL/sycl/detail/kernel_desc.hpp>
-#include <CL/sycl/detail/kernel_impl.hpp>
-#include <CL/sycl/detail/scheduler/scheduler.hpp>
 #include <CL/sycl/event.hpp>
 #include <CL/sycl/handler.hpp>
 #include <CL/sycl/info/info_desc.hpp>
+#include <detail/kernel_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/kernel.cpp b/sycl/source/kernel.cpp
index 7ec86691e5e22..2b4b2631d9e9f 100644
--- a/sycl/source/kernel.cpp
+++ b/sycl/source/kernel.cpp
@@ -6,9 +6,9 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/kernel_impl.hpp>
 #include <CL/sycl/kernel.hpp>
 #include <CL/sycl/program.hpp>
+#include <detail/kernel_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/ordered_queue.cpp b/sycl/source/ordered_queue.cpp
index 80d3b130550e1..228bd27a3f890 100644
--- a/sycl/source/ordered_queue.cpp
+++ b/sycl/source/ordered_queue.cpp
@@ -6,9 +6,12 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/queue_impl.hpp>
+#include <CL/sycl/device.hpp>
+#include <CL/sycl/event.hpp>
 #include <CL/sycl/exception_list.hpp>
+#include <CL/sycl/handler.hpp>
 #include <CL/sycl/ordered_queue.hpp>
+#include <detail/queue_impl.hpp>
 
 #include <algorithm>
 
diff --git a/sycl/source/platform.cpp b/sycl/source/platform.cpp
index ec2412684456e..077a95531df66 100644
--- a/sycl/source/platform.cpp
+++ b/sycl/source/platform.cpp
@@ -7,11 +7,11 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/force_device.hpp>
-#include <CL/sycl/detail/platform_impl.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/device_selector.hpp>
 #include <CL/sycl/info/info_desc.hpp>
 #include <CL/sycl/platform.hpp>
+#include <detail/platform_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/source/program.cpp b/sycl/source/program.cpp
index ddcbe5260cbcb..bb3011d614f3f 100644
--- a/sycl/source/program.cpp
+++ b/sycl/source/program.cpp
@@ -6,8 +6,8 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/program_impl.hpp>
 #include <CL/sycl/program.hpp>
+#include <detail/program_impl.hpp>
 
 #include <vector>
 
diff --git a/sycl/source/queue.cpp b/sycl/source/queue.cpp
index b50e5ed8e9aed..c47eba506958d 100644
--- a/sycl/source/queue.cpp
+++ b/sycl/source/queue.cpp
@@ -6,12 +6,12 @@
 //
 //===----------------------------------------------------------------------===//
 
-#include <CL/sycl/detail/queue_impl.hpp>
 #include <CL/sycl/event.hpp>
 #include <CL/sycl/exception_list.hpp>
 #include <CL/sycl/handler.hpp>
 #include <CL/sycl/queue.hpp>
 #include <CL/sycl/stl.hpp>
+#include <detail/queue_impl.hpp>
 
 #include <algorithm>
 
@@ -74,6 +74,12 @@ queue::queue(const context &syclContext, const device_selector &deviceSelector,
             detail::getSyclObjImpl(syclContext)->get_async_handler(),
             propList) {}
 
+queue::queue(const context &SyclContext, const device &SyclDevice,
+             const property_list &PropList)
+    : queue(SyclContext, SyclDevice,
+            detail::getSyclObjImpl(SyclContext)->get_async_handler(),
+            PropList) {}
+
 cl_command_queue queue::get() const { return impl->get(); }
 
 context queue::get_context() const { return impl->get_context(); }
@@ -134,5 +140,8 @@ template bool queue::has_property<property::queue::enable_profiling>() const;
 template property::queue::enable_profiling
 queue::get_property<property::queue::enable_profiling>() const;
 
+bool queue::is_in_order() const {
+  return impl->has_property<property::queue::in_order>();
+}
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
diff --git a/sycl/source/spirv_ops.cpp b/sycl/source/spirv_ops.cpp
index a4e1dba7e9004..9dc66ede1bb71 100644
--- a/sycl/source/spirv_ops.cpp
+++ b/sycl/source/spirv_ops.cpp
@@ -7,8 +7,9 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/__spirv/spirv_ops.hpp>
-#include <CL/sycl/detail/platform_util.hpp>
 #include <CL/sycl/exception.hpp>
+#include <detail/platform_util.hpp>
+
 #include <atomic>
 
 // This operation is NOP on HOST as all operations there are blocking and
diff --git a/sycl/source/stream.cpp b/sycl/source/stream.cpp
index c947246f56a2b..6bd8df572d267 100644
--- a/sycl/source/stream.cpp
+++ b/sycl/source/stream.cpp
@@ -7,6 +7,7 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/stream.hpp>
+#include <detail/queue_impl.hpp>
 
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
diff --git a/sycl/test/CMakeLists.txt b/sycl/test/CMakeLists.txt
index d2ec0c46cdced..e5e19b44e7a72 100644
--- a/sycl/test/CMakeLists.txt
+++ b/sycl/test/CMakeLists.txt
@@ -12,7 +12,6 @@ set(CLANGXX_IN_DEPLOY "${CMAKE_INSTALL_PREFIX}/bin/clang++")
 set(CLANGCL_IN_DEPLOY "${CMAKE_INSTALL_PREFIX}/bin/clang-cl")
 
 get_target_property(SYCL_BINARY_DIR sycl-toolchain BINARY_DIR)
-get_target_property(SYCL_SOURCE_DIR sycl-toolchain SOURCE_DIR)
 
 set(SYCL_INCLUDE "${dst_dir}")
 set(SYCL_DEPLOY_INCLUDE "${dst_deploy_dir}")
diff --git a/sycl/test/basic_tests/image_api.cpp b/sycl/test/basic_tests/image_api.cpp
index 9fc4406395abd..de8ffb1024eac 100644
--- a/sycl/test/basic_tests/image_api.cpp
+++ b/sycl/test/basic_tests/image_api.cpp
@@ -1,5 +1,5 @@
-// RUN: %clangxx -fsycl %s -o %t1.out
-// RUN: %clangxx %s -o %t3.out -lsycl
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t1.out
+// RUN: %clangxx -I %sycl_source_dir %s -o %t3.out -lsycl
 // RUN: env SYCL_DEVICE_TYPE=HOST %t1.out
 // RUN: env SYCL_DEVICE_TYPE=HOST %t3.out
 // RUN: %CPU_RUN_PLACEHOLDER %t1.out
@@ -9,7 +9,8 @@
 #include <CL/sycl.hpp>
 // FIXME do not use internal methods in tests.
 #include <CL/sycl/detail/cg.hpp>
-#include <CL/sycl/detail/kernel_impl.hpp>
+#include <detail/kernel_impl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <algorithm>
 #include <array>
diff --git a/sycl/test/kernel-and-program/cache.cpp b/sycl/test/kernel-and-program/cache.cpp
index 83fd39d493e65..8c89ebe7e6278 100644
--- a/sycl/test/kernel-and-program/cache.cpp
+++ b/sycl/test/kernel-and-program/cache.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %CPU_RUN_PLACEHOLDER %t.out
 //==---------------- cache.cpp - SYCL kernel/program test ------------------==//
 //
@@ -10,7 +10,7 @@
 
 #include <CL/sycl.hpp>
 // FIXME do not use internal methods in tests.
-#include <CL/sycl/detail/program_impl.hpp>
+#include <detail/program_impl.hpp>
 
 namespace RT = cl::sycl::RT;
 namespace detail = cl::sycl::detail;
diff --git a/sycl/test/lit.cfg.py b/sycl/test/lit.cfg.py
index 60437bfd03c73..fc039cb55bf5d 100644
--- a/sycl/test/lit.cfg.py
+++ b/sycl/test/lit.cfg.py
@@ -74,6 +74,7 @@
 config.substitutions.append( ('%llvm_build_libs_dir',  config.llvm_build_libs_dir ) )
 config.substitutions.append( ('%sycl_include',  config.sycl_include ) )
 config.substitutions.append( ('%opencl_libs_dir',  config.opencl_libs_dir) )
+config.substitutions.append( ('%sycl_source_dir', config.sycl_source_dir) )
 
 tools = ['llvm-spirv']
 tool_dirs = [config.llvm_tools_dir]
diff --git a/sycl/test/lit.site.cfg.py.in b/sycl/test/lit.site.cfg.py.in
index da09c97616aac..f1a66894b1f0f 100644
--- a/sycl/test/lit.site.cfg.py.in
+++ b/sycl/test/lit.site.cfg.py.in
@@ -12,6 +12,7 @@ config.llvm_build_bins_dir = "@LLVM_BUILD_BINARY_DIRS@"
 config.llvm_binary_dir = "@LLVM_BINARY_DIR@"
 config.sycl_include = "@SYCL_INCLUDE@"
 config.sycl_obj_root = "@SYCL_BINARY_DIR@"
+config.sycl_source_dir = "@SYCL_SOURCE_DIR@/source"
 config.opencl_libs_dir = os.path.dirname("@OpenCL_LIBRARIES@")
 
 config.llvm_enable_projects = "@LLVM_ENABLE_PROJECTS@"
diff --git a/sycl/test/program_manager/program_manager.cpp b/sycl/test/program_manager/program_manager.cpp
index 6d816bb5b8afa..7f0a63eb0a416 100644
--- a/sycl/test/program_manager/program_manager.cpp
+++ b/sycl/test/program_manager/program_manager.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %CPU_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 // RUN: %ACC_RUN_PLACEHOLDER %t.out
@@ -14,9 +14,10 @@
 #include <CL/sycl.hpp>
 #include <CL/sycl/detail/os_util.hpp>
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/program_manager/program_manager.hpp>
+#include <detail/program_manager/program_manager.hpp>
 
 #include <cassert>
+#include <map>
 
 using namespace cl::sycl;
 
diff --git a/sycl/test/scheduler/BasicSchedulerTests.cpp b/sycl/test/scheduler/BasicSchedulerTests.cpp
index 3f25578d58249..2015b2b9b131f 100644
--- a/sycl/test/scheduler/BasicSchedulerTests.cpp
+++ b/sycl/test/scheduler/BasicSchedulerTests.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: env SYCL_DEVICE_TYPE=HOST %t.out
 // RUN: %CPU_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
diff --git a/sycl/test/scheduler/BlockedCommands.cpp b/sycl/test/scheduler/BlockedCommands.cpp
index 6f5e05c8ec1b8..0cba98584a8c1 100644
--- a/sycl/test/scheduler/BlockedCommands.cpp
+++ b/sycl/test/scheduler/BlockedCommands.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: env SYCL_DEVICE_TYPE=HOST %t.out
 //==------------------- BlockedCommands.cpp --------------------------------==//
 //
@@ -12,6 +12,7 @@
 
 #include <CL/cl.h>
 #include <CL/sycl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 using namespace cl::sycl;
 
diff --git a/sycl/test/scheduler/DataMovement.cpp b/sycl/test/scheduler/DataMovement.cpp
index 6b0edcf9a1faf..d0fd1dd43d86b 100644
--- a/sycl/test/scheduler/DataMovement.cpp
+++ b/sycl/test/scheduler/DataMovement.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %t.out
 //
 //==-------------------------- DataMovement.cpp ----------------------------==//
diff --git a/sycl/test/scheduler/FinishedCmdCleanup.cpp b/sycl/test/scheduler/FinishedCmdCleanup.cpp
index 8546212b2d129..e0f736886b040 100644
--- a/sycl/test/scheduler/FinishedCmdCleanup.cpp
+++ b/sycl/test/scheduler/FinishedCmdCleanup.cpp
@@ -1,6 +1,7 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %t.out
 #include <CL/sycl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <algorithm>
 #include <vector>
diff --git a/sycl/test/scheduler/GetWaitList.cpp b/sycl/test/scheduler/GetWaitList.cpp
index 096f028467469..95c11993ea825 100644
--- a/sycl/test/scheduler/GetWaitList.cpp
+++ b/sycl/test/scheduler/GetWaitList.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %t.out
 //==------------------- GetWaitList.cpp ----------------------------==//
 //
diff --git a/sycl/test/scheduler/HostAcc.cpp b/sycl/test/scheduler/HostAcc.cpp
index dbc6527ae7833..c6d33e465a7db 100644
--- a/sycl/test/scheduler/HostAcc.cpp
+++ b/sycl/test/scheduler/HostAcc.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: env SYCL_THROW_ON_BLOCK=1 %t.out
 //==--------------------------- HostAcc.cpp --------------------------------==//
 //
diff --git a/sycl/test/scheduler/HostAccDestruction.cpp b/sycl/test/scheduler/HostAccDestruction.cpp
index 38a94161d9727..e812816ead5ac 100644
--- a/sycl/test/scheduler/HostAccDestruction.cpp
+++ b/sycl/test/scheduler/HostAccDestruction.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: env SYCL_PI_TRACE=1 %CPU_RUN_PLACEHOLDER %t.out 2>&1 %CPU_CHECK_PLACEHOLDER
 //==---------------------- HostAccDestruction.cpp --------------------------==//
 //
diff --git a/sycl/test/scheduler/LeafLimit.cpp b/sycl/test/scheduler/LeafLimit.cpp
index 39fc6005d2b11..220d846ea6455 100644
--- a/sycl/test/scheduler/LeafLimit.cpp
+++ b/sycl/test/scheduler/LeafLimit.cpp
@@ -1,6 +1,7 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %t.out
 #include <CL/sycl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <algorithm>
 #include <cstddef>
diff --git a/sycl/test/scheduler/MemObjCommandCleanup.cpp b/sycl/test/scheduler/MemObjCommandCleanup.cpp
index e96c771f95c79..8e32985944838 100644
--- a/sycl/test/scheduler/MemObjCommandCleanup.cpp
+++ b/sycl/test/scheduler/MemObjCommandCleanup.cpp
@@ -1,6 +1,7 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %t.out
 #include <CL/sycl.hpp>
+#include <detail/scheduler/scheduler.hpp>
 
 #include <memory>
 #include <utility>
diff --git a/sycl/test/scheduler/MultipleDevices.cpp b/sycl/test/scheduler/MultipleDevices.cpp
index 143c67c479ae8..2e5e965c338bb 100644
--- a/sycl/test/scheduler/MultipleDevices.cpp
+++ b/sycl/test/scheduler/MultipleDevices.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: %t.out
 //===- MultipleDevices.cpp - Test checking multi-device execution --------===//
 //
diff --git a/sycl/test/scheduler/ReleaseResourcesTest.cpp b/sycl/test/scheduler/ReleaseResourcesTest.cpp
index 4e9bf62f1804b..069a25892e534 100644
--- a/sycl/test/scheduler/ReleaseResourcesTest.cpp
+++ b/sycl/test/scheduler/ReleaseResourcesTest.cpp
@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %clangxx -fsycl -I %sycl_source_dir %s -o %t.out
 // RUN: env SYCL_DEVICE_TYPE=HOST %t.out
 // RUN: env SYCL_PI_TRACE=1 %CPU_RUN_PLACEHOLDER %t.out 2>&1 %CPU_CHECK_PLACEHOLDER
 // RUN: env SYCL_PI_TRACE=1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER
diff --git a/sycl/unittests/CMakeLists.txt b/sycl/unittests/CMakeLists.txt
index 4498448871e40..ba6452ee4edcc 100644
--- a/sycl/unittests/CMakeLists.txt
+++ b/sycl/unittests/CMakeLists.txt
@@ -22,7 +22,10 @@ function(add_sycl_unittest test_dirname)
       LLVMTestingSupport
       OpenCL-Headers
     )
-  target_include_directories(${test_dirname} PRIVATE SYSTEM ${sycl_inc_dir})
+  target_include_directories(${test_dirname} PRIVATE SYSTEM
+      ${sycl_inc_dir}
+      ${SYCL_SOURCE_DIR}/source/
+    )
   # LLVM gtest uses LLVM utilities that require C++-14
   # CXX_STANDARD_REQUIRED makes CXX_STANDARD a hard requirement.
   set_target_properties(${test_dirname}
diff --git a/sycl/unittests/pi/PlatformTest.cpp b/sycl/unittests/pi/PlatformTest.cpp
index e100ba50060b6..33a480d53716c 100644
--- a/sycl/unittests/pi/PlatformTest.cpp
+++ b/sycl/unittests/pi/PlatformTest.cpp
@@ -7,7 +7,7 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl/detail/pi.hpp>
-#include <CL/sycl/detail/plugin.hpp>
+#include <detail/plugin.hpp>
 #include <gtest/gtest.h>
 #include <vector>
 
diff --git a/sycl/unittests/thread_safety/ThreadUtils.h b/sycl/unittests/thread_safety/ThreadUtils.h
index 3f9d91b41ea20..18a50bb37a7b9 100644
--- a/sycl/unittests/thread_safety/ThreadUtils.h
+++ b/sycl/unittests/thread_safety/ThreadUtils.h
@@ -1,3 +1,5 @@
+#include <detail/scheduler/scheduler.hpp>
+
 #include <thread>
 #include <vector>