24 #include <sycl/version.hpp>
35 #ifdef XPTI_ENABLE_INSTRUMENTATION
38 #include "xpti/xpti_trace_framework.h"
44 #ifdef XPTI_ENABLE_INSTRUMENTATION
48 xpti_td *GSYCLGraphEvent =
nullptr;
50 xpti_td *GPICallEvent =
nullptr;
52 xpti_td *GPIArgCallEvent =
nullptr;
53 #endif // XPTI_ENABLE_INSTRUMENTATION
56 void *ReturnOpaqueData =
nullptr;
57 const sycl::detail::plugin &Plugin = sycl::detail::pi::getPlugin<BE>();
60 OpaqueDataParam, &ReturnOpaqueData);
62 return ReturnOpaqueData;
65 template __SYCL_EXPORT
void *
66 getPluginOpaqueData<sycl::backend::ext_intel_esimd_emulator>(
void *);
77 uint64_t CorrelationID = 0;
78 #ifdef XPTI_ENABLE_INSTRUMENTATION
109 if (xptiTraceEnabled()) {
111 CorrelationID = xptiGetUniqueId();
112 xptiNotifySubscribers(
113 StreamID, (uint16_t)xpti::trace_point_type_t::function_begin,
114 GPICallEvent,
nullptr, CorrelationID,
static_cast<const void *
>(FName));
116 #endif // XPTI_ENABLE_INSTRUMENTATION
117 return CorrelationID;
121 #ifdef XPTI_ENABLE_INSTRUMENTATION
122 if (xptiTraceEnabled()) {
128 xptiNotifySubscribers(
129 StreamID, (uint16_t)xpti::trace_point_type_t::function_end,
130 GPICallEvent,
nullptr, CorrelationID,
static_cast<const void *
>(FName));
132 #endif // XPTI_ENABLE_INSTRUMENTATION
136 unsigned char *ArgsData,
138 uint64_t CorrelationID = 0;
139 #ifdef XPTI_ENABLE_INSTRUMENTATION
140 if (xptiTraceEnabled()) {
142 CorrelationID = xptiGetUniqueId();
144 xpti::function_with_args_t Payload{FuncID, FuncName, ArgsData,
nullptr,
147 xptiNotifySubscribers(
148 StreamID, (uint16_t)xpti::trace_point_type_t::function_with_args_begin,
149 GPIArgCallEvent,
nullptr, CorrelationID, &Payload);
152 return CorrelationID;
156 const char *FuncName,
unsigned char *ArgsData,
158 #ifdef XPTI_ENABLE_INSTRUMENTATION
159 if (xptiTraceEnabled()) {
162 xpti::function_with_args_t Payload{FuncID, FuncName, ArgsData, &Result,
165 xptiNotifySubscribers(
166 StreamID, (uint16_t)xpti::trace_point_type_t::function_with_args_end,
167 GPIArgCallEvent,
nullptr, CorrelationID, &Payload);
176 auto contextHandle =
reinterpret_cast<pi_context>(impl->getHandleRef());
177 auto plugin = impl->getPlugin();
185 return "PI_PLATFORM_INFO_PROFILE";
187 return "PI_PLATFORM_INFO_VERSION";
189 return "PI_PLATFORM_INFO_NAME";
191 return "PI_PLATFORM_INFO_VENDOR";
193 return "PI_PLATFORM_INFO_EXTENSIONS";
195 die(
"Unknown pi_platform_info value passed to "
196 "sycl::detail::pi::platformInfoToString");
200 assertion(((Flag == 0u) || ((Flag & (Flag - 1)) == 0)) &&
201 "More than one bit set");
203 std::stringstream Sstream;
207 Sstream <<
"pi_mem_flags(0)";
210 Sstream <<
"PI_MEM_FLAGS_ACCESS_RW";
213 Sstream <<
"PI_MEM_FLAGS_HOST_PTR_USE";
216 Sstream <<
"PI_MEM_FLAGS_HOST_PTR_COPY";
219 Sstream <<
"unknown pi_mem_flags bit == " << Flag;
222 return Sstream.str();
226 std::stringstream Sstream;
227 bool FoundFlag =
false;
229 auto FlagSeparator = [](
bool FoundFlag) {
return FoundFlag ?
"|" :
""; };
236 Sstream <<
"pi_mem_flags(0)";
238 for (
const auto Flag : ValidFlags) {
248 if (UnkownBits.any()) {
249 Sstream << FlagSeparator(FoundFlag)
250 <<
"unknown pi_mem_flags bits == " << UnkownBits;
254 return Sstream.str();
264 std::vector<std::pair<std::string, backend>> PluginNames;
277 if (OdsTargetList && FilterList) {
279 "ONEAPI_DEVICE_SELECTOR cannot be used in "
280 "conjunction with SYCL_DEVICE_FILTER");
281 }
else if (!FilterList && !OdsTargetList) {
282 PluginNames.emplace_back(__SYCL_OPENCL_PLUGIN_NAME, backend::opencl);
283 PluginNames.emplace_back(__SYCL_UNIFIED_RUNTIME_PLUGIN_NAME,
284 backend::ext_oneapi_unified_runtime);
285 PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
286 backend::ext_oneapi_level_zero);
287 PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME, backend::ext_oneapi_cuda);
288 PluginNames.emplace_back(__SYCL_HIP_PLUGIN_NAME, backend::ext_oneapi_hip);
289 }
else if (FilterList) {
290 std::vector<device_filter> Filters = FilterList->
get();
291 bool OpenCLFound =
false;
292 bool LevelZeroFound =
false;
293 bool CudaFound =
false;
294 bool EsimdCpuFound =
false;
295 bool HIPFound =
false;
297 backend Backend = Filter.Backend ? Filter.Backend.value() : backend::all;
299 (Backend == backend::opencl || Backend == backend::all)) {
300 PluginNames.emplace_back(__SYCL_OPENCL_PLUGIN_NAME, backend::opencl);
303 if (!LevelZeroFound && (Backend == backend::ext_oneapi_level_zero ||
304 Backend == backend::all)) {
305 PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
306 backend::ext_oneapi_level_zero);
307 LevelZeroFound =
true;
310 (Backend == backend::ext_oneapi_cuda || Backend == backend::all)) {
311 PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME,
312 backend::ext_oneapi_cuda);
315 if (!EsimdCpuFound && Backend == backend::ext_intel_esimd_emulator) {
316 PluginNames.emplace_back(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME,
317 backend::ext_intel_esimd_emulator);
318 EsimdCpuFound =
true;
321 (Backend == backend::ext_oneapi_hip || Backend == backend::all)) {
322 PluginNames.emplace_back(__SYCL_HIP_PLUGIN_NAME,
323 backend::ext_oneapi_hip);
330 PluginNames.emplace_back(__SYCL_OPENCL_PLUGIN_NAME, backend::opencl);
333 PluginNames.emplace_back(__SYCL_UNIFIED_RUNTIME_PLUGIN_NAME,
334 backend::ext_oneapi_unified_runtime);
337 PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
338 backend::ext_oneapi_level_zero);
341 PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME,
342 backend::ext_oneapi_cuda);
345 PluginNames.emplace_back(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME,
346 backend::ext_intel_esimd_emulator);
349 PluginNames.emplace_back(__SYCL_HIP_PLUGIN_NAME, backend::ext_oneapi_hip);
372 const std::shared_ptr<PiPlugin> &PluginInformation) {
377 if (PluginInitializeFunction ==
nullptr)
380 int Err = PluginInitializeFunction(PluginInformation.get());
384 assert((Err == PI_SUCCESS) &&
"Unexpected error when binding to Plugin.");
393 return (TraceLevelMask & Level) == Level;
398 static std::once_flag PluginsInitDone;
401 std::call_once(PluginsInitDone, [&]() {
404 return GlobalHandler::instance().getPlugins();
408 std::vector<std::pair<std::string, backend>> PluginNames =
findPlugins();
412 <<
"No Plugins Found." << std::endl;
414 const std::string LibSYCLDir =
415 sycl::detail::OSUtil::getCurrentDSODir() + sycl::detail::OSUtil::DirSep;
417 for (
unsigned int I = 0; I < PluginNames.size(); I++) {
418 std::shared_ptr<PiPlugin> PluginInformation = std::make_shared<PiPlugin>(
422 void *Library =
loadPlugin(LibSYCLDir + PluginNames[I].first);
427 <<
"Check if plugin is present. "
428 <<
"Failed to load plugin: " << PluginNames[I].first
434 if (!
bindPlugin(Library, PluginInformation)) {
437 <<
"Failed to bind PI APIs to the plugin: "
438 << PluginNames[I].first << std::endl;
442 plugin &NewPlugin = Plugins.emplace_back(
443 plugin(PluginInformation, PluginNames[I].second, Library));
446 <<
"Plugin found and successfully loaded: "
447 << PluginNames[I].first
448 <<
" [ PluginVersion: " << NewPlugin.
getPiPlugin().PluginVersion
449 <<
" ]" << std::endl;
452 #ifdef XPTI_ENABLE_INSTRUMENTATION
453 GlobalHandler::instance().getXPTIRegistry().initializeFrameworkOnce();
470 GlobalHandler::instance().getXPTIRegistry().initializeStream(
473 xpti::payload_t GraphPayload(
"application_graph");
474 uint64_t GraphInstanceNo;
476 xptiMakeEvent(
"application_graph", &GraphPayload, xpti::trace_graph_event,
477 xpti_at::active, &GraphInstanceNo);
478 if (GSYCLGraphEvent) {
481 xptiNotifySubscribers(StreamID, xpti::trace_graph_create,
nullptr,
482 GSYCLGraphEvent, GraphInstanceNo,
nullptr);
486 GlobalHandler::instance().getXPTIRegistry().initializeStream(
488 xpti::payload_t PIPayload(
"Plugin Interface Layer");
489 uint64_t PiInstanceNo;
491 xptiMakeEvent(
"PI Layer", &PIPayload, xpti::trace_algorithm_event,
492 xpti_at::active, &PiInstanceNo);
494 GlobalHandler::instance().getXPTIRegistry().initializeStream(
496 xpti::payload_t PIArgPayload(
497 "Plugin Interface Layer (with function arguments)");
498 uint64_t PiArgInstanceNo;
499 GPIArgCallEvent = xptiMakeEvent(
"PI Layer with arguments", &PIArgPayload,
500 xpti::trace_algorithm_event, xpti_at::active,
507 static const plugin *Plugin =
nullptr;
512 for (
const auto &P : Plugins)
513 if (P.getBackend() == BE) {
518 throw runtime_error(
"pi::getPlugin couldn't find plugin",
519 PI_ERROR_INVALID_OPERATION);
522 template __SYCL_EXPORT
const plugin &getPlugin<backend::opencl>();
523 template __SYCL_EXPORT
const plugin &
524 getPlugin<backend::ext_oneapi_level_zero>();
525 template __SYCL_EXPORT
const plugin &
526 getPlugin<backend::ext_intel_esimd_emulator>();
527 template __SYCL_EXPORT
const plugin &getPlugin<backend::ext_oneapi_cuda>();
533 [[noreturn]]
void die(
const char *Message) {
534 std::cerr <<
"pi_die: " << Message << std::endl;
544 template <
typename ResT>
547 assert(NumBytes <=
sizeof(ResT));
550 for (
size_t I = 0; I < NumBytes; ++I) {
551 Result = (Result << 8) | static_cast<ResT>(Data[I]);
554 std::copy(Data, Data + NumBytes,
reinterpret_cast<char *
>(&Result));
561 const unsigned char *ImgData,
564 bool Is64bit = ImgData[4] == 2;
565 bool IsBigEndian = ImgData[5] == 2;
568 size_t SectionHeaderOffsetInfoOffset = Is64bit ? 0x28 : 0x20;
569 size_t SectionHeaderSizeInfoOffset = Is64bit ? 0x3A : 0x2E;
570 size_t SectionHeaderNumInfoOffset = Is64bit ? 0x3C : 0x30;
571 size_t SectionStringsHeaderIndexInfoOffset = Is64bit ? 0x3E : 0x32;
574 if (ImgSize < SectionStringsHeaderIndexInfoOffset + 2)
579 uint64_t SectionHeaderOffset = readELFValue<uint64_t>(
580 ImgData + SectionHeaderOffsetInfoOffset, Is64bit ? 8 : 4, IsBigEndian);
581 uint16_t SectionHeaderSize = readELFValue<uint16_t>(
582 ImgData + SectionHeaderSizeInfoOffset, 2, IsBigEndian);
583 uint16_t SectionHeaderNum = readELFValue<uint16_t>(
584 ImgData + SectionHeaderNumInfoOffset, 2, IsBigEndian);
585 uint16_t SectionStringsHeaderIndex = readELFValue<uint16_t>(
586 ImgData + SectionStringsHeaderIndexInfoOffset, 2, IsBigEndian);
590 if (ImgSize < SectionHeaderOffset + SectionHeaderNum * SectionHeaderSize ||
591 SectionStringsHeaderIndex >= SectionHeaderNum)
595 size_t SectionStringsInfoOffset = Is64bit ? 0x18 : 0x10;
596 const unsigned char *SectionStringsHeaderData =
597 ImgData + SectionHeaderOffset +
598 SectionStringsHeaderIndex * SectionHeaderSize;
599 uint64_t SectionStrings = readELFValue<uint64_t>(
600 SectionStringsHeaderData + SectionStringsInfoOffset, Is64bit ? 8 : 4,
602 const unsigned char *SectionStringsData = ImgData + SectionStrings;
606 for (
size_t I = 0; I < SectionHeaderNum; ++I) {
608 const unsigned char *HeaderData =
609 ImgData + SectionHeaderOffset + I * SectionHeaderSize;
610 uint32_t SectionNameOffset =
611 readELFValue<uint32_t>(HeaderData, 4, IsBigEndian);
615 const char *SectionName =
616 reinterpret_cast<const char *
>(SectionStringsData + SectionNameOffset);
617 if (SectionName == ExpectedSectionName)
626 assert(ImgSize >= 18 &&
"Not enough bytes to have an ELF header type.");
628 bool IsBigEndian = ImgData[5] == 2;
629 return readELFValue<uint16_t>(ImgData + 16, 2, IsBigEndian);
637 const uint32_t Magic;
643 if (ImgSize >=
sizeof(Fmts[0].Magic)) {
645 std::copy(ImgData, ImgData +
sizeof(Hdr),
reinterpret_cast<char *
>(&Hdr));
648 for (
const auto &Fmt : Fmts) {
649 if (Hdr == Fmt.Magic)
656 const uint16_t Magic;
662 if (Hdr == 0x464c457F && ImgSize >= 18) {
664 for (
const auto &ELFFmt : ELFFmts) {
665 if (HdrType == ELFFmt.Magic)