Bump UR and adjust use of urKernelSuggestMaxCooperativeGroupCountExp

Author: igchor

sycl/cmake/modules/FetchUnifiedRuntime.cmake

@@ -116,7 +116,7 @@ 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")
+  set(UNIFIED_RUNTIME_REPO "https://github.com/igchor/unified-runtime.git")
   include(${CMAKE_CURRENT_SOURCE_DIR}/cmake/modules/UnifiedRuntimeTag.cmake)
 
   set(UMF_BUILD_EXAMPLES OFF CACHE INTERNAL "EXAMPLES")

sycl/cmake/modules/UnifiedRuntimeTag.cmake

@@ -1,7 +1 @@
-# commit 39df0317814c164f5242eda8d6f08550f6268492
-# Merge: 68d93efd be27d8f0
-# Author: Kenneth Benzie (Benie) <[email protected]>
-# Date:   Mon Dec 16 13:53:13 2024 +0000
-#     Merge pull request #2467 from nrspruit/fix_external_import_function_call
-#     [L0] Fix external semaphore import function calls to match the header
-set(UNIFIED_RUNTIME_TAG 39df0317814c164f5242eda8d6f08550f6268492)
+set(UNIFIED_RUNTIME_TAG 4a89e1c69a65acd4f2792743584dfc704086da5e)

sycl/source/detail/kernel_impl.hpp

@@ -320,6 +320,7 @@ kernel_impl::queryMaxNumWorkGroups(queue Queue,
   const auto &Adapter = getAdapter();
   const auto &Handle = getHandleRef();
   auto Device = Queue.get_device();
+  auto DeviceHandleRef = sycl::detail::getSyclObjImpl(Device)->getHandleRef();
 
   size_t WG[Dimensions];
   WG[0] = WorkGroupSize[0];
@@ -331,7 +332,8 @@ kernel_impl::queryMaxNumWorkGroups(queue Queue,
   uint32_t GroupCount{0};
   if (auto Result = Adapter->call_nocheck<
                     UrApiKind::urKernelSuggestMaxCooperativeGroupCountExp>(
-          Handle, Dimensions, WG, DynamicLocalMemorySize, &GroupCount);
+          Handle, DeviceHandleRef, Dimensions, WG, DynamicLocalMemorySize,
+          &GroupCount);
       Result != UR_RESULT_ERROR_UNSUPPORTED_FEATURE &&
       Result != UR_RESULT_ERROR_INVALID_WORK_GROUP_SIZE) {
     // The feature is supported and the group size is valid. Check for other