From b5fe80cbb4ab2d38d95a4846757b6e0fe1510437 Mon Sep 17 00:00:00 2001
From: "Agarwal, Udit" <udit.agarwal@intel.com>
Date: Tue, 1 Sep 2026 04:15:35 +0200
Subject: [PATCH] [SYCL][UR] Don't pass addresses of local variables as UR
 handles in unit tests

Three unit tests handed the address of a stack variable to the runtime as a
ur_event_handle_t / native cl_kernel. The mock adapter treats every handle as a
mock::dummy_handle_t_ and increments or decrements its reference counter, so the
runtime's release calls were writing into the tests' stack frames. Use real
dummy handles instead.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
---
 .../Extensions/BindlessImages/Semaphores.cpp  |  8 +++--
 .../handler/SetArgForLocalAccessor.cpp        | 26 +++++++++-------
 .../thread_safety/InteropKernelEnqueue.cpp    | 30 +++++++++++--------
 3 files changed, 39 insertions(+), 25 deletions(-)

diff --git a/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp b/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp
index 16f6aa2917920..0a95bb83d6419 100644
--- a/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp
+++ b/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp
@@ -120,7 +120,7 @@ TEST(BindlessImagesExtensionTests, ExternalSemaphoreSignal) {
 
   // Create a dummy external semaphore and set the raw handle to some dummy.
   // The mock implementation should never access the handle, so this is safe.
-  int DummyInt1 = 0, DummyInt2 = 0;
+  int DummyInt1 = 0;
   syclexp::external_semaphore DummySemaphore{};
   DummySemaphore.raw_handle =
       reinterpret_cast<ur_exp_external_semaphore_handle_t>(&DummyInt1);
@@ -131,8 +131,10 @@ TEST(BindlessImagesExtensionTests, ExternalSemaphoreSignal) {
       *sycl::detail::getSyclObjImpl(Q));
   auto DummyEventImpl2 = sycl::detail::event_impl::create_device_event(
       *sycl::detail::getSyclObjImpl(Q));
-  DummyEventImpl1->setHandle(reinterpret_cast<ur_event_handle_t>(&DummyInt1));
-  DummyEventImpl2->setHandle(reinterpret_cast<ur_event_handle_t>(&DummyInt2));
+  // The runtime releases these handles when the events die, so they must be
+  // real dummy handles rather than arbitrary pointers.
+  DummyEventImpl1->setHandle(mock::createDummyHandle<ur_event_handle_t>());
+  DummyEventImpl2->setHandle(mock::createDummyHandle<ur_event_handle_t>());
   sycl::event DummyEvent1 =
       sycl::detail::createSyclObjFromImpl<sycl::event>(DummyEventImpl1);
   sycl::event DummyEvent2 =
diff --git a/sycl/unittests/handler/SetArgForLocalAccessor.cpp b/sycl/unittests/handler/SetArgForLocalAccessor.cpp
index aef26577bc50b..7aeb1656ee94f 100644
--- a/sycl/unittests/handler/SetArgForLocalAccessor.cpp
+++ b/sycl/unittests/handler/SetArgForLocalAccessor.cpp
@@ -44,17 +44,23 @@ TEST(HandlerSetArg, LocalAccessor) {
   constexpr size_t Size = 128;
   sycl::queue Q;
 
-  ur_native_handle_t handle = mock::createDummyHandle<ur_native_handle_t>();
+  // The mock treats the native handle as a dummy handle and retains/releases
+  // it, so it has to be passed by value and the reference owned by the test
+  // released once the kernel is gone.
+  ur_native_handle_t Handle = mock::createDummyHandle<ur_native_handle_t>();
   auto KernelCL = reinterpret_cast<typename sycl::backend_traits<
-      sycl::backend::opencl>::template input_type<sycl::kernel>>(&handle);
-  auto Kernel =
-      sycl::make_kernel<sycl::backend::opencl>(KernelCL, Q.get_context());
-
-  Q.submit([&](sycl::handler &CGH) {
-     sycl::local_accessor<float, 1> Acc(Size, CGH);
-     CGH.set_arg(0, Acc);
-     CGH.single_task(Kernel);
-   }).wait();
+      sycl::backend::opencl>::template input_type<sycl::kernel>>(Handle);
+  {
+    auto Kernel =
+        sycl::make_kernel<sycl::backend::opencl>(KernelCL, Q.get_context());
+
+    Q.submit([&](sycl::handler &CGH) {
+       sycl::local_accessor<float, 1> Acc(Size, CGH);
+       CGH.set_arg(0, Acc);
+       CGH.single_task(Kernel);
+     }).wait();
+  }
+  mock::releaseDummyHandle(Handle);
 
   ASSERT_EQ(LocalBufferArgSize, Size * sizeof(float));
 }
diff --git a/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp b/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp
index 935ca1eba18ac..de0ae3dd6fe2b 100644
--- a/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp
+++ b/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp
@@ -55,22 +55,28 @@ TEST(KernelEnqueue, InteropKernel) {
   platform Plt = sycl::platform();
   queue Q;
 
+  // The mock treats the native handle as a dummy handle and retains/releases
+  // it, so it has to be passed by value and the reference owned by the test
+  // released once the kernel is gone.
   ur_native_handle_t Handle = mock::createDummyHandle<ur_native_handle_t>();
   auto KernelCL = reinterpret_cast<typename sycl::backend_traits<
-      sycl::backend::opencl>::template input_type<sycl::kernel>>(&Handle);
-  auto Kernel =
-      sycl::make_kernel<sycl::backend::opencl>(KernelCL, Q.get_context());
+      sycl::backend::opencl>::template input_type<sycl::kernel>>(Handle);
+  {
+    auto Kernel =
+        sycl::make_kernel<sycl::backend::opencl>(KernelCL, Q.get_context());
 
-  auto TestLambda = [&](std::size_t ThreadId) {
-    Q.submit([&](sycl::handler &CGH) {
-       for (std::size_t I = 0; I < NArgs; ++I)
-         CGH.set_arg(I, ThreadId);
-       CGH.single_task(Kernel);
-     }).wait();
-  };
+    auto TestLambda = [&](std::size_t ThreadId) {
+      Q.submit([&](sycl::handler &CGH) {
+         for (std::size_t I = 0; I < NArgs; ++I)
+           CGH.set_arg(I, ThreadId);
+         CGH.single_task(Kernel);
+       }).wait();
+    };
 
-  for (std::size_t I = 0; I < LaunchCount; ++I) {
-    ThreadPool Pool(ThreadCount, TestLambda);
+    for (std::size_t I = 0; I < LaunchCount; ++I) {
+      ThreadPool Pool(ThreadCount, TestLambda);
+    }
   }
+  mock::releaseDummyHandle(Handle);
 }
 } // namespace