diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt
index 335bae656c3e6..dee5dacd4db9c 100644
--- a/sycl/plugins/unified_runtime/CMakeLists.txt
+++ b/sycl/plugins/unified_runtime/CMakeLists.txt
@@ -99,14 +99,8 @@ if(SYCL_PI_UR_USE_FETCH_CONTENT)
       CACHE PATH "Path to external '${name}' adapter source dir" FORCE)
   endfunction()
 
-  set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git")
-  # commit 905804c2e93dd046140057fd07a5d6191063bedc
-  # Merge: 0a11fb44 d3d3f6e5
-  # Author: Kenneth Benzie (Benie) <k.benzie@codeplay.com>
-  # Date:   Mon May 27 10:34:13 2024 +0100
-  #     Merge pull request #1581 from 0x12CC/l0_cooperative_kernels
-  #     Implement L0 cooperative kernel functions
-  set(UNIFIED_RUNTIME_TAG 905804c2e93dd046140057fd07a5d6191063bedc)
+  set(UNIFIED_RUNTIME_REPO "https://github.com/lbushi25/unified-runtime.git")
+  set(UNIFIED_RUNTIME_TAG 8a90ad7e15deb9fd78027b2c30fe7bce01c649c6)
 
   fetch_adapter_source(level_zero
     ${UNIFIED_RUNTIME_REPO}
diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp
index 4f105a079d3e6..37272f581d3ce 100755
--- a/sycl/test-e2e/USM/align.cpp
+++ b/sycl/test-e2e/USM/align.cpp
@@ -1,14 +1,11 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-// UNSUPPORTED: gpu
-
-// E2E tests for annotated USM allocation functions with alignment arguments
-// that are not powers of 2. Note this test does not work on gpu because some
-// tests expect non-templated aligned_alloc_xxx functions to return nullptr,
-// e.g. when the alignment argument is not a power of 2, while they fail to do
-// so when run on gpu. This maybe because the gpu runtime has different
-// behavior. Therefore, GPU is unsupported until issue #12638 gets resolved.
+// E2E tests for aligned USM allocation functions with different alignment
+// arguments. Depending on the backend the alignment may or may not be supported
+// but either way, according to the SYCL spec, we should see a nullptr if it is
+// not supported or we should verify that the pointer returned is indeed
+// aligned if it is supported.
 
 #include <sycl/detail/core.hpp>
 #include <sycl/usm.hpp>
@@ -29,69 +26,41 @@ template <typename T> void testAlign(sycl::queue &q, unsigned align) {
   assert(align > 0 || (align & (align - 1)) == 0);
 
   auto ADevice = [&](size_t align, auto... args) {
-    return aligned_alloc_device(align, N, args...);
+    auto ptr = aligned_alloc_device(align, N, args...);
+    assert(!ptr || !(reinterpret_cast<std::uintptr_t>(ptr) % align));
+    return 0;
   };
   auto AHost = [&](size_t align, auto... args) {
-    return aligned_alloc_host(align, N, args...);
+    auto ptr = aligned_alloc_host(align, N, args...);
+    assert(!ptr || !(reinterpret_cast<std::uintptr_t>(ptr) % align));
+    return 0;
   };
   auto AShared = [&](size_t align, auto... args) {
-    return aligned_alloc_shared(align, N, args...);
+    void *ptr = nullptr;
+    if (dev.has(aspect::usm_shared_allocations)) {
+      ptr = aligned_alloc_shared(align, N, args...);
+    }
+    assert(!ptr || !(reinterpret_cast<std::uintptr_t>(ptr) % align));
+    return 0;
   };
   auto AAnnotated = [&](size_t align, auto... args) {
-    return aligned_alloc(align, N, args...);
-  };
-
-  auto ATDevice = [&](size_t align, auto... args) {
-    return aligned_alloc_device<T>(align, N, args...);
-  };
-  auto ATHost = [&](size_t align, auto... args) {
-    return aligned_alloc_host<T>(align, N, args...);
-  };
-  auto ATShared = [&](size_t align, auto... args) {
-    return aligned_alloc_shared<T>(align, N, args...);
-  };
-  auto ATAnnotated = [&](size_t align, auto... args) {
-    return aligned_alloc<T>(align, N, args...);
-  };
-
-  // Test cases that are expected to return null
-  auto check_null = [&q](auto AllocFn, int Line, int Case) {
-    decltype(AllocFn()) Ptr = AllocFn();
-    if (Ptr != nullptr) {
-      free(Ptr, q);
-      std::cout << "Failed at line " << Line << ", case " << Case << std::endl;
-      assert(false && "The return is not null!");
-    }
+    auto ptr = aligned_alloc(align, N, args...);
+    assert(!ptr || !(reinterpret_cast<std::uintptr_t>(ptr) % align));
+    return 0;
   };
 
-  auto CheckNullAll = [&](auto Funcs, int Line = __builtin_LINE()) {
-    std::apply(
-        [&](auto... Fs) {
-          int Case = 0;
-          (void)std::initializer_list<int>{
-              (check_null(Fs, Line, Case++), 0)...};
-        },
-        Funcs);
+  auto CheckNullOrAlignedAll = [&](auto Funcs) {
+    std::apply([&](auto... Fs) { (void)std::initializer_list<int>{Fs()...}; },
+               Funcs);
   };
 
-  CheckNullAll(std::tuple{
-      // Case: aligned_alloc_xxx with no alignment property, and the alignment
-      // argument is not a power of 2, the result is nullptr
+  CheckNullOrAlignedAll(std::tuple{
       [&]() { return ADevice(3, q); }, [&]() { return ADevice(5, dev, Ctx); },
       [&]() { return AHost(7, q); }, [&]() { return AHost(9, Ctx); },
       [&]() { return AShared(114, q); },
       [&]() { return AShared(1023, dev, Ctx); },
       [&]() { return AAnnotated(15, q, alloc::device); },
-      [&]() { return AAnnotated(17, dev, Ctx, alloc::host); }
-      // Case: aligned_alloc_xxx<T> with no alignment property, and the
-      // alignment argument is not a power of 2, the result is nullptr
-      ,
-      [&]() { return ATDevice(3, q); }, [&]() { return ATDevice(5, dev, Ctx); },
-      [&]() { return ATHost(7, q); }, [&]() { return ATHost(9, Ctx); },
-      [&]() { return ATShared(1919, q); },
-      [&]() { return ATShared(11, dev, Ctx); },
-      [&]() { return ATAnnotated(15, q, alloc::device); },
-      [&]() { return ATAnnotated(17, dev, Ctx, alloc::host); }});
+      [&]() { return AAnnotated(17, dev, Ctx, alloc::host); }});
 }
 
 int main() {