[llvm] [libsycl] add heuristics to prefer devices with compatible images (PR #203248)
Kseniya Tikhomirova via llvm-commits
llvm-commits at lists.llvm.org
Thu Jun 11 04:21:43 PDT 2026
https://github.com/KseniyaTikhomirova created https://github.com/llvm/llvm-project/pull/203248
None
>From d797ec4b7ec2a71f9d4b540f50dea8a4e074bc68 Mon Sep 17 00:00:00 2001
From: "Tikhomirova, Kseniya" <kseniya.tikhomirova at intel.com>
Date: Tue, 26 May 2026 06:05:13 -0700
Subject: [PATCH] [libsycl] add heuristics to prefer devices with compatible
images
Signed-off-by: Tikhomirova, Kseniya <kseniya.tikhomirova at intel.com>
---
libsycl/src/detail/program_manager.cpp | 13 ++
libsycl/src/detail/program_manager.hpp | 4 +
libsycl/src/device_selector.cpp | 5 +-
libsycl/unittests/CMakeLists.txt | 1 +
.../unittests/device_selector/CMakeLists.txt | 3 +
.../device_selector/get_device_preference.cpp | 177 ++++++++++++++++++
libsycl/unittests/mock/helpers.hpp | 4 +
7 files changed, 206 insertions(+), 1 deletion(-)
create mode 100644 libsycl/unittests/device_selector/CMakeLists.txt
create mode 100644 libsycl/unittests/device_selector/get_device_preference.cpp
diff --git a/libsycl/src/detail/program_manager.cpp b/libsycl/src/detail/program_manager.cpp
index 79134ffa7a0b0..459860384b6fb 100644
--- a/libsycl/src/detail/program_manager.cpp
+++ b/libsycl/src/detail/program_manager.cpp
@@ -152,6 +152,19 @@ ProgramAndKernelManager::getOrCreateKernel(DeviceKernelInfo &KernelInfo,
return Kernel;
}
+bool ProgramAndKernelManager::hasCompatibleImage(const DeviceImpl &Device) {
+ std::lock_guard<std::mutex> Guard(MDataCollectionMutex);
+
+ for (const auto &BinaryImagesPair : MDeviceImageManagers) {
+ for (const auto &Image : BinaryImagesPair.second) {
+ if (isImageCompatible(*Image, Device))
+ return true;
+ }
+ }
+
+ return false;
+}
+
} // namespace detail
_LIBSYCL_END_NAMESPACE_SYCL
diff --git a/libsycl/src/detail/program_manager.hpp b/libsycl/src/detail/program_manager.hpp
index 9afd6b49bbfd6..828ef9b93aff1 100644
--- a/libsycl/src/detail/program_manager.hpp
+++ b/libsycl/src/detail/program_manager.hpp
@@ -92,6 +92,10 @@ class ProgramAndKernelManager {
/// Release device image managers and corresponding resources.
void releaseResources();
+ /// \return true if at least one registered device image is compatible with
+ /// the given device.
+ bool hasCompatibleImage(const DeviceImpl &Device);
+
protected:
ProgramAndKernelManager() = default;
~ProgramAndKernelManager() = default;
diff --git a/libsycl/src/device_selector.cpp b/libsycl/src/device_selector.cpp
index 86e5f5657c6b5..2744cf7afe35c 100644
--- a/libsycl/src/device_selector.cpp
+++ b/libsycl/src/device_selector.cpp
@@ -10,6 +10,7 @@
#include <sycl/__impl/device_selector.hpp>
#include <detail/device_impl.hpp>
+#include <detail/program_manager.hpp>
#include <algorithm>
@@ -25,7 +26,9 @@ static int getDevicePreference(const device &Device) {
int Score = 0;
const auto &DeviceImpl = detail::getSyclObjImpl(Device);
- // TODO: increase score for devices with compatible program images.
+ auto &ProgramManager = detail::ProgramAndKernelManager::getInstance();
+ if (ProgramManager.hasCompatibleImage(*DeviceImpl))
+ Score += 1000;
if (DeviceImpl->getBackend() == backend::level_zero)
Score += 50;
diff --git a/libsycl/unittests/CMakeLists.txt b/libsycl/unittests/CMakeLists.txt
index 5fa7b6ada0cb4..56bb28161737c 100644
--- a/libsycl/unittests/CMakeLists.txt
+++ b/libsycl/unittests/CMakeLists.txt
@@ -5,6 +5,7 @@ add_custom_target(LibsyclUnitTests)
add_custom_target(check-sycl-unittests)
add_subdirectory(mock)
+add_subdirectory(device_selector)
add_subdirectory(platform)
add_subdirectory(program_manager)
add_subdirectory(queue)
diff --git a/libsycl/unittests/device_selector/CMakeLists.txt b/libsycl/unittests/device_selector/CMakeLists.txt
new file mode 100644
index 0000000000000..195ba15dca1e8
--- /dev/null
+++ b/libsycl/unittests/device_selector/CMakeLists.txt
@@ -0,0 +1,3 @@
+add_sycl_unittest(DeviceSelectorTests
+ get_device_preference.cpp
+)
diff --git a/libsycl/unittests/device_selector/get_device_preference.cpp b/libsycl/unittests/device_selector/get_device_preference.cpp
new file mode 100644
index 0000000000000..a23ecadba0f98
--- /dev/null
+++ b/libsycl/unittests/device_selector/get_device_preference.cpp
@@ -0,0 +1,177 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <mock/helpers.hpp>
+
+#include <common/device_images.hpp>
+
+#include <detail/program_manager.hpp>
+
+#include <detail/device_impl.hpp>
+
+#include <sycl/__impl/detail/obj_utils.hpp>
+#include <sycl/__impl/device_selector.hpp>
+#include <sycl/sycl.hpp>
+
+#include <gmock/gmock.h>
+#include <gtest/gtest.h>
+
+using namespace sycl;
+using namespace ::testing;
+
+namespace {
+
+class ScopedBinaryRegistration {
+public:
+ explicit ScopedBinaryRegistration(llvm::ArrayRef<llvm::StringRef> KernelNames)
+ : MBinary(sycl::unittest::createSYCLDeviceBinary(KernelNames)) {
+ sycl::detail::ProgramAndKernelManager::getInstance().registerFatBin(
+ MBinary.data(), MBinary.size());
+ }
+
+ ~ScopedBinaryRegistration() {
+ sycl::detail::ProgramAndKernelManager::getInstance().unregisterFatBin(
+ MBinary.data(), MBinary.size());
+ }
+
+private:
+ llvm::SmallString<0> MBinary;
+};
+
+class DeviceSelectorScoreTest : public ::testing::Test {
+protected:
+ void SetUp() override {
+ // In reality gpu and cpu devices relate to different platforms. These
+ // tests don't have to follow this rule since selectors work with types.
+ Platform = mock::createDummyHandle<ol_platform_handle_t>();
+ Device1 = mock::createDummyHandleWithData<ol_device_handle_t>(
+ reinterpret_cast<unsigned char *>(&Platform), sizeof(Platform));
+ Device2 = mock::createDummyHandleWithData<ol_device_handle_t>(
+ reinterpret_cast<unsigned char *>(&Platform), sizeof(Platform));
+
+ EXPECT_CALL(Mock.get(), olIterateDevices(_, _))
+ .WillRepeatedly([this](ol_device_iterate_cb_t Callback,
+ void *UserData) -> ol_result_t {
+ std::ignore = Callback(Device1, UserData);
+ std::ignore = Callback(Device2, UserData);
+ return OL_SUCCESS;
+ });
+
+ EXPECT_CALL(Mock.get(), olGetDeviceInfo(_, OL_DEVICE_INFO_PLATFORM, _, _))
+ .WillRepeatedly([this](ol_device_handle_t Device,
+ ol_device_info_t /*PropName*/, size_t PropSize,
+ void *PropValue) -> ol_result_t {
+ *static_cast<ol_platform_handle_t *>(PropValue) = Platform;
+ return OL_SUCCESS;
+ });
+ }
+
+ void TearDown() override {
+ mock::releaseDummyHandles(Platform, Device1, Device2);
+ }
+
+ mock::MockWrapper Mock;
+ ol_platform_handle_t Platform{};
+ ol_device_handle_t Device1{};
+ ol_device_handle_t Device2{};
+};
+
+TEST_F(DeviceSelectorScoreTest, CPUAndGPU) {
+ EXPECT_CALL(Mock.get(), olGetDeviceInfo(_, OL_DEVICE_INFO_TYPE, _, _))
+ .WillRepeatedly([this](ol_device_handle_t Device,
+ ol_device_info_t /*PropName*/, size_t PropSize,
+ void *PropValue) -> ol_result_t {
+ if (Device == Device1)
+ *static_cast<ol_device_type_t *>(PropValue) = OL_DEVICE_TYPE_GPU;
+ else if (Device == Device2)
+ *static_cast<ol_device_type_t *>(PropValue) = OL_DEVICE_TYPE_CPU;
+ else
+ return mock::getMockLiboffload().makeEmptyStrError(
+ OL_ERRC_INVALID_NULL_HANDLE);
+
+ return OL_SUCCESS;
+ });
+
+ auto Devices = sycl::device::get_devices();
+ ASSERT_EQ(Devices.size(), 2u);
+
+ for (const auto &Dev : Devices) {
+ if (Dev.is_gpu()) {
+ EXPECT_EQ(sycl::default_selector_v(Dev), 550);
+ EXPECT_EQ(sycl::gpu_selector_v(Dev), 1050);
+ EXPECT_EQ(sycl::cpu_selector_v(Dev), -1);
+ EXPECT_EQ(sycl::accelerator_selector_v(Dev), -1);
+ } else if (Dev.is_cpu()) {
+ EXPECT_EQ(sycl::default_selector_v(Dev), 350);
+ EXPECT_EQ(sycl::gpu_selector_v(Dev), -1);
+ EXPECT_EQ(sycl::cpu_selector_v(Dev), 1050);
+ EXPECT_EQ(sycl::accelerator_selector_v(Dev), -1);
+ } else
+ FAIL() << "Unexpected device type";
+ }
+}
+
+TEST_F(DeviceSelectorScoreTest, TwoGpusOneCompatibleImage) {
+ EXPECT_CALL(Mock.get(), olGetDeviceInfo(_, OL_DEVICE_INFO_TYPE, _, _))
+ .WillRepeatedly([](ol_device_handle_t Device,
+ ol_device_info_t /*PropName*/, size_t PropSize,
+ void *PropValue) -> ol_result_t {
+ *static_cast<ol_device_type_t *>(PropValue) = OL_DEVICE_TYPE_GPU;
+ return OL_SUCCESS;
+ });
+
+ EXPECT_CALL(Mock.get(), olIsValidBinary(_, _, _, _))
+ .WillRepeatedly([this](ol_device_handle_t Device,
+ const void * /*ProgData*/, size_t /*ProgDataSize*/,
+ bool *Valid) -> ol_result_t {
+ *Valid = (Device == Device2);
+ return OL_SUCCESS;
+ });
+
+ std::array<llvm::StringRef, 1> KernelNames = {"kernel"};
+ ScopedBinaryRegistration Registration{KernelNames};
+
+ auto Devices = sycl::device::get_devices();
+ ASSERT_EQ(Devices.size(), 2u);
+
+ auto DeviceNative = sycl::detail::getSyclObjImpl(Devices[0])->getOLHandle();
+ int Score = sycl::default_selector_v(Devices[0]);
+ if (DeviceNative == Device1)
+ EXPECT_EQ(Score, 550);
+ else if (DeviceNative == Device2)
+ EXPECT_EQ(Score, 1550);
+ else
+ FAIL() << "Unexpected device handle: ";
+
+ sycl::device DefaultDevice{sycl::default_selector_v};
+ auto DeviceDefaultNative =
+ sycl::detail::getSyclObjImpl(DefaultDevice)->getOLHandle();
+ EXPECT_EQ(DeviceDefaultNative, Device2);
+}
+
+TEST(DeviceSelector, AspectSelector) {
+ auto Devices = sycl::device::get_devices();
+ ASSERT_FALSE(Devices.empty());
+
+ const sycl::device &Dev = Devices.front();
+
+ const std::vector<sycl::aspect> EmptyAspects{};
+ const std::vector<sycl::aspect> RequireGpu{sycl::aspect::gpu};
+ const std::vector<sycl::aspect> DenyGpu{sycl::aspect::gpu};
+
+ auto FallbackSelector = sycl::aspect_selector(EmptyAspects, EmptyAspects);
+ EXPECT_EQ(FallbackSelector(Dev), sycl::default_selector_v(Dev));
+
+ auto RequireGpuSelector = sycl::aspect_selector(RequireGpu, EmptyAspects);
+ EXPECT_EQ(RequireGpuSelector(Dev), 1050);
+
+ auto DenyGpuSelector = sycl::aspect_selector(EmptyAspects, DenyGpu);
+ EXPECT_EQ(DenyGpuSelector(Dev), -1);
+}
+
+} // namespace
diff --git a/libsycl/unittests/mock/helpers.hpp b/libsycl/unittests/mock/helpers.hpp
index 77a4b39ca8ced..5de2ac0098f13 100644
--- a/libsycl/unittests/mock/helpers.hpp
+++ b/libsycl/unittests/mock/helpers.hpp
@@ -65,6 +65,10 @@ template <class T> void releaseDummyHandle(T Handle) {
delete DummyHandlePtr;
}
+template <typename... HandleT> void releaseDummyHandles(HandleT... Handles) {
+ (releaseDummyHandle(Handles), ...);
+}
+
class MockLiboffload {
public:
MockLiboffload() { initDefault(); }
More information about the llvm-commits
mailing list