Skip to content

Commit 8335e17

Browse files
romanovvladromanovvlad
authored
[SYCL] Add implementation of kernel_bundle. Part 4 (#3464)
The patch: 1. Adds minimal interfaces for SYCL2020 style exceptions 2. Adds diagnostics in kernel_bundle related APIs 3. Adds hash functions for new classes 4. Refactors has_kernel_bundle API
1 parent f110dd4 commit 8335e17

File tree

9 files changed

+432
-33
lines changed

9 files changed

+432
-33
lines changed

‎sycl/include/CL/sycl/exception.hpp

Lines changed: 40 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -30,6 +30,12 @@ class __SYCL_EXPORT exception : public std::exception {
3030
public:
3131
exception() = default;
3232

33+
exception(std::error_code, const char *Msg)
34+
: exception(Msg, PI_INVALID_VALUE) {}
35+
36+
exception(std::error_code, const std::string &Msg)
37+
: exception(Msg, PI_INVALID_VALUE) {}
38+
3339
const char *what() const noexcept final;
3440

3541
bool has_context() const;
@@ -111,5 +117,39 @@ class feature_not_supported : public device_error {
111117
using device_error::device_error;
112118
};
113119

120+
enum class errc : unsigned int {
121+
runtime = 0,
122+
kernel = 1,
123+
accessor = 2,
124+
nd_range = 3,
125+
event = 4,
126+
kernel_argument = 5,
127+
build = 6,
128+
invalid = 7,
129+
memory_allocation = 8,
130+
platform = 9,
131+
profiling = 10,
132+
feature_not_supported = 11,
133+
kernel_not_supported = 12,
134+
backend_mismatch = 13,
135+
};
136+
137+
/// Constructs an error code using e and sycl_category()
138+
__SYCL_EXPORT std::error_code make_error_code(sycl::errc E) noexcept;
139+
140+
__SYCL_EXPORT const std::error_category &sycl_category() noexcept;
141+
142+
namespace detail {
143+
class __SYCL_EXPORT SYCLCategory : public std::error_category {
144+
public:
145+
const char *name() const noexcept override { return "SYCL"; }
146+
std::string message(int) const override { return "SYCL Error"; }
147+
};
148+
} // namespace detail
149+
114150
} // namespace sycl
115151
} // __SYCL_INLINE_NAMESPACE(cl)
152+
153+
namespace std {
154+
template <> struct is_error_condition_enum<cl::sycl::errc> : true_type {};
155+
} // namespace std

‎sycl/include/CL/sycl/kernel_bundle.hpp

Lines changed: 28 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -553,7 +553,7 @@ compile(const kernel_bundle<bundle_state::input> &InputBundle,
553553
/////////////////////////
554554

555555
namespace detail {
556-
std::vector<sycl::device> find_device_intersection(
556+
__SYCL_EXPORT std::vector<sycl::device> find_device_intersection(
557557
const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles);
558558

559559
__SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
@@ -628,3 +628,30 @@ build(const kernel_bundle<bundle_state::input> &InputBundle,
628628

629629
} // namespace sycl
630630
} // __SYCL_INLINE_NAMESPACE(cl)
631+
632+
namespace std {
633+
template <> struct hash<cl::sycl::kernel_id> {
634+
size_t operator()(const cl::sycl::kernel_id &KernelID) const {
635+
return hash<cl::sycl::shared_ptr_class<cl::sycl::detail::kernel_id_impl>>()(
636+
cl::sycl::detail::getSyclObjImpl(KernelID));
637+
}
638+
};
639+
640+
template <cl::sycl::bundle_state State>
641+
struct hash<cl::sycl::device_image<State>> {
642+
size_t operator()(const cl::sycl::device_image<State> &DeviceImage) const {
643+
return hash<
644+
cl::sycl::shared_ptr_class<cl::sycl::detail::device_image_impl>>()(
645+
cl::sycl::detail::getSyclObjImpl(DeviceImage));
646+
}
647+
};
648+
649+
template <cl::sycl::bundle_state State>
650+
struct hash<cl::sycl::kernel_bundle<State>> {
651+
size_t operator()(const cl::sycl::kernel_bundle<State> &KernelBundle) const {
652+
return hash<
653+
cl::sycl::shared_ptr_class<cl::sycl::detail::kernel_bundle_impl>>()(
654+
cl::sycl::detail::getSyclObjImpl(KernelBundle));
655+
}
656+
};
657+
} // namespace std

‎sycl/source/detail/kernel_bundle_impl.hpp

Lines changed: 116 additions & 16 deletions
Original file line numberDiff line numberDiff line change
@@ -28,20 +28,58 @@ namespace sycl {
2828
namespace detail {
2929

3030
template <class T> struct LessByHash {
31-
bool operator()(const T &LHS, const T &RHS) {
31+
bool operator()(const T &LHS, const T &RHS) const {
3232
return getSyclObjImpl(LHS) < getSyclObjImpl(RHS);
3333
}
3434
};
3535

36+
static bool checkAllDevicesAreInContext(const std::vector<device> &Devices,
37+
const context &Context) {
38+
const std::vector<device> &ContextDevices = Context.get_devices();
39+
return std::all_of(
40+
Devices.begin(), Devices.end(), [&ContextDevices](const device &Dev) {
41+
return ContextDevices.end() !=
42+
std::find(ContextDevices.begin(), ContextDevices.end(), Dev);
43+
});
44+
}
45+
46+
static bool checkAllDevicesHaveAspect(const std::vector<device> &Devices,
47+
aspect Aspect) {
48+
return std::all_of(Devices.begin(), Devices.end(),
49+
[&Aspect](const device &Dev) { return Dev.has(Aspect); });
50+
}
51+
3652
// The class is an impl counterpart of the sycl::kernel_bundle.
3753
// It provides an access and utilities to manage set of sycl::device_images
3854
// objects.
3955
class kernel_bundle_impl {
4056

57+
void common_ctor_checks(bundle_state State) {
58+
const bool AllDevicesInTheContext =
59+
checkAllDevicesAreInContext(MDevices, MContext);
60+
if (MDevices.empty() || !AllDevicesInTheContext)
61+
throw sycl::exception(
62+
make_error_code(errc::invalid),
63+
"Not all devices are associated with the context or "
64+
"vector of devices is empty");
65+
66+
if (bundle_state::input == State &&
67+
!checkAllDevicesHaveAspect(MDevices, aspect::online_compiler))
68+
throw sycl::exception(make_error_code(errc::invalid),
69+
"Not all devices have aspect::online_compiler");
70+
71+
if (bundle_state::object == State &&
72+
!checkAllDevicesHaveAspect(MDevices, aspect::online_linker))
73+
throw sycl::exception(make_error_code(errc::invalid),
74+
"Not all devices have aspect::online_linker");
75+
}
76+
4177
public:
4278
kernel_bundle_impl(context Ctx, std::vector<device> Devs, bundle_state State)
4379
: MContext(std::move(Ctx)), MDevices(std::move(Devs)) {
4480

81+
common_ctor_checks(State);
82+
4583
MDeviceImages = detail::ProgramManager::getInstance().getSYCLDeviceImages(
4684
MContext, MDevices, State);
4785
}
@@ -54,6 +92,21 @@ class kernel_bundle_impl {
5492
bundle_state TargetState)
5593
: MContext(InputBundle.get_context()), MDevices(std::move(Devs)) {
5694

95+
const std::vector<device> &InputBundleDevices =
96+
getSyclObjImpl(InputBundle)->get_devices();
97+
const bool AllDevsAssociatedWithInputBundle =
98+
std::all_of(MDevices.begin(), MDevices.end(),
99+
[&InputBundleDevices](const device &Dev) {
100+
return InputBundleDevices.end() !=
101+
std::find(InputBundleDevices.begin(),
102+
InputBundleDevices.end(), Dev);
103+
});
104+
if (MDevices.empty() || !AllDevsAssociatedWithInputBundle)
105+
throw sycl::exception(
106+
make_error_code(errc::invalid),
107+
"Not all devices are in the set of associated "
108+
"devices for input bundle or vector of devices is empty");
109+
57110
for (const device_image_plain &DeviceImage : InputBundle) {
58111
// Skip images which are not compatible with devices provided
59112
if (std::none_of(
@@ -85,7 +138,39 @@ class kernel_bundle_impl {
85138
kernel_bundle_impl(
86139
const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
87140
std::vector<device> Devs, const property_list &PropList)
88-
: MContext(ObjectBundles[0].get_context()), MDevices(std::move(Devs)) {
141+
: MDevices(std::move(Devs)) {
142+
143+
if (ObjectBundles.empty())
144+
return;
145+
146+
MContext = ObjectBundles[0].get_context();
147+
for (size_t I = 1; I < ObjectBundles.size(); ++I) {
148+
if (ObjectBundles[I].get_context() != MContext)
149+
throw sycl::exception(
150+
make_error_code(errc::invalid),
151+
"Not all input bundles have the same associated context");
152+
}
153+
154+
// Check if any of the devices in devs are not in the set of associated
155+
// devices for any of the bundles in ObjectBundles
156+
const bool AllDevsAssociatedWithInputBundles = std::all_of(
157+
MDevices.begin(), MDevices.end(), [&ObjectBundles](const device &Dev) {
158+
// Number of devices is expected to be small
159+
return std::all_of(
160+
ObjectBundles.begin(), ObjectBundles.end(),
161+
[&Dev](const kernel_bundle<bundle_state::object> &KernelBundle) {
162+
const std::vector<device> &BundleDevices =
163+
getSyclObjImpl(KernelBundle)->get_devices();
164+
return BundleDevices.end() != std::find(BundleDevices.begin(),
165+
BundleDevices.end(),
166+
Dev);
167+
});
168+
});
169+
if (MDevices.empty() || !AllDevsAssociatedWithInputBundles)
170+
throw sycl::exception(
171+
make_error_code(errc::invalid),
172+
"Not all devices are in the set of associated "
173+
"devices for input bundles or vector of devices is empty");
89174

90175
// TODO: Unify with c'tor for sycl::comile and sycl::build by calling
91176
// sycl::join on vector of kernel_bundles
@@ -116,6 +201,10 @@ class kernel_bundle_impl {
116201
bundle_state State)
117202
: MContext(std::move(Ctx)), MDevices(std::move(Devs)) {
118203

204+
// TODO: Add a check that all kernel ids are compatible with at least one
205+
// device in Devs
206+
common_ctor_checks(State);
207+
119208
MDeviceImages = detail::ProgramManager::getInstance().getSYCLDeviceImages(
120209
MContext, MDevices, KernelIDs, State);
121210
}
@@ -124,24 +213,36 @@ class kernel_bundle_impl {
124213
const DevImgSelectorImpl &Selector, bundle_state State)
125214
: MContext(std::move(Ctx)), MDevices(std::move(Devs)) {
126215

216+
common_ctor_checks(State);
217+
127218
MDeviceImages = detail::ProgramManager::getInstance().getSYCLDeviceImages(
128219
MContext, MDevices, Selector, State);
129220
}
130221

131222
// C'tor matches sycl::join API
132223
kernel_bundle_impl(const std::vector<detail::KernelBundleImplPtr> &Bundles) {
224+
if (Bundles.empty())
225+
return;
226+
133227
MContext = Bundles[0]->MContext;
228+
MDevices = Bundles[0]->MDevices;
229+
for (size_t I = 1; I < Bundles.size(); ++I) {
230+
if (Bundles[I]->MContext != MContext)
231+
throw sycl::exception(
232+
make_error_code(errc::invalid),
233+
"Not all input bundles have the same associated context.");
234+
if (Bundles[I]->MDevices != MDevices)
235+
throw sycl::exception(
236+
make_error_code(errc::invalid),
237+
"Not all input bundles have the same set of associated devices.");
238+
}
239+
134240
for (const detail::KernelBundleImplPtr &Bundle : Bundles) {
135-
MDevices.insert(MDevices.end(), Bundle->MDevices.begin(),
136-
Bundle->MDevices.end());
241+
137242
MDeviceImages.insert(MDeviceImages.end(), Bundle->MDeviceImages.begin(),
138243
Bundle->MDeviceImages.end());
139244
}
140245

141-
std::sort(MDevices.begin(), MDevices.end(), LessByHash<device>{});
142-
const auto DevIt = std::unique(MDevices.begin(), MDevices.end());
143-
MDevices.erase(DevIt, MDevices.end());
144-
145246
std::sort(MDeviceImages.begin(), MDeviceImages.end(),
146247
LessByHash<device_image_plain>{});
147248
const auto DevImgIt =
@@ -171,14 +272,7 @@ class kernel_bundle_impl {
171272
}
172273
std::sort(Result.begin(), Result.end(), LessByNameComp{});
173274

174-
auto NewIt =
175-
std::unique(Result.begin(), Result.end(),
176-
[](const sycl::kernel_id &LHS, const sycl::kernel_id &RHS) {
177-
return strcmp(LHS.get_name(), RHS.get_name()) == 0;
178-
}
179-
180-
);
181-
275+
auto NewIt = std::unique(Result.begin(), Result.end(), EqualByNameComp{});
182276
Result.erase(NewIt, Result.end());
183277

184278
return Result;
@@ -192,6 +286,12 @@ class kernel_bundle_impl {
192286
[&KernelID](const device_image_plain &DeviceImage) {
193287
return DeviceImage.has_kernel(KernelID);
194288
});
289+
290+
if (MDeviceImages.end() == It)
291+
throw sycl::exception(make_error_code(errc::invalid),
292+
"The kernel bundle does not contain the kernel "
293+
"identified by kernelId.");
294+
195295
const std::shared_ptr<detail::device_image_impl> &DeviceImageImpl =
196296
detail::getSyclObjImpl(*It);
197297

‎sycl/source/detail/kernel_id_impl.hpp

Lines changed: 9 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -14,11 +14,19 @@ namespace detail {
1414

1515
// Used for sorting vector of kernel_id's
1616
struct LessByNameComp {
17-
bool operator()(const sycl::kernel_id &LHS, const sycl::kernel_id &RHS) {
17+
bool operator()(const sycl::kernel_id &LHS,
18+
const sycl::kernel_id &RHS) const {
1819
return std::strcmp(LHS.get_name(), RHS.get_name()) < 0;
1920
}
2021
};
2122

23+
struct EqualByNameComp {
24+
bool operator()(const sycl::kernel_id &LHS,
25+
const sycl::kernel_id &RHS) const {
26+
return strcmp(LHS.get_name(), RHS.get_name()) == 0;
27+
}
28+
};
29+
2230
// The class is impl counterpart for sycl::kernel_id which represent a kernel
2331
// identificator
2432
class kernel_id_impl {

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

Lines changed: 16 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -1394,13 +1394,18 @@ ProgramManager::compile(const device_image_plain &DeviceImage,
13941394
// TODO: Set spec constatns here.
13951395

13961396
// TODO: Handle zero sized Device list.
1397-
Plugin.call<PiApiKind::piProgramCompile>(
1397+
RT::PiResult Error = Plugin.call_nocheck<PiApiKind::piProgramCompile>(
13981398
ObjectImpl->get_program_ref(), /*num devices=*/Devs.size(),
13991399
PIDevices.data(),
14001400
/*options=*/nullptr,
14011401
/*num_input_headers=*/0, /*input_headers=*/nullptr,
14021402
/*header_include_names=*/nullptr,
14031403
/*pfn_notify=*/nullptr, /*user_data*/ nullptr);
1404+
if (Error != PI_SUCCESS)
1405+
throw sycl::exception(
1406+
make_error_code(errc::build),
1407+
getProgramBuildLog(ObjectImpl->get_program_ref(),
1408+
getSyclObjImpl(ObjectImpl->get_context())));
14041409

14051410
return createSyclObjFromImpl<device_image_plain>(ObjectImpl);
14061411
}
@@ -1422,19 +1427,23 @@ ProgramManager::link(const std::vector<device_image_plain> &DeviceImages,
14221427
PIDevices.push_back(getSyclObjImpl(Dev)->getHandleRef());
14231428

14241429
const context &Context = getSyclObjImpl(DeviceImages[0])->get_context();
1430+
const ContextImplPtr ContextImpl = getSyclObjImpl(Context);
14251431

1426-
const detail::plugin &Plugin = getSyclObjImpl(Context)->getPlugin();
1432+
const detail::plugin &Plugin = ContextImpl->getPlugin();
14271433

14281434
RT::PiProgram LinkedProg = nullptr;
14291435
RT::PiResult Error = Plugin.call_nocheck<PiApiKind::piProgramLink>(
1430-
getSyclObjImpl(Context)->getHandleRef(), PIDevices.size(),
1431-
PIDevices.data(),
1436+
ContextImpl->getHandleRef(), PIDevices.size(), PIDevices.data(),
14321437
/*options=*/nullptr, PIPrograms.size(), PIPrograms.data(),
14331438
/*pfn_notify=*/nullptr,
14341439
/*user_data=*/nullptr, &LinkedProg);
14351440

1436-
(void)Error;
1437-
// TODO: Add error handling
1441+
if (Error != PI_SUCCESS) {
1442+
const string_class ErrorMsg =
1443+
LinkedProg ? getProgramBuildLog(LinkedProg, ContextImpl)
1444+
: "Online link operation failed";
1445+
throw sycl::exception(make_error_code(errc::build), ErrorMsg);
1446+
}
14381447

14391448
std::vector<kernel_id> KernelIDs;
14401449
for (const device_image_plain &DeviceImage : DeviceImages) {
@@ -1582,6 +1591,7 @@ device_image_plain ProgramManager::build(const device_image_plain &DeviceImage,
15821591
SerializedObj SpecConsts = InputImpl->get_spec_const_blob_ref();
15831592

15841593
const RT::PiDevice PiDevice = getRawSyclObjImpl(Devs[0])->getHandleRef();
1594+
// TODO: Throw SYCL2020 style exception
15851595
auto BuildResult = getOrBuild<PiProgramT, compile_program_error>(
15861596
Cache,
15871597
std::make_pair(std::make_pair(std::move(SpecConsts), (size_t)ImgPtr),

‎sycl/source/exception.cpp

Lines changed: 9 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -26,5 +26,14 @@ context exception::get_context() const {
2626

2727
cl_int exception::get_cl_code() const { return MCLErr; }
2828

29+
const std::error_category &sycl_category() noexcept {
30+
static const detail::SYCLCategory SYCLCategoryObj;
31+
return SYCLCategoryObj;
32+
}
33+
34+
std::error_code make_error_code(sycl::errc Err) noexcept {
35+
return {static_cast<int>(Err), sycl_category()};
36+
}
37+
2938
} // namespace sycl
3039
} // __SYCL_INLINE_NAMESPACE(cl)

0 commit comments

Comments
 (0)