Skip to content

Commit 1a84537

Browse files
aelovikov-intelaelovikov-intel
committed
WIP
1 parent ed3767c commit 1a84537

File tree

12 files changed

+99
-104
lines changed

12 files changed

+99
-104
lines changed

‎sycl/source/detail/cg.hpp‎

Lines changed: 7 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -20,6 +20,8 @@
2020
#include <sycl/kernel.hpp> // for kernel_impl
2121
#include <sycl/kernel_bundle.hpp> // for kernel_bundle_impl
2222

23+
#include <detail/device_kernel_info.hpp>
24+
2325
#include <assert.h> // for assert
2426
#include <memory> // for shared_ptr, unique_ptr
2527
#include <stddef.h> // for size_t
@@ -253,7 +255,6 @@ class CGExecKernel : public CG {
253255
std::shared_ptr<detail::kernel_impl> MSyclKernel;
254256
std::shared_ptr<detail::kernel_bundle_impl> MKernelBundle;
255257
std::vector<ArgDesc> MArgs;
256-
KernelNameStrT MKernelName;
257258
DeviceKernelInfo &MDeviceKernelInfo;
258259
std::vector<std::shared_ptr<detail::stream_impl>> MStreams;
259260
std::vector<std::shared_ptr<const void>> MAuxiliaryResources;
@@ -269,7 +270,7 @@ class CGExecKernel : public CG {
269270
std::shared_ptr<detail::kernel_impl> SyclKernel,
270271
std::shared_ptr<detail::kernel_bundle_impl> KernelBundle,
271272
CG::StorageInitHelper CGData, std::vector<ArgDesc> Args,
272-
KernelNameStrT KernelName, DeviceKernelInfo &DeviceKernelInfo,
273+
DeviceKernelInfo &DeviceKernelInfo,
273274
std::vector<std::shared_ptr<detail::stream_impl>> Streams,
274275
std::vector<std::shared_ptr<const void>> AuxiliaryResources,
275276
CGType Type, ur_kernel_cache_config_t KernelCacheConfig,
@@ -278,8 +279,7 @@ class CGExecKernel : public CG {
278279
: CG(Type, std::move(CGData), std::move(loc)), MNDRDesc(NDRDesc),
279280
MHostKernel(std::move(HKernel)), MSyclKernel(std::move(SyclKernel)),
280281
MKernelBundle(std::move(KernelBundle)), MArgs(std::move(Args)),
281-
MKernelName(std::move(KernelName)), MDeviceKernelInfo(DeviceKernelInfo),
282-
MStreams(std::move(Streams)),
282+
MDeviceKernelInfo(DeviceKernelInfo), MStreams(std::move(Streams)),
283283
MAuxiliaryResources(std::move(AuxiliaryResources)),
284284
MAlternativeKernels{}, MKernelCacheConfig(std::move(KernelCacheConfig)),
285285
MKernelIsCooperative(KernelIsCooperative),
@@ -291,7 +291,9 @@ class CGExecKernel : public CG {
291291
CGExecKernel(const CGExecKernel &CGExec) = default;
292292

293293
const std::vector<ArgDesc> &getArguments() const { return MArgs; }
294-
KernelNameStrRefT getKernelName() const { return MKernelName; }
294+
std::string_view getKernelName() const {
295+
return static_cast<std::string_view>(MDeviceKernelInfo.Name);
296+
}
295297
const std::vector<std::shared_ptr<detail::stream_impl>> &getStreams() const {
296298
return MStreams;
297299
}

‎sycl/source/detail/graph/graph_impl.cpp‎

Lines changed: 9 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -739,9 +739,8 @@ ur_exp_command_buffer_sync_point_t exec_graph_impl::enqueueNodeDirect(
739739
CGExec->MLine, CGExec->MColumn);
740740
std::tie(CmdTraceEvent, InstanceID) = emitKernelInstrumentationData(
741741
sycl::detail::GSYCLStreamID, CGExec->MSyclKernel, CodeLoc,
742-
CGExec->MIsTopCodeLoc, CGExec->MKernelName.data(),
743-
CGExec->MDeviceKernelInfo, nullptr, CGExec->MNDRDesc,
744-
CGExec->MKernelBundle.get(), CGExec->MArgs);
742+
CGExec->MIsTopCodeLoc, CGExec->MDeviceKernelInfo, nullptr,
743+
CGExec->MNDRDesc, CGExec->MKernelBundle.get(), CGExec->MArgs);
745744
if (CmdTraceEvent)
746745
sycl::detail::emitInstrumentationGeneral(sycl::detail::GSYCLStreamID,
747746
InstanceID, CmdTraceEvent,
@@ -1401,14 +1400,14 @@ void exec_graph_impl::update(std::shared_ptr<graph_impl> GraphImpl) {
14011400
sycl::detail::CGExecKernel *TargetCGExec =
14021401
static_cast<sycl::detail::CGExecKernel *>(
14031402
MNodeStorage[i]->MCommandGroup.get());
1404-
KernelNameStrRefT TargetKernelName = TargetCGExec->getKernelName();
1403+
std::string_view TargetKernelName = TargetCGExec->getKernelName();
14051404

14061405
sycl::detail::CGExecKernel *SourceCGExec =
14071406
static_cast<sycl::detail::CGExecKernel *>(
14081407
GraphImpl->MNodeStorage[i]->MCommandGroup.get());
1409-
KernelNameStrRefT SourceKernelName = SourceCGExec->getKernelName();
1408+
std::string_view SourceKernelName = SourceCGExec->getKernelName();
14101409

1411-
if (TargetKernelName.compare(SourceKernelName) != 0) {
1410+
if (TargetKernelName != SourceKernelName) {
14121411
std::stringstream ErrorStream(
14131412
"Cannot update using a graph with mismatched kernel "
14141413
"types. Source node type ");
@@ -1568,14 +1567,14 @@ void exec_graph_impl::populateURKernelUpdateStructs(
15681567
UrKernel = Kernel->getHandleRef();
15691568
EliminatedArgMask = Kernel->getKernelArgMask();
15701569
} else if (auto SyclKernelImpl =
1571-
KernelBundleImplPtr
1572-
? KernelBundleImplPtr->tryGetKernel(ExecCG.MKernelName)
1573-
: std::shared_ptr<kernel_impl>{nullptr}) {
1570+
KernelBundleImplPtr ? KernelBundleImplPtr->tryGetKernel(
1571+
ExecCG.MDeviceKernelInfo.Name)
1572+
: std::shared_ptr<kernel_impl>{nullptr}) {
15741573
UrKernel = SyclKernelImpl->getHandleRef();
15751574
EliminatedArgMask = SyclKernelImpl->getKernelArgMask();
15761575
} else {
15771576
BundleObjs = sycl::detail::ProgramManager::getInstance().getOrCreateKernel(
1578-
ContextImpl, DeviceImpl, ExecCG.MKernelName, ExecCG.MDeviceKernelInfo);
1577+
ContextImpl, DeviceImpl, ExecCG.MDeviceKernelInfo);
15791578
UrKernel = BundleObjs->MKernelHandle;
15801579
EliminatedArgMask = BundleObjs->MKernelArgMask;
15811580
}

‎sycl/source/detail/graph/node_impl.hpp‎

Lines changed: 5 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -341,7 +341,8 @@ class node_impl : public std::enable_shared_from_this<node_impl> {
341341
static_cast<sycl::detail::CGExecKernel *>(MCommandGroup.get());
342342
sycl::detail::CGExecKernel *ExecKernelB =
343343
static_cast<sycl::detail::CGExecKernel *>(Node.MCommandGroup.get());
344-
return ExecKernelA->MKernelName.compare(ExecKernelB->MKernelName) == 0;
344+
return std::string_view{ExecKernelA->MDeviceKernelInfo.Name} ==
345+
std::string_view{ExecKernelB->MDeviceKernelInfo.Name};
345346
}
346347
case sycl::detail::CGType::CopyUSM: {
347348
sycl::detail::CGCopyUSM *CopyA =
@@ -543,7 +544,9 @@ class node_impl : public std::enable_shared_from_this<node_impl> {
543544
Stream << "CGExecKernel \\n";
544545
sycl::detail::CGExecKernel *Kernel =
545546
static_cast<sycl::detail::CGExecKernel *>(MCommandGroup.get());
546-
Stream << "NAME = " << Kernel->MKernelName << "\\n";
547+
Stream << "NAME = "
548+
<< static_cast<std::string_view>(Kernel->MDeviceKernelInfo.Name)
549+
<< "\\n";
547550
if (Verbose) {
548551
Stream << "ARGS = \\n";
549552
for (size_t i = 0; i < Kernel->MArgs.size(); i++) {

‎sycl/source/detail/handler_impl.hpp‎

Lines changed: 5 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -60,6 +60,11 @@ class handler_impl {
6060
HandlerSubmissionState::EXPLICIT_KERNEL_BUNDLE_STATE;
6161
}
6262

63+
KernelNameStrRefT getKernelName() const {
64+
assert(MDeviceKernelInfoPtr);
65+
return static_cast<KernelNameStrRefT>(MDeviceKernelInfoPtr->Name);
66+
}
67+
6368
/// Registers mutually exclusive submission states.
6469
HandlerSubmissionState MSubmissionState = HandlerSubmissionState::NO_STATE;
6570

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

Lines changed: 17 additions & 9 deletions
Original file line numberDiff line numberDiff line change
@@ -1084,25 +1084,32 @@ ProgramManager::getBuiltURProgram(const BinImgWithDeps &ImgWithDeps,
10841084

10851085
FastKernelCacheValPtr ProgramManager::getOrCreateKernel(
10861086
context_impl &ContextImpl, device_impl &DeviceImpl,
1087-
KernelNameStrRefT KernelName, DeviceKernelInfo &DeviceKernelInfo,
1088-
const NDRDescT &NDRDesc) {
1087+
DeviceKernelInfo &DeviceKernelInfo, const NDRDescT &NDRDesc) {
10891088
if constexpr (DbgProgMgr > 0) {
10901089
std::cerr << ">>> ProgramManager::getOrCreateKernel(" << &ContextImpl
1091-
<< ", " << &DeviceImpl << ", " << KernelName << ")\n";
1090+
<< ", " << &DeviceImpl << ", "
1091+
<< static_cast<std::string_view>(DeviceKernelInfo.Name) << ")\n";
10921092
}
10931093

10941094
KernelProgramCache &Cache = ContextImpl.getKernelProgramCache();
10951095
ur_device_handle_t UrDevice = DeviceImpl.getHandleRef();
10961096
if (SYCLConfig<SYCL_CACHE_IN_MEM>::get()) {
1097-
if (auto KernelCacheValPtr = Cache.tryToGetKernelFast(
1098-
KernelName, UrDevice, DeviceKernelInfo.getKernelSubcache())) {
1097+
if (auto KernelCacheValPtr =
1098+
Cache.tryToGetKernelFast(DeviceKernelInfo.Name, UrDevice,
1099+
DeviceKernelInfo.getKernelSubcache())) {
10991100
return KernelCacheValPtr;
11001101
}
11011102
}
11021103

1103-
Managed<ur_program_handle_t> Program =
1104-
getBuiltURProgram(ContextImpl, DeviceImpl, KernelName, NDRDesc);
1104+
Managed<ur_program_handle_t> Program = getBuiltURProgram(
1105+
ContextImpl, DeviceImpl, DeviceKernelInfo.Name, NDRDesc);
11051106

1107+
#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
1108+
// Simplify this once `DeviceKernelInfo.Name`'s type is known.
1109+
// Using `decltype(auto)` insteado of just `auto` to get reference when
1110+
// possible.
1111+
#endif
1112+
decltype(auto) KernelName = KernelNameStrRefT{DeviceKernelInfo.Name};
11061113
auto BuildF = [this, &Program, &KernelName, &ContextImpl] {
11071114
adapter_impl &Adapter = ContextImpl.getAdapter();
11081115
Managed<ur_kernel_handle_t> Kernel{Adapter};
@@ -1125,7 +1132,8 @@ FastKernelCacheValPtr ProgramManager::getOrCreateKernel(
11251132
return std::make_pair(std::move(Kernel), ArgMask);
11261133
};
11271134

1128-
auto GetCachedBuildF = [&Cache, &KernelName, &Program]() {
1135+
auto GetCachedBuildF = [&Cache, &KernelName = DeviceKernelInfo.Name,
1136+
&Program]() {
11291137
return Cache.getOrInsertKernel(Program, KernelName);
11301138
};
11311139

@@ -1147,7 +1155,7 @@ FastKernelCacheValPtr ProgramManager::getOrCreateKernel(
11471155
auto ret_val = std::make_shared<FastKernelCacheVal>(
11481156
KernelArgMaskPair.first.retain(), &(BuildResult->MBuildResultMutex),
11491157
KernelArgMaskPair.second, std::move(Program), ContextImpl.getAdapter());
1150-
Cache.saveKernel(KernelName, UrDevice, ret_val,
1158+
Cache.saveKernel(DeviceKernelInfo.Name, UrDevice, ret_val,
11511159
DeviceKernelInfo.getKernelSubcache());
11521160
return ret_val;
11531161
}

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

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -200,7 +200,6 @@ class ProgramManager {
200200

201201
FastKernelCacheValPtr getOrCreateKernel(context_impl &ContextImpl,
202202
device_impl &DeviceImpl,
203-
KernelNameStrRefT KernelName,
204203
DeviceKernelInfo &DeviceKernelInfo,
205204
const NDRDescT &NDRDesc = {});
206205

0 commit comments

Comments
 (0)