24 #include <sycl/version.hpp>
36 #ifdef XPTI_ENABLE_INSTRUMENTATION
39 #include "xpti/xpti_trace_framework.h"
43 inline namespace _V1 {
45 #ifdef XPTI_ENABLE_INSTRUMENTATION
49 xpti_td *GSYCLGraphEvent =
nullptr;
51 xpti_td *GPICallEvent =
nullptr;
53 xpti_td *GPIArgCallEvent =
nullptr;
54 xpti_td *GPIArgCallActiveEvent =
nullptr;
56 uint8_t PiCallStreamID = 0;
57 uint8_t PiDebugCallStreamID = 0;
59 #endif // XPTI_ENABLE_INSTRUMENTATION
62 void *ReturnOpaqueData =
nullptr;
63 const PluginPtr &Plugin = pi::getPlugin<BE>();
66 OpaqueDataParam, &ReturnOpaqueData);
68 return ReturnOpaqueData;
71 template __SYCL_EXPORT
void *
72 getPluginOpaqueData<sycl::backend::ext_intel_esimd_emulator>(
void *);
83 uint64_t CorrelationID = 0;
84 #ifdef XPTI_ENABLE_INSTRUMENTATION
115 constexpr uint16_t NotificationTraceType =
116 (uint16_t)xpti::trace_point_type_t::function_begin;
117 if (xptiCheckTraceEnabled(PiCallStreamID, NotificationTraceType)) {
118 CorrelationID = xptiGetUniqueId();
119 xptiNotifySubscribers(PiCallStreamID, NotificationTraceType, GPICallEvent,
120 nullptr, CorrelationID,
121 static_cast<const void *
>(FName));
123 #endif // XPTI_ENABLE_INSTRUMENTATION
124 return CorrelationID;
128 #ifdef XPTI_ENABLE_INSTRUMENTATION
129 constexpr uint16_t NotificationTraceType =
130 (uint16_t)xpti::trace_point_type_t::function_end;
131 if (xptiCheckTraceEnabled(PiCallStreamID, NotificationTraceType)) {
136 xptiNotifySubscribers(PiCallStreamID, NotificationTraceType, GPICallEvent,
137 nullptr, CorrelationID,
138 static_cast<const void *
>(FName));
140 #endif // XPTI_ENABLE_INSTRUMENTATION
144 unsigned char *ArgsData,
146 uint64_t CorrelationID = 0;
147 #ifdef XPTI_ENABLE_INSTRUMENTATION
148 constexpr uint16_t NotificationTraceType =
149 (uint16_t)xpti::trace_point_type_t::function_with_args_begin;
150 if (xptiCheckTraceEnabled(PiDebugCallStreamID, NotificationTraceType)) {
151 xpti::function_with_args_t Payload{FuncID, FuncName, ArgsData,
nullptr,
155 auto CodeLoc = Tls.
query();
156 xpti::payload_t PL = xpti::payload_t(
157 CodeLoc.functionName(), CodeLoc.fileName(), CodeLoc.lineNumber(),
158 CodeLoc.columnNumber(),
nullptr);
159 uint64_t InstanceNumber{};
160 assert(GPIArgCallActiveEvent ==
nullptr);
161 GPIArgCallActiveEvent =
162 xptiMakeEvent(
"Plugin interface call", &PL, xpti::trace_graph_event,
163 xpti_at::active, &InstanceNumber);
166 CorrelationID = xptiGetUniqueId();
167 xptiNotifySubscribers(PiDebugCallStreamID, NotificationTraceType,
168 GPIArgCallEvent, GPIArgCallActiveEvent, CorrelationID,
172 return CorrelationID;
176 const char *FuncName,
unsigned char *ArgsData,
178 #ifdef XPTI_ENABLE_INSTRUMENTATION
179 constexpr uint16_t NotificationTraceType =
180 (uint16_t)xpti::trace_point_type_t::function_with_args_end;
181 if (xptiCheckTraceEnabled(PiDebugCallStreamID, NotificationTraceType)) {
182 xpti::function_with_args_t Payload{FuncID, FuncName, ArgsData, &Result,
185 xptiNotifySubscribers(PiDebugCallStreamID, NotificationTraceType,
186 GPIArgCallEvent, GPIArgCallActiveEvent, CorrelationID,
188 GPIArgCallActiveEvent =
nullptr;
197 auto contextHandle =
reinterpret_cast<pi_context>(impl->getHandleRef());
198 const auto &Plugin = impl->getPlugin();
206 return "PI_PLATFORM_INFO_PROFILE";
208 return "PI_PLATFORM_INFO_VERSION";
210 return "PI_PLATFORM_INFO_NAME";
212 return "PI_PLATFORM_INFO_VENDOR";
214 return "PI_PLATFORM_INFO_EXTENSIONS";
216 return "PI_EXT_PLATFORM_INFO_BACKEND";
218 die(
"Unknown pi_platform_info value passed to "
219 "sycl::detail::pi::platformInfoToString");
223 assertion(((Flag == 0u) || ((Flag & (Flag - 1)) == 0)) &&
224 "More than one bit set");
226 std::stringstream Sstream;
230 Sstream <<
"pi_mem_flags(0)";
233 Sstream <<
"PI_MEM_FLAGS_ACCESS_RW";
236 Sstream <<
"PI_MEM_FLAGS_HOST_PTR_USE";
239 Sstream <<
"PI_MEM_FLAGS_HOST_PTR_COPY";
242 Sstream <<
"unknown pi_mem_flags bit == " << Flag;
245 return Sstream.str();
249 std::stringstream Sstream;
250 bool FoundFlag =
false;
252 auto FlagSeparator = [](
bool FoundFlag) {
return FoundFlag ?
"|" :
""; };
259 Sstream <<
"pi_mem_flags(0)";
261 for (
const auto Flag : ValidFlags) {
271 if (UnkownBits.any()) {
272 Sstream << FlagSeparator(FoundFlag)
273 <<
"unknown pi_mem_flags bits == " << UnkownBits;
277 return Sstream.str();
287 std::vector<std::pair<std::string, backend>> PluginNames;
300 if (OdsTargetList && FilterList) {
302 "ONEAPI_DEVICE_SELECTOR cannot be used in "
303 "conjunction with SYCL_DEVICE_FILTER");
304 }
else if (!FilterList && !OdsTargetList) {
306 PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
310 PluginNames.emplace_back(__SYCL_UR_PLUGIN_NAME,
backend::all);
311 PluginNames.emplace_back(__SYCL_NATIVE_CPU_PLUGIN_NAME,
313 }
else if (FilterList) {
314 std::vector<device_filter> Filters = FilterList->
get();
315 bool OpenCLFound =
false;
316 bool LevelZeroFound =
false;
317 bool CudaFound =
false;
318 bool EsimdCpuFound =
false;
319 bool HIPFound =
false;
320 bool NativeCPUFound =
false;
330 PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
332 LevelZeroFound =
true;
336 PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME,
340 if (!EsimdCpuFound && Backend == backend::ext_intel_esimd_emulator) {
341 PluginNames.emplace_back(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME,
342 backend::ext_intel_esimd_emulator);
343 EsimdCpuFound =
true;
347 PluginNames.emplace_back(__SYCL_HIP_PLUGIN_NAME,
351 if (!NativeCPUFound &&
353 PluginNames.emplace_back(__SYCL_NATIVE_CPU_PLUGIN_NAME,
356 PluginNames.emplace_back(__SYCL_UR_PLUGIN_NAME,
backend::all);
364 PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
368 PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME,
372 PluginNames.emplace_back(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME,
373 backend::ext_intel_esimd_emulator);
379 PluginNames.emplace_back(__SYCL_NATIVE_CPU_PLUGIN_NAME,
382 PluginNames.emplace_back(__SYCL_UR_PLUGIN_NAME,
backend::all);
404 const std::shared_ptr<PiPlugin> &PluginInformation) {
409 if (PluginInitializeFunction ==
nullptr)
412 int Err = PluginInitializeFunction(PluginInformation.get());
416 assert((Err == PI_SUCCESS) &&
"Unexpected error when binding to Plugin.");
425 return (TraceLevelMask & Level) == Level;
430 static std::once_flag PluginsInitDone;
433 std::call_once(PluginsInitDone, [&]() {
443 std::vector<std::tuple<std::string, backend, void *>>
444 loadPlugins(
const std::vector<std::pair<std::string, backend>> &&PluginNames);
447 const std::vector<std::pair<std::string, backend>> PluginNames =
452 <<
"No Plugins Found." << std::endl;
455 std::vector<std::tuple<std::string, backend, void *>> LoadedPlugins =
458 for (
auto &[Name, Backend, Library] : LoadedPlugins) {
459 std::shared_ptr<PiPlugin> PluginInformation = std::make_shared<PiPlugin>(
466 <<
"Check if plugin is present. "
467 <<
"Failed to load plugin: " << Name << std::endl;
472 if (!
bindPlugin(Library, PluginInformation)) {
475 <<
"Failed to bind PI APIs to the plugin: " << Name
480 PluginPtr &NewPlugin = Plugins.emplace_back(
481 std::make_shared<plugin>(PluginInformation, Backend, Library));
484 <<
"Plugin found and successfully loaded: " << Name
485 <<
" [ PluginVersion: "
486 << NewPlugin->getPiPlugin().PluginVersion <<
" ]" << std::endl;
489 #ifdef XPTI_ENABLE_INSTRUMENTATION
510 xpti::payload_t GraphPayload(
"application_graph");
511 uint64_t GraphInstanceNo;
513 xptiMakeEvent(
"application_graph", &GraphPayload, xpti::trace_graph_event,
514 xpti_at::active, &GraphInstanceNo);
515 if (GSYCLGraphEvent) {
518 xptiNotifySubscribers(StreamID, xpti::trace_graph_create,
nullptr,
519 GSYCLGraphEvent, GraphInstanceNo,
nullptr);
525 xpti::payload_t PIPayload(
"Plugin Interface Layer");
526 uint64_t PiInstanceNo;
528 xptiMakeEvent(
"PI Layer", &PIPayload, xpti::trace_algorithm_event,
529 xpti_at::active, &PiInstanceNo);
533 xpti::payload_t PIArgPayload(
534 "Plugin Interface Layer (with function arguments)");
535 uint64_t PiArgInstanceNo;
536 GPIArgCallEvent = xptiMakeEvent(
"PI Layer with arguments", &PIArgPayload,
537 xpti::trace_algorithm_event, xpti_at::active,
552 for (
auto &P : Plugins)
553 if (P->hasBackend(BE)) {
558 throw runtime_error(
"pi::getPlugin couldn't find plugin",
559 PI_ERROR_INVALID_OPERATION);
562 template __SYCL_EXPORT
const PluginPtr &getPlugin<backend::opencl>();
564 getPlugin<backend::ext_oneapi_level_zero>();
566 getPlugin<backend::ext_intel_esimd_emulator>();
567 template __SYCL_EXPORT
const PluginPtr &getPlugin<backend::ext_oneapi_cuda>();
568 template __SYCL_EXPORT
const PluginPtr &getPlugin<backend::ext_oneapi_hip>();
574 [[noreturn]]
void die(
const char *Message) {
575 std::cerr <<
"pi_die: " << Message << std::endl;
585 template <
typename ResT>
588 assert(NumBytes <=
sizeof(ResT));
591 for (
size_t I = 0; I < NumBytes; ++I) {
592 Result = (Result << 8) | static_cast<ResT>(Data[I]);
595 std::copy(Data, Data + NumBytes,
reinterpret_cast<char *
>(&Result));
602 const unsigned char *ImgData,
605 bool Is64bit = ImgData[4] == 2;
606 bool IsBigEndian = ImgData[5] == 2;
609 size_t SectionHeaderOffsetInfoOffset = Is64bit ? 0x28 : 0x20;
610 size_t SectionHeaderSizeInfoOffset = Is64bit ? 0x3A : 0x2E;
611 size_t SectionHeaderNumInfoOffset = Is64bit ? 0x3C : 0x30;
612 size_t SectionStringsHeaderIndexInfoOffset = Is64bit ? 0x3E : 0x32;
615 if (ImgSize < SectionStringsHeaderIndexInfoOffset + 2)
620 uint64_t SectionHeaderOffset = readELFValue<uint64_t>(
621 ImgData + SectionHeaderOffsetInfoOffset, Is64bit ? 8 : 4, IsBigEndian);
622 uint16_t SectionHeaderSize = readELFValue<uint16_t>(
623 ImgData + SectionHeaderSizeInfoOffset, 2, IsBigEndian);
624 uint16_t SectionHeaderNum = readELFValue<uint16_t>(
625 ImgData + SectionHeaderNumInfoOffset, 2, IsBigEndian);
626 uint16_t SectionStringsHeaderIndex = readELFValue<uint16_t>(
627 ImgData + SectionStringsHeaderIndexInfoOffset, 2, IsBigEndian);
631 if (ImgSize < SectionHeaderOffset + SectionHeaderNum * SectionHeaderSize ||
632 SectionStringsHeaderIndex >= SectionHeaderNum)
636 size_t SectionStringsInfoOffset = Is64bit ? 0x18 : 0x10;
637 const unsigned char *SectionStringsHeaderData =
638 ImgData + SectionHeaderOffset +
639 SectionStringsHeaderIndex * SectionHeaderSize;
640 uint64_t SectionStrings = readELFValue<uint64_t>(
641 SectionStringsHeaderData + SectionStringsInfoOffset, Is64bit ? 8 : 4,
643 const unsigned char *SectionStringsData = ImgData + SectionStrings;
647 for (
size_t I = 0; I < SectionHeaderNum; ++I) {
649 const unsigned char *HeaderData =
650 ImgData + SectionHeaderOffset + I * SectionHeaderSize;
651 uint32_t SectionNameOffset =
652 readELFValue<uint32_t>(HeaderData, 4, IsBigEndian);
656 const char *SectionName =
657 reinterpret_cast<const char *
>(SectionStringsData + SectionNameOffset);
658 if (SectionName == ExpectedSectionName)
667 assert(ImgSize >= 18 &&
"Not enough bytes to have an ELF header type.");
669 bool IsBigEndian = ImgData[5] == 2;
670 return readELFValue<uint16_t>(ImgData + 16, 2, IsBigEndian);
678 const uint32_t Magic;
684 if (ImgSize >=
sizeof(Fmts[0].Magic)) {
685 std::remove_const_t<decltype(Fmts[0].Magic)> Hdr = 0;
686 std::copy(ImgData, ImgData +
sizeof(Hdr),
reinterpret_cast<char *
>(&Hdr));
689 for (
const auto &Fmt : Fmts) {
690 if (Hdr == Fmt.Magic)
697 const uint16_t Magic;
703 if (Hdr == 0x464c457F && ImgSize >= 18) {
705 for (
const auto &ELFFmt : ELFFmts) {
706 if (HdrType == ELFFmt.Magic)