diff --git a/common/sys/alloc.cpp b/common/sys/alloc.cpp
index b9b36e82f9..71ea84074a 100644
--- a/common/sys/alloc.cpp
+++ b/common/sys/alloc.cpp
@@ -12,50 +12,6 @@
   
 namespace embree
 {
-  size_t total_allocations = 0;
-
-#if defined(EMBREE_SYCL_SUPPORT)
-  
-  __thread sycl::context* tls_context_tutorial = nullptr;
-  __thread sycl::device* tls_device_tutorial = nullptr;
-  
-  __thread sycl::context* tls_context_embree = nullptr;
-  __thread sycl::device* tls_device_embree = nullptr;
-  
-  void enableUSMAllocEmbree(sycl::context* context, sycl::device* device)
-  {
-    if (tls_context_embree != nullptr) throw std::runtime_error("USM allocation already enabled");
-    if (tls_device_embree != nullptr) throw std::runtime_error("USM allocation already enabled");
-    tls_context_embree = context;
-    tls_device_embree = device;
-  }
-
-  void disableUSMAllocEmbree()
-  {
-    if (tls_context_embree  == nullptr) throw std::runtime_error("USM allocation not enabled");
-    if (tls_device_embree  == nullptr) throw std::runtime_error("USM allocation not enabled");
-    tls_context_embree = nullptr;
-    tls_device_embree = nullptr;
-  }
-
-  void enableUSMAllocTutorial(sycl::context* context, sycl::device* device)
-  {
-    //if (tls_context_tutorial != nullptr) throw std::runtime_error("USM allocation already enabled");
-    //if (tls_device_tutorial != nullptr) throw std::runtime_error("USM allocation already enabled");
-    tls_context_tutorial = context;
-    tls_device_tutorial = device;
-  }
-
-  void disableUSMAllocTutorial()
-  {
-    //if (tls_context_tutorial  == nullptr) throw std::runtime_error("USM allocation not enabled");
-    //if (tls_device_tutorial  == nullptr) throw std::runtime_error("USM allocation not enabled");
-    tls_context_tutorial = nullptr;
-    tls_device_tutorial = nullptr;
-  }
-
-#endif
-  
   void* alignedMalloc(size_t size, size_t align)
   {
     if (size == 0)
@@ -86,7 +42,6 @@ namespace embree
       return nullptr;
 
     assert((align & (align-1)) == 0);
-    total_allocations++;    
 
     void* ptr = nullptr;
     if (mode == EMBREE_USM_SHARED_DEVICE_READ_ONLY)
@@ -109,7 +64,6 @@ namespace embree
       return nullptr;
 
     assert((align & (align-1)) == 0);
-    total_allocations++;
 
     void* ptr = nullptr;
     if (type == EmbreeMemoryType::SHARED) {
@@ -134,22 +88,6 @@ namespace embree
     return ptr;
   }
   
-  static MutexSys g_alloc_mutex;
-  
-  void* alignedSYCLMalloc(size_t size, size_t align, EmbreeUSMMode mode)
-  {
-    if (tls_context_tutorial) return alignedSYCLMalloc(tls_context_tutorial, tls_device_tutorial, size, align, mode);
-    if (tls_context_embree  ) return alignedSYCLMalloc(tls_context_embree,   tls_device_embree,   size, align, mode);
-    return nullptr;
-  }
-
-  void* alignedSYCLMalloc(size_t size, size_t align, EmbreeUSMMode mode, EmbreeMemoryType type)
-  {
-    if (tls_context_tutorial) return alignedSYCLMalloc(tls_context_tutorial, tls_device_tutorial, size, align, mode, type);
-    if (tls_context_embree  ) return alignedSYCLMalloc(tls_context_embree,   tls_device_embree,   size, align, mode, type);
-    return nullptr;
-  }
-
   void alignedSYCLFree(sycl::context* context, void* ptr)
   {
     assert(context);
@@ -163,33 +101,7 @@ namespace embree
     }
   }
 
-  void alignedSYCLFree(void* ptr)
-  {
-    if (tls_context_tutorial) return alignedSYCLFree(tls_context_tutorial, ptr);
-    if (tls_context_embree  ) return alignedSYCLFree(tls_context_embree, ptr);
-  }
-
-#endif
-
-  void* alignedUSMMalloc(size_t size, size_t align, EmbreeUSMMode mode)
-  {
-#if defined(EMBREE_SYCL_SUPPORT)
-    if (tls_context_embree || tls_context_tutorial)
-      return alignedSYCLMalloc(size,align,mode);
-    else
-#endif
-      return alignedMalloc(size,align);
-  }
-
-  void alignedUSMFree(void* ptr)
-  {
-#if defined(EMBREE_SYCL_SUPPORT)
-    if (tls_context_embree || tls_context_tutorial)
-      return alignedSYCLFree(ptr);
-    else
 #endif
-      return alignedFree(ptr);
-  }
 
   static bool huge_pages_enabled = false;
   static MutexSys os_init_mutex;
diff --git a/common/sys/alloc.h b/common/sys/alloc.h
index 26da09f896..22fafabec5 100644
--- a/common/sys/alloc.h
+++ b/common/sys/alloc.h
@@ -9,35 +9,22 @@
 
 namespace embree
 {
-#if defined(EMBREE_SYCL_SUPPORT)
-
-  /* enables SYCL USM allocation */
-  void enableUSMAllocEmbree(sycl::context* context, sycl::device* device);
-  void enableUSMAllocTutorial(sycl::context* context, sycl::device* device);
-
-  /* disables SYCL USM allocation */
-  void disableUSMAllocEmbree();
-  void disableUSMAllocTutorial();
-
-#endif
-  
 #define ALIGNED_STRUCT_(align)                                            \
   void* operator new(size_t size) { return alignedMalloc(size,align); }   \
   void operator delete(void* ptr) { alignedFree(ptr); }                   \
   void* operator new[](size_t size) { return alignedMalloc(size,align); } \
   void operator delete[](void* ptr) { alignedFree(ptr); }
   
-#define ALIGNED_STRUCT_USM_(align)                                          \
-  void* operator new(size_t size) { return alignedUSMMalloc(size,align); }   \
-  void operator delete(void* ptr) { alignedUSMFree(ptr); }                   \
-  void* operator new[](size_t size) { return alignedUSMMalloc(size,align); } \
-  void operator delete[](void* ptr) { alignedUSMFree(ptr); }
-  
 #define ALIGNED_CLASS_(align)                                          \
  public:                                                               \
     ALIGNED_STRUCT_(align)                                             \
  private:
 
+  /*! aligned allocation */
+  void* alignedMalloc(size_t size, size_t align);
+  void alignedFree(void* ptr);
+
+
   enum EmbreeUSMMode {
     EMBREE_USM_SHARED = 0,
     EMBREE_USM_SHARED_DEVICE_READ_WRITE = 0,
@@ -50,17 +37,9 @@ namespace embree
     SHARED = 2,
     UNKNOWN = 3
   };
-  
-  /*! aligned allocation */
-  void* alignedMalloc(size_t size, size_t align);
-  void alignedFree(void* ptr);
-
-  /*! aligned allocation using SYCL USM */
-  void* alignedUSMMalloc(size_t size, size_t align = 16, EmbreeUSMMode mode = EMBREE_USM_SHARED_DEVICE_READ_ONLY);
-  void alignedUSMFree(void* ptr);
 
 #if defined(EMBREE_SYCL_SUPPORT)
-  
+
   /*! aligned allocation using SYCL USM */
   void* alignedSYCLMalloc(sycl::context* context, sycl::device* device, size_t size, size_t align, EmbreeUSMMode mode);
   void* alignedSYCLMalloc(sycl::context* context, sycl::device* device, size_t size, size_t align, EmbreeUSMMode mode, EmbreeMemoryType type);
diff --git a/kernels/common/device.cpp b/kernels/common/device.cpp
index e6c9176350..4aa5bd75f7 100644
--- a/kernels/common/device.cpp
+++ b/kernels/common/device.cpp
@@ -720,11 +720,9 @@ namespace embree
   }
 
   void DeviceGPU::enter() {
-    enableUSMAllocEmbree(&gpu_context,&gpu_device);
   }
 
   void DeviceGPU::leave() {
-    disableUSMAllocEmbree();
   }
 
   void* DeviceGPU::malloc(size_t size, size_t align) {
diff --git a/tutorials/common/CMakeLists.txt b/tutorials/common/CMakeLists.txt
index 28e16eaf82..6e39a0707d 100644
--- a/tutorials/common/CMakeLists.txt
+++ b/tutorials/common/CMakeLists.txt
@@ -5,6 +5,7 @@ IF (EMBREE_TUTORIALS_GLFW)
   ADD_SUBDIRECTORY(imgui)
 ENDIF()
 
+ADD_SUBDIRECTORY(alloc)
 ADD_SUBDIRECTORY(tutorial)
 ADD_SUBDIRECTORY(scenegraph)
 ADD_SUBDIRECTORY(lights)
diff --git a/tutorials/common/alloc/CMakeLists.txt b/tutorials/common/alloc/CMakeLists.txt
new file mode 100644
index 0000000000..884fcbec8b
--- /dev/null
+++ b/tutorials/common/alloc/CMakeLists.txt
@@ -0,0 +1,15 @@
+## Copyright 2009-2021 Intel Corporation
+## SPDX-License-Identifier: Apache-2.0
+
+ADD_LIBRARY(alloc_tutorial STATIC alloc.cpp)
+TARGET_LINK_LIBRARIES(alloc_tutorial sys)
+SET_PROPERTY(TARGET alloc_tutorial PROPERTY FOLDER tutorials/common)
+SET_PROPERTY(TARGET alloc_tutorial APPEND PROPERTY COMPILE_FLAGS " ${FLAGS_LOWEST}")
+
+IF (EMBREE_SYCL_SUPPORT)
+  ADD_LIBRARY(alloc_tutorial_sycl STATIC alloc.cpp)
+  TARGET_LINK_LIBRARIES(alloc_tutorial_sycl sys)
+  SET_PROPERTY(TARGET alloc_tutorial_sycl PROPERTY FOLDER tutorials/common)
+  SET_PROPERTY(TARGET alloc_tutorial_sycl APPEND PROPERTY COMPILE_FLAGS " ${FLAGS_LOWEST} ${CMAKE_CXX_FLAGS_SYCL}")
+  TARGET_COMPILE_DEFINITIONS(alloc_tutorial_sycl PUBLIC EMBREE_SYCL_TUTORIAL)
+ENDIF()
\ No newline at end of file
diff --git a/tutorials/common/alloc/alloc.cpp b/tutorials/common/alloc/alloc.cpp
new file mode 100644
index 0000000000..2918c2a381
--- /dev/null
+++ b/tutorials/common/alloc/alloc.cpp
@@ -0,0 +1,51 @@
+// Copyright 2009-2021 Intel Corporation
+// SPDX-License-Identifier: Apache-2.0
+
+#include "alloc.h"
+
+////////////////////////////////////////////////////////////////////////////////
+/// All Platforms
+////////////////////////////////////////////////////////////////////////////////
+  
+namespace embree
+{
+#if defined(EMBREE_SYCL_SUPPORT)
+  
+  __thread sycl::context* tls_context = nullptr;
+  __thread sycl::device* tls_device = nullptr;
+  
+  void enableUSMAllocTutorial(sycl::context* context, sycl::device* device)
+  {
+    tls_context = context;
+    tls_device = device;
+  }
+
+  void disableUSMAllocTutorial()
+  {
+    tls_context = nullptr;
+    tls_device = nullptr;
+  }
+
+#endif
+
+  void* alignedUSMMalloc(size_t size, size_t align, EmbreeUSMMode mode)
+  {
+#if defined(EMBREE_SYCL_SUPPORT)
+    if (tls_context)
+      return alignedSYCLMalloc(tls_context,tls_device,size,align,mode);
+    else
+#endif
+      return alignedMalloc(size,align);
+  }
+
+  void alignedUSMFree(void* ptr)
+  {
+#if defined(EMBREE_SYCL_SUPPORT)
+    if (tls_context)
+      return alignedSYCLFree(tls_context,ptr);
+    else
+#endif
+      return alignedFree(ptr);
+  }
+
+}
\ No newline at end of file
diff --git a/tutorials/common/alloc/alloc.h b/tutorials/common/alloc/alloc.h
new file mode 100644
index 0000000000..0ef037890c
--- /dev/null
+++ b/tutorials/common/alloc/alloc.h
@@ -0,0 +1,34 @@
+// Copyright 2009-2021 Intel Corporation
+// SPDX-License-Identifier: Apache-2.0
+
+#pragma once
+
+#include "../../../common/sys/alloc.h"
+
+#if defined(EMBREE_SYCL_SUPPORT)
+#include <sycl/sycl.hpp>
+#endif
+
+namespace embree
+{
+#if defined(EMBREE_SYCL_SUPPORT)
+
+  /* enables SYCL USM allocation */
+  void enableUSMAllocTutorial(sycl::context* context, sycl::device* device);
+
+  /* disables SYCL USM allocation */
+  void disableUSMAllocTutorial();
+
+#endif
+
+#define ALIGNED_STRUCT_USM_(align)                                          \
+  void* operator new(size_t size) { return alignedUSMMalloc(size,align); }   \
+  void operator delete(void* ptr) { alignedUSMFree(ptr); }                   \
+  void* operator new[](size_t size) { return alignedUSMMalloc(size,align); } \
+  void operator delete[](void* ptr) { alignedUSMFree(ptr); }
+
+  /*! aligned allocation using SYCL USM */
+  void* alignedUSMMalloc(size_t size, size_t align = 16, EmbreeUSMMode mode = EMBREE_USM_SHARED_DEVICE_READ_ONLY);
+  void alignedUSMFree(void* ptr);
+
+}
\ No newline at end of file
diff --git a/tutorials/common/default.h b/tutorials/common/default.h
index d46f25e4b5..cbe69dddce 100644
--- a/tutorials/common/default.h
+++ b/tutorials/common/default.h
@@ -3,6 +3,7 @@
 
 #pragma once
 
+#include "alloc/alloc.h"
 #include "../../kernels/config.h"
 #include "../../common/sys/platform.h"
 #include "../../common/sys/sysinfo.h"
diff --git a/tutorials/common/device_default.h b/tutorials/common/device_default.h
index 94d953f8a6..323449ff0d 100644
--- a/tutorials/common/device_default.h
+++ b/tutorials/common/device_default.h
@@ -29,6 +29,8 @@
 RTC_NAMESPACE_USE
 #include "../../kernels/config.h"
 
+#include "alloc/alloc.h"
+
 namespace embree
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && defined(EMBREE_SYCL_SUPPORT)
diff --git a/tutorials/common/scenegraph/CMakeLists.txt b/tutorials/common/scenegraph/CMakeLists.txt
index adda0d90cf..4d71ffb2ac 100644
--- a/tutorials/common/scenegraph/CMakeLists.txt
+++ b/tutorials/common/scenegraph/CMakeLists.txt
@@ -12,6 +12,6 @@ ADD_LIBRARY(scenegraph STATIC
     scenegraph.cpp
     geometry_creation.cpp)
 
-TARGET_LINK_LIBRARIES(scenegraph sys math lexers image embree)
+TARGET_LINK_LIBRARIES(scenegraph alloc_tutorial sys math lexers image embree)
 SET_PROPERTY(TARGET scenegraph PROPERTY FOLDER tutorials/common)
 SET_PROPERTY(TARGET scenegraph APPEND PROPERTY COMPILE_FLAGS " ${FLAGS_LOWEST}")
diff --git a/tutorials/common/texture/texture2d.cpp b/tutorials/common/texture/texture2d.cpp
index 657e7fe3ed..90b2d99c1e 100644
--- a/tutorials/common/texture/texture2d.cpp
+++ b/tutorials/common/texture/texture2d.cpp
@@ -3,6 +3,8 @@
 
 #include "texture2d.h"
 
+#include "../alloc/alloc.h"
+
 namespace embree {
 
 
diff --git a/tutorials/common/tutorial/CMakeLists.txt b/tutorials/common/tutorial/CMakeLists.txt
index 09e8e01c08..3d5a09f012 100644
--- a/tutorials/common/tutorial/CMakeLists.txt
+++ b/tutorials/common/tutorial/CMakeLists.txt
@@ -6,13 +6,13 @@ IF (EMBREE_TUTORIALS_GLFW)
 ENDIF()
 
 ADD_LIBRARY(tutorial STATIC tutorial.cpp application.cpp scene.cpp tutorial_device.cpp scene_device.cpp)
-TARGET_LINK_LIBRARIES(tutorial sys math lexers scenegraph lights embree tasking ${GUI_LIBRARIES})
+TARGET_LINK_LIBRARIES(tutorial alloc_tutorial sys math lexers scenegraph lights embree tasking ${GUI_LIBRARIES})
 SET_PROPERTY(TARGET tutorial PROPERTY FOLDER tutorials/common)
 SET_PROPERTY(TARGET tutorial APPEND PROPERTY COMPILE_FLAGS " ${FLAGS_LOWEST}")
 
 IF (EMBREE_SYCL_SUPPORT)
   ADD_LIBRARY(tutorial_sycl STATIC tutorial.cpp application.cpp scene.cpp tutorial_device.cpp scene_device.cpp)
-  TARGET_LINK_LIBRARIES(tutorial_sycl sys math lexers scenegraph lights_sycl embree tasking ze_wrapper ${GUI_LIBRARIES})
+  TARGET_LINK_LIBRARIES(tutorial_sycl alloc_tutorial_sycl sys math lexers scenegraph lights_sycl embree tasking ze_wrapper ${GUI_LIBRARIES})
   SET_PROPERTY(TARGET tutorial_sycl PROPERTY FOLDER tutorials/common)
   SET_PROPERTY(TARGET tutorial_sycl APPEND PROPERTY COMPILE_FLAGS " ${FLAGS_LOWEST} ${CMAKE_CXX_FLAGS_SYCL}")
   TARGET_COMPILE_DEFINITIONS(tutorial_sycl PUBLIC EMBREE_SYCL_TUTORIAL)