diff --git a/sycl/plugins/CMakeLists.txt b/sycl/plugins/CMakeLists.txt
old mode 100644
new mode 100755
index 091b7d6b57fa1..85fe986249cb3
--- a/sycl/plugins/CMakeLists.txt
+++ b/sycl/plugins/CMakeLists.txt
@@ -8,3 +8,8 @@ foreach(plugin ${SYCL_ENABLE_PLUGINS})
   add_subdirectory(${plugin})
 endforeach()
 
+# level_zero plugin depends today on unified_runtime plugin
+# and unified_runtime plugin is not an independent plugin, adding it explicitly
+if ("level_zero" IN_LIST SYCL_ENABLE_PLUGINS)
+  add_subdirectory(unified_runtime)
+endif()
diff --git a/sycl/plugins/level_zero/CMakeLists.txt b/sycl/plugins/level_zero/CMakeLists.txt
index 5c599380aaf01..4988a348f2b9a 100755
--- a/sycl/plugins/level_zero/CMakeLists.txt
+++ b/sycl/plugins/level_zero/CMakeLists.txt
@@ -117,6 +117,7 @@ add_sycl_plugin(level_zero
   LIBRARIES
     "${LEVEL_ZERO_LOADER}"
     Threads::Threads
+    unified_runtime
 )
 
 find_package(Python3 REQUIRED)
diff --git a/sycl/plugins/level_zero/pi_level_zero.cpp b/sycl/plugins/level_zero/pi_level_zero.cpp
old mode 100644
new mode 100755
index 708b49fd6e6a3..8a179382a04b2
--- a/sycl/plugins/level_zero/pi_level_zero.cpp
+++ b/sycl/plugins/level_zero/pi_level_zero.cpp
@@ -42,22 +42,6 @@ void disableZeTracing();
 
 namespace {
 
-// Controls Level Zero calls serialization to w/a Level Zero driver being not MT
-// ready. Recognized values (can be used as a bit mask):
-enum {
-  ZeSerializeNone =
-      0, // no locking or blocking (except when SYCL RT requested blocking)
-  ZeSerializeLock = 1, // locking around each ZE_CALL
-  ZeSerializeBlock =
-      2, // blocking ZE calls, where supported (usually in enqueue commands)
-};
-static const pi_uint32 ZeSerialize = [] {
-  const char *SerializeMode = std::getenv("ZE_SERIALIZE");
-  const pi_uint32 SerializeModeValue =
-      SerializeMode ? std::atoi(SerializeMode) : 0;
-  return SerializeModeValue;
-}();
-
 // This is an experimental option to test performance of device to device copy
 // operations on copy engines (versus compute engine)
 static const bool UseCopyEngineForD2DCopy = [] {
@@ -106,30 +90,6 @@ static const bool ReuseDiscardedEvents = [] {
   return std::stoi(ReuseDiscardedEventsFlag) > 0;
 }();
 
-// This class encapsulates actions taken along with a call to Level Zero API.
-class ZeCall {
-private:
-  // The global mutex that is used for total serialization of Level Zero calls.
-  static std::mutex GlobalLock;
-
-public:
-  ZeCall() {
-    if ((ZeSerialize & ZeSerializeLock) != 0) {
-      GlobalLock.lock();
-    }
-  }
-  ~ZeCall() {
-    if ((ZeSerialize & ZeSerializeLock) != 0) {
-      GlobalLock.unlock();
-    }
-  }
-
-  // The non-static version just calls static one.
-  ze_result_t doCall(ze_result_t ZeResult, const char *ZeName,
-                     const char *ZeArgs, bool TraceError = true);
-};
-std::mutex ZeCall::GlobalLock;
-
 // Controls PI level tracing prints.
 static bool PrintPiTrace = false;
 
@@ -139,45 +99,14 @@ static const bool IndirectAccessTrackingEnabled = [] {
          nullptr;
 }();
 
-// Map Level Zero runtime error code to PI error code.
-static pi_result mapError(ze_result_t ZeResult) {
-  // TODO: these mapping need to be clarified and synced with the PI API return
-  // values, which is TBD.
-  static std::unordered_map<ze_result_t, pi_result> ErrorMapping = {
-      {ZE_RESULT_SUCCESS, PI_SUCCESS},
-      {ZE_RESULT_ERROR_DEVICE_LOST, PI_ERROR_DEVICE_NOT_FOUND},
-      {ZE_RESULT_ERROR_INSUFFICIENT_PERMISSIONS, PI_ERROR_INVALID_OPERATION},
-      {ZE_RESULT_ERROR_NOT_AVAILABLE, PI_ERROR_INVALID_OPERATION},
-      {ZE_RESULT_ERROR_UNINITIALIZED, PI_ERROR_INVALID_PLATFORM},
-      {ZE_RESULT_ERROR_INVALID_ARGUMENT, PI_ERROR_INVALID_ARG_VALUE},
-      {ZE_RESULT_ERROR_INVALID_NULL_POINTER, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_INVALID_SIZE, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_UNSUPPORTED_SIZE, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_UNSUPPORTED_ALIGNMENT, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT, PI_ERROR_INVALID_EVENT},
-      {ZE_RESULT_ERROR_INVALID_ENUMERATION, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_UNSUPPORTED_ENUMERATION, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT, PI_ERROR_INVALID_VALUE},
-      {ZE_RESULT_ERROR_INVALID_NATIVE_BINARY, PI_ERROR_INVALID_BINARY},
-      {ZE_RESULT_ERROR_INVALID_KERNEL_NAME, PI_ERROR_INVALID_KERNEL_NAME},
-      {ZE_RESULT_ERROR_INVALID_FUNCTION_NAME, PI_ERROR_BUILD_PROGRAM_FAILURE},
-      {ZE_RESULT_ERROR_OVERLAPPING_REGIONS, PI_ERROR_INVALID_OPERATION},
-      {ZE_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION,
-       PI_ERROR_INVALID_WORK_GROUP_SIZE},
-      {ZE_RESULT_ERROR_MODULE_BUILD_FAILURE, PI_ERROR_BUILD_PROGRAM_FAILURE},
-      {ZE_RESULT_ERROR_OUT_OF_DEVICE_MEMORY, PI_ERROR_OUT_OF_RESOURCES},
-      {ZE_RESULT_ERROR_OUT_OF_HOST_MEMORY, PI_ERROR_OUT_OF_HOST_MEMORY}};
-
-  auto It = ErrorMapping.find(ZeResult);
-  if (It == ErrorMapping.end()) {
-    return PI_ERROR_UNKNOWN;
-  }
-  return It->second;
-}
-
 // This will count the calls to Level-Zero
 static std::map<const char *, int> *ZeCallCount = nullptr;
 
+// Map from L0 to PI result
+static inline pi_result mapError(ze_result_t Result) {
+  return ur2piResult(ze2urResult(Result));
+}
+
 // Trace a call to Level-Zero RT
 #define ZE_CALL(ZeName, ZeArgs)                                                \
   {                                                                            \
@@ -186,9 +115,6 @@ static std::map<const char *, int> *ZeCallCount = nullptr;
       return mapError(Result);                                                 \
   }
 
-#define ZE_CALL_NOCHECK(ZeName, ZeArgs)                                        \
-  ZeCall().doCall(ZeName ZeArgs, #ZeName, #ZeArgs, false)
-
 // Trace an internal PI call; returns in case of an error.
 #define PI_CALL(Call)                                                          \
   {                                                                            \
@@ -199,29 +125,6 @@ static std::map<const char *, int> *ZeCallCount = nullptr;
       return Result;                                                           \
   }
 
-enum DebugLevel {
-  ZE_DEBUG_NONE = 0x0,
-  ZE_DEBUG_BASIC = 0x1,
-  ZE_DEBUG_VALIDATION = 0x2,
-  ZE_DEBUG_CALL_COUNT = 0x4,
-  ZE_DEBUG_ALL = -1
-};
-
-// Controls Level Zero calls tracing.
-static const int ZeDebug = [] {
-  const char *DebugMode = std::getenv("ZE_DEBUG");
-  return DebugMode ? std::atoi(DebugMode) : ZE_DEBUG_NONE;
-}();
-
-static void zePrint(const char *Format, ...) {
-  if (ZeDebug & ZE_DEBUG_BASIC) {
-    va_list Args;
-    va_start(Args, Format);
-    vfprintf(stderr, Format, Args);
-    va_end(Args);
-  }
-}
-
 // Controls if we should choose doing eager initialization
 // to make it happen on warmup paths and have the reportable
 // paths be less likely affected.
@@ -459,10 +362,6 @@ static sycl::detail::SpinLock *PiPlatformsCacheMutex =
     new sycl::detail::SpinLock;
 static bool PiPlatformCachePopulated = false;
 
-// Flags which tell whether various Level Zero extensions are available.
-static bool PiDriverGlobalOffsetExtensionFound = false;
-static bool PiDriverModuleProgramExtensionFound = false;
-
 pi_result
 _pi_context::getFreeSlotInExistingOrNewPool(ze_event_pool_handle_t &Pool,
                                             size_t &Index, bool HostVisible,
@@ -2313,123 +2212,6 @@ static ze_result_t
 checkUnresolvedSymbols(ze_module_handle_t ZeModule,
                        ze_module_build_log_handle_t *ZeBuildLog);
 
-// This function will ensure compatibility with both Linux and Windows for
-// setting environment variables.
-static bool setEnvVar(const char *name, const char *value) {
-#ifdef _WIN32
-  int Res = _putenv_s(name, value);
-#else
-  int Res = setenv(name, value, 1);
-#endif
-  if (Res != 0) {
-    zePrint(
-        "Level Zero plugin was unable to set the environment variable: %s\n",
-        name);
-    return false;
-  }
-  return true;
-}
-
-static class ZeUSMImportExtension {
-  // Pointers to functions that import/release host memory into USM
-  ze_result_t (*zexDriverImportExternalPointer)(ze_driver_handle_t hDriver,
-                                                void *, size_t);
-  ze_result_t (*zexDriverReleaseImportedPointer)(ze_driver_handle_t, void *);
-
-public:
-  // Whether user has requested Import/Release, and platform supports it.
-  bool Enabled;
-
-  ZeUSMImportExtension() : Enabled{false} {}
-
-  void setZeUSMImport(pi_platform Platform) {
-    // Whether env var SYCL_USM_HOSTPTR_IMPORT has been set requesting
-    // host ptr import during buffer creation.
-    const char *USMHostPtrImportStr = std::getenv("SYCL_USM_HOSTPTR_IMPORT");
-    if (!USMHostPtrImportStr || std::atoi(USMHostPtrImportStr) == 0)
-      return;
-
-    // Check if USM hostptr import feature is available.
-    ze_driver_handle_t driverHandle = Platform->ZeDriver;
-    if (ZE_CALL_NOCHECK(zeDriverGetExtensionFunctionAddress,
-                        (driverHandle, "zexDriverImportExternalPointer",
-                         reinterpret_cast<void **>(
-                             &zexDriverImportExternalPointer))) == 0) {
-      ZE_CALL_NOCHECK(
-          zeDriverGetExtensionFunctionAddress,
-          (driverHandle, "zexDriverReleaseImportedPointer",
-           reinterpret_cast<void **>(&zexDriverReleaseImportedPointer)));
-      // Hostptr import/release is turned on because it has been requested
-      // by the env var, and this platform supports the APIs.
-      Enabled = true;
-      // Hostptr import is only possible if piMemBufferCreate receives a
-      // hostptr as an argument. The SYCL runtime passes a host ptr
-      // only when SYCL_HOST_UNIFIED_MEMORY is enabled. Therefore we turn it on.
-      setEnvVar("SYCL_HOST_UNIFIED_MEMORY", "1");
-    }
-  }
-  void doZeUSMImport(ze_driver_handle_t driverHandle, void *HostPtr,
-                     size_t Size) {
-    ZE_CALL_NOCHECK(zexDriverImportExternalPointer,
-                    (driverHandle, HostPtr, Size));
-  }
-  void doZeUSMRelease(ze_driver_handle_t driverHandle, void *HostPtr) {
-    ZE_CALL_NOCHECK(zexDriverReleaseImportedPointer, (driverHandle, HostPtr));
-  }
-} ZeUSMImport;
-
-pi_result _pi_platform::initialize() {
-  // Cache driver properties
-  ZeStruct<ze_driver_properties_t> ZeDriverProperties;
-  ZE_CALL(zeDriverGetProperties, (ZeDriver, &ZeDriverProperties));
-  uint32_t DriverVersion = ZeDriverProperties.driverVersion;
-  // Intel Level-Zero GPU driver stores version as:
-  // | 31 - 24 | 23 - 16 | 15 - 0 |
-  // |  Major  |  Minor  | Build  |
-  auto VersionMajor = std::to_string((DriverVersion & 0xFF000000) >> 24);
-  auto VersionMinor = std::to_string((DriverVersion & 0x00FF0000) >> 16);
-  auto VersionBuild = std::to_string(DriverVersion & 0x0000FFFF);
-  ZeDriverVersion = VersionMajor + "." + VersionMinor + "." + VersionBuild;
-
-  ZE_CALL(zeDriverGetApiVersion, (ZeDriver, &ZeApiVersion));
-  ZeDriverApiVersion = std::to_string(ZE_MAJOR_VERSION(ZeApiVersion)) + "." +
-                       std::to_string(ZE_MINOR_VERSION(ZeApiVersion));
-
-  // Cache driver extension properties
-  uint32_t Count = 0;
-  ZE_CALL(zeDriverGetExtensionProperties, (ZeDriver, &Count, nullptr));
-
-  std::vector<ze_driver_extension_properties_t> zeExtensions(Count);
-
-  ZE_CALL(zeDriverGetExtensionProperties,
-          (ZeDriver, &Count, zeExtensions.data()));
-
-  for (auto extension : zeExtensions) {
-    // Check if global offset extension is available
-    if (strncmp(extension.name, ZE_GLOBAL_OFFSET_EXP_NAME,
-                strlen(ZE_GLOBAL_OFFSET_EXP_NAME) + 1) == 0) {
-      if (extension.version == ZE_GLOBAL_OFFSET_EXP_VERSION_1_0) {
-        PiDriverGlobalOffsetExtensionFound = true;
-      }
-    }
-    // Check if extension is available for "static linking" (compiling multiple
-    // SPIR-V modules together into one Level Zero module).
-    if (strncmp(extension.name, ZE_MODULE_PROGRAM_EXP_NAME,
-                strlen(ZE_MODULE_PROGRAM_EXP_NAME) + 1) == 0) {
-      if (extension.version == ZE_MODULE_PROGRAM_EXP_VERSION_1_0) {
-        PiDriverModuleProgramExtensionFound = true;
-      }
-    }
-    zeDriverExtensionMap[extension.name] = extension.version;
-  }
-
-  // Check if import user ptr into USM feature has been requested.
-  // If yes, then set up L0 API pointers if the platform supports it.
-  ZeUSMImport.setZeUSMImport(this);
-
-  return PI_SUCCESS;
-}
-
 pi_result piPlatformsGet(pi_uint32 NumEntries, pi_platform *Platforms,
                          pi_uint32 *NumPlatforms) {
 
@@ -4886,7 +4668,8 @@ pi_result piProgramLink(pi_context Context, pi_uint32 NumDevices,
     // input module.
     //
     // TODO: Remove this workaround when the driver is fixed.
-    if (!PiDriverModuleProgramExtensionFound || (NumInputPrograms == 1)) {
+    if (!DeviceList[0]->Platform->ZeDriverModuleProgramExtensionFound ||
+        (NumInputPrograms == 1)) {
       if (NumInputPrograms == 1) {
         ZeModuleDesc.pNext = nullptr;
         ZeModuleDesc.inputSize = ZeExtModuleDesc.inputSizes[0];
@@ -5525,7 +5308,7 @@ piEnqueueKernelLaunch(pi_queue Queue, pi_kernel Kernel, pi_uint32 WorkDim,
   std::scoped_lock<pi_shared_mutex, pi_shared_mutex, pi_shared_mutex> Lock(
       Queue->Mutex, Kernel->Mutex, Kernel->Program->Mutex);
   if (GlobalWorkOffset != NULL) {
-    if (!PiDriverGlobalOffsetExtensionFound) {
+    if (!Queue->Device->Platform->ZeDriverGlobalOffsetExtensionFound) {
       zePrint("No global offset extension found on this driver\n");
       return PI_ERROR_INVALID_VALUE;
     }
diff --git a/sycl/plugins/level_zero/pi_level_zero.hpp b/sycl/plugins/level_zero/pi_level_zero.hpp
index 733fc43cea223..7a34d86c9e7ed 100644
--- a/sycl/plugins/level_zero/pi_level_zero.hpp
+++ b/sycl/plugins/level_zero/pi_level_zero.hpp
@@ -44,6 +44,10 @@
 #include <level_zero/zes_api.h>
 #include <sycl/detail/iostream_proxy.hpp>
 
+// Share code between this PI L0 Plugin and UR L0 Adapter
+#include <adapters/level_zero/ur_level_zero.hpp>
+#include <pi2ur.hpp>
+
 #include "usm_allocator.hpp"
 
 template <class To, class From> To pi_cast(From Value) {
@@ -59,191 +63,6 @@ template <> uint32_t pi_cast(uint64_t Value) {
   return CastedValue;
 }
 
-// TODO: Currently die is defined in each plugin. Probably some
-// common header file with utilities should be created.
-[[noreturn]] void die(const char *Message) {
-  std::cerr << "die: " << Message << std::endl;
-  std::terminate();
-}
-
-// Returns the ze_structure_type_t to use in .stype of a structured descriptor.
-// Intentionally not defined; will give an error if no proper specialization
-template <class T> ze_structure_type_t getZeStructureType();
-template <class T> zes_structure_type_t getZesStructureType();
-
-template <> ze_structure_type_t getZeStructureType<ze_event_pool_desc_t>() {
-  return ZE_STRUCTURE_TYPE_EVENT_POOL_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_fence_desc_t>() {
-  return ZE_STRUCTURE_TYPE_FENCE_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_command_list_desc_t>() {
-  return ZE_STRUCTURE_TYPE_COMMAND_LIST_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_context_desc_t>() {
-  return ZE_STRUCTURE_TYPE_CONTEXT_DESC;
-}
-template <>
-ze_structure_type_t
-getZeStructureType<ze_relaxed_allocation_limits_exp_desc_t>() {
-  return ZE_STRUCTURE_TYPE_RELAXED_ALLOCATION_LIMITS_EXP_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_host_mem_alloc_desc_t>() {
-  return ZE_STRUCTURE_TYPE_HOST_MEM_ALLOC_DESC;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_mem_alloc_desc_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_MEM_ALLOC_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_command_queue_desc_t>() {
-  return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_image_desc_t>() {
-  return ZE_STRUCTURE_TYPE_IMAGE_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_module_desc_t>() {
-  return ZE_STRUCTURE_TYPE_MODULE_DESC;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_module_program_exp_desc_t>() {
-  return ZE_STRUCTURE_TYPE_MODULE_PROGRAM_EXP_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_kernel_desc_t>() {
-  return ZE_STRUCTURE_TYPE_KERNEL_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_event_desc_t>() {
-  return ZE_STRUCTURE_TYPE_EVENT_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_sampler_desc_t>() {
-  return ZE_STRUCTURE_TYPE_SAMPLER_DESC;
-}
-template <> ze_structure_type_t getZeStructureType<ze_driver_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DRIVER_PROPERTIES;
-}
-template <> ze_structure_type_t getZeStructureType<ze_device_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_compute_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_COMPUTE_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_command_queue_group_properties_t>() {
-  return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_GROUP_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_image_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_IMAGE_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_module_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_MODULE_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_cache_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_CACHE_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_memory_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_device_memory_access_properties_t>() {
-  return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_ACCESS_PROPERTIES;
-}
-template <> ze_structure_type_t getZeStructureType<ze_module_properties_t>() {
-  return ZE_STRUCTURE_TYPE_MODULE_PROPERTIES;
-}
-template <> ze_structure_type_t getZeStructureType<ze_kernel_properties_t>() {
-  return ZE_STRUCTURE_TYPE_KERNEL_PROPERTIES;
-}
-template <>
-ze_structure_type_t getZeStructureType<ze_memory_allocation_properties_t>() {
-  return ZE_STRUCTURE_TYPE_MEMORY_ALLOCATION_PROPERTIES;
-}
-
-template <> zes_structure_type_t getZesStructureType<zes_pci_properties_t>() {
-  return ZES_STRUCTURE_TYPE_PCI_PROPERTIES;
-}
-
-template <> zes_structure_type_t getZesStructureType<zes_mem_state_t>() {
-  return ZES_STRUCTURE_TYPE_MEM_STATE;
-}
-
-template <> zes_structure_type_t getZesStructureType<zes_mem_properties_t>() {
-  return ZES_STRUCTURE_TYPE_MEM_PROPERTIES;
-}
-
-// The helpers to properly default initialize Level-Zero descriptor and
-// properties structures.
-template <class T> struct ZeStruct : public T {
-  ZeStruct() : T{} { // zero initializes base struct
-    this->stype = getZeStructureType<T>();
-    this->pNext = nullptr;
-  }
-};
-template <class T> struct ZesStruct : public T {
-  ZesStruct() : T{} { // zero initializes base struct
-    this->stype = getZesStructureType<T>();
-    this->pNext = nullptr;
-  }
-};
-
-// A single-threaded app has an opportunity to enable this mode to avoid
-// overhead from mutex locking. Default value is 0 which means that single
-// thread mode is disabled.
-static const bool SingleThreadMode = [] {
-  const char *Ret = std::getenv("SYCL_PI_LEVEL_ZERO_SINGLE_THREAD_MODE");
-  const bool RetVal = Ret ? std::stoi(Ret) : 0;
-  return RetVal;
-}();
-
-// Class which acts like shared_mutex if SingleThreadMode variable is not set.
-// If SingleThreadMode variable is set then mutex operations are turned into
-// nop.
-class pi_shared_mutex : public std::shared_mutex {
-public:
-  void lock() {
-    if (!SingleThreadMode)
-      std::shared_mutex::lock();
-  }
-  bool try_lock() {
-    return SingleThreadMode ? true : std::shared_mutex::try_lock();
-  }
-  void unlock() {
-    if (!SingleThreadMode)
-      std::shared_mutex::unlock();
-  }
-
-  void lock_shared() {
-    if (!SingleThreadMode)
-      std::shared_mutex::lock_shared();
-  }
-  bool try_lock_shared() {
-    return SingleThreadMode ? true : std::shared_mutex::try_lock_shared();
-  }
-  void unlock_shared() {
-    if (!SingleThreadMode)
-      std::shared_mutex::unlock_shared();
-  }
-};
-
-// Class which acts like std::mutex if SingleThreadMode variable is not set.
-// If SingleThreadMode variable is set then mutex operations are turned into
-// nop.
-class pi_mutex : public std::mutex {
-public:
-  void lock() {
-    if (!SingleThreadMode)
-      std::mutex::lock();
-  }
-  bool try_lock() { return SingleThreadMode ? true : std::mutex::try_lock(); }
-  void unlock() {
-    if (!SingleThreadMode)
-      std::mutex::unlock();
-  }
-};
-
 // The wrapper for immutable Level-Zero data.
 // The data is initialized only once at first access (via ->) with the
 // initialization function provided in Init. All subsequent access to
@@ -359,22 +178,13 @@ struct MemAllocRecord : _pi_object {
 // Define the types that are opaque in pi.h in a manner suitabale for Level Zero
 // plugin
 
-struct _pi_platform {
-  _pi_platform(ze_driver_handle_t Driver) : ZeDriver{Driver} {}
-  // Performs initialization of a newly constructed PI platform.
-  pi_result initialize();
-
-  // Level Zero lacks the notion of a platform, but there is a driver, which is
-  // a pretty good fit to keep here.
-  ze_driver_handle_t ZeDriver;
+struct _pi_platform : public _ur_level_zero_platform {
+  _pi_platform(ze_driver_handle_t Driver) : _ur_level_zero_platform{Driver} {}
 
-  // Cache versions info from zeDriverGetProperties.
-  std::string ZeDriverVersion;
-  std::string ZeDriverApiVersion;
-  ze_api_version_t ZeApiVersion{};
-
-  // Cache driver extensions
-  std::unordered_map<std::string, uint32_t> zeDriverExtensionMap;
+  // Performs initialization of a newly constructed PI platform.
+  pi_result initialize() {
+    return ur2piResult(_ur_level_zero_platform::initialize());
+  }
 
   // Cache pi_devices for reuse
   std::vector<std::unique_ptr<_pi_device>> PiDevicesCache;
diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt
new file mode 100755
index 0000000000000..f442db4a4f942
--- /dev/null
+++ b/sycl/plugins/unified_runtime/CMakeLists.txt
@@ -0,0 +1,78 @@
+# PI Unified Runtime plugin library
+#
+
+include(FetchContent)
+
+set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git")
+set(UNIFIED_RUNTIME_TAG fd711c920acc4434cb52ff18b078c082d9d7f44d)
+
+message(STATUS "Will fetch Unified Runtime from ${UNIFIED_RUNTIME_REPO}")
+FetchContent_Declare(unified-runtime
+  GIT_REPOSITORY    ${UNIFIED_RUNTIME_REPO}
+  GIT_TAG           ${UNIFIED_RUNTIME_TAG}
+)
+
+FetchContent_MakeAvailable(unified-runtime)
+FetchContent_GetProperties(unified-runtime)
+
+set(UNIFIED_RUNTIME_SOURCE_DIR
+  ${unified-runtime_SOURCE_DIR} CACHE PATH "Path to Unified Runtime Headers")
+set(UNIFIED_RUNTIME_INCLUDE_DIR "${UNIFIED_RUNTIME_SOURCE_DIR}/include")
+
+
+#include_directories("${LEVEL_ZERO_INCLUDE_DIR}")
+include_directories("${UNIFIED_RUNTIME_INCLUDE_DIR}")
+
+add_library (UnifiedRuntime-Headers INTERFACE)
+target_include_directories(UnifiedRuntime-Headers
+  INTERFACE "${UNIFIED_RUNTIME_INCLUDE_DIR}"
+)
+
+list(APPEND SYCL_TOOLCHAIN_DEPLOY_COMPONENTS unified-runtime)
+
+find_package(Threads REQUIRED)
+
+#
+# NOTE: the Unified Runtime doesn't have the loader [yet].
+# So what we really build is the Unified Runtime with Level Zero Adapter
+# together.
+#
+# TODO: begin
+#       Unified Runtime today is not a plugin yet. It's experimental static
+#       library that's linked into level_zero plugin. As soon as it's ready to
+#       become a plugin code below should be replaced with the following:
+#add_sycl_plugin(unified_runtime
+#  SOURCES
+#    "${sycl_inc_dir}/sycl/detail/pi.h"
+#    "${CMAKE_CURRENT_SOURCE_DIR}/pi2ur.cpp"
+#    "${CMAKE_CURRENT_SOURCE_DIR}/pi2ur.hpp"
+#    "${CMAKE_CURRENT_SOURCE_DIR}/ur.hpp"
+#    "${CMAKE_CURRENT_SOURCE_DIR}/adapters/level_zero/ur_level_zero.hpp"
+#    "${CMAKE_CURRENT_SOURCE_DIR}/adapters/level_zero/ur_level_zero.cpp"
+#   LIBRARIES
+#    Threads::Threads
+#    UnifiedRuntime-Headers
+#    LevelZeroLoader-Headers
+#)
+
+add_library(unified_runtime STATIC
+  "${sycl_inc_dir}/sycl/detail/pi.h"
+  "${CMAKE_CURRENT_SOURCE_DIR}/ur.hpp"
+  "${CMAKE_CURRENT_SOURCE_DIR}/adapters/level_zero/ur_level_zero.hpp"
+  "${CMAKE_CURRENT_SOURCE_DIR}/adapters/level_zero/ur_level_zero.cpp"
+)
+
+target_include_directories(unified_runtime INTERFACE
+  "${UNIFIED_RUNTIME_INCLUDE_DIR}"
+  "${CMAKE_CURRENT_SOURCE_DIR}"
+)
+
+target_link_libraries(unified_runtime PRIVATE
+  UnifiedRuntime-Headers
+  LevelZeroLoader-Headers
+)
+
+# TODO: end
+
+add_dependencies(unified_runtime ze-api)
+
diff --git a/sycl/plugins/unified_runtime/adapters/level_zero/ur_level_zero.cpp b/sycl/plugins/unified_runtime/adapters/level_zero/ur_level_zero.cpp
new file mode 100644
index 0000000000000..71d184d29a8a8
--- /dev/null
+++ b/sycl/plugins/unified_runtime/adapters/level_zero/ur_level_zero.cpp
@@ -0,0 +1,206 @@
+//===--------- ur_level_zero.hpp - Level Zero Adapter -----------------===//
+//
+// 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 <string.h>
+
+#include "ur_level_zero.hpp"
+
+// Define the static class field
+std::mutex ZeCall::GlobalLock;
+
+ZeUSMImportExtension ZeUSMImport;
+
+void zePrint(const char *Format, ...) {
+  if (ZeDebug & ZE_DEBUG_BASIC) {
+    va_list Args;
+    va_start(Args, Format);
+    vfprintf(stderr, Format, Args);
+    va_end(Args);
+  }
+}
+
+// This function will ensure compatibility with both Linux and Windows for
+// setting environment variables.
+bool setEnvVar(const char *name, const char *value) {
+#ifdef _WIN32
+  int Res = _putenv_s(name, value);
+#else
+  int Res = setenv(name, value, 1);
+#endif
+  if (Res != 0) {
+    zePrint(
+        "Level Zero plugin was unable to set the environment variable: %s\n",
+        name);
+    return false;
+  }
+  return true;
+}
+
+// Trace a call to Level-Zero RT
+#define ZE_CALL(ZeName, ZeArgs)                                                \
+  {                                                                            \
+    ze_result_t ZeResult = ZeName ZeArgs;                                      \
+    if (auto Result = ZeCall().doCall(ZeResult, #ZeName, #ZeArgs, true))       \
+      return ze2urResult(Result);                                              \
+  }
+
+// Specializations for various L0 structures
+template <> ze_structure_type_t getZeStructureType<ze_event_pool_desc_t>() {
+  return ZE_STRUCTURE_TYPE_EVENT_POOL_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_fence_desc_t>() {
+  return ZE_STRUCTURE_TYPE_FENCE_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_command_list_desc_t>() {
+  return ZE_STRUCTURE_TYPE_COMMAND_LIST_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_context_desc_t>() {
+  return ZE_STRUCTURE_TYPE_CONTEXT_DESC;
+}
+template <>
+ze_structure_type_t
+getZeStructureType<ze_relaxed_allocation_limits_exp_desc_t>() {
+  return ZE_STRUCTURE_TYPE_RELAXED_ALLOCATION_LIMITS_EXP_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_host_mem_alloc_desc_t>() {
+  return ZE_STRUCTURE_TYPE_HOST_MEM_ALLOC_DESC;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_mem_alloc_desc_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_MEM_ALLOC_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_command_queue_desc_t>() {
+  return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_image_desc_t>() {
+  return ZE_STRUCTURE_TYPE_IMAGE_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_module_desc_t>() {
+  return ZE_STRUCTURE_TYPE_MODULE_DESC;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_module_program_exp_desc_t>() {
+  return ZE_STRUCTURE_TYPE_MODULE_PROGRAM_EXP_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_kernel_desc_t>() {
+  return ZE_STRUCTURE_TYPE_KERNEL_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_event_desc_t>() {
+  return ZE_STRUCTURE_TYPE_EVENT_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_sampler_desc_t>() {
+  return ZE_STRUCTURE_TYPE_SAMPLER_DESC;
+}
+template <> ze_structure_type_t getZeStructureType<ze_driver_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DRIVER_PROPERTIES;
+}
+template <> ze_structure_type_t getZeStructureType<ze_device_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_compute_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_COMPUTE_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_command_queue_group_properties_t>() {
+  return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_GROUP_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_image_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_IMAGE_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_module_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_MODULE_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_cache_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_CACHE_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_memory_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_device_memory_access_properties_t>() {
+  return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_ACCESS_PROPERTIES;
+}
+template <> ze_structure_type_t getZeStructureType<ze_module_properties_t>() {
+  return ZE_STRUCTURE_TYPE_MODULE_PROPERTIES;
+}
+template <> ze_structure_type_t getZeStructureType<ze_kernel_properties_t>() {
+  return ZE_STRUCTURE_TYPE_KERNEL_PROPERTIES;
+}
+template <>
+ze_structure_type_t getZeStructureType<ze_memory_allocation_properties_t>() {
+  return ZE_STRUCTURE_TYPE_MEMORY_ALLOCATION_PROPERTIES;
+}
+
+template <> zes_structure_type_t getZesStructureType<zes_pci_properties_t>() {
+  return ZES_STRUCTURE_TYPE_PCI_PROPERTIES;
+}
+
+template <> zes_structure_type_t getZesStructureType<zes_mem_state_t>() {
+  return ZES_STRUCTURE_TYPE_MEM_STATE;
+}
+
+template <> zes_structure_type_t getZesStructureType<zes_mem_properties_t>() {
+  return ZES_STRUCTURE_TYPE_MEM_PROPERTIES;
+}
+
+zer_result_t _ur_level_zero_platform::initialize() {
+  // Cache driver properties
+  ZeStruct<ze_driver_properties_t> ZeDriverProperties;
+  ZE_CALL(zeDriverGetProperties, (ZeDriver, &ZeDriverProperties));
+  uint32_t DriverVersion = ZeDriverProperties.driverVersion;
+  // Intel Level-Zero GPU driver stores version as:
+  // | 31 - 24 | 23 - 16 | 15 - 0 |
+  // |  Major  |  Minor  | Build  |
+  auto VersionMajor = std::to_string((DriverVersion & 0xFF000000) >> 24);
+  auto VersionMinor = std::to_string((DriverVersion & 0x00FF0000) >> 16);
+  auto VersionBuild = std::to_string(DriverVersion & 0x0000FFFF);
+  ZeDriverVersion = VersionMajor + "." + VersionMinor + "." + VersionBuild;
+
+  ZE_CALL(zeDriverGetApiVersion, (ZeDriver, &ZeApiVersion));
+  ZeDriverApiVersion = std::to_string(ZE_MAJOR_VERSION(ZeApiVersion)) + "." +
+                       std::to_string(ZE_MINOR_VERSION(ZeApiVersion));
+
+  // Cache driver extension properties
+  uint32_t Count = 0;
+  ZE_CALL(zeDriverGetExtensionProperties, (ZeDriver, &Count, nullptr));
+
+  std::vector<ze_driver_extension_properties_t> ZeExtensions(Count);
+
+  ZE_CALL(zeDriverGetExtensionProperties,
+          (ZeDriver, &Count, ZeExtensions.data()));
+
+  for (auto extension : ZeExtensions) {
+    // Check if global offset extension is available
+    if (strncmp(extension.name, ZE_GLOBAL_OFFSET_EXP_NAME,
+                strlen(ZE_GLOBAL_OFFSET_EXP_NAME) + 1) == 0) {
+      if (extension.version == ZE_GLOBAL_OFFSET_EXP_VERSION_1_0) {
+        ZeDriverGlobalOffsetExtensionFound = true;
+      }
+    }
+    // Check if extension is available for "static linking" (compiling multiple
+    // SPIR-V modules together into one Level Zero module).
+    if (strncmp(extension.name, ZE_MODULE_PROGRAM_EXP_NAME,
+                strlen(ZE_MODULE_PROGRAM_EXP_NAME) + 1) == 0) {
+      if (extension.version == ZE_MODULE_PROGRAM_EXP_VERSION_1_0) {
+        ZeDriverModuleProgramExtensionFound = true;
+      }
+    }
+    zeDriverExtensionMap[extension.name] = extension.version;
+  }
+
+  // Check if import user ptr into USM feature has been requested.
+  // If yes, then set up L0 API pointers if the platform supports it.
+  ZeUSMImport.setZeUSMImport(this);
+
+  return ZER_RESULT_SUCCESS;
+}
diff --git a/sycl/plugins/unified_runtime/adapters/level_zero/ur_level_zero.hpp b/sycl/plugins/unified_runtime/adapters/level_zero/ur_level_zero.hpp
new file mode 100755
index 0000000000000..9593d475af33a
--- /dev/null
+++ b/sycl/plugins/unified_runtime/adapters/level_zero/ur_level_zero.hpp
@@ -0,0 +1,218 @@
+//===--------- ur_level_zero.hpp - Level Zero Adapter -----------------===//
+//
+// 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
+
+#include <list>
+#include <map>
+#include <stdarg.h>
+#include <string>
+#include <unordered_map>
+#include <vector>
+
+#include <level_zero/ze_api.h>
+#include <level_zero/zes_api.h>
+#include <ur.hpp>
+#include <zer_api.h>
+
+// Returns the ze_structure_type_t to use in .stype of a structured descriptor.
+// Intentionally not defined; will give an error if no proper specialization
+template <class T> ze_structure_type_t getZeStructureType();
+template <class T> zes_structure_type_t getZesStructureType();
+
+// The helpers to properly default initialize Level-Zero descriptor and
+// properties structures.
+template <class T> struct ZeStruct : public T {
+  ZeStruct() : T{} { // zero initializes base struct
+    this->stype = getZeStructureType<T>();
+    this->pNext = nullptr;
+  }
+};
+
+template <class T> struct ZesStruct : public T {
+  ZesStruct() : T{} { // zero initializes base struct
+    this->stype = getZesStructureType<T>();
+    this->pNext = nullptr;
+  }
+};
+
+// Controls Level Zero calls serialization to w/a Level Zero driver being not MT
+// ready. Recognized values (can be used as a bit mask):
+enum {
+  ZeSerializeNone =
+      0, // no locking or blocking (except when SYCL RT requested blocking)
+  ZeSerializeLock = 1, // locking around each ZE_CALL
+  ZeSerializeBlock =
+      2, // blocking ZE calls, where supported (usually in enqueue commands)
+};
+static const uint32_t ZeSerialize = [] {
+  const char *SerializeMode = std::getenv("ZE_SERIALIZE");
+  const uint32_t SerializeModeValue =
+      SerializeMode ? std::atoi(SerializeMode) : 0;
+  return SerializeModeValue;
+}();
+
+// This class encapsulates actions taken along with a call to Level Zero API.
+class ZeCall {
+private:
+  // The global mutex that is used for total serialization of Level Zero calls.
+  static std::mutex GlobalLock;
+
+public:
+  ZeCall() {
+    if ((ZeSerialize & ZeSerializeLock) != 0) {
+      GlobalLock.lock();
+    }
+  }
+  ~ZeCall() {
+    if ((ZeSerialize & ZeSerializeLock) != 0) {
+      GlobalLock.unlock();
+    }
+  }
+
+  // The non-static version just calls static one.
+  ze_result_t doCall(ze_result_t ZeResult, const char *ZeName,
+                     const char *ZeArgs, bool TraceError = true);
+};
+
+// Map Level Zero runtime error code to UR error code.
+static zer_result_t ze2urResult(ze_result_t ZeResult) {
+  static std::unordered_map<ze_result_t, zer_result_t> ErrorMapping = {
+      {ZE_RESULT_SUCCESS, ZER_RESULT_SUCCESS},
+      {ZE_RESULT_ERROR_DEVICE_LOST, ZER_RESULT_ERROR_DEVICE_LOST},
+      {ZE_RESULT_ERROR_INSUFFICIENT_PERMISSIONS, ZER_RESULT_INVALID_OPERATION},
+      {ZE_RESULT_ERROR_NOT_AVAILABLE, ZER_RESULT_INVALID_OPERATION},
+      {ZE_RESULT_ERROR_UNINITIALIZED, ZER_RESULT_INVALID_PLATFORM},
+      {ZE_RESULT_ERROR_INVALID_ARGUMENT, ZER_RESULT_ERROR_INVALID_ARGUMENT},
+      {ZE_RESULT_ERROR_INVALID_NULL_POINTER, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_INVALID_SIZE, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_UNSUPPORTED_SIZE, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_UNSUPPORTED_ALIGNMENT, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT,
+       ZER_RESULT_INVALID_EVENT},
+      {ZE_RESULT_ERROR_INVALID_ENUMERATION, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_UNSUPPORTED_ENUMERATION, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT, ZER_RESULT_INVALID_VALUE},
+      {ZE_RESULT_ERROR_INVALID_NATIVE_BINARY, ZER_RESULT_INVALID_BINARY},
+      {ZE_RESULT_ERROR_INVALID_KERNEL_NAME, ZER_RESULT_INVALID_KERNEL_NAME},
+      {ZE_RESULT_ERROR_INVALID_FUNCTION_NAME,
+       ZER_RESULT_ERROR_INVALID_FUNCTION_NAME},
+      {ZE_RESULT_ERROR_OVERLAPPING_REGIONS, ZER_RESULT_INVALID_OPERATION},
+      {ZE_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION,
+       ZER_RESULT_INVALID_WORK_GROUP_SIZE},
+      {ZE_RESULT_ERROR_MODULE_BUILD_FAILURE,
+       ZER_RESULT_ERROR_MODULE_BUILD_FAILURE},
+      {ZE_RESULT_ERROR_OUT_OF_DEVICE_MEMORY,
+       ZER_RESULT_ERROR_OUT_OF_DEVICE_MEMORY},
+      {ZE_RESULT_ERROR_OUT_OF_HOST_MEMORY,
+       ZER_RESULT_ERROR_OUT_OF_HOST_MEMORY}};
+
+  auto It = ErrorMapping.find(ZeResult);
+  if (It == ErrorMapping.end()) {
+    return ZER_RESULT_ERROR_UNKNOWN;
+  }
+  return It->second;
+}
+
+// Controls Level Zero calls tracing.
+enum DebugLevel {
+  ZE_DEBUG_NONE = 0x0,
+  ZE_DEBUG_BASIC = 0x1,
+  ZE_DEBUG_VALIDATION = 0x2,
+  ZE_DEBUG_CALL_COUNT = 0x4,
+  ZE_DEBUG_ALL = -1
+};
+
+const int ZeDebug = [] {
+  const char *DebugMode = std::getenv("ZE_DEBUG");
+  return DebugMode ? std::atoi(DebugMode) : ZE_DEBUG_NONE;
+}();
+
+// Prints to stderr if ZE_DEBUG allows it
+void zePrint(const char *Format, ...);
+
+// This function will ensure compatibility with both Linux and Windows for
+// setting environment variables.
+bool setEnvVar(const char *name, const char *value);
+
+// Perform traced call to L0 without checking for errors
+#define ZE_CALL_NOCHECK(ZeName, ZeArgs)                                        \
+  ZeCall().doCall(ZeName ZeArgs, #ZeName, #ZeArgs, false)
+
+struct _ur_level_zero_platform : public _ur_platform {
+  _ur_level_zero_platform(ze_driver_handle_t Driver) : ZeDriver{Driver} {}
+  // Performs initialization of a newly constructed PI platform.
+  zer_result_t initialize();
+
+  // Level Zero lacks the notion of a platform, but there is a driver, which is
+  // a pretty good fit to keep here.
+  ze_driver_handle_t ZeDriver;
+
+  // Cache versions info from zeDriverGetProperties.
+  std::string ZeDriverVersion;
+  std::string ZeDriverApiVersion;
+  ze_api_version_t ZeApiVersion;
+
+  // Cache driver extensions
+  std::unordered_map<std::string, uint32_t> zeDriverExtensionMap;
+
+  // Flags to tell whether various Level Zero platform extensions are available.
+  bool ZeDriverGlobalOffsetExtensionFound{false};
+  bool ZeDriverModuleProgramExtensionFound{false};
+};
+
+using ur_level_zero_platform = _ur_level_zero_platform *;
+
+class ZeUSMImportExtension {
+  // Pointers to functions that import/release host memory into USM
+  ze_result_t (*zexDriverImportExternalPointer)(ze_driver_handle_t hDriver,
+                                                void *, size_t);
+  ze_result_t (*zexDriverReleaseImportedPointer)(ze_driver_handle_t, void *);
+
+public:
+  // Whether user has requested Import/Release, and platform supports it.
+  bool Enabled;
+
+  ZeUSMImportExtension() : Enabled{false} {}
+
+  void setZeUSMImport(ur_level_zero_platform Platform) {
+    // Whether env var SYCL_USM_HOSTPTR_IMPORT has been set requesting
+    // host ptr import during buffer creation.
+    const char *USMHostPtrImportStr = std::getenv("SYCL_USM_HOSTPTR_IMPORT");
+    if (!USMHostPtrImportStr || std::atoi(USMHostPtrImportStr) == 0)
+      return;
+
+    // Check if USM hostptr import feature is available.
+    ze_driver_handle_t DriverHandle = Platform->ZeDriver;
+    if (ZE_CALL_NOCHECK(zeDriverGetExtensionFunctionAddress,
+                        (DriverHandle, "zexDriverImportExternalPointer",
+                         reinterpret_cast<void **>(
+                             &zexDriverImportExternalPointer))) == 0) {
+      ZE_CALL_NOCHECK(
+          zeDriverGetExtensionFunctionAddress,
+          (DriverHandle, "zexDriverReleaseImportedPointer",
+           reinterpret_cast<void **>(&zexDriverReleaseImportedPointer)));
+      // Hostptr import/release is turned on because it has been requested
+      // by the env var, and this platform supports the APIs.
+      Enabled = true;
+      // Hostptr import is only possible if piMemBufferCreate receives a
+      // hostptr as an argument. The SYCL runtime passes a host ptr
+      // only when SYCL_HOST_UNIFIED_MEMORY is enabled. Therefore we turn it on.
+      setEnvVar("SYCL_HOST_UNIFIED_MEMORY", "1");
+    }
+  }
+  void doZeUSMImport(ze_driver_handle_t DriverHandle, void *HostPtr,
+                     size_t Size) {
+    ZE_CALL_NOCHECK(zexDriverImportExternalPointer,
+                    (DriverHandle, HostPtr, Size));
+  }
+  void doZeUSMRelease(ze_driver_handle_t DriverHandle, void *HostPtr) {
+    ZE_CALL_NOCHECK(zexDriverReleaseImportedPointer, (DriverHandle, HostPtr));
+  }
+};
+
+extern ZeUSMImportExtension ZeUSMImport;
diff --git a/sycl/plugins/unified_runtime/pi2ur.cpp b/sycl/plugins/unified_runtime/pi2ur.cpp
new file mode 100644
index 0000000000000..a10ca7534ef14
--- /dev/null
+++ b/sycl/plugins/unified_runtime/pi2ur.cpp
@@ -0,0 +1,41 @@
+//===---------------- pi2ur.cpp - PI API to UR API  --------------------==//
+//
+// 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
+//
+//===------------------------------------------------------------------===//
+
+// This thin layer performs conversion from PI API to Unified Runtime API
+// TODO: remove when SYCL RT is changed to talk in UR directly
+
+#include <pi2ur.hpp>
+#include <ur.hpp>
+
+// Early exits on any error
+#define HANDLE_ERRORS(urCall)                                                  \
+  if (auto Result = urCall)                                                    \
+    return ur2piResult(Result);
+
+__SYCL_EXPORT pi_result piPlatformsGet(pi_uint32 num_entries,
+                                       pi_platform *platforms,
+                                       pi_uint32 *num_platforms) {
+
+  // https://spec.oneapi.io/unified-runtime/latest/core/api.html#zerplatformget
+
+  uint32_t Count = num_entries;
+  auto phPlatforms = reinterpret_cast<zer_platform_handle_t *>(platforms);
+  HANDLE_ERRORS(zerPlatformGet(&Count, phPlatforms));
+  if (*num_platforms) {
+    *num_platforms = Count;
+  }
+  return PI_SUCCESS;
+}
+
+__SYCL_EXPORT pi_result piPlatformGetInfo(pi_platform platform,
+                                          pi_platform_info param_name,
+                                          size_t param_value_size,
+                                          void *param_value,
+                                          size_t *param_value_size_ret) {
+  die("Unified Runtime: piPlatformGetInfo is not implemented");
+}
diff --git a/sycl/plugins/unified_runtime/pi2ur.hpp b/sycl/plugins/unified_runtime/pi2ur.hpp
new file mode 100755
index 0000000000000..d461bac4a2cfb
--- /dev/null
+++ b/sycl/plugins/unified_runtime/pi2ur.hpp
@@ -0,0 +1,44 @@
+//===---------------- pi2ur.cpp - PI API to UR API  --------------------==//
+//
+// 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
+
+#include <unordered_map>
+
+#include "zer_api.h"
+#include <sycl/detail/pi.h>
+
+// Map of UR error codes to PI error codes
+static pi_result ur2piResult(zer_result_t urResult) {
+
+  // TODO: replace "global lifetime" objects with a non-trivial d'tor with
+  // either pointers to such objects (which would be allocated and dealocated
+  // during init and teardown) or objects with trivial d'tor.
+  // E.g. for this case we could have an std::array with sorted values.
+  //
+  static std::unordered_map<zer_result_t, pi_result> ErrorMapping = {
+      {ZER_RESULT_SUCCESS, PI_SUCCESS},
+      {ZER_RESULT_ERROR_DEVICE_LOST, PI_ERROR_DEVICE_NOT_FOUND},
+      {ZER_RESULT_INVALID_OPERATION, PI_ERROR_INVALID_OPERATION},
+      {ZER_RESULT_INVALID_PLATFORM, PI_ERROR_INVALID_PLATFORM},
+      {ZER_RESULT_ERROR_INVALID_ARGUMENT, PI_ERROR_INVALID_ARG_VALUE},
+      {ZER_RESULT_INVALID_VALUE, PI_ERROR_INVALID_VALUE},
+      {ZER_RESULT_INVALID_EVENT, PI_ERROR_INVALID_EVENT},
+      {ZER_RESULT_INVALID_BINARY, PI_ERROR_INVALID_BINARY},
+      {ZER_RESULT_INVALID_KERNEL_NAME, PI_ERROR_INVALID_KERNEL_NAME},
+      {ZER_RESULT_ERROR_INVALID_FUNCTION_NAME, PI_ERROR_BUILD_PROGRAM_FAILURE},
+      {ZER_RESULT_INVALID_WORK_GROUP_SIZE, PI_ERROR_INVALID_WORK_GROUP_SIZE},
+      {ZER_RESULT_ERROR_MODULE_BUILD_FAILURE, PI_ERROR_BUILD_PROGRAM_FAILURE},
+      {ZER_RESULT_ERROR_OUT_OF_DEVICE_MEMORY, PI_ERROR_OUT_OF_RESOURCES},
+      {ZER_RESULT_ERROR_OUT_OF_HOST_MEMORY, PI_ERROR_OUT_OF_HOST_MEMORY}};
+
+  auto It = ErrorMapping.find(urResult);
+  if (It == ErrorMapping.end()) {
+    return PI_ERROR_UNKNOWN;
+  }
+  return It->second;
+}
diff --git a/sycl/plugins/unified_runtime/ur.hpp b/sycl/plugins/unified_runtime/ur.hpp
new file mode 100755
index 0000000000000..63b4a54bb1ffd
--- /dev/null
+++ b/sycl/plugins/unified_runtime/ur.hpp
@@ -0,0 +1,80 @@
+//===--------- ur.hpp - Unified Runtime  -----------------------------===//
+//
+// 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
+
+#include <iostream>
+#include <mutex>
+#include <shared_mutex>
+
+#include <zer_api.h>
+
+// Terminates the process with a catastrophic error message.
+[[noreturn]] inline void die(const char *Message) {
+  std::cerr << "die: " << Message << std::endl;
+  std::terminate();
+}
+
+// A single-threaded app has an opportunity to enable this mode to avoid
+// overhead from mutex locking. Default value is 0 which means that single
+// thread mode is disabled.
+static const bool SingleThreadMode = [] {
+  const char *Ret = std::getenv("SYCL_PI_LEVEL_ZERO_SINGLE_THREAD_MODE");
+  const bool RetVal = Ret ? std::stoi(Ret) : 0;
+  return RetVal;
+}();
+
+// Class which acts like shared_mutex if SingleThreadMode variable is not set.
+// If SingleThreadMode variable is set then mutex operations are turned into
+// nop.
+class pi_shared_mutex {
+    std::shared_mutex Mutex;
+public:
+  void lock() {
+    if (!SingleThreadMode)
+      Mutex.lock();
+  }
+  bool try_lock() {
+    return SingleThreadMode ? true : Mutex.try_lock();
+  }
+  void unlock() {
+    if (!SingleThreadMode)
+      Mutex.unlock();
+  }
+
+  void lock_shared() {
+    if (!SingleThreadMode)
+      Mutex.lock_shared();
+  }
+  bool try_lock_shared() {
+    return SingleThreadMode ? true : Mutex.try_lock_shared();
+  }
+  void unlock_shared() {
+    if (!SingleThreadMode)
+      Mutex.unlock_shared();
+  }
+};
+
+// Class which acts like std::mutex if SingleThreadMode variable is not set.
+// If SingleThreadMode variable is set then mutex operations are turned into
+// nop.
+class pi_mutex {
+  std::mutex Mutex;
+public:
+  void lock() {
+    if (!SingleThreadMode)
+      Mutex.lock();
+  }
+  bool try_lock() { return SingleThreadMode ? true : Mutex.try_lock(); }
+  void unlock() {
+    if (!SingleThreadMode)
+      Mutex.unlock();
+  }
+};
+
+// TODO: populate with target agnostic handling of UR platforms
+struct _ur_platform {};