Skip to content

Commit b4aaca7

Browse files
KornevNikitaKornevNikita
authored
[SYCL] Add a unittest for is_compatible (#7619)
1 parent 3cf2bfb commit b4aaca7

File tree

6 files changed

+220
-18
lines changed

6 files changed

+220
-18
lines changed

‎sycl/source/detail/program_manager/program_manager.cpp

Lines changed: 13 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -1605,6 +1605,17 @@ void ProgramManager::addOrInitDeviceGlobalEntry(const void *DeviceGlobalPtr,
16051605
m_Ptr2DeviceGlobal.insert({DeviceGlobalPtr, NewEntry.first->second.get()});
16061606
}
16071607

1608+
void ProgramManager::getRawDeviceImages(
1609+
const std::vector<kernel_id> &KernelIDs,
1610+
std::set<RTDeviceBinaryImage *> &BinImages) {
1611+
std::lock_guard<std::mutex> KernelIDsGuard(m_KernelIDsMutex);
1612+
for (const kernel_id &KID : KernelIDs) {
1613+
auto Range = m_KernelIDs2BinImage.equal_range(KID);
1614+
for (auto It = Range.first, End = Range.second; It != End; ++It)
1615+
BinImages.insert(It->second);
1616+
}
1617+
}
1618+
16081619
std::vector<device_image_plain>
16091620
ProgramManager::getSYCLDeviceImagesWithCompatibleState(
16101621
const context &Ctx, const std::vector<device> &Devs,
@@ -1614,12 +1625,7 @@ ProgramManager::getSYCLDeviceImagesWithCompatibleState(
16141625
// TODO: Can we avoid repacking?
16151626
std::set<RTDeviceBinaryImage *> BinImages;
16161627
if (!KernelIDs.empty()) {
1617-
std::lock_guard<std::mutex> KernelIDsGuard(m_KernelIDsMutex);
1618-
for (const kernel_id &KID : KernelIDs) {
1619-
auto Range = m_KernelIDs2BinImage.equal_range(KID);
1620-
for (auto It = Range.first, End = Range.second; It != End; ++It)
1621-
BinImages.insert(It->second);
1622-
}
1628+
getRawDeviceImages(KernelIDs, BinImages);
16231629
} else {
16241630
std::lock_guard<std::mutex> Guard(Sync::getGlobalLock());
16251631
for (auto &ImagesSets : m_DeviceImages) {
@@ -1628,7 +1634,7 @@ ProgramManager::getSYCLDeviceImagesWithCompatibleState(
16281634
BinImages.insert(ImageUPtr.get());
16291635
}
16301636
}
1631-
assert(BinImages.size() > 0 && "Expected to find at least on device image");
1637+
assert(BinImages.size() > 0 && "Expected to find at least one device image");
16321638

16331639
std::vector<device_image_plain> SYCLDeviceImages;
16341640
for (RTDeviceBinaryImage *BinImage : BinImages) {

‎sycl/source/detail/program_manager/program_manager.hpp

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -255,6 +255,9 @@ class ProgramManager {
255255

256256
bool kernelUsesAssert(OSModuleHandle M, const std::string &KernelName) const;
257257

258+
void getRawDeviceImages(const std::vector<kernel_id> &KernelIDs,
259+
std::set<RTDeviceBinaryImage *> &BinImages);
260+
258261
private:
259262
ProgramManager(ProgramManager const &) = delete;
260263
ProgramManager &operator=(ProgramManager const &) = delete;

‎sycl/source/kernel_bundle.cpp

Lines changed: 8 additions & 11 deletions
Original file line numberDiff line numberDiff line change
@@ -296,20 +296,17 @@ std::vector<kernel_id> get_kernel_ids() {
296296
}
297297

298298
bool is_compatible(const std::vector<kernel_id> &KernelIDs, const device &Dev) {
299-
for (const auto &KernelId : KernelIDs) {
300-
const detail::RTDeviceBinaryImage &Img =
301-
detail::ProgramManager::getInstance().getDeviceImage(
302-
detail::OSUtil::ExeModuleHandle, KernelId.get_name(), context(Dev),
303-
Dev);
304-
const detail::RTDeviceBinaryImage::PropertyRange &ARange =
305-
Img.getDeviceRequirements();
306-
for (detail::RTDeviceBinaryImage::PropertyRange::ConstIterator It :
307-
ARange) {
299+
using namespace detail;
300+
std::set<RTDeviceBinaryImage *> BinImages;
301+
ProgramManager::getInstance().getRawDeviceImages(KernelIDs, BinImages);
302+
for (RTDeviceBinaryImage *Img : BinImages) {
303+
const RTDeviceBinaryImage::PropertyRange &PropRange =
304+
Img->getDeviceRequirements();
305+
for (RTDeviceBinaryImage::PropertyRange::ConstIterator It : PropRange) {
308306
using namespace std::literals;
309307
if ((*It)->Name != "aspects"sv)
310308
continue;
311-
detail::ByteArray Aspects =
312-
detail::DeviceBinaryProperty(*It).asByteArray();
309+
ByteArray Aspects = DeviceBinaryProperty(*It).asByteArray();
313310
// Drop 8 bytes describing the size of the byte array
314311
Aspects.dropBytes(8);
315312
while (!Aspects.empty()) {

‎sycl/unittests/SYCL2020/CMakeLists.txt

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -6,5 +6,6 @@ add_sycl_unittest(SYCL2020Tests OBJECT
66
KernelBundle.cpp
77
KernelID.cpp
88
HasExtension.cpp
9+
IsCompatible.cpp
910
)
1011

‎sycl/unittests/SYCL2020/IsCompatible.cpp

Lines changed: 177 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,177 @@
1+
#include <sycl/sycl.hpp>
2+
3+
#include <helpers/PiImage.hpp>
4+
#include <helpers/PiMock.hpp>
5+
6+
#include <gtest/gtest.h>
7+
8+
class TestKernelCPU;
9+
class TestKernelGPU;
10+
class TestKernelACC;
11+
12+
namespace sycl {
13+
__SYCL_INLINE_VER_NAMESPACE(_V1) {
14+
namespace detail {
15+
template <> struct KernelInfo<TestKernelCPU> {
16+
static constexpr unsigned getNumParams() { return 0; }
17+
static const kernel_param_desc_t &getParamDesc(int) {
18+
static kernel_param_desc_t Dummy;
19+
return Dummy;
20+
}
21+
static constexpr const char *getName() { return "TestKernelCPU"; }
22+
static constexpr bool isESIMD() { return false; }
23+
static constexpr bool callsThisItem() { return false; }
24+
static constexpr bool callsAnyThisFreeFunction() { return false; }
25+
static constexpr int64_t getKernelSize() { return 1; }
26+
};
27+
28+
} // namespace detail
29+
} // __SYCL_INLINE_VER_NAMESPACE(_V1)
30+
} // namespace sycl
31+
32+
namespace sycl {
33+
__SYCL_INLINE_VER_NAMESPACE(_V1) {
34+
namespace detail {
35+
template <> struct KernelInfo<TestKernelGPU> {
36+
static constexpr unsigned getNumParams() { return 0; }
37+
static const kernel_param_desc_t &getParamDesc(int) {
38+
static kernel_param_desc_t Dummy;
39+
return Dummy;
40+
}
41+
static constexpr const char *getName() { return "TestKernelGPU"; }
42+
static constexpr bool isESIMD() { return false; }
43+
static constexpr bool callsThisItem() { return false; }
44+
static constexpr bool callsAnyThisFreeFunction() { return false; }
45+
static constexpr int64_t getKernelSize() { return 1; }
46+
};
47+
48+
} // namespace detail
49+
} // __SYCL_INLINE_VER_NAMESPACE(_V1)
50+
} // namespace sycl
51+
52+
namespace sycl {
53+
__SYCL_INLINE_VER_NAMESPACE(_V1) {
54+
namespace detail {
55+
template <> struct KernelInfo<TestKernelACC> {
56+
static constexpr unsigned getNumParams() { return 0; }
57+
static const kernel_param_desc_t &getParamDesc(int) {
58+
static kernel_param_desc_t Dummy;
59+
return Dummy;
60+
}
61+
static constexpr const char *getName() { return "TestKernelACC"; }
62+
static constexpr bool isESIMD() { return false; }
63+
static constexpr bool callsThisItem() { return false; }
64+
static constexpr bool callsAnyThisFreeFunction() { return false; }
65+
static constexpr int64_t getKernelSize() { return 1; }
66+
};
67+
68+
} // namespace detail
69+
} // __SYCL_INLINE_VER_NAMESPACE(_V1)
70+
} // namespace sycl
71+
72+
static sycl::unittest::PiImage
73+
generateDefaultImage(std::initializer_list<std::string> KernelNames,
74+
const std::vector<sycl::aspect> &Aspects) {
75+
using namespace sycl::unittest;
76+
77+
PiPropertySet PropSet;
78+
addAspects(PropSet, Aspects);
79+
80+
std::vector<unsigned char> Bin{0, 1, 2, 3, 4, 5}; // Random data
81+
82+
PiArray<PiOffloadEntry> Entries = makeEmptyKernels(KernelNames);
83+
84+
PiImage Img{PI_DEVICE_BINARY_TYPE_SPIRV, // Format
85+
__SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64, // DeviceTargetSpec
86+
"", // Compile options
87+
"", // Link options
88+
std::move(Bin),
89+
std::move(Entries),
90+
std::move(PropSet)};
91+
92+
return Img;
93+
}
94+
95+
static sycl::unittest::PiImage Imgs[3] = {
96+
generateDefaultImage({"TestKernelCPU"}, {sycl::aspect::cpu}),
97+
generateDefaultImage({"TestKernelGPU"}, {sycl::aspect::gpu}),
98+
generateDefaultImage({"TestKernelACC"}, {sycl::aspect::accelerator})};
99+
100+
static sycl::unittest::PiImageArray<3> ImgArray{Imgs};
101+
102+
static pi_result redefinedDeviceGetInfoCPU(pi_device device,
103+
pi_device_info param_name,
104+
size_t param_value_size,
105+
void *param_value,
106+
size_t *param_value_size_ret) {
107+
if (param_name == PI_DEVICE_INFO_TYPE) {
108+
auto *Result = reinterpret_cast<_pi_device_type *>(param_value);
109+
*Result = PI_DEVICE_TYPE_CPU;
110+
}
111+
return PI_SUCCESS;
112+
}
113+
114+
// Mock device is "GPU" by default, but we need to redefine it just in case
115+
// if there are some changes in the future
116+
static pi_result redefinedDeviceGetInfoGPU(pi_device device,
117+
pi_device_info param_name,
118+
size_t param_value_size,
119+
void *param_value,
120+
size_t *param_value_size_ret) {
121+
if (param_name == PI_DEVICE_INFO_TYPE) {
122+
auto *Result = reinterpret_cast<_pi_device_type *>(param_value);
123+
*Result = PI_DEVICE_TYPE_GPU;
124+
}
125+
return PI_SUCCESS;
126+
}
127+
128+
static pi_result redefinedDeviceGetInfoACC(pi_device device,
129+
pi_device_info param_name,
130+
size_t param_value_size,
131+
void *param_value,
132+
size_t *param_value_size_ret) {
133+
if (param_name == PI_DEVICE_INFO_TYPE) {
134+
auto *Result = reinterpret_cast<_pi_device_type *>(param_value);
135+
*Result = PI_DEVICE_TYPE_ACC;
136+
}
137+
return PI_SUCCESS;
138+
}
139+
140+
TEST(IsCompatible, CPU) {
141+
sycl::unittest::PiMock Mock;
142+
Mock.redefineAfter<sycl::detail::PiApiKind::piDeviceGetInfo>(
143+
redefinedDeviceGetInfoCPU);
144+
sycl::platform Plt = Mock.getPlatform();
145+
const sycl::device Dev = Plt.get_devices()[0];
146+
147+
EXPECT_TRUE(Dev.is_cpu());
148+
EXPECT_TRUE(sycl::is_compatible<TestKernelCPU>(Dev));
149+
EXPECT_FALSE(sycl::is_compatible<TestKernelGPU>(Dev));
150+
EXPECT_FALSE(sycl::is_compatible<TestKernelACC>(Dev));
151+
}
152+
153+
TEST(IsCompatible, GPU) {
154+
sycl::unittest::PiMock Mock;
155+
Mock.redefineAfter<sycl::detail::PiApiKind::piDeviceGetInfo>(
156+
redefinedDeviceGetInfoGPU);
157+
sycl::platform Plt = Mock.getPlatform();
158+
const sycl::device Dev = Plt.get_devices()[0];
159+
160+
EXPECT_TRUE(Dev.is_gpu());
161+
EXPECT_FALSE(sycl::is_compatible<TestKernelCPU>(Dev));
162+
EXPECT_TRUE(sycl::is_compatible<TestKernelGPU>(Dev));
163+
EXPECT_FALSE(sycl::is_compatible<TestKernelACC>(Dev));
164+
}
165+
166+
TEST(IsCompatible, ACC) {
167+
sycl::unittest::PiMock Mock;
168+
Mock.redefineAfter<sycl::detail::PiApiKind::piDeviceGetInfo>(
169+
redefinedDeviceGetInfoACC);
170+
sycl::platform Plt = Mock.getPlatform();
171+
const sycl::device Dev = Plt.get_devices()[0];
172+
173+
EXPECT_TRUE(Dev.is_accelerator());
174+
EXPECT_FALSE(sycl::is_compatible<TestKernelCPU>(Dev));
175+
EXPECT_FALSE(sycl::is_compatible<TestKernelGPU>(Dev));
176+
EXPECT_TRUE(sycl::is_compatible<TestKernelACC>(Dev));
177+
}

‎sycl/unittests/helpers/PiImage.hpp

Lines changed: 18 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -455,6 +455,24 @@ makeKernelParamOptInfo(const std::string &Name, const size_t NumArgs,
455455
return Prop;
456456
}
457457

458+
/// Utility function to add aspects to property set.
459+
inline void addAspects(PiPropertySet &Props,
460+
const std::vector<sycl::aspect> &Aspects) {
461+
const size_t BYTES_FOR_SIZE = 8;
462+
std::vector<char> ValData(BYTES_FOR_SIZE +
463+
Aspects.size() * sizeof(sycl::aspect));
464+
uint64_t ValDataSize = ValData.size();
465+
std::uninitialized_copy(&ValDataSize, &ValDataSize + sizeof(uint64_t),
466+
ValData.data());
467+
auto *AspectsPtr = reinterpret_cast<const unsigned char *>(&Aspects[0]);
468+
std::uninitialized_copy(AspectsPtr, AspectsPtr + Aspects.size(),
469+
ValData.data() + BYTES_FOR_SIZE);
470+
PiProperty Prop{"aspects", ValData, PI_PROPERTY_TYPE_BYTE_ARRAY};
471+
PiArray<PiProperty> Value{std::move(Prop)};
472+
Props.insert(__SYCL_PI_PROPERTY_SET_SYCL_DEVICE_REQUIREMENTS,
473+
std::move(Value));
474+
}
475+
458476
} // namespace unittest
459477
} // __SYCL_INLINE_VER_NAMESPACE(_V1)
460478
} // namespace sycl

0 commit comments

Comments
 (0)