Skip to content

Commit b92cc90

Browse files
[NFC][SYCL] std::shared_ptr<device_image_impl> cleanups
* Avoid unnecessary copies * Use rvalue-reference if param is getting moved from (except `device_image` ctor) * Remove `DeviceImageImplPtr` type alias (not too many uses remaining, doesn't bring much value anymore) * Inline some temporaries so that explicit `std::move` wouldn't be needed * Switch some sets to use raw `device_image_impl *` ptr * `kernel_impl::getDeviceImage` to return raw reference
1 parent 8861f6b commit b92cc90

File tree

10 files changed

+63
-75
lines changed

10 files changed

+63
-75
lines changed

sycl/include/sycl/kernel_bundle.hpp

Lines changed: 5 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -100,13 +100,12 @@ class __SYCL_EXPORT kernel_id : public detail::OwnerLessBase<kernel_id> {
100100

101101
namespace detail {
102102
class device_image_impl;
103-
using DeviceImageImplPtr = std::shared_ptr<device_image_impl>;
104103

105104
// The class is used as a base for device_image for "untemplating" public
106105
// methods.
107106
class __SYCL_EXPORT device_image_plain {
108107
public:
109-
device_image_plain(const detail::DeviceImageImplPtr &Impl)
108+
device_image_plain(std::shared_ptr<device_image_impl> &&Impl)
110109
: impl(std::move(Impl)) {}
111110

112111
bool operator==(const device_image_plain &RHS) const {
@@ -124,7 +123,7 @@ class __SYCL_EXPORT device_image_plain {
124123
ur_native_handle_t getNative() const;
125124

126125
protected:
127-
detail::DeviceImageImplPtr impl;
126+
std::shared_ptr<device_image_impl> impl;
128127

129128
template <class Obj>
130129
friend const decltype(Obj::impl) &
@@ -191,7 +190,7 @@ class device_image : public detail::device_image_plain,
191190
#endif // _HAS_STD_BYTE
192191

193192
private:
194-
device_image(detail::DeviceImageImplPtr Impl)
193+
device_image(std::shared_ptr<detail::device_image_impl> Impl)
195194
: device_image_plain(std::move(Impl)) {}
196195

197196
template <class Obj>
@@ -736,7 +735,7 @@ namespace detail {
736735

737736
// Stable selector function type for passing thru library boundaries
738737
using DevImgSelectorImpl =
739-
std::function<bool(const detail::DeviceImageImplPtr &DevImgImpl)>;
738+
std::function<bool(const std::shared_ptr<device_image_impl> &DevImgImpl)>;
740739

741740
// Internal non-template versions of get_kernel_bundle API which is used by
742741
// public onces
@@ -769,7 +768,7 @@ kernel_bundle<State> get_kernel_bundle(const context &Ctx,
769768
std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
770769

771770
detail::DevImgSelectorImpl SelectorWrapper =
772-
[Selector](const detail::DeviceImageImplPtr &DevImg) {
771+
[Selector](const std::shared_ptr<detail::device_image_impl> &DevImg) {
773772
return Selector(
774773
detail::createSyclObjFromImpl<sycl::device_image<State>>(DevImg));
775774
};

sycl/source/backend.cpp

Lines changed: 6 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -300,13 +300,12 @@ make_kernel_bundle(ur_native_handle_t NativeHandle,
300300
// this by pre-building the device image and extracting kernel info. We can't
301301
// do the same to user images, since they may contain references to undefined
302302
// symbols (e.g. when kernel_bundle is supposed to be joined with another).
303-
auto KernelIDs = std::make_shared<std::vector<kernel_id>>();
304-
auto DevImgImpl =
305-
device_image_impl::create(nullptr, TargetContext, Devices, State,
306-
KernelIDs, UrProgram, ImageOriginInterop);
307-
device_image_plain DevImg{DevImgImpl};
308-
309-
return kernel_bundle_impl::create(TargetContext, Devices, DevImg);
303+
return kernel_bundle_impl::create(
304+
TargetContext, Devices,
305+
device_image_plain{
306+
device_image_impl::create(nullptr, TargetContext, Devices, State,
307+
std::make_shared<std::vector<kernel_id>>(),
308+
UrProgram, ImageOriginInterop)});
310309
}
311310

312311
// TODO: Unused. Remove when allowed.

sycl/source/detail/helpers.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -64,12 +64,12 @@ const RTDeviceBinaryImage *retrieveKernelBinary(queue_impl &Queue,
6464
}
6565

6666
if (KernelCG->MSyclKernel != nullptr)
67-
return KernelCG->MSyclKernel->getDeviceImage()->get_bin_image_ref();
67+
return KernelCG->MSyclKernel->getDeviceImage().get_bin_image_ref();
6868

6969
if (auto KernelBundleImpl = KernelCG->getKernelBundle())
7070
if (auto SyclKernelImpl = KernelBundleImpl->tryGetKernel(KernelName))
7171
// Retrieve the device image from the kernel bundle.
72-
return SyclKernelImpl->getDeviceImage()->get_bin_image_ref();
72+
return SyclKernelImpl->getDeviceImage().get_bin_image_ref();
7373

7474
context_impl &ContextImpl = Queue.getContextImpl();
7575
return &detail::ProgramManager::getInstance().getDeviceImage(

sycl/source/detail/kernel_bundle_impl.hpp

Lines changed: 10 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -124,9 +124,9 @@ class kernel_bundle_impl
124124

125125
// Interop constructor
126126
kernel_bundle_impl(context Ctx, devices_range Devs,
127-
device_image_plain &DevImage, private_tag Tag)
127+
device_image_plain &&DevImage, private_tag Tag)
128128
: kernel_bundle_impl(std::move(Ctx), Devs, Tag) {
129-
MDeviceImages.emplace_back(DevImage);
129+
MDeviceImages.emplace_back(std::move(DevImage));
130130
MUniqueDeviceImages.emplace_back(DevImage);
131131
}
132132

@@ -162,9 +162,9 @@ class kernel_bundle_impl
162162
InputBundleImpl.MDeviceImages) {
163163
// Skip images which are not compatible with devices provided
164164
if (std::none_of(get_devices().begin(), get_devices().end(),
165-
[&DevImgWithDeps](device_impl &Dev) {
166-
return getSyclObjImpl(DevImgWithDeps.getMain())
167-
->compatible_with_device(Dev);
165+
[&MainImg = *getSyclObjImpl(DevImgWithDeps.getMain())](
166+
device_impl &Dev) {
167+
return MainImg.compatible_with_device(Dev);
168168
}))
169169
continue;
170170

@@ -396,9 +396,9 @@ class kernel_bundle_impl
396396
ImagesWithSpecConsts) {
397397
// Skip images which are not compatible with devices provided
398398
if (std::none_of(get_devices().begin(), get_devices().end(),
399-
[DeviceImageWithDeps](device_impl &Dev) {
400-
return getSyclObjImpl(DeviceImageWithDeps->getMain())
401-
->compatible_with_device(Dev);
399+
[&MainImg = *getSyclObjImpl(
400+
DeviceImageWithDeps->getMain())](device_impl &Dev) {
401+
return MainImg.compatible_with_device(Dev);
402402
}))
403403
continue;
404404

@@ -994,8 +994,8 @@ class kernel_bundle_impl
994994
SelectedImage->get_ur_program());
995995

996996
return std::make_shared<kernel_impl>(
997-
Kernel, *detail::getSyclObjImpl(MContext), SelectedImage, *this,
998-
ArgMask, SelectedImage->get_ur_program(), CacheMutex);
997+
Kernel, *detail::getSyclObjImpl(MContext), std::move(SelectedImage),
998+
*this, ArgMask, SelectedImage->get_ur_program(), CacheMutex);
999999
}
10001000

10011001
std::shared_ptr<kernel_impl>

sycl/source/detail/kernel_impl.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -40,7 +40,7 @@ kernel_impl::kernel_impl(ur_kernel_handle_t Kernel, context_impl &Context,
4040
}
4141

4242
kernel_impl::kernel_impl(ur_kernel_handle_t Kernel, context_impl &ContextImpl,
43-
DeviceImageImplPtr DeviceImageImpl,
43+
std::shared_ptr<device_image_impl> &&DeviceImageImpl,
4444
const kernel_bundle_impl &KernelBundleImpl,
4545
const KernelArgMask *ArgMask,
4646
ur_program_handle_t Program, std::mutex *CacheMutex)

sycl/source/detail/kernel_impl.hpp

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -50,7 +50,7 @@ class kernel_impl {
5050
/// \param ContextImpl is a valid SYCL context
5151
/// \param KernelBundleImpl is a valid instance of kernel_bundle_impl
5252
kernel_impl(ur_kernel_handle_t Kernel, context_impl &ContextImpl,
53-
DeviceImageImplPtr DeviceImageImpl,
53+
std::shared_ptr<device_image_impl> &&DeviceImageImpl,
5454
const kernel_bundle_impl &KernelBundleImpl,
5555
const KernelArgMask *ArgMask, ur_program_handle_t Program,
5656
std::mutex *CacheMutex);
@@ -213,7 +213,7 @@ class kernel_impl {
213213
bool isInteropOrSourceBased() const noexcept;
214214
bool hasSYCLMetadata() const noexcept;
215215

216-
const DeviceImageImplPtr &getDeviceImage() const { return MDeviceImageImpl; }
216+
device_image_impl &getDeviceImage() const { return *MDeviceImageImpl; }
217217

218218
ur_native_handle_t getNative() const {
219219
adapter_impl &Adapter = MContext->getAdapter();
@@ -247,7 +247,7 @@ class kernel_impl {
247247
const std::shared_ptr<context_impl> MContext;
248248
const ur_program_handle_t MProgram = nullptr;
249249
bool MCreatedFromSource = true;
250-
const DeviceImageImplPtr MDeviceImageImpl;
250+
const std::shared_ptr<device_image_impl> MDeviceImageImpl;
251251
const KernelBundleImplPtr MKernelBundleImpl;
252252
bool MIsInterop = false;
253253
mutable std::mutex MNoncacheableEnqueueMutex;

sycl/source/detail/program_manager/program_manager.cpp

Lines changed: 14 additions & 19 deletions
Original file line numberDiff line numberDiff line change
@@ -2476,11 +2476,9 @@ device_image_plain ProgramManager::getDeviceImageFromBinaryImage(
24762476
KernelIDs = m_BinImg2KernelIDs[BinImage];
24772477
}
24782478

2479-
DeviceImageImplPtr Impl =
2479+
return createSyclObjFromImpl<device_image_plain>(
24802480
device_image_impl::create(BinImage, Ctx, Dev, ImgState, KernelIDs,
2481-
/*PIProgram=*/nullptr, ImageOriginSYCLOffline);
2482-
2483-
return createSyclObjFromImpl<device_image_plain>(std::move(Impl));
2481+
/*PIProgram=*/nullptr, ImageOriginSYCLOffline));
24842482
}
24852483

24862484
std::vector<DevImgPlainWithDeps>
@@ -2638,7 +2636,7 @@ ProgramManager::getSYCLDeviceImagesWithCompatibleState(
26382636
if (ImgInfoPair.second.RequirementCounter == 0)
26392637
continue;
26402638

2641-
DeviceImageImplPtr MainImpl = device_image_impl::create(
2639+
std::shared_ptr<device_image_impl> MainImpl = device_image_impl::create(
26422640
ImgInfoPair.first, Ctx, Devs, ImgInfoPair.second.State,
26432641
ImgInfoPair.second.KernelIDs, /*PIProgram=*/nullptr,
26442642
ImageOriginSYCLOffline);
@@ -2673,11 +2671,10 @@ ProgramManager::createDependencyImage(const context &Ctx, devices_range Devs,
26732671

26742672
assert(DepState == getBinImageState(DepImage) &&
26752673
"State mismatch between main image and its dependency");
2676-
DeviceImageImplPtr DepImpl = device_image_impl::create(
2677-
DepImage, Ctx, Devs, DepState, std::move(DepKernelIDs),
2678-
/*PIProgram=*/nullptr, ImageOriginSYCLOffline);
26792674

2680-
return createSyclObjFromImpl<device_image_plain>(std::move(DepImpl));
2675+
return createSyclObjFromImpl<device_image_plain>(device_image_impl::create(
2676+
DepImage, Ctx, Devs, DepState, std::move(DepKernelIDs),
2677+
/*PIProgram=*/nullptr, ImageOriginSYCLOffline));
26812678
}
26822679

26832680
void ProgramManager::bringSYCLDeviceImageToState(
@@ -2846,7 +2843,7 @@ ProgramManager::compile(const DevImgPlainWithDeps &ImgWithDeps,
28462843

28472844
std::optional<detail::KernelCompilerBinaryInfo> RTCInfo =
28482845
InputImpl.getRTCInfo();
2849-
DeviceImageImplPtr ObjectImpl = device_image_impl::create(
2846+
std::shared_ptr<device_image_impl> ObjectImpl = device_image_impl::create(
28502847
InputImpl.get_bin_image_ref(), InputImpl.get_context(), Devs,
28512848
bundle_state::object, InputImpl.get_kernel_ids_ptr(), Prog,
28522849
InputImpl.get_spec_const_data_ref(),
@@ -3044,15 +3041,14 @@ ProgramManager::link(const std::vector<device_image_plain> &Imgs,
30443041
}
30453042
auto MergedRTCInfo = detail::KernelCompilerBinaryInfo::Merge(RTCInfoPtrs);
30463043

3047-
DeviceImageImplPtr ExecutableImpl = device_image_impl::create(
3044+
// TODO: Make multiple sets of device images organized by devices they are
3045+
// compiled for.
3046+
return {createSyclObjFromImpl<device_image_plain>(device_image_impl::create(
30483047
NewBinImg, Context, Devs, bundle_state::executable, std::move(KernelIDs),
30493048
LinkedProg, std::move(NewSpecConstMap), std::move(NewSpecConstBlob),
30503049
CombinedOrigins, std::move(MergedRTCInfo), std::move(MergedKernelNames),
3051-
std::move(MergedEliminatedKernelArgMasks), std::move(MergedImageStorage));
3052-
3053-
// TODO: Make multiple sets of device images organized by devices they are
3054-
// compiled for.
3055-
return {createSyclObjFromImpl<device_image_plain>(std::move(ExecutableImpl))};
3050+
std::move(MergedEliminatedKernelArgMasks),
3051+
std::move(MergedImageStorage)))};
30563052
}
30573053

30583054
// The function duplicates most of the code from existing getBuiltPIProgram.
@@ -3126,13 +3122,12 @@ ProgramManager::build(const DevImgPlainWithDeps &DevImgWithDeps,
31263122
}
31273123
auto MergedRTCInfo = detail::KernelCompilerBinaryInfo::Merge(RTCInfoPtrs);
31283124

3129-
DeviceImageImplPtr ExecImpl = device_image_impl::create(
3125+
return createSyclObjFromImpl<device_image_plain>(device_image_impl::create(
31303126
ResultBinImg, Context, Devs, bundle_state::executable,
31313127
std::move(KernelIDs), ResProgram, std::move(SpecConstMap),
31323128
std::move(SpecConstBlob), CombinedOrigins, std::move(MergedRTCInfo),
31333129
std::move(MergedKernelNames), std::move(MergedEliminatedKernelArgMasks),
3134-
std::move(MergedImageStorage));
3135-
return createSyclObjFromImpl<device_image_plain>(std::move(ExecImpl));
3130+
std::move(MergedImageStorage)));
31363131
}
31373132

31383133
// When caching is enabled, the returned UrKernel will already have

sycl/source/detail/scheduler/commands.cpp

Lines changed: 17 additions & 21 deletions
Original file line numberDiff line numberDiff line change
@@ -2400,10 +2400,9 @@ static void SetArgBasedOnType(
24002400

24012401
static ur_result_t SetKernelParamsAndLaunch(
24022402
queue_impl &Queue, std::vector<ArgDesc> &Args,
2403-
const std::shared_ptr<device_image_impl> &DeviceImageImpl,
2404-
ur_kernel_handle_t Kernel, NDRDescT &NDRDesc,
2405-
std::vector<ur_event_handle_t> &RawEvents, detail::event_impl *OutEventImpl,
2406-
const KernelArgMask *EliminatedArgMask,
2403+
device_image_impl *DeviceImageImpl, ur_kernel_handle_t Kernel,
2404+
NDRDescT &NDRDesc, std::vector<ur_event_handle_t> &RawEvents,
2405+
detail::event_impl *OutEventImpl, const KernelArgMask *EliminatedArgMask,
24072406
const std::function<void *(Requirement *Req)> &getMemAllocationFunc,
24082407
bool IsCooperative, bool KernelUsesClusterLaunch,
24092408
uint32_t WorkGroupMemorySize, const RTDeviceBinaryImage *BinImage,
@@ -2418,8 +2417,7 @@ static ur_result_t SetKernelParamsAndLaunch(
24182417
std::vector<unsigned char> Empty;
24192418
Kernel = Scheduler::getInstance().completeSpecConstMaterialization(
24202419
Queue, BinImage, KernelName,
2421-
DeviceImageImpl.get() ? DeviceImageImpl->get_spec_const_blob_ref()
2422-
: Empty);
2420+
DeviceImageImpl ? DeviceImageImpl->get_spec_const_blob_ref() : Empty);
24232421
}
24242422

24252423
if (KernelFuncPtr && !KernelHasSpecialCaptures) {
@@ -2449,9 +2447,8 @@ static ur_result_t SetKernelParamsAndLaunch(
24492447
} else {
24502448
auto setFunc = [&Adapter, Kernel, &DeviceImageImpl, &getMemAllocationFunc,
24512449
&Queue](detail::ArgDesc &Arg, size_t NextTrueIndex) {
2452-
SetArgBasedOnType(Adapter, Kernel, DeviceImageImpl.get(),
2453-
getMemAllocationFunc, Queue.getContextImpl(), Arg,
2454-
NextTrueIndex);
2450+
SetArgBasedOnType(Adapter, Kernel, DeviceImageImpl, getMemAllocationFunc,
2451+
Queue.getContextImpl(), Arg, NextTrueIndex);
24552452
};
24562453
applyFuncOnFilteredArgs(EliminatedArgMask, Args, setFunc);
24572454
}
@@ -2537,14 +2534,14 @@ static ur_result_t SetKernelParamsAndLaunch(
25372534
return Error;
25382535
}
25392536

2540-
static std::tuple<ur_kernel_handle_t, std::shared_ptr<device_image_impl>,
2537+
static std::tuple<ur_kernel_handle_t, device_image_impl *,
25412538
const KernelArgMask *>
25422539
getCGKernelInfo(const CGExecKernel &CommandGroup, context_impl &ContextImpl,
25432540
device_impl &DeviceImpl,
25442541
std::vector<FastKernelCacheValPtr> &KernelCacheValsToRelease) {
25452542

25462543
ur_kernel_handle_t UrKernel = nullptr;
2547-
std::shared_ptr<device_image_impl> DeviceImageImpl = nullptr;
2544+
device_image_impl *DeviceImageImpl = nullptr;
25482545
const KernelArgMask *EliminatedArgMask = nullptr;
25492546
kernel_bundle_impl *KernelBundleImplPtr = CommandGroup.MKernelBundle.get();
25502547

@@ -2556,7 +2553,7 @@ getCGKernelInfo(const CGExecKernel &CommandGroup, context_impl &ContextImpl,
25562553
CommandGroup.MKernelName)
25572554
: std::shared_ptr<kernel_impl>{nullptr}) {
25582555
UrKernel = SyclKernelImpl->getHandleRef();
2559-
DeviceImageImpl = SyclKernelImpl->getDeviceImage();
2556+
DeviceImageImpl = &SyclKernelImpl->getDeviceImage();
25602557
EliminatedArgMask = SyclKernelImpl->getKernelArgMask();
25612558
} else {
25622559
FastKernelCacheValPtr FastKernelCacheVal =
@@ -2568,8 +2565,7 @@ getCGKernelInfo(const CGExecKernel &CommandGroup, context_impl &ContextImpl,
25682565
// To keep UrKernel valid, we return FastKernelCacheValPtr.
25692566
KernelCacheValsToRelease.push_back(std::move(FastKernelCacheVal));
25702567
}
2571-
return std::make_tuple(UrKernel, std::move(DeviceImageImpl),
2572-
EliminatedArgMask);
2568+
return std::make_tuple(UrKernel, DeviceImageImpl, EliminatedArgMask);
25732569
}
25742570

25752571
ur_result_t enqueueImpCommandBufferKernel(
@@ -2586,7 +2582,7 @@ ur_result_t enqueueImpCommandBufferKernel(
25862582
std::vector<FastKernelCacheValPtr> FastKernelCacheValsToRelease;
25872583

25882584
ur_kernel_handle_t UrKernel = nullptr;
2589-
std::shared_ptr<device_image_impl> DeviceImageImpl = nullptr;
2585+
device_image_impl *DeviceImageImpl = nullptr;
25902586
const KernelArgMask *EliminatedArgMask = nullptr;
25912587

25922588
context_impl &ContextImpl = *sycl::detail::getSyclObjImpl(Ctx);
@@ -2610,10 +2606,10 @@ ur_result_t enqueueImpCommandBufferKernel(
26102606
}
26112607

26122608
adapter_impl &Adapter = ContextImpl.getAdapter();
2613-
auto SetFunc = [&Adapter, &UrKernel, &DeviceImageImpl, &ContextImpl,
2614-
&getMemAllocationFunc](sycl::detail::ArgDesc &Arg,
2615-
size_t NextTrueIndex) {
2616-
sycl::detail::SetArgBasedOnType(Adapter, UrKernel, DeviceImageImpl.get(),
2609+
auto SetFunc = [&Adapter, &UrKernel, &ContextImpl, &getMemAllocationFunc,
2610+
DeviceImageImpl](sycl::detail::ArgDesc &Arg,
2611+
size_t NextTrueIndex) {
2612+
sycl::detail::SetArgBasedOnType(Adapter, UrKernel, DeviceImageImpl,
26172613
getMemAllocationFunc, ContextImpl, Arg,
26182614
NextTrueIndex);
26192615
};
@@ -2695,7 +2691,7 @@ void enqueueImpKernel(
26952691
const KernelArgMask *EliminatedArgMask;
26962692

26972693
std::shared_ptr<kernel_impl> SyclKernelImpl;
2698-
std::shared_ptr<device_image_impl> DeviceImageImpl;
2694+
device_image_impl *DeviceImageImpl = nullptr;
26992695
FastKernelCacheValPtr KernelCacheVal;
27002696

27012697
if (nullptr != MSyclKernel) {
@@ -2717,7 +2713,7 @@ void enqueueImpKernel(
27172713
? KernelBundleImplPtr->tryGetKernel(KernelName)
27182714
: std::shared_ptr<kernel_impl>{nullptr})) {
27192715
Kernel = SyclKernelImpl->getHandleRef();
2720-
DeviceImageImpl = SyclKernelImpl->getDeviceImage();
2716+
DeviceImageImpl = &SyclKernelImpl->getDeviceImage();
27212717

27222718
Program = DeviceImageImpl->get_ur_program();
27232719

sycl/source/kernel_bundle.cpp

Lines changed: 3 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -285,11 +285,10 @@ bool has_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
285285
std::set<kernel_id, LessByNameComp> CombinedKernelIDs;
286286
for (const DevImgPlainWithDeps &DeviceImageWithDeps : DeviceImagesWithDeps) {
287287
for (const device_image_plain &DeviceImage : DeviceImageWithDeps) {
288-
const std::shared_ptr<device_image_impl> &DeviceImageImpl =
289-
getSyclObjImpl(DeviceImage);
288+
device_image_impl &DeviceImageImpl = *getSyclObjImpl(DeviceImage);
290289

291-
CombinedKernelIDs.insert(DeviceImageImpl->get_kernel_ids().begin(),
292-
DeviceImageImpl->get_kernel_ids().end());
290+
CombinedKernelIDs.insert(DeviceImageImpl.get_kernel_ids().begin(),
291+
DeviceImageImpl.get_kernel_ids().end());
293292
}
294293
}
295294

sycl/unittests/program_manager/arg_mask/EliminatedArgMask.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -185,9 +185,9 @@ const sycl::detail::KernelArgMask *getKernelArgMaskFromBundle(
185185
auto SyclKernelImpl =
186186
KernelBundleImplPtr->tryGetKernel(ExecKernel->MKernelName);
187187
EXPECT_TRUE(SyclKernelImpl != nullptr);
188-
std::shared_ptr<sycl::detail::device_image_impl> DeviceImageImpl =
188+
sycl::detail::device_image_impl &DeviceImageImpl =
189189
SyclKernelImpl->getDeviceImage();
190-
ur_program_handle_t Program = DeviceImageImpl->get_ur_program();
190+
ur_program_handle_t Program = DeviceImageImpl.get_ur_program();
191191

192192
EXPECT_TRUE(nullptr == ExecKernel->MSyclKernel ||
193193
!ExecKernel->MSyclKernel->isCreatedFromSource());

0 commit comments

Comments
 (0)