diff --git a/sycl/include/sycl/ext/oneapi/experimental/detail/ipc_common.hpp b/sycl/include/sycl/ext/oneapi/experimental/detail/ipc_common.hpp
new file mode 100644
index 0000000000000..69e52714deb3b
--- /dev/null
+++ b/sycl/include/sycl/ext/oneapi/experimental/detail/ipc_common.hpp
@@ -0,0 +1,67 @@
+//==------- ipc_common.hpp ------- SYCL inter-process common ---------------==//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+#pragma once
+
+#if (!defined(_HAS_STD_BYTE) || _HAS_STD_BYTE != 0)
+
+#include <sycl/context.hpp>
+#include <sycl/detail/defines_elementary.hpp>
+#include <sycl/detail/export.hpp>
+
+#include <cstddef>
+
+#if __has_include(<span>)
+#include <span>
+#endif
+
+namespace sycl {
+inline namespace _V1 {
+
+namespace ext::oneapi::experimental::ipc {
+struct handle;
+}
+
+namespace ext::oneapi::experimental::ipc::memory {
+__SYCL_EXPORT handle get(void *Ptr, const sycl::context &Ctx);
+__SYCL_EXPORT void put(handle &HandleData, const sycl::context &Ctx);
+} // namespace ext::oneapi::experimental::ipc::memory
+
+namespace ext::oneapi::experimental::ipc {
+
+using handle_data_t = std::vector<std::byte>;
+
+#if __cpp_lib_span
+using handle_data_view_t = std::span<const std::byte, std::dynamic_extent>;
+#endif
+
+struct handle {
+public:
+  handle_data_t data() const { return {MData, MData + MSize}; }
+
+#if __cpp_lib_span
+  handle_data_view_t data_view() const { return {MData, MSize}; }
+#endif
+
+private:
+  handle(void *Data, size_t Size)
+      : MData{reinterpret_cast<std::byte *>(Data)}, MSize{Size} {}
+
+  std::byte *MData;
+  size_t MSize;
+
+  friend __SYCL_EXPORT handle memory::get(void *Ptr, const sycl::context &Ctx);
+  friend __SYCL_EXPORT void memory::put(handle &HandleData,
+                                        const sycl::context &Ctx);
+};
+
+} // namespace ext::oneapi::experimental::ipc
+} // namespace _V1
+} // namespace sycl
+
+#endif
diff --git a/sycl/include/sycl/ext/oneapi/experimental/ipc_memory.hpp b/sycl/include/sycl/ext/oneapi/experimental/ipc_memory.hpp
index 3e931f1418bb7..28b23d4a7f50a 100644
--- a/sycl/include/sycl/ext/oneapi/experimental/ipc_memory.hpp
+++ b/sycl/include/sycl/ext/oneapi/experimental/ipc_memory.hpp
@@ -16,6 +16,10 @@
 #include <sycl/device.hpp>
 #include <sycl/platform.hpp>
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+#include "detail/ipc_common.hpp"
+#endif
+
 #include <cstddef>
 
 #if __has_include(<span>)
@@ -32,7 +36,11 @@ __SYCL_EXPORT void *openIPCMemHandle(const std::byte *HandleData,
                                      const sycl::device &Dev);
 }
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+namespace ext::oneapi::experimental::ipc::memory {
+#else
 namespace ext::oneapi::experimental::ipc_memory {
+#endif
 
 using handle_data_t = std::vector<std::byte>;
 
@@ -40,6 +48,7 @@ using handle_data_t = std::vector<std::byte>;
 using handle_data_view_t = std::span<const std::byte, std::dynamic_extent>;
 #endif
 
+#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
 struct handle {
 public:
   handle_data_t data() const { return {MData, MData + MSize}; }
@@ -58,13 +67,14 @@ struct handle {
   friend __SYCL_EXPORT handle get(void *Ptr, const sycl::context &Ctx);
   friend __SYCL_EXPORT void put(handle &HandleData, const sycl::context &Ctx);
 };
+#endif
 
 __SYCL_EXPORT handle get(void *Ptr, const sycl::context &Ctx);
 
 inline handle get(void *Ptr) {
   sycl::device Dev;
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  return ipc_memory::get(Ptr, Ctx);
+  return get(Ptr, Ctx);
 }
 
 __SYCL_EXPORT void put(handle &HandleData, const sycl::context &Ctx);
@@ -72,7 +82,7 @@ __SYCL_EXPORT void put(handle &HandleData, const sycl::context &Ctx);
 inline void put(handle &HandleData) {
   sycl::device Dev;
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  ipc_memory::put(HandleData, Ctx);
+  put(HandleData, Ctx);
 }
 
 inline void *open(const handle_data_t &HandleData, const sycl::context &Ctx,
@@ -83,13 +93,13 @@ inline void *open(const handle_data_t &HandleData, const sycl::context &Ctx,
 
 inline void *open(handle_data_t HandleData, const sycl::device &Dev) {
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  return ipc_memory::open(HandleData, Ctx, Dev);
+  return open(HandleData, Ctx, Dev);
 }
 
 inline void *open(handle_data_t HandleData) {
   sycl::device Dev;
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  return ipc_memory::open(HandleData, Ctx, Dev);
+  return open(HandleData, Ctx, Dev);
 }
 
 #if __cpp_lib_span
@@ -101,13 +111,13 @@ inline void *open(const handle_data_view_t &HandleDataView,
 
 inline void *open(handle_data_view_t HandleDataView, const sycl::device &Dev) {
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  return ipc_memory::open(HandleDataView, Ctx, Dev);
+  return open(HandleDataView, Ctx, Dev);
 }
 
 inline void *open(handle_data_view_t HandleDataView) {
   sycl::device Dev;
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  return ipc_memory::open(HandleDataView, Ctx, Dev);
+  return open(HandleDataView, Ctx, Dev);
 }
 #endif
 
@@ -116,10 +126,14 @@ __SYCL_EXPORT void close(void *Ptr, const sycl::context &Ctx);
 inline void close(void *Ptr) {
   sycl::device Dev;
   sycl::context Ctx = Dev.get_platform().khr_get_default_context();
-  ipc_memory::close(Ptr, Ctx);
+  close(Ptr, Ctx);
 }
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+} // namespace ext::oneapi::experimental::ipc::memory
+#else
 } // namespace ext::oneapi::experimental::ipc_memory
+#endif
 } // namespace _V1
 } // namespace sycl
 
diff --git a/sycl/source/ipc_memory.cpp b/sycl/source/ipc_memory.cpp
index ade8ab4180c67..186d4f0d2dfb0 100644
--- a/sycl/source/ipc_memory.cpp
+++ b/sycl/source/ipc_memory.cpp
@@ -57,7 +57,11 @@ __SYCL_EXPORT void *openIPCMemHandle(const std::byte *HandleData,
 
 } // namespace detail
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+namespace ext::oneapi::experimental::ipc::memory {
+#else
 namespace ext::oneapi::experimental::ipc_memory {
+#endif
 
 __SYCL_EXPORT handle get(void *Ptr, const sycl::context &Ctx) {
   auto CtxImpl = sycl::detail::getSyclObjImpl(Ctx);
@@ -97,6 +101,10 @@ __SYCL_EXPORT void close(void *Ptr, const sycl::context &Ctx) {
       CtxImpl->getHandleRef(), Ptr);
 }
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+} // namespace ext::oneapi::experimental::ipc::memory
+#else
 } // namespace ext::oneapi::experimental::ipc_memory
+#endif
 } // namespace _V1
 } // namespace sycl
diff --git a/sycl/test-e2e/Experimental/ipc_memory.cpp b/sycl/test-e2e/Experimental/ipc_memory.cpp
index 7366935a0cba4..ad26b020df8e0 100644
--- a/sycl/test-e2e/Experimental/ipc_memory.cpp
+++ b/sycl/test-e2e/Experimental/ipc_memory.cpp
@@ -26,6 +26,14 @@
 
 namespace syclexp = sycl::ext::oneapi::experimental;
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+namespace ipc_memory = syclexp::ipc::memory;
+namespace ipc_memory_common = syclexp::ipc;
+#else
+namespace ipc_memory = syclexp::ipc_memory;
+namespace ipc_memory_common = syclexp::ipc_memory;
+#endif
+
 constexpr size_t N = 32;
 constexpr const char *CommsFile = "ipc_comms.txt";
 
@@ -52,12 +60,12 @@ int spawner(int argc, char *argv[]) {
   {
     // Write handle data to file.
     {
-      syclexp::ipc_memory::handle Handle =
-          syclexp::ipc_memory::get(DataPtr, Q.get_context());
+      ipc_memory_common::handle Handle =
+          ipc_memory::get(DataPtr, Q.get_context());
 #ifdef USE_VIEW
-      syclexp::ipc_memory::handle_data_view_t HandleData = Handle.data_view();
+      ipc_memory_common::handle_data_view_t HandleData = Handle.data_view();
 #else
-      syclexp::ipc_memory::handle_data_t HandleData = Handle.data();
+      ipc_memory_common::handle_data_t HandleData = Handle.data();
 #endif
       size_t HandleDataSize = HandleData.size();
       std::fstream FS(CommsFile, std::ios_base::out | std::ios_base::binary);
@@ -98,13 +106,13 @@ int consumer() {
 
   // Open IPC handle.
 #ifdef USE_VIEW
-  syclexp::ipc_memory::handle_data_view_t Handle{HandleData.get(), HandleSize};
+  ipc_memory_common::handle_data_view_t Handle{HandleData.get(), HandleSize};
 #else
-  syclexp::ipc_memory::handle_data_t Handle{HandleData.get(),
-                                            HandleData.get() + HandleSize};
+  ipc_memory_common::handle_data_t Handle{HandleData.get(),
+                                          HandleData.get() + HandleSize};
 #endif
   int *DataPtr = reinterpret_cast<int *>(
-      syclexp::ipc_memory::open(Handle, Q.get_context(), Q.get_device()));
+      ipc_memory::open(Handle, Q.get_context(), Q.get_device()));
 
   // Test the data already in the USM pointer.
   int Failures = 0;
@@ -123,7 +131,7 @@ int consumer() {
    }).wait();
 
   // Close the IPC pointer.
-  syclexp::ipc_memory::close(DataPtr, Q.get_context());
+  ipc_memory::close(DataPtr, Q.get_context());
 
   return Failures;
 }
diff --git a/sycl/test-e2e/Experimental/ipc_put_after_free.cpp b/sycl/test-e2e/Experimental/ipc_put_after_free.cpp
index 2851a2589eb7f..9425b9278575c 100644
--- a/sycl/test-e2e/Experimental/ipc_put_after_free.cpp
+++ b/sycl/test-e2e/Experimental/ipc_put_after_free.cpp
@@ -21,6 +21,14 @@
 
 namespace syclexp = sycl::ext::oneapi::experimental;
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+namespace ipc_memory = syclexp::ipc::memory;
+namespace ipc_memory_common = syclexp::ipc;
+#else
+namespace ipc_memory = syclexp::ipc_memory;
+namespace ipc_memory_common = syclexp::ipc_memory;
+#endif
+
 int main() {
   sycl::queue Q;
 
@@ -36,14 +44,13 @@ int main() {
 #endif // defined(__linux__)
 
   int *DataPtr = sycl::malloc_device<int>(32, Q);
-  syclexp::ipc_memory::handle Handle =
-      syclexp::ipc_memory::get(DataPtr, Q.get_context());
+  ipc_memory_common::handle Handle = ipc_memory::get(DataPtr, Q.get_context());
 
   // Free data before put.
   sycl::free(DataPtr, Q);
 
   // Try calling put after free.
-  syclexp::ipc_memory::put(Handle, Q.get_context());
+  ipc_memory::put(Handle, Q.get_context());
 
   return 0;
 }
diff --git a/sycl/unittests/Extensions/InterProcessCommunication/Memory.cpp b/sycl/unittests/Extensions/InterProcessCommunication/Memory.cpp
index afcb5d80c4e6d..99c863f93eaa9 100644
--- a/sycl/unittests/Extensions/InterProcessCommunication/Memory.cpp
+++ b/sycl/unittests/Extensions/InterProcessCommunication/Memory.cpp
@@ -16,6 +16,14 @@
 
 namespace syclexp = sycl::ext::oneapi::experimental;
 
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+namespace ipc_memory = syclexp::ipc::memory;
+namespace ipc_memory_common = syclexp::ipc;
+#else
+namespace ipc_memory = syclexp::ipc_memory;
+namespace ipc_memory_common = syclexp::ipc_memory;
+#endif
+
 namespace {
 
 int DummyInt = 42;
@@ -109,9 +117,8 @@ class IPCTests : public ::testing::Test {
 };
 
 TEST_F(IPCTests, IPCGetPutImplicit) {
-  syclexp::ipc_memory::handle IPCMemHandle =
-      syclexp::ipc_memory::get(DummyPtr, Ctxt);
-  syclexp::ipc_memory::handle_data_t IPCMemHandleData = IPCMemHandle.data();
+  ipc_memory_common::handle IPCMemHandle = ipc_memory::get(DummyPtr, Ctxt);
+  ipc_memory_common::handle_data_t IPCMemHandleData = IPCMemHandle.data();
   ASSERT_EQ(IPCMemHandleData.size(), DummyHandleDataSize);
   EXPECT_EQ(std::memcmp(IPCMemHandleData.data(), DummyHandleData,
                         DummyHandleDataSize),
@@ -125,9 +132,8 @@ TEST_F(IPCTests, IPCGetPutImplicit) {
 }
 
 TEST_F(IPCTests, IPCGetPutExplicit) {
-  syclexp::ipc_memory::handle IPCMemHandle =
-      syclexp::ipc_memory::get(DummyPtr, Ctxt);
-  syclexp::ipc_memory::handle_data_t IPCMemHandleData = IPCMemHandle.data();
+  ipc_memory_common::handle IPCMemHandle = ipc_memory::get(DummyPtr, Ctxt);
+  ipc_memory_common::handle_data_t IPCMemHandleData = IPCMemHandle.data();
   ASSERT_EQ(IPCMemHandleData.size(), DummyHandleDataSize);
   EXPECT_EQ(std::memcmp(IPCMemHandleData.data(), DummyHandleData,
                         DummyHandleDataSize),
@@ -139,7 +145,7 @@ TEST_F(IPCTests, IPCGetPutExplicit) {
   EXPECT_EQ(urIPCOpenMemHandleExp_counter, 0);
   EXPECT_EQ(urIPCCloseMemHandleExp_counter, 0);
 
-  syclexp::ipc_memory::put(IPCMemHandle, Ctxt);
+  ipc_memory::put(IPCMemHandle, Ctxt);
 
   // Calling "put" explicitly should call the UR function.
   EXPECT_EQ(urIPCGetMemHandleExp_counter, 1);
@@ -149,10 +155,9 @@ TEST_F(IPCTests, IPCGetPutExplicit) {
 }
 
 TEST_F(IPCTests, IPCOpenClose) {
-  syclexp::ipc_memory::handle_data_t HandleData{
+  ipc_memory_common::handle_data_t HandleData{
       DummyHandleData, DummyHandleData + DummyHandleDataSize};
-  void *Ptr =
-      syclexp::ipc_memory::open(HandleData, Ctxt, Ctxt.get_devices()[0]);
+  void *Ptr = ipc_memory::open(HandleData, Ctxt, Ctxt.get_devices()[0]);
   EXPECT_EQ(Ptr, DummyPtr);
 
   // Opening an IPC handle should call open.
@@ -161,7 +166,7 @@ TEST_F(IPCTests, IPCOpenClose) {
   EXPECT_EQ(urIPCOpenMemHandleExp_counter, 1);
   EXPECT_EQ(urIPCCloseMemHandleExp_counter, 0);
 
-  syclexp::ipc_memory::close(Ptr, Ctxt);
+  ipc_memory::close(Ptr, Ctxt);
 
   // When we close an IPC memory pointer, it should call close.
   EXPECT_EQ(urIPCGetMemHandleExp_counter, 0);
@@ -171,10 +176,9 @@ TEST_F(IPCTests, IPCOpenClose) {
 }
 
 TEST_F(IPCTests, IPCOpenCloseView) {
-  syclexp::ipc_memory::handle_data_view_t HandleDataView{DummyHandleData,
-                                                         DummyHandleDataSize};
-  void *Ptr =
-      syclexp::ipc_memory::open(HandleDataView, Ctxt, Ctxt.get_devices()[0]);
+  ipc_memory_common::handle_data_view_t HandleDataView{DummyHandleData,
+                                                       DummyHandleDataSize};
+  void *Ptr = ipc_memory::open(HandleDataView, Ctxt, Ctxt.get_devices()[0]);
   EXPECT_EQ(Ptr, DummyPtr);
 
   // Opening an IPC handle should call open.
@@ -183,7 +187,7 @@ TEST_F(IPCTests, IPCOpenCloseView) {
   EXPECT_EQ(urIPCOpenMemHandleExp_counter, 1);
   EXPECT_EQ(urIPCCloseMemHandleExp_counter, 0);
 
-  syclexp::ipc_memory::close(Ptr, Ctxt);
+  ipc_memory::close(Ptr, Ctxt);
 
   // When we close an IPC memory pointer, it should call close.
   EXPECT_EQ(urIPCGetMemHandleExp_counter, 0);