[DeviceSanitizer] Memory overhead statistics

sycl/cmake/modules/FetchUnifiedRuntime.cmake

@@ -116,14 +116,14 @@ if(SYCL_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 b766009add8dc41ad03e2a63be9f6ed40eea4c55 (HEAD, origin/main, origin/HEAD)
-  # Merge: 608e9184 c3c57284
-  # Author: Omar Ahmed <omar.ahmed@codeplay.com>
-  # Date:   Tue Aug 27 16:30:04 2024 +0100
-  #   Merge pull request #2011 from nrspruit/fix_opencl_usm_align_check
-  #   [OpenCL] Fix USM alignment error check to occur always and return nulllptr
-  set(UNIFIED_RUNTIME_TAG b766009add8dc41ad03e2a63be9f6ed40eea4c55)
+  set(UNIFIED_RUNTIME_REPO "https://github.com/AllanZyne/unified-runtime.git")
+  # commit 7e38af77174f92bf4025122b1fa8e3fc49f3ba29
+  # Merge: ddafd29b 6c7d7a76
+  # Author: Kenneth Benzie (Benie) <k.benzie@codeplay.com>
+  # Date:   Mon Jul 15 15:14:47 2024 +0100
+  #     Merge pull request #1863 from kbenzie/benie/main-reverts
+  #     Revert "Merge pull request #1849 from ayylol/cl-subgroupsizes"
+  set(UNIFIED_RUNTIME_TAG review/yang/statistics)
 
   set(UMF_BUILD_EXAMPLES OFF CACHE INTERNAL "EXAMPLES")
   # Due to the use of dependentloadflag and no installer for UMF and hwloc we need

sycl/test-e2e/AddressSanitizer/common/statistics.cpp

@@ -0,0 +1,31 @@
+// REQUIRES: linux
+// RUN: %{build} %device_asan_flags -O2 -g -o %t
+// RUN: %{run} %t 2>&1 | FileCheck %s
+// RUN: env UR_LAYER_ASAN_OPTIONS=print_stats:1 %{run} %t 2>&1 | FileCheck --check-prefixes CHECK-STATS %s
+#include <sycl/usm.hpp>
+
+/// This test is used to check enabling/disabling kernel debug message
+/// We always use "[kernel]" prefix in kernel debug message
+
+constexpr std::size_t N = 4;
+constexpr std::size_t group_size = 1;
+
+int main() {
+  sycl::queue Q;
+  int *array = sycl::malloc_device<int>(1024 * 1024, Q);
+
+  Q.submit([&](sycl::handler &cgh) {
+    auto acc = sycl::local_accessor<int>(group_size, cgh);
+    cgh.parallel_for<class MyKernel>(
+        sycl::nd_range<1>(N, group_size), [=](sycl::nd_item<1> item) {
+          array[item.get_global_id()] = acc[item.get_local_id()];
+        });
+  });
+  Q.wait();
+  // CHECK-STATS: Stats
+  // CHECK-NOT: Stats
+
+  sycl::free(array, Q);
+  std::cout << "PASS" << std::endl;
+  return 0;
+}