DPC++ Runtime
Runtime libraries for oneAPI Data Parallel C++
pi.cpp
Go to the documentation of this file.
1 //===-- pi.cpp - PI utilities implementation -------------------*- C++ -*--===//
2 //
3 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4 // See https://llvm.org/LICENSE.txt for license information.
5 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6 //
7 //===----------------------------------------------------------------------===//
8 
13 
14 #include "context_impl.hpp"
15 #include <CL/sycl/context.hpp>
18 #include <CL/sycl/detail/pi.hpp>
20 #include <CL/sycl/version.hpp>
21 #include <detail/config.hpp>
23 #include <detail/plugin.hpp>
24 #include <detail/xpti_registry.hpp>
25 
26 #include <bitset>
27 #include <cstdarg>
28 #include <cstring>
29 #include <iostream>
30 #include <map>
31 #include <sstream>
32 #include <stddef.h>
33 #include <string>
34 
35 #ifdef XPTI_ENABLE_INSTRUMENTATION
36 // Include the headers necessary for emitting
37 // traces using the trace framework
38 #include "xpti/xpti_trace_framework.h"
39 #endif
40 
41 #define STR(x) #x
42 #define SYCL_VERSION_STR \
43  "sycl " STR(__LIBSYCL_MAJOR_VERSION) "." STR(__LIBSYCL_MINOR_VERSION)
44 
46 namespace sycl {
47 namespace detail {
48 #ifdef XPTI_ENABLE_INSTRUMENTATION
49 // Global (to the SYCL runtime) graph handle that all command groups are a
50 // child of
52 xpti_td *GSYCLGraphEvent = nullptr;
54 xpti_td *GPICallEvent = nullptr;
56 xpti_td *GPIArgCallEvent = nullptr;
59 constexpr uint32_t GMajVer = __LIBSYCL_MAJOR_VERSION;
60 constexpr uint32_t GMinVer = __LIBSYCL_MINOR_VERSION;
61 constexpr const char *GVerStr = SYCL_VERSION_STR;
62 #endif // XPTI_ENABLE_INSTRUMENTATION
63 
64 template <cl::sycl::backend BE>
65 void *getPluginOpaqueData(void *OpaqueDataParam) {
66  void *ReturnOpaqueData = nullptr;
67  const cl::sycl::detail::plugin &Plugin =
68  cl::sycl::detail::pi::getPlugin<BE>();
69 
71  OpaqueDataParam, &ReturnOpaqueData);
72 
73  return ReturnOpaqueData;
74 }
75 
76 template __SYCL_EXPORT void *
77 getPluginOpaqueData<cl::sycl::backend::ext_intel_esimd_emulator>(void *);
78 
79 namespace pi {
80 
81 static void initializePlugins(std::vector<plugin> &Plugins);
82 
83 bool XPTIInitDone = false;
84 
85 // Implementation of the SYCL PI API call tracing methods that use XPTI
86 // framework to emit these traces that will be used by tools.
87 uint64_t emitFunctionBeginTrace(const char *FName) {
88  uint64_t CorrelationID = 0;
89 #ifdef XPTI_ENABLE_INSTRUMENTATION
90  // The function_begin and function_end trace point types are defined to
91  // trace library API calls and they are currently enabled here for support
92  // tools that need the API scope. The methods emitFunctionBeginTrace() and
93  // emitFunctionEndTrace() can be extended to also trace the arguments of the
94  // PI API call using a trace point type the extends the predefined trace
95  // point types.
96  //
97  // You can use the sample collector in llvm/xptifw/samples/syclpi_collector
98  // to print the API traces and also extend them to support arguments that
99  // may be traced later.
100  //
120  if (xptiTraceEnabled()) {
121  uint8_t StreamID = xptiRegisterStream(SYCL_PICALL_STREAM_NAME);
122  CorrelationID = xptiGetUniqueId();
123  xptiNotifySubscribers(
124  StreamID, (uint16_t)xpti::trace_point_type_t::function_begin,
125  GPICallEvent, nullptr, CorrelationID, static_cast<const void *>(FName));
126  }
127 #endif // XPTI_ENABLE_INSTRUMENTATION
128  return CorrelationID;
129 }
130 
131 void emitFunctionEndTrace(uint64_t CorrelationID, const char *FName) {
132 #ifdef XPTI_ENABLE_INSTRUMENTATION
133  if (xptiTraceEnabled()) {
134  // CorrelationID is the unique ID that ties together a function_begin and
135  // function_end pair of trace calls. The splitting of a scoped_notify into
136  // two function calls incurs an additional overhead as the StreamID must
137  // be looked up twice.
138  uint8_t StreamID = xptiRegisterStream(SYCL_PICALL_STREAM_NAME);
139  xptiNotifySubscribers(
140  StreamID, (uint16_t)xpti::trace_point_type_t::function_end,
141  GPICallEvent, nullptr, CorrelationID, static_cast<const void *>(FName));
142  }
143 #endif // XPTI_ENABLE_INSTRUMENTATION
144 }
145 
146 uint64_t emitFunctionWithArgsBeginTrace(uint32_t FuncID, const char *FuncName,
147  unsigned char *ArgsData,
148  pi_plugin Plugin) {
149  uint64_t CorrelationID = 0;
150 #ifdef XPTI_ENABLE_INSTRUMENTATION
151  if (xptiTraceEnabled()) {
152  uint8_t StreamID = xptiRegisterStream(SYCL_PIDEBUGCALL_STREAM_NAME);
153  CorrelationID = xptiGetUniqueId();
154 
155  xpti::function_with_args_t Payload{FuncID, FuncName, ArgsData, nullptr,
156  &Plugin};
157 
158  xptiNotifySubscribers(
159  StreamID, (uint16_t)xpti::trace_point_type_t::function_with_args_begin,
160  GPIArgCallEvent, nullptr, CorrelationID, &Payload);
161  }
162 #endif
163  return CorrelationID;
164 }
165 
166 void emitFunctionWithArgsEndTrace(uint64_t CorrelationID, uint32_t FuncID,
167  const char *FuncName, unsigned char *ArgsData,
168  pi_result Result, pi_plugin Plugin) {
169 #ifdef XPTI_ENABLE_INSTRUMENTATION
170  if (xptiTraceEnabled()) {
171  uint8_t StreamID = xptiRegisterStream(SYCL_PIDEBUGCALL_STREAM_NAME);
172 
173  xpti::function_with_args_t Payload{FuncID, FuncName, ArgsData, &Result,
174  &Plugin};
175 
176  xptiNotifySubscribers(
177  StreamID, (uint16_t)xpti::trace_point_type_t::function_with_args_end,
178  GPIArgCallEvent, nullptr, CorrelationID, &Payload);
179  }
180 #endif
181 }
182 
185  void *user_data) {
186  auto impl = getSyclObjImpl(context);
187  auto contextHandle = reinterpret_cast<pi_context>(impl->getHandleRef());
188  auto plugin = impl->getPlugin();
190  user_data);
191 }
192 
194  switch (info) {
196  return "PI_PLATFORM_INFO_PROFILE";
198  return "PI_PLATFORM_INFO_VERSION";
200  return "PI_PLATFORM_INFO_NAME";
202  return "PI_PLATFORM_INFO_VENDOR";
204  return "PI_PLATFORM_INFO_EXTENSIONS";
205  }
206  die("Unknown pi_platform_info value passed to "
207  "cl::sycl::detail::pi::platformInfoToString");
208 }
209 
210 std::string memFlagToString(pi_mem_flags Flag) {
211  assertion(((Flag == 0u) || ((Flag & (Flag - 1)) == 0)) &&
212  "More than one bit set");
213 
214  std::stringstream Sstream;
215 
216  switch (Flag) {
217  case pi_mem_flags{0}:
218  Sstream << "pi_mem_flags(0)";
219  break;
221  Sstream << "PI_MEM_FLAGS_ACCESS_RW";
222  break;
224  Sstream << "PI_MEM_FLAGS_HOST_PTR_USE";
225  break;
227  Sstream << "PI_MEM_FLAGS_HOST_PTR_COPY";
228  break;
229  default:
230  Sstream << "unknown pi_mem_flags bit == " << Flag;
231  }
232 
233  return Sstream.str();
234 }
235 
236 std::string memFlagsToString(pi_mem_flags Flags) {
237  std::stringstream Sstream;
238  bool FoundFlag = false;
239 
240  auto FlagSeparator = [](bool FoundFlag) { return FoundFlag ? "|" : ""; };
241 
242  pi_mem_flags ValidFlags[] = {PI_MEM_FLAGS_ACCESS_RW,
245 
246  if (Flags == 0u) {
247  Sstream << "pi_mem_flags(0)";
248  } else {
249  for (const auto Flag : ValidFlags) {
250  if (Flag & Flags) {
251  Sstream << FlagSeparator(FoundFlag) << memFlagToString(Flag);
252  FoundFlag = true;
253  }
254  }
255 
256  std::bitset<64> UnkownBits(Flags & ~(PI_MEM_FLAGS_ACCESS_RW |
259  if (UnkownBits.any()) {
260  Sstream << FlagSeparator(FoundFlag)
261  << "unknown pi_mem_flags bits == " << UnkownBits;
262  }
263  }
264 
265  return Sstream.str();
266 }
267 
268 // GlobalPlugin is a global Plugin used with Interoperability constructors that
269 // use OpenCL objects to construct SYCL class objects.
270 std::shared_ptr<plugin> GlobalPlugin;
271 
272 // Find the plugin at the appropriate location and return the location.
273 std::vector<std::pair<std::string, backend>> findPlugins() {
274  std::vector<std::pair<std::string, backend>> PluginNames;
275 
276  // TODO: Based on final design discussions, change the location where the
277  // plugin must be searched; how to identify the plugins etc. Currently the
278  // search is done for libpi_opencl.so/pi_opencl.dll file in LD_LIBRARY_PATH
279  // env only.
280  //
282  if (!FilterList) {
283  PluginNames.emplace_back(__SYCL_OPENCL_PLUGIN_NAME, backend::opencl);
284  PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
285  backend::ext_oneapi_level_zero);
286  PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME, backend::ext_oneapi_cuda);
287  PluginNames.emplace_back(__SYCL_HIP_PLUGIN_NAME, backend::ext_oneapi_hip);
288  PluginNames.emplace_back(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME,
289  backend::ext_intel_esimd_emulator);
290  } else {
291  std::vector<device_filter> Filters = FilterList->get();
292  bool OpenCLFound = false;
293  bool LevelZeroFound = false;
294  bool CudaFound = false;
295  bool EsimdCpuFound = false;
296  bool HIPFound = false;
297  for (const device_filter &Filter : Filters) {
298  backend Backend = Filter.Backend;
299  if (!OpenCLFound &&
300  (Backend == backend::opencl || Backend == backend::all)) {
301  PluginNames.emplace_back(__SYCL_OPENCL_PLUGIN_NAME, backend::opencl);
302  OpenCLFound = true;
303  }
304  if (!LevelZeroFound && (Backend == backend::ext_oneapi_level_zero ||
305  Backend == backend::all)) {
306  PluginNames.emplace_back(__SYCL_LEVEL_ZERO_PLUGIN_NAME,
307  backend::ext_oneapi_level_zero);
308  LevelZeroFound = true;
309  }
310  if (!CudaFound &&
311  (Backend == backend::ext_oneapi_cuda || Backend == backend::all)) {
312  PluginNames.emplace_back(__SYCL_CUDA_PLUGIN_NAME,
313  backend::ext_oneapi_cuda);
314  CudaFound = true;
315  }
316  if (!EsimdCpuFound && (Backend == backend::ext_intel_esimd_emulator ||
317  Backend == backend::all)) {
318  PluginNames.emplace_back(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME,
319  backend::ext_intel_esimd_emulator);
320  EsimdCpuFound = true;
321  }
322  if (!HIPFound &&
323  (Backend == backend::ext_oneapi_hip || Backend == backend::all)) {
324  PluginNames.emplace_back(__SYCL_HIP_PLUGIN_NAME,
325  backend::ext_oneapi_hip);
326  HIPFound = true;
327  }
328  }
329  }
330  return PluginNames;
331 }
332 
333 // Load the Plugin by calling the OS dependent library loading call.
334 // Return the handle to the Library.
335 void *loadPlugin(const std::string &PluginPath) {
336  return loadOsLibrary(PluginPath);
337 }
338 
339 // Unload the given plugin by calling teh OS-specific library unloading call.
340 // \param Library OS-specific library handle created when loading.
341 int unloadPlugin(void *Library) { return unloadOsLibrary(Library); }
342 
343 // Binds all the PI Interface APIs to Plugin Library Function Addresses.
344 // TODO: Remove the 'OclPtr' extension to PI_API.
345 // TODO: Change the functionality such that a single getOsLibraryFuncAddress
346 // call is done to get all Interface API mapping. The plugin interface also
347 // needs to setup infrastructure to route PI_CALLs to the appropriate plugins.
348 // Currently, we bind to a singe plugin.
349 bool bindPlugin(void *Library,
350  const std::shared_ptr<PiPlugin> &PluginInformation) {
351 
352  decltype(::piPluginInit) *PluginInitializeFunction = (decltype(
353  &::piPluginInit))(getOsLibraryFuncAddress(Library, "piPluginInit"));
354  if (PluginInitializeFunction == nullptr)
355  return false;
356 
357  int Err = PluginInitializeFunction(PluginInformation.get());
358 
359  // TODO: Compare Supported versions and check for backward compatibility.
360  // Make sure err is PI_SUCCESS.
361  assert((Err == PI_SUCCESS) && "Unexpected error when binding to Plugin.");
362  (void)Err;
363 
364  // TODO: Return a more meaningful value/enum.
365  return true;
366 }
367 
368 bool trace(TraceLevel Level) {
369  auto TraceLevelMask = SYCLConfig<SYCL_PI_TRACE>::get();
370  return (TraceLevelMask & Level) == Level;
371 }
372 
373 // Initializes all available Plugins.
374 std::vector<plugin> &initialize() {
375  static std::once_flag PluginsInitDone;
376  // std::call_once is blocking all other threads if a thread is already
377  // creating a vector of plugins. So, no additional lock is needed.
378  std::call_once(PluginsInitDone, [&]() {
379  initializePlugins(GlobalHandler::instance().getPlugins());
380  });
381  return GlobalHandler::instance().getPlugins();
382 }
383 
384 static void initializePlugins(std::vector<plugin> &Plugins) {
385  std::vector<std::pair<std::string, backend>> PluginNames = findPlugins();
386 
387  if (PluginNames.empty() && trace(PI_TRACE_ALL))
388  std::cerr << "SYCL_PI_TRACE[all]: "
389  << "No Plugins Found." << std::endl;
390 
391  for (unsigned int I = 0; I < PluginNames.size(); I++) {
392  std::shared_ptr<PiPlugin> PluginInformation = std::make_shared<PiPlugin>(
394  /*Targets=*/nullptr, /*FunctionPointers=*/{}});
395 
396  void *Library = loadPlugin(PluginNames[I].first);
397 
398  if (!Library) {
399  if (trace(PI_TRACE_ALL)) {
400  std::cerr << "SYCL_PI_TRACE[all]: "
401  << "Check if plugin is present. "
402  << "Failed to load plugin: " << PluginNames[I].first
403  << std::endl;
404  }
405  continue;
406  }
407 
408  if (!bindPlugin(Library, PluginInformation)) {
409  if (trace(PI_TRACE_ALL)) {
410  std::cerr << "SYCL_PI_TRACE[all]: "
411  << "Failed to bind PI APIs to the plugin: "
412  << PluginNames[I].first << std::endl;
413  }
414  continue;
415  }
417  // Use OpenCL as the default interoperability plugin.
418  // This will go away when we make backend interoperability selection
419  // explicit in SYCL-2020.
420  backend InteropBE = BE ? *BE : backend::opencl;
421 
422  if (InteropBE == backend::opencl &&
423  PluginNames[I].first.find("opencl") != std::string::npos) {
424  // Use the OpenCL plugin as the GlobalPlugin
425  GlobalPlugin =
426  std::make_shared<plugin>(PluginInformation, backend::opencl, Library);
427  } else if (InteropBE == backend::ext_oneapi_cuda &&
428  PluginNames[I].first.find("cuda") != std::string::npos) {
429  // Use the CUDA plugin as the GlobalPlugin
430  GlobalPlugin = std::make_shared<plugin>(
431  PluginInformation, backend::ext_oneapi_cuda, Library);
432  } else if (InteropBE == backend::ext_oneapi_hip &&
433  PluginNames[I].first.find("hip") != std::string::npos) {
434  // Use the HIP plugin as the GlobalPlugin
435  GlobalPlugin = std::make_shared<plugin>(PluginInformation,
436  backend::ext_oneapi_hip, Library);
437  } else if (InteropBE == backend::ext_oneapi_level_zero &&
438  PluginNames[I].first.find("level_zero") != std::string::npos) {
439  // Use the LEVEL_ZERO plugin as the GlobalPlugin
440  GlobalPlugin = std::make_shared<plugin>(
441  PluginInformation, backend::ext_oneapi_level_zero, Library);
442  } else if (InteropBE == backend::ext_intel_esimd_emulator &&
443  PluginNames[I].first.find("esimd_emulator") !=
444  std::string::npos) {
445  // Use the ESIMD_EMULATOR plugin as the GlobalPlugin
446  GlobalPlugin = std::make_shared<plugin>(
447  PluginInformation, backend::ext_intel_esimd_emulator, Library);
448  }
449  Plugins.emplace_back(
450  plugin(PluginInformation, PluginNames[I].second, Library));
452  std::cerr << "SYCL_PI_TRACE[basic]: "
453  << "Plugin found and successfully loaded: "
454  << PluginNames[I].first << std::endl;
455  }
456 
457 #ifdef XPTI_ENABLE_INSTRUMENTATION
458  GlobalHandler::instance().getXPTIRegistry().initializeFrameworkOnce();
459 
460  if (!(xptiTraceEnabled() && !XPTIInitDone))
461  return;
462  // Not sure this is the best place to initialize the framework; SYCL runtime
463  // team needs to advise on the right place, until then we piggy-back on the
464  // initialization of the PI layer.
465 
466  // Initialize the global events just once, in the case pi::initialize() is
467  // called multiple times
468  XPTIInitDone = true;
469  // Registers a new stream for 'sycl' and any plugin that wants to listen to
470  // this stream will register itself using this string or stream ID for this
471  // string.
472  uint8_t StreamID = xptiRegisterStream(SYCL_STREAM_NAME);
473  // Let all tool plugins know that a stream by the name of 'sycl' has been
474  // initialized and will be generating the trace stream.
475  GlobalHandler::instance().getXPTIRegistry().initializeStream(
476  SYCL_STREAM_NAME, GMajVer, GMinVer, GVerStr);
477  // Create a tracepoint to indicate the graph creation
478  xpti::payload_t GraphPayload("application_graph");
479  uint64_t GraphInstanceNo;
480  GSYCLGraphEvent =
481  xptiMakeEvent("application_graph", &GraphPayload, xpti::trace_graph_event,
482  xpti_at::active, &GraphInstanceNo);
483  if (GSYCLGraphEvent) {
484  // The graph event is a global event and will be used as the parent for
485  // all nodes (command groups)
486  xptiNotifySubscribers(StreamID, xpti::trace_graph_create, nullptr,
487  GSYCLGraphEvent, GraphInstanceNo, nullptr);
488  }
489 
490  // Let subscribers know a new stream is being initialized
491  GlobalHandler::instance().getXPTIRegistry().initializeStream(
492  SYCL_PICALL_STREAM_NAME, GMajVer, GMinVer, GVerStr);
493  xpti::payload_t PIPayload("Plugin Interface Layer");
494  uint64_t PiInstanceNo;
495  GPICallEvent =
496  xptiMakeEvent("PI Layer", &PIPayload, xpti::trace_algorithm_event,
497  xpti_at::active, &PiInstanceNo);
498 
499  GlobalHandler::instance().getXPTIRegistry().initializeStream(
500  SYCL_PIDEBUGCALL_STREAM_NAME, GMajVer, GMinVer, GVerStr);
501  xpti::payload_t PIArgPayload(
502  "Plugin Interface Layer (with function arguments)");
503  uint64_t PiArgInstanceNo;
504  GPIArgCallEvent = xptiMakeEvent("PI Layer with arguments", &PIArgPayload,
505  xpti::trace_algorithm_event, xpti_at::active,
506  &PiArgInstanceNo);
507 #endif
508 }
509 
510 // Get the plugin serving given backend.
511 template <backend BE> const plugin &getPlugin() {
512  static const plugin *Plugin = nullptr;
513  if (Plugin)
514  return *Plugin;
515 
516  const std::vector<plugin> &Plugins = pi::initialize();
517  for (const auto &P : Plugins)
518  if (P.getBackend() == BE) {
519  Plugin = &P;
520  return *Plugin;
521  }
522 
523  throw runtime_error("pi::getPlugin couldn't find plugin",
525 }
526 
527 template __SYCL_EXPORT const plugin &getPlugin<backend::opencl>();
528 template __SYCL_EXPORT const plugin &
529 getPlugin<backend::ext_oneapi_level_zero>();
530 template __SYCL_EXPORT const plugin &
531 getPlugin<backend::ext_intel_esimd_emulator>();
532 
533 // Report error and no return (keeps compiler from printing warnings).
534 // TODO: Probably change that to throw a catchable exception,
535 // but for now it is useful to see every failure.
536 //
537 [[noreturn]] void die(const char *Message) {
538  std::cerr << "pi_die: " << Message << std::endl;
539  std::terminate();
540 }
541 
542 void assertion(bool Condition, const char *Message) {
543  if (!Condition)
544  die(Message);
545 }
546 
547 std::ostream &operator<<(std::ostream &Out, const DeviceBinaryProperty &P) {
548  switch (P.Prop->Type) {
550  Out << "[UINT32] ";
551  break;
553  Out << "[Byte array] ";
554  break;
556  Out << "[String] ";
557  break;
558  default:
559  assert(false && "unsupported property");
560  return Out;
561  }
562  Out << P.Prop->Name << "=";
563 
564  switch (P.Prop->Type) {
566  Out << P.asUint32();
567  break;
569  ByteArray BA = P.asByteArray();
570  std::ios_base::fmtflags FlagsBackup = Out.flags();
571  Out << std::hex;
572  for (const auto &Byte : BA) {
573  Out << "0x" << static_cast<unsigned>(Byte) << " ";
574  }
575  Out.flags(FlagsBackup);
576  break;
577  }
579  Out << P.asCString();
580  break;
581  default:
582  assert(false && "Unsupported property");
583  return Out;
584  }
585  return Out;
586 }
587 
589  std::cerr << " --- Image " << Bin << "\n";
590  if (!Bin)
591  return;
592  std::cerr << " Version : " << (int)Bin->Version << "\n";
593  std::cerr << " Kind : " << (int)Bin->Kind << "\n";
594  std::cerr << " Format : " << (int)Bin->Format << "\n";
595  std::cerr << " Target : " << Bin->DeviceTargetSpec << "\n";
596  std::cerr << " Bin size : "
597  << ((intptr_t)Bin->BinaryEnd - (intptr_t)Bin->BinaryStart) << "\n";
598  std::cerr << " Compile options : "
599  << (Bin->CompileOptions ? Bin->CompileOptions : "NULL") << "\n";
600  std::cerr << " Link options : "
601  << (Bin->LinkOptions ? Bin->LinkOptions : "NULL") << "\n";
602  std::cerr << " Entries : ";
603  for (_pi_offload_entry EntriesIt = Bin->EntriesBegin;
604  EntriesIt != Bin->EntriesEnd; ++EntriesIt)
605  std::cerr << EntriesIt->name << " ";
606  std::cerr << "\n";
607  std::cerr << " Properties [" << Bin->PropertySetsBegin << "-"
608  << Bin->PropertySetsEnd << "]:\n";
609 
610  for (pi_device_binary_property_set PS = Bin->PropertySetsBegin;
611  PS != Bin->PropertySetsEnd; ++PS) {
612  std::cerr << " Category " << PS->Name << " [" << PS->PropertiesBegin
613  << "-" << PS->PropertiesEnd << "]:\n";
614 
615  for (pi_device_binary_property P = PS->PropertiesBegin;
616  P != PS->PropertiesEnd; ++P) {
617  std::cerr << " " << DeviceBinaryProperty(P) << "\n";
618  }
619  }
620 }
621 
622 void DeviceBinaryImage::dump(std::ostream &Out) const {
623  size_t ImgSize = getSize();
624  Out.write(reinterpret_cast<const char *>(Bin->BinaryStart), ImgSize);
625 }
626 
627 static pi_uint32 asUint32(const void *Addr) {
628  assert(Addr && "Addr is NULL");
629  const auto *P = reinterpret_cast<const unsigned char *>(Addr);
630  return (*P) | (*(P + 1) << 8) | (*(P + 2) << 16) | (*(P + 3) << 24);
631 }
632 
634  assert(Prop->Type == PI_PROPERTY_TYPE_UINT32 && "property type mismatch");
635  // if type fits into the ValSize - it is used to store the property value
636  assert(Prop->ValAddr == nullptr && "primitive types must be stored inline");
637  return sycl::detail::pi::asUint32(&Prop->ValSize);
638 }
639 
640 ByteArray DeviceBinaryProperty::asByteArray() const {
641  assert(Prop->Type == PI_PROPERTY_TYPE_BYTE_ARRAY && "property type mismatch");
642  assert(Prop->ValSize > 0 && "property size mismatch");
643  const auto *Data = pi::cast<const std::uint8_t *>(Prop->ValAddr);
644  return {Data, Prop->ValSize};
645 }
646 
647 const char *DeviceBinaryProperty::asCString() const {
648  assert(Prop->Type == PI_PROPERTY_TYPE_STRING && "property type mismatch");
649  assert(Prop->ValSize > 0 && "property size mismatch");
650  return pi::cast<const char *>(Prop->ValAddr);
651 }
652 
653 void DeviceBinaryImage::PropertyRange::init(pi_device_binary Bin,
654  const char *PropSetName) {
655  assert(!this->Begin && !this->End && "already initialized");
656  pi_device_binary_property_set PS = nullptr;
657 
658  for (PS = Bin->PropertySetsBegin; PS != Bin->PropertySetsEnd; ++PS) {
659  assert(PS->Name && "nameless property set - bug in the offload wrapper?");
660  if (!strcmp(PropSetName, PS->Name))
661  break;
662  }
663  if (PS == Bin->PropertySetsEnd) {
664  Begin = End = nullptr;
665  return;
666  }
667  Begin = PS->PropertiesBegin;
668  End = Begin ? PS->PropertiesEnd : nullptr;
669 }
670 
672 DeviceBinaryImage::getProperty(const char *PropName) const {
674  BoolProp.init(Bin, __SYCL_PI_PROPERTY_SET_SYCL_MISC_PROP);
675  if (!BoolProp.isAvailable())
676  return nullptr;
677  auto It = std::find_if(BoolProp.begin(), BoolProp.end(),
678  [=](pi_device_binary_property Prop) {
679  return !strcmp(PropName, Prop->Name);
680  });
681  if (It == BoolProp.end())
682  return nullptr;
683 
684  return *It;
685 }
686 
687 RT::PiDeviceBinaryType getBinaryImageFormat(const unsigned char *ImgData,
688  size_t ImgSize) {
689  struct {
691  const uint32_t Magic;
692  } Fmts[] = {{PI_DEVICE_BINARY_TYPE_SPIRV, 0x07230203},
694 
695  if (ImgSize >= sizeof(Fmts[0].Magic)) {
696  detail::remove_const_t<decltype(Fmts[0].Magic)> Hdr = 0;
697  std::copy(ImgData, ImgData + sizeof(Hdr), reinterpret_cast<char *>(&Hdr));
698 
699  for (const auto &Fmt : Fmts) {
700  if (Hdr == Fmt.Magic)
701  return Fmt.Fmt;
702  }
703  }
705 }
706 
707 void DeviceBinaryImage::init(pi_device_binary Bin) {
708  this->Bin = Bin;
709  // If device binary image format wasn't set by its producer, then can't change
710  // now, because 'Bin' data is part of the executable image loaded into memory
711  // which can't be modified (easily).
712  // TODO clang driver + ClangOffloadWrapper can figure out the format and set
713  // it when invoking the offload wrapper job
714  Format = static_cast<pi::PiDeviceBinaryType>(Bin->Format);
715 
716  if (Format == PI_DEVICE_BINARY_TYPE_NONE)
717  // try to determine the format; may remain "NONE"
718  Format = getBinaryImageFormat(Bin->BinaryStart, getSize());
719 
720  SpecConstIDMap.init(Bin, __SYCL_PI_PROPERTY_SET_SPEC_CONST_MAP);
721  DeviceLibReqMask.init(Bin, __SYCL_PI_PROPERTY_SET_DEVICELIB_REQ_MASK);
722  KernelParamOptInfo.init(Bin, __SYCL_PI_PROPERTY_SET_KERNEL_PARAM_OPT_INFO);
723  ProgramMetadata.init(Bin, __SYCL_PI_PROPERTY_SET_PROGRAM_METADATA);
724 }
725 
726 } // namespace pi
727 } // namespace detail
728 } // namespace sycl
729 } // __SYCL_INLINE_NAMESPACE(cl)
PI_DEVICE_BINARY_TYPE_NONE
static constexpr pi_device_binary_type PI_DEVICE_BINARY_TYPE_NONE
Definition: pi.h:669
cl::sycl::backend
backend
Definition: backend_types.hpp:21
cl::sycl::detail::SYCL_PIDEBUGCALL_STREAM_NAME
constexpr const char * SYCL_PIDEBUGCALL_STREAM_NAME
Definition: xpti_registry.hpp:31
_pi_offload_entry_struct
Definition: pi.h:626
cl::sycl::detail::pi::getPlugin
const plugin & getPlugin()
Definition: pi.cpp:511
PI_SUCCESS
@ PI_SUCCESS
Definition: pi.h:82
cl::sycl::detail::pi::TraceLevel
TraceLevel
Definition: pi.hpp:55
cl::sycl::detail::pi::print
std::enable_if<!std::is_pointer< T >::value, void >::type print(T val)
Definition: plugin_printers.hpp:24
cl::sycl::detail::pi::findPlugins
std::vector< std::pair< std::string, backend > > findPlugins()
Definition: pi.cpp:273
context_impl.hpp
device_filter.hpp
PI_INVALID_OPERATION
@ PI_INVALID_OPERATION
Definition: pi.h:84
PI_DEVICE_BINARY_TYPE_SPIRV
static constexpr pi_device_binary_type PI_DEVICE_BINARY_TYPE_SPIRV
Definition: pi.h:674
cl::sycl::detail::pi::initialize
std::vector< plugin > & initialize()
Definition: pi.cpp:374
_pi_device_binary_property_set_struct
Definition: pi.h:658
cl::sycl::detail::device_filter
Definition: device_filter.hpp:22
config.hpp
cl::sycl::detail::device_filter_list::get
std::vector< device_filter > & get()
Definition: device_filter.hpp:45
PI_DEVICE_BINARY_TYPE_LLVMIR_BITCODE
static constexpr pi_device_binary_type PI_DEVICE_BINARY_TYPE_LLVMIR_BITCODE
Definition: pi.h:676
PI_MEM_FLAGS_HOST_PTR_COPY
constexpr pi_mem_flags PI_MEM_FLAGS_HOST_PTR_COPY
Definition: pi.h:553
__SYCL_OPENCL_PLUGIN_NAME
#define __SYCL_OPENCL_PLUGIN_NAME
Definition: pi.hpp:71
cl::sycl::detail::pi::PI_TRACE_ALL
@ PI_TRACE_ALL
Definition: pi.hpp:58
PI_PROPERTY_TYPE_STRING
@ PI_PROPERTY_TYPE_STRING
Definition: pi.h:641
cl::sycl::detail::pi::assertion
void assertion(bool Condition, const char *Message=nullptr)
Definition: pi.cpp:542
cl::sycl::detail::pi::getBinaryImageFormat
PiDeviceBinaryType getBinaryImageFormat(const unsigned char *ImgData, size_t ImgSize)
Tries to determine the device binary image foramat.
Definition: pi.cpp:687
_pi_plugin
Definition: pi.h:1739
xpti_registry.hpp
cl::sycl::detail::pi::DeviceBinaryImage::PropertyRange::end
ConstIterator end() const
Definition: pi.hpp:278
_pi_device_binary_property_set_struct::Name
char * Name
Definition: pi.h:659
cl::sycl::detail::pi::memFlagToString
std::string memFlagToString(pi_mem_flags Flag)
Definition: pi.cpp:210
pi_device_binary_struct::PropertySetsEnd
pi_device_binary_property_set PropertySetsEnd
Definition: pi.h:784
pi_device_binary_struct::BinaryStart
const unsigned char * BinaryStart
Pointer to the target code start.
Definition: pi.h:775
_pi_result
_pi_result
Definition: pi.h:81
__SYCL_CUDA_PLUGIN_NAME
#define __SYCL_CUDA_PLUGIN_NAME
Definition: pi.hpp:73
piPluginInit
pi_result piPluginInit(pi_plugin *plugin_info)
Definition: pi_cuda.cpp:4841
cl::sycl::operator<<
std::ostream & operator<<(std::ostream &Out, backend be)
Definition: backend_types.hpp:46
__SYCL_PI_PROPERTY_SET_KERNEL_PARAM_OPT_INFO
#define __SYCL_PI_PROPERTY_SET_KERNEL_PARAM_OPT_INFO
PropertySetRegistry::SYCL_KERNEL_PARAM_OPT_INFO defined in PropertySetIO.h.
Definition: pi.h:723
context.hpp
cl::sycl::detail::pi::initializePlugins
static void initializePlugins(std::vector< plugin > &Plugins)
Definition: pi.cpp:384
cl::sycl::detail::pi::PI_TRACE_BASIC
@ PI_TRACE_BASIC
Definition: pi.hpp:56
__SYCL_ESIMD_EMULATOR_PLUGIN_NAME
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME
Definition: pi.hpp:74
cl::sycl::detail::pi::PiDeviceBinaryType
::pi_device_binary_type PiDeviceBinaryType
Definition: pi.hpp:105
__SYCL_LEVEL_ZERO_PLUGIN_NAME
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME
Definition: pi.hpp:72
cl::sycl::detail::pi::getOsLibraryFuncAddress
void * getOsLibraryFuncAddress(void *Library, const std::string &FunctionName)
Definition: posix_pi.cpp:27
cl::sycl::detail::SYCL_PICALL_STREAM_NAME
constexpr const char * SYCL_PICALL_STREAM_NAME
Definition: xpti_registry.hpp:28
version.hpp
cl::sycl::detail::pi::unloadPlugin
int unloadPlugin(void *Library)
Definition: pi.cpp:341
plugin.hpp
cl::sycl::detail::pi::emitFunctionEndTrace
void emitFunctionEndTrace(uint64_t CorrelationID, const char *FName)
Emits an XPTI trace after the PI API call has been made.
Definition: pi.cpp:131
__LIBSYCL_MAJOR_VERSION
#define __LIBSYCL_MAJOR_VERSION
Definition: version.hpp:10
PI_PLATFORM_INFO_NAME
@ PI_PLATFORM_INFO_NAME
Definition: pi.h:131
pi.hpp
cl::sycl::detail::SYCL_STREAM_NAME
constexpr const char * SYCL_STREAM_NAME
Definition: xpti_registry.hpp:26
cl::sycl::detail::pi::emitFunctionWithArgsBeginTrace
uint64_t emitFunctionWithArgsBeginTrace(uint32_t FuncID, const char *FName, unsigned char *ArgsData, pi_plugin Plugin)
Notifies XPTI subscribers about PI function calls and packs call arguments.
Definition: pi.cpp:146
__LIBSYCL_MINOR_VERSION
#define __LIBSYCL_MINOR_VERSION
Definition: version.hpp:11
cl::sycl::detail::pi::emitFunctionBeginTrace
uint64_t emitFunctionBeginTrace(const char *FName)
Emits an XPTI trace before a PI API call is made.
Definition: pi.cpp:87
stl_type_traits.hpp
_pi_device_binary_property_set_struct::PropertiesBegin
pi_device_binary_property PropertiesBegin
Definition: pi.h:660
cl::sycl::detail::pi::asUint32
static pi_uint32 asUint32(const void *Addr)
Definition: pi.cpp:627
cl::sycl::detail::pi::DeviceBinaryImage::PropertyRange::begin
ConstIterator begin() const
Definition: pi.hpp:277
cl::sycl::detail::pi::loadPlugin
void * loadPlugin(const std::string &PluginPath)
Definition: pi.cpp:335
cl::sycl::detail::pi::platformInfoToString
std::string platformInfoToString(pi_platform_info info)
Definition: pi.cpp:193
pi_uint32
uint32_t pi_uint32
Definition: pi.h:68
pi_device_binary_struct
This struct is a record of the device binary information.
Definition: pi.h:743
piextContextSetExtendedDeleter
pi_result piextContextSetExtendedDeleter(pi_context context, pi_context_extended_deleter func, void *user_data)
Definition: pi_esimd_emulator.cpp:650
__SYCL_PI_PROPERTY_SET_DEVICELIB_REQ_MASK
#define __SYCL_PI_PROPERTY_SET_DEVICELIB_REQ_MASK
PropertySetRegistry::SYCL_DEVICELIB_REQ_MASK defined in PropertySetIO.h.
Definition: pi.h:721
cl::sycl::detail::pi::GlobalPlugin
std::shared_ptr< plugin > GlobalPlugin
Definition: pi.cpp:270
cl::sycl::detail::pi::die
void die(const char *Message)
Definition: pi.cpp:537
pi_mem_flags
pi_bitfield pi_mem_flags
Definition: pi.h:547
cl::sycl::detail::pi::unloadOsLibrary
int unloadOsLibrary(void *Library)
Definition: posix_pi.cpp:25
cl::sycl::detail::pi::XPTIInitDone
bool XPTIInitDone
Definition: pi.cpp:83
pi_context_extended_deleter
void(* pi_context_extended_deleter)(void *user_data)
Definition: pi.h:1028
pi_device_binary_struct::Format
uint8_t Format
format of the binary data - SPIR-V, LLVM IR bitcode,...
Definition: pi.h:751
PI_PROPERTY_TYPE_BYTE_ARRAY
@ PI_PROPERTY_TYPE_BYTE_ARRAY
Definition: pi.h:640
cl::sycl::detail::device_filter_list
Definition: device_filter.hpp:37
cl::sycl::detail::plugin::call
void call(ArgsT... Args) const
Calls the API, traces the call, checks the result.
Definition: plugin.hpp:187
__SYCL_PI_PROPERTY_SET_SYCL_MISC_PROP
#define __SYCL_PI_PROPERTY_SET_SYCL_MISC_PROP
PropertySetRegistry::SYCL_MISC_PROP defined in PropertySetIO.h.
Definition: pi.h:727
cl::sycl::detail::pi::DeviceBinaryProperty
Definition: pi.hpp:227
cl
We provide new interfaces for matrix muliply in this patch:
Definition: access.hpp:13
piextPluginGetOpaqueData
pi_result piextPluginGetOpaqueData(void *opaque_data_param, void **opaque_data_return)
API to get Plugin internal data, opaque to SYCL RT.
Definition: pi_esimd_emulator.cpp:1527
PI_MEM_FLAGS_HOST_PTR_USE
constexpr pi_mem_flags PI_MEM_FLAGS_HOST_PTR_USE
Definition: pi.h:552
cl::sycl::detail::pi::DeviceBinaryImage::PropertyRange
Definition: pi.hpp:249
global_handler.hpp
PI_PLATFORM_INFO_VERSION
@ PI_PLATFORM_INFO_VERSION
Definition: pi.h:134
cl::sycl::detail::pi::bindPlugin
bool bindPlugin(void *Library, const std::shared_ptr< PiPlugin > &PluginInformation)
Definition: pi.cpp:349
_pi_device_binary_property_struct
Definition: pi.h:648
cl::sycl::detail::plugin
The plugin class provides a unified interface to the underlying low-level runtimes for the device-agn...
Definition: plugin.hpp:89
PI_PLATFORM_INFO_PROFILE
@ PI_PLATFORM_INFO_PROFILE
Definition: pi.h:132
PI_PROPERTY_TYPE_UINT32
@ PI_PROPERTY_TYPE_UINT32
Definition: pi.h:639
cl::sycl::detail::getPluginOpaqueData
void * getPluginOpaqueData(void *opaquedata_arg)
Definition: pi.cpp:65
PI_MEM_FLAGS_ACCESS_RW
constexpr pi_mem_flags PI_MEM_FLAGS_ACCESS_RW
Definition: pi.h:549
PI_PLATFORM_INFO_VENDOR
@ PI_PLATFORM_INFO_VENDOR
Definition: pi.h:133
std::get
constexpr tuple_element< I, tuple< Types... > >::type & get(cl::sycl::detail::tuple< Types... > &Arg) noexcept
Definition: tuple.hpp:199
cl::sycl::detail::pi::emitFunctionWithArgsEndTrace
void emitFunctionWithArgsEndTrace(uint64_t CorrelationID, uint32_t FuncID, const char *FName, unsigned char *ArgsData, pi_result Result, pi_plugin Plugin)
Notifies XPTI subscribers about PI function call result.
Definition: pi.cpp:166
__SYCL_HIP_PLUGIN_NAME
#define __SYCL_HIP_PLUGIN_NAME
Definition: pi.hpp:75
_PI_H_VERSION_STRING
#define _PI_H_VERSION_STRING
Definition: pi.h:51
cl::sycl::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:198
_pi_platform_info
_pi_platform_info
Definition: pi.h:129
SYCL_VERSION_STR
#define SYCL_VERSION_STR
Definition: pi.cpp:42
uint16_t
cl::sycl::detail::pi::contextSetExtendedDeleter
void contextSetExtendedDeleter(const cl::sycl::context &constext, pi_context_extended_deleter func, void *user_data)
Definition: pi.cpp:183
cl::sycl::context
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:35
cl::sycl::detail::pi::ByteArray
Definition: pi.hpp:211
cl::sycl::all
detail::enable_if_t< detail::is_sigeninteger< T >::value, int > all(T x) __NOEXC
Definition: builtins.hpp:1282
P
#define P(n)
cl::sycl::detail::remove_const_t
typename std::remove_const< T >::type remove_const_t
Definition: stl_type_traits.hpp:30
__SYCL_PI_PROPERTY_SET_PROGRAM_METADATA
#define __SYCL_PI_PROPERTY_SET_PROGRAM_METADATA
PropertySetRegistry::SYCL_KERNEL_PROGRAM_METADATA defined in PropertySetIO.h.
Definition: pi.h:725
common.hpp
_pi_device_binary_property_set_struct::PropertiesEnd
pi_device_binary_property PropertiesEnd
Definition: pi.h:661
__SYCL_PI_PROPERTY_SET_SPEC_CONST_MAP
#define __SYCL_PI_PROPERTY_SET_SPEC_CONST_MAP
Device binary image property set names recognized by the SYCL runtime.
Definition: pi.h:715
cl::sycl::detail::pi::DeviceBinaryImage::PropertyRange::isAvailable
bool isAvailable() const
Definition: pi.hpp:280
cl::sycl::detail::pi::loadOsLibrary
void * loadOsLibrary(const std::string &Library)
Definition: posix_pi.cpp:19
cl::sycl::detail::pi::memFlagsToString
std::string memFlagsToString(pi_mem_flags Flags)
Definition: pi.cpp:236
cl::sycl::detail::pi::trace
bool trace(TraceLevel level)
Definition: pi.cpp:368
_pi_context
PI context mapping to a CUDA context object.
Definition: pi_cuda.hpp:148
PI_PLATFORM_INFO_EXTENSIONS
@ PI_PLATFORM_INFO_EXTENSIONS
Definition: pi.h:130
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition: defines_elementary.hpp:12
pi_device_binary_struct::PropertySetsBegin
pi_device_binary_property_set PropertySetsBegin
Definition: pi.h:783