DPC++ Runtime
Runtime libraries for oneAPI Data Parallel C++
pi_esimd_emulator.cpp
Go to the documentation of this file.
1 //===---------- pi_esimd_emulator.cpp - CM Emulation Plugin ---------------===//
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 
14 
15 #include <stdint.h>
16 
25 #include <CL/sycl/group.hpp>
26 #include <CL/sycl/id.hpp>
27 #include <CL/sycl/kernel.hpp>
28 #include <CL/sycl/nd_item.hpp>
29 #include <CL/sycl/range.hpp>
30 
31 #include <esimdemu_support.h>
32 
33 #include <cstdarg>
34 #include <cstdio>
35 #include <cstring>
36 #include <functional>
37 #include <map>
38 #include <memory>
39 #include <string>
40 #include <thread>
41 #include <utility>
42 
43 #include "pi_esimd_emulator.hpp"
44 
45 namespace {
46 
47 // Helper functions for unified 'Return' type declaration - imported
48 // from pi_level_zero.cpp
49 template <typename T, typename Assign>
50 pi_result getInfoImpl(size_t ParamValueSize, void *ParamValue,
51  size_t *ParamValueSizeRet, T Value, size_t ValueSize,
52  Assign &&AssignFunc) {
53  if (ParamValue != nullptr) {
54  if (ParamValueSize < ValueSize) {
55  return PI_INVALID_VALUE;
56  }
57  AssignFunc(ParamValue, Value, ValueSize);
58  }
59  if (ParamValueSizeRet != nullptr) {
60  *ParamValueSizeRet = ValueSize;
61  }
62  return PI_SUCCESS;
63 }
64 
65 template <typename T>
66 pi_result getInfo(size_t ParamValueSize, void *ParamValue,
67  size_t *ParamValueSizeRet, T Value) {
68  auto assignment = [](void *ParamValue, T Value, size_t ValueSize) {
69  *static_cast<T *>(ParamValue) = Value;
70  };
71  return getInfoImpl(ParamValueSize, ParamValue, ParamValueSizeRet, Value,
72  sizeof(T), assignment);
73 }
74 
75 template <typename T>
76 pi_result getInfoArray(size_t ArrayLength, size_t ParamValueSize,
77  void *ParamValue, size_t *ParamValueSizeRet, T *Value) {
78  return getInfoImpl(ParamValueSize, ParamValue, ParamValueSizeRet, Value,
79  ArrayLength * sizeof(T), memcpy);
80 }
81 
82 template <>
83 pi_result getInfo<const char *>(size_t ParamValueSize, void *ParamValue,
84  size_t *ParamValueSizeRet, const char *Value) {
85  return getInfoArray(strlen(Value) + 1, ParamValueSize, ParamValue,
86  ParamValueSizeRet, Value);
87 }
88 
89 class ReturnHelper {
90 public:
91  ReturnHelper(size_t ArgParamValueSize, void *ArgParamValue,
92  size_t *ArgParamValueSizeRet)
93  : ParamValueSize(ArgParamValueSize), ParamValue(ArgParamValue),
94  ParamValueSizeRet(ArgParamValueSizeRet) {}
95 
96  template <class T> pi_result operator()(const T &t) {
97  return getInfo(ParamValueSize, ParamValue, ParamValueSizeRet, t);
98  }
99 
100 private:
101  size_t ParamValueSize;
102  void *ParamValue;
103  size_t *ParamValueSizeRet;
104 };
105 
106 } // anonymous namespace
107 
108 // Controls PI level tracing prints.
109 static bool PrintPiTrace = false;
110 
111 // Global variables used in PI_esimd_emulator
112 // Note we only create a simple pointer variables such that C++ RT won't
113 // deallocate them automatically at the end of the main program.
114 // The heap memory allocated for this global variable reclaimed only when
115 // Sycl RT calls piTearDown().
116 static sycl::detail::ESIMDEmuPluginOpaqueData *PiESimdDeviceAccess;
117 
118 // To be compared with ESIMD_EMULATOR_PLUGIN_OPAQUE_DATA_VERSION in device
119 // interface header file
120 #define ESIMDEmuPluginDataVersion 0
121 
122 // To be compared with ESIMD_DEVICE_INTERFACE_VERSION in device
123 // interface header file
124 #define ESIMDEmuPluginInterfaceVersion 1
125 
126 using IDBuilder = sycl::detail::Builder;
127 
128 template <int NDims>
129 using KernelFunc = std::function<void(const sycl::nd_item<NDims> &)>;
130 
131 // Struct to wrap dimension info and lambda function to be invoked by
132 // CM Kernel launcher that only accepts raw function pointer for
133 // kernel execution. Function instances of 'InvokeLambda' un-wrap this
134 // struct instance and invoke lambda function ('Func')
135 template <int NDims> struct LambdaWrapper {
137  const sycl::range<NDims> &LocalSize;
138  const sycl::range<NDims> &GlobalSize;
139  const sycl::id<NDims> &GlobalOffset;
141  const sycl::range<NDims> &ArgLocalSize,
142  const sycl::range<NDims> &ArgGlobalSize,
143  const sycl::id<NDims> &ArgGlobalOffset)
144  : Func(ArgFunc), LocalSize(ArgLocalSize), GlobalSize(ArgGlobalSize),
145  GlobalOffset(ArgGlobalOffset) {}
146 };
147 
148 // Function to generate a lambda wrapper object above
149 template <int NDims>
151  const sycl::range<NDims> &LocalSize,
152  const sycl::range<NDims> &GlobalSize,
153  const sycl::id<NDims> &GlobalOffset) {
154  std::unique_ptr<LambdaWrapper<NDims>> Wrapper =
155  std::make_unique<LambdaWrapper<NDims>>(LambdaWrapper<NDims>(
156  KernelFunc<NDims>(ArgFunc), LocalSize, GlobalSize, GlobalOffset));
157  return Wrapper;
158 }
159 
160 // A helper structure to create multi-dimensional range when
161 // dimensionality is given as a template parameter. `create` function
162 // in specializations accepts a template `Gen` function which
163 // generates range extent for a dimension given as an argument.
164 template <int NDims> struct RangeBuilder;
165 
166 template <> struct RangeBuilder<1> {
167  template <typename Gen> static sycl::range<1> create(Gen G) {
168  return sycl::range<1>{G(0)};
169  }
170 };
171 template <> struct RangeBuilder<2> {
172  template <typename Gen> static sycl::range<2> create(Gen G) {
173  return sycl::range<2>{G(0), G(1)};
174  }
175 };
176 template <> struct RangeBuilder<3> {
177  template <typename Gen> static sycl::range<3> create(Gen G) {
178  return sycl::range<3>{G(0), G(1), G(2)};
179  }
180 };
181 
182 // Function template to generate entry point of kernel execution as
183 // raw function pointer. CM kernel launcher executes one instance of
184 // this function per 'NDims'
185 template <int NDims> void InvokeLambda(void *Wrapper) {
186  auto *WrappedLambda = reinterpret_cast<LambdaWrapper<NDims> *>(Wrapper);
187  sycl::range<NDims> GroupSize(
188  sycl::detail::InitializedVal<NDims, sycl::range>::template get<0>());
189 
190  for (int I = 0; I < NDims /*Dims*/; ++I) {
191  GroupSize[I] = WrappedLambda->GlobalSize[I] / WrappedLambda->LocalSize[I];
192  }
193 
194  const sycl::id<NDims> LocalID = RangeBuilder<NDims>::create(
195  [](int i) { return cm_support::get_thread_idx(i); });
196 
197  const sycl::id<NDims> GroupID = RangeBuilder<NDims>::create(
198  [](int Id) { return cm_support::get_group_idx(Id); });
199 
200  const sycl::group<NDims> Group = IDBuilder::createGroup<NDims>(
201  WrappedLambda->GlobalSize, WrappedLambda->LocalSize, GroupSize, GroupID);
202 
203  const sycl::id<NDims> GlobalID = GroupID * WrappedLambda->LocalSize +
204  LocalID + WrappedLambda->GlobalOffset;
205  const sycl::item<NDims, /*Offset=*/true> GlobalItem =
206  IDBuilder::createItem<NDims, true>(WrappedLambda->GlobalSize, GlobalID,
207  WrappedLambda->GlobalOffset);
208  const sycl::item<NDims, /*Offset=*/false> LocalItem =
209  IDBuilder::createItem<NDims, false>(WrappedLambda->LocalSize, LocalID);
210 
211  const sycl::nd_item<NDims> NDItem =
212  IDBuilder::createNDItem<NDims>(GlobalItem, LocalItem, Group);
213 
214  WrappedLambda->Func(NDItem);
215 }
216 
217 // libCMBatch class defines interface for lauching kernels with
218 // software multi-threads
219 template <int DIMS> class libCMBatch {
220 private:
221  // Kernel function
222  KernelFunc<DIMS> MKernel;
223 
224  // Space-dimension info
225  std::vector<uint32_t> GroupDim;
226  std::vector<uint32_t> SpaceDim;
227 
228 public:
230  : MKernel(Kernel), GroupDim{1, 1, 1}, SpaceDim{1, 1, 1} {}
231 
234  void runIterationSpace(const sycl::range<DIMS> &LocalSize,
235  const sycl::range<DIMS> &GlobalSize,
236  const sycl::id<DIMS> &GlobalOffset) {
237  auto WrappedLambda =
238  MakeLambdaWrapper<DIMS>(MKernel, LocalSize, GlobalSize, GlobalOffset);
239 
240  for (int I = 0; I < DIMS; I++) {
241  SpaceDim[I] = (uint32_t)LocalSize[I];
242  GroupDim[I] = (uint32_t)(GlobalSize[I] / LocalSize[I]);
243  }
244 
245  EsimdemuKernel Esimdemu((fptrVoid)InvokeLambda<DIMS>, GroupDim, SpaceDim);
246 
247  Esimdemu.launchMT(sizeof(struct LambdaWrapper<DIMS>), WrappedLambda.get());
248  }
249 };
250 
251 // Function to provide buffer info for kernel compilation without
252 // dependency on '_pi_buffer' definition
253 void sycl_get_cm_buffer_params(void *PtrInput, char **BaseAddr, uint32_t *Width,
254  std::mutex **MtxLock) {
255  _pi_buffer *Buf = static_cast<_pi_buffer *>(PtrInput);
256 
257  *BaseAddr = cm_support::get_surface_base_addr(Buf->SurfaceIndex);
258  *Width = static_cast<uint32_t>(Buf->Size);
259 
260  *MtxLock = &(Buf->mutexLock);
261 }
262 
263 // Function to provide image info for kernel compilation without
264 // dependency on '_pi_image' definition
265 void sycl_get_cm_image_params(void *PtrInput, char **BaseAddr, uint32_t *Width,
266  uint32_t *Height, uint32_t *Bpp,
267  std::mutex **MtxLock) {
268  _pi_image *Img = static_cast<_pi_image *>(PtrInput);
269 
270  *BaseAddr = cm_support::get_surface_base_addr(Img->SurfaceIndex);
271 
272  *Bpp = static_cast<uint32_t>(Img->BytesPerPixel);
273  *Width = static_cast<uint32_t>(Img->Width) * (*Bpp);
274  *Height = static_cast<uint32_t>(Img->Height);
275 
276  *MtxLock = &(Img->mutexLock);
277 }
278 
281 sycl::detail::ESIMDDeviceInterface::ESIMDDeviceInterface() {
283  reserved = nullptr;
284 
285  /* From 'esimd_emulator_functions_v1.h' : Start */
287  cm_sbarrier_ptr = cm_support::split_barrier;
289 
290  sycl_get_surface_base_addr_ptr = cm_support::get_surface_base_addr;
291  __cm_emu_get_slm_ptr = cm_support::get_slm_base;
292  cm_slm_init_ptr = cm_support::init_slm;
293 
296  /* From 'esimd_emulator_functions_v1.h' : End */
297 }
298 
301 
302 static bool isNull(int NDims, const size_t *R) {
303  return ((0 == R[0]) && (NDims < 2 || 0 == R[1]) && (NDims < 3 || 0 == R[2]));
304 }
305 
306 // NDims is the number of dimensions in the ND-range. Kernels are
307 // normalized in the handler so that all kernels take an sycl::nd_item
308 // as argument (see StoreLambda in CL/sycl/handler.hpp). For kernels
309 // whose workgroup size (LocalWorkSize) is unspecified, InvokeImpl
310 // sets LocalWorkSize to {1, 1, 1}, i.e. each workgroup contains just
311 // one work item. CM emulator will run several workgroups in parallel
312 // depending on environment settings.
313 
314 template <int NDims> struct InvokeImpl {
315 
316  static sycl::range<NDims> get_range(const size_t *Array) {
317  if constexpr (NDims == 1)
318  return sycl::range<NDims>{Array[0]};
319  else if constexpr (NDims == 2)
320  return sycl::range<NDims>{Array[0], Array[1]};
321  else if constexpr (NDims == 3)
322  return sycl::range<NDims>{Array[0], Array[1], Array[2]};
323  }
324 
325  static void invoke(void *Fptr, const size_t *GlobalWorkOffset,
326  const size_t *GlobalWorkSize,
327  const size_t *LocalWorkSize) {
328  auto GlobalSize = get_range(GlobalWorkSize);
329  auto LocalSize = get_range(LocalWorkSize);
330  sycl::id<NDims> GlobalOffset = get_range(GlobalWorkOffset);
331 
332  auto KFunc = reinterpret_cast<KernelFunc<NDims> *>(Fptr);
333  libCMBatch<NDims> CmThreading(*KFunc);
334 
335  CmThreading.runIterationSpace(LocalSize, GlobalSize, GlobalOffset);
336  }
337 };
338 
339 extern "C" {
340 
341 #define DIE_NO_IMPLEMENTATION \
342  if (PrintPiTrace) { \
343  std::cerr << "Not Implemented : " << __FUNCTION__ \
344  << " - File : " << __FILE__; \
345  std::cerr << " / Line : " << __LINE__ << std::endl; \
346  } \
347  return PI_INVALID_OPERATION;
348 
349 #define CONTINUE_NO_IMPLEMENTATION \
350  if (PrintPiTrace) { \
351  std::cerr << "Warning : Not Implemented : " << __FUNCTION__ \
352  << " - File : " << __FILE__; \
353  std::cerr << " / Line : " << __LINE__ << std::endl; \
354  } \
355  return PI_SUCCESS;
356 
358  pi_uint32 *NumPlatforms) {
359 
360  static const char *PiTrace = std::getenv("SYCL_PI_TRACE");
361  static const int PiTraceValue = PiTrace ? std::stoi(PiTrace) : 0;
362  if (PiTraceValue == -1) { // Means print all PI traces
363  PrintPiTrace = true;
364  }
365 
366  if (NumEntries == 0 && Platforms != nullptr) {
367  return PI_INVALID_VALUE;
368  }
369  if (Platforms == nullptr && NumPlatforms == nullptr) {
370  return PI_INVALID_VALUE;
371  }
372 
373  if (Platforms && NumEntries > 0) {
374  *Platforms = new _pi_platform();
375  Platforms[0]->CmEmuVersion = std::string("0.0.1");
376  }
377 
378  if (NumPlatforms) {
379  *NumPlatforms = 1;
380  }
381 
382  return PI_SUCCESS;
383 }
384 
386  size_t ParamValueSize, void *ParamValue,
387  size_t *ParamValueSizeRet) {
388  if (Platform == nullptr) {
389  return PI_INVALID_PLATFORM;
390  }
391  ReturnHelper ReturnValue(ParamValueSize, ParamValue, ParamValueSizeRet);
392 
393  switch (ParamName) {
395  return ReturnValue("Intel(R) ESIMD_EMULATOR/GPU");
396 
398  return ReturnValue("Intel(R) Corporation");
399 
401  return ReturnValue(Platform->CmEmuVersion);
402 
404  return ReturnValue("FULL_PROFILE");
405 
407  return ReturnValue("");
408 
409  default:
410  // TODO: implement other parameters
411  die("Unsupported ParamName in piPlatformGetInfo");
412  }
413 
414  return PI_SUCCESS;
415 }
416 
419 }
420 
423 }
424 
426  pi_uint32 NumEntries, pi_device *Devices,
427  pi_uint32 *NumDevices) {
428  if (Platform == nullptr) {
429  return PI_INVALID_PLATFORM;
430  }
431 
432  // CM has single-root-device without sub-device support.
433  if (NumDevices) {
434  *NumDevices = 1;
435  }
436 
437  cm_support::CmDevice *CmDevice = nullptr;
438  // TODO FIXME Implement proper version checking and reporting:
439  // - version passed to cm_support::CreateCmDevice
440  // - CmEmuVersion
441  // - PluginVersion
442  // - ESIMDEmuPluginOpaqueData::version
443  //
444  // PI_DEVICE_INFO_DRIVER_VERSION could report the ESIMDDeviceInterface
445  // version, PI_PLATFORM_INFO_VERSION - the underlying libCM library version.
446  unsigned int Version = 0;
447 
448  int Result = cm_support::CreateCmDevice(CmDevice, Version);
449 
450  if (Result != cm_support::CM_SUCCESS) {
451  return PI_INVALID_DEVICE;
452  }
453 
454  // FIXME / TODO : piDevicesGet always must return same pointer for
455  // 'Devices[0]' from cached entry. Reference : level-zero
456  // platform/device implementation with PiDevicesCache and
457  // PiDevicesCache
458  if (Devices) {
459  Devices[0] = new _pi_device(Platform, CmDevice);
460  }
461 
462  return PI_SUCCESS;
463 }
464 
466  if (Device == nullptr) {
467  return PI_INVALID_DEVICE;
468  }
469 
470  // CM supports only single device, which is root-device. 'Retain' is
471  // No-op.
472  return PI_SUCCESS;
473 }
474 
476  if (Device == nullptr) {
477  return PI_INVALID_DEVICE;
478  }
479 
480  // CM supports only single device, which is root-device. 'Release'
481  // is No-op.
482  return PI_SUCCESS;
483 }
484 
486  size_t ParamValueSize, void *ParamValue,
487  size_t *ParamValueSizeRet) {
488  ReturnHelper ReturnValue(ParamValueSize, ParamValue, ParamValueSizeRet);
489 
490  switch (ParamName) {
491  case PI_DEVICE_INFO_TYPE:
492  return ReturnValue(PI_DEVICE_TYPE_GPU);
494  return ReturnValue(pi_device{0});
496  return ReturnValue(Device->Platform);
497  case PI_DEVICE_INFO_NAME:
498  return ReturnValue("ESIMD_EMULATOR");
500  return ReturnValue(pi_bool{true});
502  return ReturnValue("0.0.1");
504  return ReturnValue("Intel(R) Corporation");
506  return ReturnValue(size_t{8192});
508  return ReturnValue(size_t{8192});
510  return ReturnValue(pi_bool{1});
512  // TODO : Populate return string accordingly - e.g. cl_khr_fp16,
513  // cl_khr_fp64, cl_khr_int64_base_atomics,
514  // cl_khr_int64_extended_atomics
515  return ReturnValue("");
516 
517 #define UNSUPPORTED_INFO(info) \
518  case info: \
519  std::cerr << std::endl \
520  << "Unsupported device info = " << #info \
521  << " from ESIMD_EMULATOR" << std::endl; \
522  DIE_NO_IMPLEMENTATION; \
523  break;
524 
596 
597 #undef UNSUPPORTED_INFO
598  default:
600  }
601  return PI_SUCCESS;
602 }
603 
605  pi_uint32, pi_device *, pi_uint32 *) {
607 }
608 
611 }
612 
614  pi_device *) {
616 }
617 
619  pi_uint32 NumDevices, const pi_device *Devices,
620  void (*PFnNotify)(const char *ErrInfo,
621  const void *PrivateInfo, size_t CB,
622  void *UserData),
623  void *UserData, pi_context *RetContext) {
624  if (NumDevices != 1) {
625  return PI_INVALID_VALUE;
626  }
627  if (Devices == nullptr) {
628  return PI_INVALID_DEVICE;
629  }
630  if (RetContext == nullptr) {
631  return PI_INVALID_VALUE;
632  }
633 
634  try {
636  *RetContext = new _pi_context(Devices[0]);
637  } catch (const std::bad_alloc &) {
638  return PI_OUT_OF_HOST_MEMORY;
639  } catch (...) {
640  return PI_ERROR_UNKNOWN;
641  }
642  return PI_SUCCESS;
643 }
644 
646  size_t *) {
648 }
649 
651  pi_context_extended_deleter, void *) {
653 }
654 
657 }
658 
660  const pi_device *, bool,
661  pi_context *) {
663 }
664 
666  if (Context == nullptr) {
667  return PI_INVALID_CONTEXT;
668  }
669 
670  ++(Context->RefCount);
671 
672  return PI_SUCCESS;
673 }
674 
676  if (Context == nullptr || (Context->RefCount <= 0)) {
677  return PI_INVALID_CONTEXT;
678  }
679 
680  if (--(Context->RefCount) == 0) {
681  for (auto &Entry : Context->Addr2CmBufferSVM) {
682  Context->Device->CmDevicePtr->DestroyBufferSVM(Entry.second);
683  }
684  delete Context;
685  }
686 
687  return PI_SUCCESS;
688 }
689 
691  pi_queue_properties Properties, pi_queue *Queue) {
692  if (Properties & PI_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE) {
693  // TODO : Support Out-of-order Queue
694  *Queue = nullptr;
696  }
697 
698  cm_support::CmQueue *CmQueue;
699 
700  int Result = Context->Device->CmDevicePtr->CreateQueue(CmQueue);
701  if (Result != cm_support::CM_SUCCESS) {
702  return PI_INVALID_CONTEXT;
703  }
704 
705  try {
706  *Queue = new _pi_queue(Context, CmQueue);
707  } catch (const std::bad_alloc &) {
708  return PI_OUT_OF_HOST_MEMORY;
709  } catch (...) {
710  return PI_ERROR_UNKNOWN;
711  }
712 
713  return PI_SUCCESS;
714 }
715 
716 pi_result piQueueGetInfo(pi_queue, pi_queue_info, size_t, void *, size_t *) {
718 }
719 
721  if (Queue == nullptr) {
722  return PI_INVALID_QUEUE;
723  }
724  ++(Queue->RefCount);
725  return PI_SUCCESS;
726 }
727 
729  if ((Queue == nullptr) || (Queue->CmQueuePtr == nullptr)) {
730  return PI_INVALID_QUEUE;
731  }
732 
733  if (--(Queue->RefCount) == 0) {
734  // CM's 'DestoryQueue' is no-op
735  // Queue->Context->Device->CmDevicePTr->DestroyQueue(Queue->CmQueuePtr);
736  delete Queue;
737  }
738 
739  return PI_SUCCESS;
740 }
741 
743  // No-op as enqueued commands with ESIMD_EMULATOR plugin are blocking
744  // ones that do not return until their completion - kernel execution
745  // and memory read.
747 }
748 
751 }
752 
754  pi_queue *, bool) {
756 }
757 
759  void *HostPtr, pi_mem *RetMem,
760  const pi_mem_properties *properties) {
761  if ((Flags & PI_MEM_FLAGS_ACCESS_RW) == 0) {
762  if (PrintPiTrace) {
763  std::cerr << "Invalid memory attribute for piMemBufferCreate"
764  << std::endl;
765  }
766  return PI_INVALID_OPERATION;
767  }
768 
769  if (Context == nullptr) {
770  return PI_INVALID_CONTEXT;
771  }
772  if (RetMem == nullptr) {
773  return PI_INVALID_VALUE;
774  }
775 
776  cm_support::CmBuffer *CmBuf = nullptr;
777  cm_support::SurfaceIndex *CmIndex;
778 
779  int Status = Context->Device->CmDevicePtr->CreateBuffer(
780  static_cast<unsigned int>(Size), CmBuf);
781 
782  if (Status != cm_support::CM_SUCCESS) {
783  return PI_OUT_OF_HOST_MEMORY;
784  }
785 
786  Status = CmBuf->GetIndex(CmIndex);
787 
788  // Initialize the buffer with user data provided with 'HostPtr'
789  if ((Flags & PI_MEM_FLAGS_HOST_PTR_USE) != 0) {
790  if (HostPtr != nullptr) {
791  Status =
792  CmBuf->WriteSurface(reinterpret_cast<const unsigned char *>(HostPtr),
793  nullptr, static_cast<unsigned int>(Size));
794  }
795  }
796 
797  auto HostPtrOrNull =
798  (Flags & PI_MEM_FLAGS_HOST_PTR_COPY) ? nullptr : pi_cast<char *>(HostPtr);
799 
800  try {
801  *RetMem =
802  new _pi_buffer(Context, HostPtrOrNull, CmBuf,
803  /* integer buffer index */ CmIndex->get_data(), Size);
804  } catch (const std::bad_alloc &) {
805  return PI_OUT_OF_HOST_MEMORY;
806  } catch (...) {
807  return PI_ERROR_UNKNOWN;
808  }
809 
810  return PI_SUCCESS;
811 }
812 
813 pi_result piMemGetInfo(pi_mem, cl_mem_info, size_t, void *, size_t *) {
815 }
816 
818  if (Mem == nullptr) {
819  return PI_INVALID_MEM_OBJECT;
820  }
821  ++(Mem->RefCount);
822  return PI_SUCCESS;
823 }
824 
826  if ((Mem == nullptr) || (Mem->RefCount == 0)) {
827  return PI_INVALID_MEM_OBJECT;
828  }
829 
830  if (--(Mem->RefCount) == 0) {
831  if (Mem->getMemType() == PI_MEM_TYPE_BUFFER) {
832  _pi_buffer *PiBuf = static_cast<_pi_buffer *>(Mem);
833  // TODO implement libCM API failure logging mechanism, so that these
834  // failures are clearly distinguishable from other EMU plugin failures.
835  int Result =
836  Mem->Context->Device->CmDevicePtr->DestroySurface(PiBuf->CmBufferPtr);
837 
838  if (Result != cm_support::CM_SUCCESS) {
839  return PI_INVALID_MEM_OBJECT;
840  }
841  } else if (Mem->getMemType() == PI_MEM_TYPE_IMAGE2D) {
842  _pi_image *PiImg = static_cast<_pi_image *>(Mem);
843  int Result = Mem->Context->Device->CmDevicePtr->DestroySurface(
844  PiImg->CmSurfacePtr);
845  if (Result != cm_support::CM_SUCCESS) {
846  return PI_INVALID_MEM_OBJECT;
847  }
848  } else {
849  return PI_INVALID_MEM_OBJECT;
850  }
851 
852  delete Mem;
853  }
854 
855  return PI_SUCCESS;
856 }
857 
858 cm_support::CM_SURFACE_FORMAT
860  using ULongPair = std::pair<unsigned long, unsigned long>;
861  using FmtMap = std::map<ULongPair, cm_support::CM_SURFACE_FORMAT>;
862  static const FmtMap pi2cm = {
864  cm_support::CM_SURFACE_FORMAT_A8R8G8B8},
865 
867  cm_support::CM_SURFACE_FORMAT_A8R8G8B8},
868 
870  cm_support::CM_SURFACE_FORMAT_A8R8G8B8},
871 
873  cm_support::CM_SURFACE_FORMAT_R32G32B32A32F},
874  };
875  auto Result = pi2cm.find(
876  {PiFormat->image_channel_data_type, PiFormat->image_channel_order});
877  if (Result != pi2cm.end()) {
878  return Result->second;
879  }
880  return cm_support::CM_SURFACE_FORMAT_UNKNOWN;
881 }
882 
884  const pi_image_format *ImageFormat,
885  const pi_image_desc *ImageDesc, void *HostPtr,
886  pi_mem *RetImage) {
887  if ((Flags & PI_MEM_FLAGS_ACCESS_RW) == 0) {
888  if (PrintPiTrace) {
889  std::cerr << "Invalid memory attribute for piMemImageCreate" << std::endl;
890  }
891  return PI_INVALID_OPERATION;
892  }
893 
894  if (ImageFormat == nullptr || ImageDesc == nullptr)
896 
897  switch (ImageDesc->image_type) {
898  case PI_MEM_TYPE_IMAGE2D:
899  break;
900  default:
901  return PI_INVALID_MEM_OBJECT;
902  }
903 
904  auto BytesPerPixel = 4;
905  switch (ImageFormat->image_channel_data_type) {
907  BytesPerPixel = 16;
908  break;
911  BytesPerPixel = 4;
912  break;
913  default:
915  }
916 
917  cm_support::CmSurface2D *CmSurface = nullptr;
918  cm_support::SurfaceIndex *CmIndex;
919  cm_support::CM_SURFACE_FORMAT CmSurfFormat =
920  ConvertPiImageFormatToCmFormat(ImageFormat);
921 
922  if (CmSurfFormat == cm_support::CM_SURFACE_FORMAT_UNKNOWN) {
924  }
925 
926  int Status = Context->Device->CmDevicePtr->CreateSurface2D(
927  static_cast<unsigned int>(ImageDesc->image_width),
928  static_cast<unsigned int>(ImageDesc->image_height), CmSurfFormat,
929  CmSurface);
930 
931  if (Status != cm_support::CM_SUCCESS) {
932  return PI_OUT_OF_HOST_MEMORY;
933  }
934 
935  Status = CmSurface->GetIndex(CmIndex);
936 
937  // Initialize the buffer with user data provided with 'HostPtr'
938  if ((Flags & PI_MEM_FLAGS_HOST_PTR_USE) != 0) {
939  if (HostPtr != nullptr) {
940  Status = CmSurface->WriteSurface(
941  reinterpret_cast<const unsigned char *>(HostPtr), nullptr,
942  static_cast<unsigned int>(ImageDesc->image_width *
943  ImageDesc->image_height * BytesPerPixel));
944  }
945  }
946 
947  auto HostPtrOrNull =
948  (Flags & PI_MEM_FLAGS_HOST_PTR_COPY) ? nullptr : pi_cast<char *>(HostPtr);
949 
950  try {
951  *RetImage = new _pi_image(Context, HostPtrOrNull, CmSurface,
952  /* integer surface index */ CmIndex->get_data(),
953  ImageDesc->image_width, ImageDesc->image_height,
954  BytesPerPixel);
955  } catch (const std::bad_alloc &) {
956  return PI_OUT_OF_HOST_MEMORY;
957  } catch (...) {
958  return PI_ERROR_UNKNOWN;
959  }
960 
961  return PI_SUCCESS;
962 }
963 
966 }
967 
970 }
971 
972 pi_result piProgramCreate(pi_context, const void *, size_t, pi_program *) {
974 }
975 
977  const size_t *, const unsigned char **,
978  size_t, const pi_device_binary_property *,
979  pi_int32 *, pi_program *) {
981 }
982 
984  const size_t *, const unsigned char **,
985  pi_int32 *, pi_program *) {
987 }
988 
990  const size_t *, pi_program *) {
992 }
993 
995  size_t *) {
997 }
998 
1000  pi_uint32, const pi_program *,
1001  void (*)(pi_program, void *), void *, pi_program *) {
1003 }
1004 
1006  const char *, pi_uint32, const pi_program *,
1007  const char **, void (*)(pi_program, void *),
1008  void *) {
1010 }
1011 
1013  void (*)(pi_program, void *), void *) {
1015 }
1016 
1018  size_t, void *, size_t *) {
1020 }
1021 
1023 
1025 
1028 }
1029 
1031  pi_program *) {
1033 }
1034 
1037 }
1038 
1039 pi_result piKernelSetArg(pi_kernel, pi_uint32, size_t, const void *) {
1041 }
1042 
1045 }
1046 
1047 // Special version of piKernelSetArg to accept pi_sampler.
1050 }
1051 
1052 pi_result piKernelGetInfo(pi_kernel, pi_kernel_info, size_t, void *, size_t *) {
1054 }
1055 
1057  size_t, void *, size_t *) {
1059 }
1060 
1062  pi_kernel_sub_group_info, size_t,
1063  const void *, size_t, void *, size_t *) {
1065 }
1066 
1068 
1070 
1072 
1073 pi_result piEventGetInfo(pi_event, pi_event_info, size_t, void *, size_t *) {
1075 }
1076 
1078  size_t ParamValueSize, void *ParamValue,
1079  size_t *ParamValueSizeRet) {
1080  if (PrintPiTrace) {
1081  std::cerr << "Warning : Profiling Not supported under PI_ESIMD_EMULATOR"
1082  << std::endl;
1083  }
1084  return PI_SUCCESS;
1085 }
1086 
1087 pi_result piEventsWait(pi_uint32 NumEvents, const pi_event *EventList) {
1088  for (int i = 0; i < (int)NumEvents; i++) {
1089  if (EventList[i]->IsDummyEvent) {
1090  // Dummy event is already completed ones done by CM. Skip
1091  // waiting.
1092  continue;
1093  }
1094  if (EventList[i]->CmEventPtr == nullptr) {
1095  return PI_INVALID_EVENT;
1096  }
1097  int Result = EventList[i]->CmEventPtr->WaitForTaskFinished();
1098  if (Result != cm_support::CM_SUCCESS) {
1099  return PI_OUT_OF_RESOURCES;
1100  }
1101  }
1102  return PI_SUCCESS;
1103 }
1104 
1106  void (*)(pi_event, pi_int32, void *), void *) {
1108 }
1109 
1111 
1113  if (Event == nullptr) {
1114  return PI_INVALID_EVENT;
1115  }
1116 
1117  ++(Event->RefCount);
1118 
1119  return PI_SUCCESS;
1120 }
1121 
1123  if (Event == nullptr || (Event->RefCount <= 0)) {
1124  return PI_INVALID_EVENT;
1125  }
1126 
1127  if (--(Event->RefCount) == 0) {
1128  if (!Event->IsDummyEvent) {
1129  if ((Event->CmEventPtr == nullptr) || (Event->OwnerQueue == nullptr)) {
1130  return PI_INVALID_EVENT;
1131  }
1132  int Result = Event->OwnerQueue->DestroyEvent(Event->CmEventPtr);
1133  if (Result != cm_support::CM_SUCCESS) {
1134  return PI_INVALID_EVENT;
1135  }
1136  }
1137  delete Event;
1138  }
1139 
1140  return PI_SUCCESS;
1141 }
1142 
1145 }
1146 
1148  pi_event *) {
1150 }
1152  pi_sampler *) {
1154 }
1155 
1157  size_t *) {
1159 }
1160 
1162 
1164 
1166  pi_event *) {
1168 }
1169 
1171  pi_event *) {
1173 }
1174 
1176  pi_bool BlockingRead, size_t Offset,
1177  size_t Size, void *Dst,
1178  pi_uint32 NumEventsInWaitList,
1179  const pi_event *EventWaitList,
1180  pi_event *Event) {
1182  if (BlockingRead) {
1183  assert(false &&
1184  "ESIMD_EMULATOR support for blocking piEnqueueMemBufferRead is NYI");
1185  }
1186  if (NumEventsInWaitList != 0) {
1188  }
1189 
1190  _pi_buffer *buf = static_cast<_pi_buffer *>(Src);
1191 
1192  std::unique_ptr<_pi_event> RetEv{nullptr};
1193  if (Event) {
1194  RetEv = std::unique_ptr<_pi_event>(new _pi_event());
1195  RetEv->IsDummyEvent = true;
1196  }
1197 
1198  int Status =
1199  buf->CmBufferPtr->ReadSurface(reinterpret_cast<unsigned char *>(Dst),
1200  nullptr, // event
1201  static_cast<uint64_t>(Size));
1202 
1203  if (Status != cm_support::CM_SUCCESS) {
1204  return PI_INVALID_MEM_OBJECT;
1205  }
1206 
1207  if (Event) {
1208  *Event = RetEv.release();
1209  }
1210 
1211  return PI_SUCCESS;
1212 }
1213 
1216  pi_buff_rect_region, size_t, size_t,
1217  size_t, size_t, void *, pi_uint32,
1218  const pi_event *, pi_event *) {
1220 }
1221 
1223  const void *, pi_uint32, const pi_event *,
1224  pi_event *) {
1226 }
1227 
1230  pi_buff_rect_region, size_t, size_t,
1231  size_t, size_t, const void *, pi_uint32,
1232  const pi_event *, pi_event *) {
1234 }
1235 
1237  size_t, pi_uint32, const pi_event *,
1238  pi_event *) {
1240 }
1241 
1244  pi_buff_rect_region, size_t, size_t,
1245  size_t, size_t, pi_uint32,
1246  const pi_event *, pi_event *) {
1248 }
1249 
1250 pi_result piEnqueueMemBufferFill(pi_queue, pi_mem, const void *, size_t, size_t,
1251  size_t, pi_uint32, const pi_event *,
1252  pi_event *) {
1254 }
1255 
1257  size_t, pi_uint32, const pi_event *, pi_event *,
1258  void **) {
1260 }
1261 
1263  const pi_event *, pi_event *) {
1265 }
1266 
1267 pi_result piMemImageGetInfo(pi_mem, pi_image_info, size_t, void *, size_t *) {
1269 }
1270 
1272  pi_bool BlockingRead, pi_image_offset Origin,
1273  pi_image_region Region, size_t RowPitch,
1274  size_t SlicePitch, void *Ptr,
1275  pi_uint32 NumEventsInWaitList,
1276  const pi_event *EventWaitList,
1277  pi_event *Event) {
1279  if (BlockingRead) {
1280  assert(false && "ESIMD_EMULATOR does not support Blocking Read");
1281  }
1282  _pi_image *PiImg = static_cast<_pi_image *>(Image);
1283 
1284  std::unique_ptr<_pi_event> RetEv{nullptr};
1285 
1286  if (Event) {
1287  RetEv = std::unique_ptr<_pi_event>(new _pi_event());
1288  RetEv->IsDummyEvent = true;
1289  }
1290 
1291  int Status =
1292  PiImg->CmSurfacePtr->ReadSurface(reinterpret_cast<unsigned char *>(Ptr),
1293  nullptr, // event
1294  RowPitch * (Region->height));
1295  if (Status != cm_support::CM_SUCCESS) {
1296  return PI_INVALID_MEM_OBJECT;
1297  }
1298 
1299  if (Event) {
1300  *Event = RetEv.release();
1301  }
1302  return PI_SUCCESS;
1303 }
1304 
1306  pi_image_region, size_t, size_t, const void *,
1307  pi_uint32, const pi_event *, pi_event *) {
1309 }
1310 
1313  const pi_event *, pi_event *) {
1315 }
1316 
1317 pi_result piEnqueueMemImageFill(pi_queue, pi_mem, const void *, const size_t *,
1318  const size_t *, pi_uint32, const pi_event *,
1319  pi_event *) {
1321 }
1322 
1324  void *, pi_mem *) {
1326 }
1327 
1328 pi_result
1330  const size_t *GlobalWorkOffset,
1331  const size_t *GlobalWorkSize, const size_t *LocalWorkSize,
1332  pi_uint32 NumEventsInWaitList,
1333  const pi_event *EventWaitList, pi_event *Event) {
1334  const size_t LocalWorkSz[] = {1, 1, 1};
1335 
1336  if (Kernel == nullptr) {
1337  return PI_INVALID_KERNEL;
1338  }
1339 
1340  if ((WorkDim > 3) || (WorkDim == 0)) {
1342  }
1343 
1344  if (isNull(WorkDim, LocalWorkSize)) {
1345  LocalWorkSize = LocalWorkSz;
1346  }
1347 
1348  for (pi_uint32 I = 0; I < WorkDim; I++) {
1349  if ((GlobalWorkSize[I] % LocalWorkSize[I]) != 0) {
1351  }
1352  }
1353 
1354  std::unique_ptr<_pi_event> RetEv{nullptr};
1355 
1356  if (Event) {
1357  RetEv = std::unique_ptr<_pi_event>(new _pi_event());
1358  RetEv->IsDummyEvent = true;
1359  }
1360 
1361  switch (WorkDim) {
1362  case 1:
1363  InvokeImpl<1>::invoke(Kernel, GlobalWorkOffset, GlobalWorkSize,
1364  LocalWorkSize);
1365  break;
1366 
1367  case 2:
1368  InvokeImpl<2>::invoke(Kernel, GlobalWorkOffset, GlobalWorkSize,
1369  LocalWorkSize);
1370  break;
1371 
1372  case 3:
1373  InvokeImpl<3>::invoke(Kernel, GlobalWorkOffset, GlobalWorkSize,
1374  LocalWorkSize);
1375  break;
1376 
1377  default:
1379  break;
1380  }
1381 
1382  if (Event) {
1383  *Event = RetEv.release();
1384  }
1385 
1386  return PI_SUCCESS;
1387 }
1388 
1390  pi_program, bool, pi_kernel *) {
1392 }
1393 
1396 }
1397 
1398 pi_result piEnqueueNativeKernel(pi_queue, void (*)(void *), void *, size_t,
1399  pi_uint32, const pi_mem *, const void **,
1400  pi_uint32, const pi_event *, pi_event *) {
1402 }
1403 
1405  pi_uint64 *) {
1407 }
1408 
1410  size_t, pi_uint32) {
1412 }
1413 
1415  pi_usm_mem_properties *, size_t, pi_uint32) {
1417 }
1418 
1419 pi_result piextUSMSharedAlloc(void **ResultPtr, pi_context Context,
1420  pi_device Device,
1421  pi_usm_mem_properties *Properties, size_t Size,
1422  pi_uint32 Alignment) {
1423  if (Context == nullptr || (Device != Context->Device)) {
1424  return PI_INVALID_CONTEXT;
1425  }
1426 
1427  if (ResultPtr == nullptr) {
1428  return PI_INVALID_OPERATION;
1429  }
1430 
1431  cm_support::CmBufferSVM *Buf = nullptr;
1432  void *SystemMemPtr = nullptr;
1433  int32_t Result = Context->Device->CmDevicePtr->CreateBufferSVM(
1434  Size, SystemMemPtr, CM_SVM_ACCESS_FLAG_DEFAULT, Buf);
1435 
1436  if (Result != cm_support::CM_SUCCESS) {
1437  return PI_OUT_OF_HOST_MEMORY;
1438  }
1439  *ResultPtr = SystemMemPtr;
1440  auto Iter = Context->Addr2CmBufferSVM.find(SystemMemPtr);
1441  if (Context->Addr2CmBufferSVM.end() != Iter) {
1442  return PI_INVALID_MEM_OBJECT;
1443  }
1444  Context->Addr2CmBufferSVM[SystemMemPtr] = Buf;
1445  return PI_SUCCESS;
1446 }
1447 
1448 pi_result piextUSMFree(pi_context Context, void *Ptr) {
1449  if (Context == nullptr) {
1450  return PI_INVALID_CONTEXT;
1451  }
1452  if (Ptr == nullptr) {
1453  return PI_INVALID_OPERATION;
1454  }
1455 
1456  cm_support::CmBufferSVM *Buf = Context->Addr2CmBufferSVM[Ptr];
1457  if (Buf == nullptr) {
1458  return PI_INVALID_MEM_OBJECT;
1459  }
1460  auto Count = Context->Addr2CmBufferSVM.erase(Ptr);
1461  if (Count != 1) {
1462  return PI_INVALID_MEM_OBJECT;
1463  }
1464  int32_t Result = Context->Device->CmDevicePtr->DestroyBufferSVM(Buf);
1465  if (cm_support::CM_SUCCESS != Result) {
1466  return PI_ERROR_UNKNOWN;
1467  }
1468  return PI_SUCCESS;
1469 }
1470 
1473 }
1474 
1476  const pi_event *, pi_event *) {
1478 }
1479 
1480 pi_result piextUSMEnqueueMemcpy(pi_queue, pi_bool, void *, const void *, size_t,
1481  pi_uint32, const pi_event *, pi_event *) {
1483 }
1484 
1486  pi_mem_advice, pi_event *) {
1488 }
1489 
1491  void *, size_t *) {
1493 }
1494 
1496  const void *) {
1498 }
1499 
1501  const void *) {
1503 }
1504 
1506  pi_uint32 RawImgSize, pi_uint32 *ImgInd) {
1509  if (RawImgSize != 1) {
1510  if (PrintPiTrace) {
1511  std::cerr
1512  << "Only single device binary image is supported in ESIMD_EMULATOR"
1513  << std::endl;
1514  }
1515  return PI_INVALID_VALUE;
1516  }
1517  *ImgInd = 0;
1518  return PI_SUCCESS;
1519 }
1520 
1523  const pi_event *, pi_event *) {
1525 }
1526 
1527 pi_result piextPluginGetOpaqueData(void *, void **OpaqueDataReturn) {
1528  *OpaqueDataReturn = reinterpret_cast<void *>(PiESimdDeviceAccess);
1529  return PI_SUCCESS;
1530 }
1531 
1533  delete reinterpret_cast<sycl::detail::ESIMDEmuPluginOpaqueData *>(
1534  PiESimdDeviceAccess->data);
1535  delete PiESimdDeviceAccess;
1536  return PI_SUCCESS;
1537 }
1538 
1540  if (PluginInit == nullptr) {
1541  return PI_INVALID_VALUE;
1542  }
1543 
1544  size_t PluginVersionSize = sizeof(PluginInit->PluginVersion);
1545  if (strlen(_PI_H_VERSION_STRING) >= PluginVersionSize) {
1546  return PI_INVALID_VALUE;
1547  }
1548  strncpy(PluginInit->PluginVersion, _PI_H_VERSION_STRING, PluginVersionSize);
1549 
1550  PiESimdDeviceAccess = new sycl::detail::ESIMDEmuPluginOpaqueData();
1551  // 'version' to be compared with 'ESIMD_EMULATOR_DEVICE_REQUIRED_VER' defined
1552  // in device interface file
1554  PiESimdDeviceAccess->data =
1555  reinterpret_cast<void *>(new sycl::detail::ESIMDDeviceInterface());
1556 
1557 #define _PI_API(api) \
1558  (PluginInit->PiFunctionTable).api = (decltype(&::api))(&api);
1559 #include <CL/sycl/detail/pi.def>
1560 
1561  return PI_SUCCESS;
1562 }
1563 
1564 } // extern C
piextDeviceSelectBinary
pi_result piextDeviceSelectBinary(pi_device, pi_device_binary *, pi_uint32 RawImgSize, pi_uint32 *ImgInd)
Selects the most appropriate device binary based on runtime information and the IR characteristics.
Definition: pi_esimd_emulator.cpp:1505
_pi_mem::mutexLock
std::mutex mutexLock
Definition: pi_esimd_emulator.hpp:99
PI_IMAGE_FORMAT_NOT_SUPPORTED
@ PI_IMAGE_FORMAT_NOT_SUPPORTED
Definition: pi.h:113
PI_DEVICE_INFO_HOST_UNIFIED_MEMORY
@ PI_DEVICE_INFO_HOST_UNIFIED_MEMORY
Definition: pi.h:247
piextUSMSharedAlloc
pi_result piextUSMSharedAlloc(void **ResultPtr, pi_context Context, pi_device Device, pi_usm_mem_properties *Properties, size_t Size, pi_uint32 Alignment)
Allocates memory accessible on both host and device.
Definition: pi_esimd_emulator.cpp:1419
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_CHAR
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_CHAR
Definition: pi.h:197
PI_DEVICE_INFO_OPENCL_C_VERSION
@ PI_DEVICE_INFO_OPENCL_C_VERSION
Definition: pi.h:267
pi_image_region_struct::height
size_t height
Definition: pi.h:829
_pi_mem
PI Mem mapping to CUDA memory allocations, both data and texture/surface.
Definition: pi_cuda.hpp:208
cl::sycl::ext::intel::experimental::esimd::SurfaceIndex
unsigned int SurfaceIndex
Surface index type.
Definition: common.hpp:250
piPluginInit
pi_result piPluginInit(pi_plugin *PluginInit)
Definition: pi_esimd_emulator.cpp:1539
PI_DEVICE_INFO_PROFILE
@ PI_DEVICE_INFO_PROFILE
Definition: pi.h:265
PI_INVALID_KERNEL
@ PI_INVALID_KERNEL
Definition: pi.h:85
piextPluginGetOpaqueData
pi_result piextPluginGetOpaqueData(void *, void **OpaqueDataReturn)
API to get Plugin internal data, opaque to SYCL RT.
Definition: pi_esimd_emulator.cpp:1527
PI_DEVICE_INFO_IMAGE3D_MAX_WIDTH
@ PI_DEVICE_INFO_IMAGE3D_MAX_WIDTH
Definition: pi.h:229
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_CHAR
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_CHAR
Definition: pi.h:211
PI_SUCCESS
@ PI_SUCCESS
Definition: pi.h:82
_pi_device::CmDevicePtr
cm_support::CmDevice * CmDevicePtr
Definition: pi_esimd_emulator.hpp:68
piProgramGetInfo
pi_result piProgramGetInfo(pi_program, pi_program_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:994
cm_barrier_ptr
void(* cm_barrier_ptr)(void)
Definition: esimd_emulator_functions_v1.h:38
PI_DEVICE_INFO_DOUBLE_FP_CONFIG
@ PI_DEVICE_INFO_DOUBLE_FP_CONFIG
Definition: pi.h:195
piDeviceGetInfo
pi_result piDeviceGetInfo(pi_device Device, pi_device_info ParamName, size_t ParamValueSize, void *ParamValue, size_t *ParamValueSizeRet)
Returns requested info for provided native device Return PI_DEVICE_INFO_EXTENSION_DEVICELIB_ASSERT fo...
Definition: pi_esimd_emulator.cpp:485
_pi_context_info
_pi_context_info
Definition: pi.h:323
pi_buff_rect_offset_struct
Definition: pi.h:800
piContextCreate
pi_result piContextCreate(const pi_context_properties *Properties, pi_uint32 NumDevices, const pi_device *Devices, void(*PFnNotify)(const char *ErrInfo, const void *PrivateInfo, size_t CB, void *UserData), void *UserData, pi_context *RetContext)
Definition: pi_esimd_emulator.cpp:618
_pi_image::Width
size_t Width
Definition: pi_esimd_emulator.hpp:138
piextProgramCreateWithNativeHandle
pi_result piextProgramCreateWithNativeHandle(pi_native_handle, pi_context, bool, pi_program *)
Creates PI program object from a native handle.
Definition: pi_esimd_emulator.cpp:1030
PI_DEVICE_INFO_DRIVER_VERSION
@ PI_DEVICE_INFO_DRIVER_VERSION
Definition: pi.h:264
piProgramCreate
pi_result piProgramCreate(pi_context, const void *, size_t, pi_program *)
Definition: pi_esimd_emulator.cpp:972
piextUSMFree
pi_result piextUSMFree(pi_context Context, void *Ptr)
Frees allocated USM memory.
Definition: pi_esimd_emulator.cpp:1448
pi_bool
pi_uint32 pi_bool
Definition: pi.h:70
IDBuilder
sycl::detail::Builder IDBuilder
Definition: pi_esimd_emulator.cpp:126
PI_DEVICE_INFO_PREFERRED_INTEROP_USER_SYNC
@ PI_DEVICE_INFO_PREFERRED_INTEROP_USER_SYNC
Definition: pi.h:270
T
piEnqueueMemUnmap
pi_result piEnqueueMemUnmap(pi_queue, pi_mem, void *, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1262
_pi_image_format::image_channel_data_type
pi_image_channel_type image_channel_data_type
Definition: pi.h:878
piextKernelSetArgSampler
pi_result piextKernelSetArgSampler(pi_kernel, pi_uint32, const pi_sampler *)
Definition: pi_esimd_emulator.cpp:1048
piQueueFinish
pi_result piQueueFinish(pi_queue)
Definition: pi_esimd_emulator.cpp:742
piEnqueueMemBufferMap
pi_result piEnqueueMemBufferMap(pi_queue, pi_mem, pi_bool, pi_map_flags, size_t, size_t, pi_uint32, const pi_event *, pi_event *, void **)
Definition: pi_esimd_emulator.cpp:1256
sycl_get_surface_base_addr_ptr
char *(* sycl_get_surface_base_addr_ptr)(int)
Definition: esimd_emulator_functions_v1.h:44
PI_INVALID_OPERATION
@ PI_INVALID_OPERATION
Definition: pi.h:84
type_traits.hpp
PI_DEVICE_INFO_NAME
@ PI_DEVICE_INFO_NAME
Definition: pi.h:262
MakeLambdaWrapper
auto MakeLambdaWrapper(KernelFunc< NDims > ArgFunc, const sycl::range< NDims > &LocalSize, const sycl::range< NDims > &GlobalSize, const sycl::id< NDims > &GlobalOffset)
Definition: pi_esimd_emulator.cpp:150
piProgramCreateWithBinary
pi_result piProgramCreateWithBinary(pi_context, pi_uint32, const pi_device *, const size_t *, const unsigned char **, size_t, const pi_device_binary_property *, pi_int32 *, pi_program *)
Creates a PI program for a context and loads the given binary into it.
Definition: pi_esimd_emulator.cpp:976
PI_DEVICE_INFO_IL_VERSION
@ PI_DEVICE_INFO_IL_VERSION
Definition: pi.h:261
PI_IMAGE_CHANNEL_TYPE_UNORM_INT8
@ PI_IMAGE_CHANNEL_TYPE_UNORM_INT8
Definition: pi.h:467
piclProgramCreateWithBinary
pi_result piclProgramCreateWithBinary(pi_context, pi_uint32, const pi_device *, const size_t *, const unsigned char **, pi_int32 *, pi_program *)
Definition: pi_esimd_emulator.cpp:983
_pi_buffer
Definition: pi_esimd_emulator.hpp:118
piextDeviceGetNativeHandle
pi_result piextDeviceGetNativeHandle(pi_device, pi_native_handle *)
Gets the native handle of a PI device object.
Definition: pi_esimd_emulator.cpp:609
PI_MEM_FLAGS_HOST_PTR_COPY
constexpr pi_mem_flags PI_MEM_FLAGS_HOST_PTR_COPY
Definition: pi.h:553
piextUSMEnqueueMemcpy
pi_result piextUSMEnqueueMemcpy(pi_queue, pi_bool, void *, const void *, size_t, pi_uint32, const pi_event *, pi_event *)
USM Memcpy API.
Definition: pi_esimd_emulator.cpp:1480
piKernelSetArg
pi_result piKernelSetArg(pi_kernel, pi_uint32, size_t, const void *)
Definition: pi_esimd_emulator.cpp:1039
_pi_image
Definition: pi_esimd_emulator.hpp:129
_pi_image_format::image_channel_order
pi_image_channel_order image_channel_order
Definition: pi.h:877
libCMBatch::libCMBatch
libCMBatch(KernelFunc< DIMS > Kernel)
Definition: pi_esimd_emulator.cpp:229
cl::sycl::ext::intel::experimental::esimd::barrier
__ESIMD_API void barrier()
Generic work-group barrier.
Definition: memory.hpp:824
PI_DEVICE_INFO_PARTITION_AFFINITY_DOMAIN
@ PI_DEVICE_INFO_PARTITION_AFFINITY_DOMAIN
Definition: pi.h:276
_pi_image_desc::image_type
pi_mem_type image_type
Definition: pi.h:882
piEventCreate
pi_result piEventCreate(pi_context, pi_event *)
Definition: pi_esimd_emulator.cpp:1071
_pi_plugin
Definition: pi.h:1739
PI_INVALID_MEM_OBJECT
@ PI_INVALID_MEM_OBJECT
Definition: pi.h:98
RangeBuilder
Definition: pi_esimd_emulator.cpp:164
PI_DEVICE_INFO_GLOBAL_MEM_CACHE_TYPE
@ PI_DEVICE_INFO_GLOBAL_MEM_CACHE_TYPE
Definition: pi.h:237
_pi_mem_advice
_pi_mem_advice
Definition: pi.h:441
piSamplerRelease
pi_result piSamplerRelease(pi_sampler)
Definition: pi_esimd_emulator.cpp:1163
piSamplerGetInfo
pi_result piSamplerGetInfo(pi_sampler, pi_sampler_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:1156
piDeviceRelease
pi_result piDeviceRelease(pi_device Device)
Definition: pi_esimd_emulator.cpp:475
RangeBuilder< 1 >::create
static sycl::range< 1 > create(Gen G)
Definition: pi_esimd_emulator.cpp:167
PI_DEVICE_INFO_MAX_WORK_ITEM_SIZES
@ PI_DEVICE_INFO_MAX_WORK_ITEM_SIZES
Definition: pi.h:191
piEnqueueEventsWaitWithBarrier
pi_result piEnqueueEventsWaitWithBarrier(pi_queue, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1170
PI_DEVICE_INFO_MAX_COMPUTE_UNITS
@ PI_DEVICE_INFO_MAX_COMPUTE_UNITS
Definition: pi.h:189
PrintPiTrace
static bool PrintPiTrace
Definition: pi_esimd_emulator.cpp:109
_pi_usm_mem_properties
_pi_usm_mem_properties
Definition: pi.h:1581
piextMemCreateWithNativeHandle
pi_result piextMemCreateWithNativeHandle(pi_native_handle, pi_mem *)
Creates PI mem object from a native handle.
Definition: pi_esimd_emulator.cpp:968
piEventSetCallback
pi_result piEventSetCallback(pi_event, pi_int32, void(*)(pi_event, pi_int32, void *), void *)
Definition: pi_esimd_emulator.cpp:1105
_pi_result
_pi_result
Definition: pi.h:81
InvokeImpl
Definition: pi_esimd_emulator.cpp:314
piProgramGetBuildInfo
pi_result piProgramGetBuildInfo(pi_program, pi_device, cl_program_build_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:1017
piextProgramGetNativeHandle
pi_result piextProgramGetNativeHandle(pi_program, pi_native_handle *)
Gets the native handle of a PI program object.
Definition: pi_esimd_emulator.cpp:1026
pi_context_properties
intptr_t pi_context_properties
Definition: pi.h:513
pi_esimd_emulator.hpp
PI_INVALID_EVENT_WAIT_LIST
@ PI_INVALID_EVENT_WAIT_LIST
Definition: pi.h:101
sycl_get_cm_image_params_ptr
void(* sycl_get_cm_image_params_ptr)(void *, char **, uint32_t *, uint32_t *, uint32_t *, std::mutex **)
Definition: esimd_emulator_functions_v1.h:49
PI_DEVICE_INFO_MAX_MEM_ALLOC_SIZE
@ PI_DEVICE_INFO_MAX_MEM_ALLOC_SIZE
Definition: pi.h:223
_pi_device_type
_pi_device_type
Definition: pi.h:162
piextPlatformCreateWithNativeHandle
pi_result piextPlatformCreateWithNativeHandle(pi_native_handle, pi_platform *)
Creates PI platform object from a native handle.
Definition: pi_esimd_emulator.cpp:421
_pi_event::CmEventPtr
cm_support::CmEvent * CmEventPtr
Definition: pi_esimd_emulator.hpp:146
piDevicePartition
pi_result piDevicePartition(pi_device, const pi_device_partition_property *, pi_uint32, pi_device *, pi_uint32 *)
Definition: pi_esimd_emulator.cpp:604
PI_DEVICE_INFO_REFERENCE_COUNT
@ PI_DEVICE_INFO_REFERENCE_COUNT
Definition: pi.h:260
PI_DEVICE_INFO_USM_CROSS_SHARED_SUPPORT
@ PI_DEVICE_INFO_USM_CROSS_SHARED_SUPPORT
Definition: pi.h:287
PI_DEVICE_INFO_PLATFORM
@ PI_DEVICE_INFO_PLATFORM
Definition: pi.h:259
piMemGetInfo
pi_result piMemGetInfo(pi_mem, cl_mem_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:813
helpers.hpp
host_profiling_info.hpp
piEventGetInfo
pi_result piEventGetInfo(pi_event, pi_event_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:1073
PI_DEVICE_INFO_ADDRESS_BITS
@ PI_DEVICE_INFO_ADDRESS_BITS
Definition: pi.h:222
_pi_image_desc::image_height
size_t image_height
Definition: pi.h:884
_pi_platform
A PI platform stores all known PI devices, in the CUDA plugin this is just a vector of available devi...
Definition: pi_cuda.hpp:62
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_DOUBLE
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_DOUBLE
Definition: pi.h:207
PI_DEVICE_INFO_MAX_SAMPLERS
@ PI_DEVICE_INFO_MAX_SAMPLERS
Definition: pi.h:234
PI_DEVICE_INFO_MAX_NUM_SUB_GROUPS
@ PI_DEVICE_INFO_MAX_NUM_SUB_GROUPS
Definition: pi.h:279
PI_DEVICE_INFO_MEM_BASE_ADDR_ALIGN
@ PI_DEVICE_INFO_MEM_BASE_ADDR_ALIGN
Definition: pi.h:236
PI_DEVICE_INFO_BUILT_IN_KERNELS
@ PI_DEVICE_INFO_BUILT_IN_KERNELS
Definition: pi.h:258
PI_DEVICE_INFO_USM_DEVICE_SUPPORT
@ PI_DEVICE_INFO_USM_DEVICE_SUPPORT
Definition: pi.h:284
LambdaWrapper
Definition: pi_esimd_emulator.cpp:135
PI_PLATFORM_INFO_NAME
@ PI_PLATFORM_INFO_NAME
Definition: pi.h:131
PI_DEVICE_INFO_IMAGE_MAX_ARRAY_SIZE
@ PI_DEVICE_INFO_IMAGE_MAX_ARRAY_SIZE
Definition: pi.h:233
_pi_device_info
_pi_device_info
Definition: pi.h:186
PI_DEVICE_INFO_MAX_CONSTANT_BUFFER_SIZE
@ PI_DEVICE_INFO_MAX_CONSTANT_BUFFER_SIZE
Definition: pi.h:242
_pi_image_info
_pi_image_info
Definition: pi.h:373
group.hpp
libCMBatch::runIterationSpace
void runIterationSpace(const sycl::range< DIMS > &LocalSize, const sycl::range< DIMS > &GlobalSize, const sycl::id< DIMS > &GlobalOffset)
Invoking kernel lambda function wrapped by 'LambdaWrapper' using 'InvokeLambda' function.
Definition: pi_esimd_emulator.cpp:234
piQueueRetain
pi_result piQueueRetain(pi_queue Queue)
Definition: pi_esimd_emulator.cpp:720
id.hpp
isNull
static bool isNull(int NDims, const size_t *R)
Implementation for Host Kernel Launch used by piEnqueueKernelLaunch.
Definition: pi_esimd_emulator.cpp:302
LambdaWrapper::GlobalSize
const sycl::range< NDims > & GlobalSize
Definition: pi_esimd_emulator.cpp:138
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_INT
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_INT
Definition: pi.h:201
_pi_kernel
Implementation of a PI Kernel for CUDA.
Definition: pi_cuda.hpp:578
PI_DEVICE_INFO_ERROR_CORRECTION_SUPPORT
@ PI_DEVICE_INFO_ERROR_CORRECTION_SUPPORT
Definition: pi.h:246
piextKernelCreateWithNativeHandle
pi_result piextKernelCreateWithNativeHandle(pi_native_handle, pi_context, pi_program, bool, pi_kernel *)
Creates PI kernel object from a native handle.
Definition: pi_esimd_emulator.cpp:1389
PI_DEVICE_INFO_IMAGE2D_MAX_HEIGHT
@ PI_DEVICE_INFO_IMAGE2D_MAX_HEIGHT
Definition: pi.h:228
PI_INVALID_EVENT
@ PI_INVALID_EVENT
Definition: pi.h:100
piEventsWait
pi_result piEventsWait(pi_uint32 NumEvents, const pi_event *EventList)
Definition: pi_esimd_emulator.cpp:1087
PI_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE
constexpr pi_queue_properties PI_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE
Definition: pi.h:571
piextQueueGetNativeHandle
pi_result piextQueueGetNativeHandle(pi_queue, pi_native_handle *)
Gets the native handle of a PI queue object.
Definition: pi_esimd_emulator.cpp:749
PI_DEVICE_INFO_GLOBAL_MEM_CACHE_SIZE
@ PI_DEVICE_INFO_GLOBAL_MEM_CACHE_SIZE
Definition: pi.h:240
piextDeviceCreateWithNativeHandle
pi_result piextDeviceCreateWithNativeHandle(pi_native_handle, pi_platform, pi_device *)
Creates PI device object from a native handle.
Definition: pi_esimd_emulator.cpp:613
cl::sycl::detail::memcpy
void memcpy(void *Dst, const void *Src, std::size_t Size)
piMemBufferPartition
pi_result piMemBufferPartition(pi_mem, pi_mem_flags, pi_buffer_create_type, void *, pi_mem *)
Definition: pi_esimd_emulator.cpp:1323
_pi_mem::Context
pi_context Context
Definition: pi_esimd_emulator.hpp:94
piEnqueueNativeKernel
pi_result piEnqueueNativeKernel(pi_queue, void(*)(void *), void *, size_t, pi_uint32, const pi_mem *, const void **, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1398
_pi_queue_info
_pi_queue_info
Definition: pi.h:332
_pi_buffer_create_type
_pi_buffer_create_type
Definition: pi.h:482
PI_ERROR_UNKNOWN
@ PI_ERROR_UNKNOWN
Definition: pi.h:119
piEnqueueMemImageRead
pi_result piEnqueueMemImageRead(pi_queue CommandQueue, pi_mem Image, pi_bool BlockingRead, pi_image_offset Origin, pi_image_region Region, size_t RowPitch, size_t SlicePitch, void *Ptr, pi_uint32 NumEventsInWaitList, const pi_event *EventWaitList, pi_event *Event)
Definition: pi_esimd_emulator.cpp:1271
piKernelGetSubGroupInfo
pi_result piKernelGetSubGroupInfo(pi_kernel, pi_device, pi_kernel_sub_group_info, size_t, const void *, size_t, void *, size_t *)
API to query information from the sub-group from a kernel.
Definition: pi_esimd_emulator.cpp:1061
piextUSMEnqueueMemAdvise
pi_result piextUSMEnqueueMemAdvise(pi_queue, const void *, size_t, pi_mem_advice, pi_event *)
USM Memadvise API.
Definition: pi_esimd_emulator.cpp:1485
export.hpp
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_HALF
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_HALF
Definition: pi.h:220
piDevicesGet
pi_result piDevicesGet(pi_platform Platform, pi_device_type DeviceType, pi_uint32 NumEntries, pi_device *Devices, pi_uint32 *NumDevices)
Definition: pi_esimd_emulator.cpp:425
piEnqueueMemBufferReadRect
pi_result piEnqueueMemBufferReadRect(pi_queue, pi_mem, pi_bool, pi_buff_rect_offset, pi_buff_rect_offset, pi_buff_rect_region, size_t, size_t, size_t, size_t, void *, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1214
PI_DEVICE_INFO_IMAGE_MAX_BUFFER_SIZE
@ PI_DEVICE_INFO_IMAGE_MAX_BUFFER_SIZE
Definition: pi.h:232
piEnqueueMemBufferWrite
pi_result piEnqueueMemBufferWrite(pi_queue, pi_mem, pi_bool, size_t, size_t, const void *, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1222
PI_DEVICE_INFO_LOCAL_MEM_SIZE
@ PI_DEVICE_INFO_LOCAL_MEM_SIZE
Definition: pi.h:245
piKernelSetExecInfo
pi_result piKernelSetExecInfo(pi_kernel, pi_kernel_exec_info, size_t, const void *)
API to set attributes controlling kernel execution.
Definition: pi_esimd_emulator.cpp:1495
piSamplerCreate
pi_result piSamplerCreate(pi_context, const pi_sampler_properties *, pi_sampler *)
Definition: pi_esimd_emulator.cpp:1151
_pi_queue
PI queue mapping on to CUstream objects.
Definition: pi_cuda.hpp:379
sycl_get_cm_buffer_params
void sycl_get_cm_buffer_params(void *PtrInput, char **BaseAddr, uint32_t *Width, std::mutex **MtxLock)
Definition: pi_esimd_emulator.cpp:253
piextKernelSetArgMemObj
pi_result piextKernelSetArgMemObj(pi_kernel, pi_uint32, const pi_mem *)
Definition: pi_esimd_emulator.cpp:1043
PI_DEVICE_TYPE_GPU
@ PI_DEVICE_TYPE_GPU
A PI device that is a GPU.
Definition: pi.h:170
pi_uint32
uint32_t pi_uint32
Definition: pi.h:68
piProgramCompile
pi_result piProgramCompile(pi_program, pi_uint32, const pi_device *, const char *, pi_uint32, const pi_program *, const char **, void(*)(pi_program, void *), void *)
Definition: pi_esimd_emulator.cpp:1005
kernel.hpp
pi_buff_rect_region_struct
Definition: pi.h:809
pi_device_binary_struct
This struct is a record of the device binary information.
Definition: pi.h:743
piextMemGetNativeHandle
pi_result piextMemGetNativeHandle(pi_mem, pi_native_handle *)
Gets the native handle of a PI mem object.
Definition: pi_esimd_emulator.cpp:964
piextContextGetNativeHandle
pi_result piextContextGetNativeHandle(pi_context, pi_native_handle *)
Gets the native handle of a PI context object.
Definition: pi_esimd_emulator.cpp:655
PI_DEVICE_INFO_LINKER_AVAILABLE
@ PI_DEVICE_INFO_LINKER_AVAILABLE
Definition: pi.h:253
_pi_plugin::PluginVersion
char PluginVersion[4]
Definition: pi.h:1749
PI_DEVICE_INFO_EXECUTION_CAPABILITIES
@ PI_DEVICE_INFO_EXECUTION_CAPABILITIES
Definition: pi.h:254
piPlatformGetInfo
pi_result piPlatformGetInfo(pi_platform Platform, pi_platform_info ParamName, size_t ParamValueSize, void *ParamValue, size_t *ParamValueSizeRet)
Definition: pi_esimd_emulator.cpp:385
PI_DEVICE_INFO_PROFILING_TIMER_RESOLUTION
@ PI_DEVICE_INFO_PROFILING_TIMER_RESOLUTION
Definition: pi.h:248
_pi_kernel_exec_info
_pi_kernel_exec_info
Definition: pi.h:1251
PI_MEM_TYPE_IMAGE2D
@ PI_MEM_TYPE_IMAGE2D
Definition: pi.h:433
piContextGetInfo
pi_result piContextGetInfo(pi_context, pi_context_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:645
LambdaWrapper::LambdaWrapper
LambdaWrapper(KernelFunc< NDims > ArgFunc, const sycl::range< NDims > &ArgLocalSize, const sycl::range< NDims > &ArgGlobalSize, const sycl::id< NDims > &ArgGlobalOffset)
Definition: pi_esimd_emulator.cpp:140
range.hpp
cl::sycl::detail::pi::die
void die(const char *Message)
Definition: pi.cpp:537
PI_DEVICE_INFO_HALF_FP_CONFIG
@ PI_DEVICE_INFO_HALF_FP_CONFIG
Definition: pi.h:194
piEnqueueMemBufferWriteRect
pi_result piEnqueueMemBufferWriteRect(pi_queue, pi_mem, pi_bool, pi_buff_rect_offset, pi_buff_rect_offset, pi_buff_rect_region, size_t, size_t, size_t, size_t, const void *, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1228
cl::sycl::ext::intel::experimental::esimd::fence
__ESIMD_API void fence(fence_mask cntl)
esimd::fence sets the memory read/write order.
Definition: memory.hpp:814
pi_mem_flags
pi_bitfield pi_mem_flags
Definition: pi.h:547
piextProgramSetSpecializationConstant
pi_result piextProgramSetSpecializationConstant(pi_program, pi_uint32, size_t, const void *)
Sets a specialization constant to a specific value.
Definition: pi_esimd_emulator.cpp:1500
PI_DEVICE_INFO_VENDOR
@ PI_DEVICE_INFO_VENDOR
Definition: pi.h:263
_pi_mem::getMemType
_pi_mem_type getMemType() const
Definition: pi_esimd_emulator.hpp:106
pi_context_extended_deleter
void(* pi_context_extended_deleter)(void *user_data)
Definition: pi.h:1028
PI_INVALID_QUEUE_PROPERTIES
@ PI_INVALID_QUEUE_PROPERTIES
Definition: pi.h:86
piextGetDeviceFunctionPointer
pi_result piextGetDeviceFunctionPointer(pi_device, pi_program, const char *, pi_uint64 *)
Retrieves a device function pointer to a user-defined function.
Definition: pi_esimd_emulator.cpp:1404
piextUSMEnqueuePrefetch
pi_result piextUSMEnqueuePrefetch(pi_queue, const void *, size_t, pi_usm_migration_flags, pi_uint32, const pi_event *, pi_event *)
Hint to migrate memory to the device.
Definition: pi_esimd_emulator.cpp:1521
PI_DEVICE_INFO_USM_SYSTEM_SHARED_SUPPORT
@ PI_DEVICE_INFO_USM_SYSTEM_SHARED_SUPPORT
Definition: pi.h:289
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_SHORT
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_SHORT
Definition: pi.h:199
PI_MEM_TYPE_BUFFER
@ PI_MEM_TYPE_BUFFER
Definition: pi.h:432
_pi_device::Platform
pi_platform Platform
Definition: pi_esimd_emulator.hpp:67
ESIMDEmuPluginDataVersion
#define ESIMDEmuPluginDataVersion
Definition: pi_esimd_emulator.cpp:120
piKernelGetInfo
pi_result piKernelGetInfo(pi_kernel, pi_kernel_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:1052
PI_DEVICE_INFO_GLOBAL_MEM_CACHELINE_SIZE
@ PI_DEVICE_INFO_GLOBAL_MEM_CACHELINE_SIZE
Definition: pi.h:238
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_LONG
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_LONG
Definition: pi.h:215
piextKernelSetArgPointer
pi_result piextKernelSetArgPointer(pi_kernel, pi_uint32, size_t, const void *)
Sets up pointer arguments for CL kernels.
Definition: pi_esimd_emulator.cpp:1471
piextEventGetNativeHandle
pi_result piextEventGetNativeHandle(pi_event, pi_native_handle *)
Gets the native handle of a PI event object.
Definition: pi_esimd_emulator.cpp:1143
_pi_platform::CmEmuVersion
std::string CmEmuVersion
Definition: pi_esimd_emulator.hpp:60
piextContextSetExtendedDeleter
pi_result piextContextSetExtendedDeleter(pi_context, pi_context_extended_deleter, void *)
Definition: pi_esimd_emulator.cpp:650
PI_DEVICE_INFO_USM_HOST_SUPPORT
@ PI_DEVICE_INFO_USM_HOST_SUPPORT
Definition: pi.h:283
PI_INVALID_QUEUE
@ PI_INVALID_QUEUE
Definition: pi.h:92
PI_MEM_FLAGS_HOST_PTR_USE
constexpr pi_mem_flags PI_MEM_FLAGS_HOST_PTR_USE
Definition: pi.h:552
_pi_image::CmSurfacePtr
cm_support::CmSurface2D * CmSurfacePtr
Definition: pi_esimd_emulator.hpp:137
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_FLOAT
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_FLOAT
Definition: pi.h:216
piEnqueueMemBufferCopyRect
pi_result piEnqueueMemBufferCopyRect(pi_queue, pi_mem, pi_mem, pi_buff_rect_offset, pi_buff_rect_offset, pi_buff_rect_region, size_t, size_t, size_t, size_t, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1242
piTearDown
pi_result piTearDown(void *)
API to notify that the plugin should clean up its resources.
Definition: pi_esimd_emulator.cpp:1532
ESIMDEmuPluginInterfaceVersion
#define ESIMDEmuPluginInterfaceVersion
Definition: pi_esimd_emulator.cpp:124
cm_slm_init_ptr
void(* cm_slm_init_ptr)(size_t)
Definition: esimd_emulator_functions_v1.h:46
piEnqueueEventsWait
pi_result piEnqueueEventsWait(pi_queue, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1165
PiESimdDeviceAccess
static sycl::detail::ESIMDEmuPluginOpaqueData * PiESimdDeviceAccess
Definition: pi_esimd_emulator.cpp:116
PI_DEVICE_INFO_PARENT_DEVICE
@ PI_DEVICE_INFO_PARENT_DEVICE
Definition: pi.h:272
PI_PLATFORM_INFO_VERSION
@ PI_PLATFORM_INFO_VERSION
Definition: pi.h:134
pi_uint64
uint64_t pi_uint64
Definition: pi.h:69
PI_DEVICE_INFO_PARTITION_TYPE
@ PI_DEVICE_INFO_PARTITION_TYPE
Definition: pi.h:278
_pi_event_info
_pi_event_info
Definition: pi.h:391
piMemBufferCreate
pi_result piMemBufferCreate(pi_context Context, pi_mem_flags Flags, size_t Size, void *HostPtr, pi_mem *RetMem, const pi_mem_properties *properties)
Definition: pi_esimd_emulator.cpp:758
_pi_image::BytesPerPixel
size_t BytesPerPixel
Definition: pi_esimd_emulator.hpp:140
_pi_sampler_info
_pi_sampler_info
Definition: pi.h:489
_pi_device_binary_property_struct
Definition: pi.h:648
pi_mem_properties
pi_bitfield pi_mem_properties
Definition: pi.h:565
_pi_program
Implementation of PI Program on CUDA Module object.
Definition: pi_cuda.hpp:523
LambdaWrapper::Func
KernelFunc< NDims > Func
Definition: pi_esimd_emulator.cpp:136
_pi_sampler
Implementation of samplers for CUDA.
Definition: pi_cuda.hpp:744
PI_DEVICE_INFO_PRINTF_BUFFER_SIZE
@ PI_DEVICE_INFO_PRINTF_BUFFER_SIZE
Definition: pi.h:269
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_DOUBLE
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_DOUBLE
Definition: pi.h:218
PI_PLATFORM_INFO_PROFILE
@ PI_PLATFORM_INFO_PROFILE
Definition: pi.h:132
PI_DEVICE_INFO_MAX_CONSTANT_ARGS
@ PI_DEVICE_INFO_MAX_CONSTANT_ARGS
Definition: pi.h:243
pi_native_handle
uintptr_t pi_native_handle
Definition: pi.h:72
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_SHORT
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_SHORT
Definition: pi.h:212
PI_DEVICE_INFO_GLOBAL_MEM_SIZE
@ PI_DEVICE_INFO_GLOBAL_MEM_SIZE
Definition: pi.h:241
InvokeImpl::get_range
static sycl::range< NDims > get_range(const size_t *Array)
Definition: pi_esimd_emulator.cpp:316
pi_sampler_properties
pi_bitfield pi_sampler_properties
Definition: pi.h:521
piKernelRetain
pi_result piKernelRetain(pi_kernel)
Definition: pi_esimd_emulator.cpp:1067
PI_MEM_FLAGS_ACCESS_RW
constexpr pi_mem_flags PI_MEM_FLAGS_ACCESS_RW
Definition: pi.h:549
accessor_impl.hpp
_pi_event::release
pi_result release()
Definition: pi_cuda.cpp:459
piclProgramCreateWithSource
pi_result piclProgramCreateWithSource(pi_context, pi_uint32, const char **, const size_t *, pi_program *)
Definition: pi_esimd_emulator.cpp:989
_pi_image::Height
size_t Height
Definition: pi_esimd_emulator.hpp:139
_pi_image_format
Definition: pi.h:876
PI_PLATFORM_INFO_VENDOR
@ PI_PLATFORM_INFO_VENDOR
Definition: pi.h:133
_pi_kernel_info
_pi_kernel_info
Definition: pi.h:341
piKernelRelease
pi_result piKernelRelease(pi_kernel)
Definition: pi_esimd_emulator.cpp:1069
PI_INVALID_CONTEXT
@ PI_INVALID_CONTEXT
Definition: pi.h:88
PI_DEVICE_INFO_USM_SINGLE_SHARED_SUPPORT
@ PI_DEVICE_INFO_USM_SINGLE_SHARED_SUPPORT
Definition: pi.h:285
InvokeImpl::invoke
static void invoke(void *Fptr, const size_t *GlobalWorkOffset, const size_t *GlobalWorkSize, const size_t *LocalWorkSize)
Definition: pi_esimd_emulator.cpp:325
PI_IMAGE_CHANNEL_ORDER_RGBA
@ PI_IMAGE_CHANNEL_ORDER_RGBA
Definition: pi.h:452
PI_DEVICE_INFO_QUEUE_PROPERTIES
@ PI_DEVICE_INFO_QUEUE_PROPERTIES
Definition: pi.h:196
PI_DEVICE_INFO_PARTITION_MAX_SUB_DEVICES
@ PI_DEVICE_INFO_PARTITION_MAX_SUB_DEVICES
Definition: pi.h:274
UNSUPPORTED_INFO
#define UNSUPPORTED_INFO(info)
piextPlatformGetNativeHandle
pi_result piextPlatformGetNativeHandle(pi_platform, pi_native_handle *)
Gets the native handle of a PI platform object.
Definition: pi_esimd_emulator.cpp:417
KernelFunc
std::function< void(const sycl::nd_item< NDims > &)> KernelFunc
Definition: pi_esimd_emulator.cpp:129
PI_INVALID_IMAGE_FORMAT_DESCRIPTOR
@ PI_INVALID_IMAGE_FORMAT_DESCRIPTOR
Definition: pi.h:112
piProgramBuild
pi_result piProgramBuild(pi_program, pi_uint32, const pi_device *, const char *, void(*)(pi_program, void *), void *)
Definition: pi_esimd_emulator.cpp:1012
_pi_image_desc::image_width
size_t image_width
Definition: pi.h:883
piextUSMEnqueueMemset
pi_result piextUSMEnqueueMemset(pi_queue, void *, pi_int32, size_t, pi_uint32, const pi_event *, pi_event *)
USM Memset API.
Definition: pi_esimd_emulator.cpp:1475
PI_DEVICE_INFO_SINGLE_FP_CONFIG
@ PI_DEVICE_INFO_SINGLE_FP_CONFIG
Definition: pi.h:193
PI_INVALID_VALUE
@ PI_INVALID_VALUE
Definition: pi.h:87
__cm_emu_get_slm_ptr
char *(* __cm_emu_get_slm_ptr)(void)
Definition: esimd_emulator_functions_v1.h:45
piextUSMGetMemAllocInfo
pi_result piextUSMGetMemAllocInfo(pi_context, const void *, pi_mem_info, size_t, void *, size_t *)
API to query information about USM allocated pointers Valid Queries: PI_MEM_ALLOC_TYPE returns host/d...
Definition: pi_esimd_emulator.cpp:1490
piextUSMDeviceAlloc
pi_result piextUSMDeviceAlloc(void **, pi_context, pi_device, pi_usm_mem_properties *, size_t, pi_uint32)
Allocates device memory.
Definition: pi_esimd_emulator.cpp:1414
_pi_context::Addr2CmBufferSVM
std::unordered_map< void *, cm_support::CmBufferSVM * > Addr2CmBufferSVM
Map SVM memory starting address to corresponding CmBufferSVM object.
Definition: pi_esimd_emulator.hpp:79
PI_DEVICE_INFO_MAX_WORK_ITEM_DIMENSIONS
@ PI_DEVICE_INFO_MAX_WORK_ITEM_DIMENSIONS
Definition: pi.h:190
PI_DEVICE_INFO_LOCAL_MEM_TYPE
@ PI_DEVICE_INFO_LOCAL_MEM_TYPE
Definition: pi.h:244
PI_DEVICE_INFO_IMAGE3D_MAX_DEPTH
@ PI_DEVICE_INFO_IMAGE3D_MAX_DEPTH
Definition: pi.h:231
PI_DEVICE_INFO_COMPILER_AVAILABLE
@ PI_DEVICE_INFO_COMPILER_AVAILABLE
Definition: pi.h:252
_PI_H_VERSION_STRING
#define _PI_H_VERSION_STRING
Definition: pi.h:51
backend_types.hpp
PI_DEVICE_INFO_VERSION
@ PI_DEVICE_INFO_VERSION
Definition: pi.h:266
sycl_get_cm_image_params
void sycl_get_cm_image_params(void *PtrInput, char **BaseAddr, uint32_t *Width, uint32_t *Height, uint32_t *Bpp, std::mutex **MtxLock)
Definition: pi_esimd_emulator.cpp:265
PI_IMAGE_CHANNEL_TYPE_UNSIGNED_INT8
@ PI_IMAGE_CHANNEL_TYPE_UNSIGNED_INT8
Definition: pi.h:475
PI_IMAGE_CHANNEL_TYPE_UNSIGNED_INT32
@ PI_IMAGE_CHANNEL_TYPE_UNSIGNED_INT32
Definition: pi.h:477
piEnqueueMemBufferFill
pi_result piEnqueueMemBufferFill(pi_queue, pi_mem, const void *, size_t, size_t, size_t, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1250
pi_queue_properties
pi_bitfield pi_queue_properties
Definition: pi.h:570
PI_DEVICE_INFO_MAX_WRITE_IMAGE_ARGS
@ PI_DEVICE_INFO_MAX_WRITE_IMAGE_ARGS
Definition: pi.h:226
_pi_platform_info
_pi_platform_info
Definition: pi.h:129
PI_INVALID_WORK_GROUP_SIZE
@ PI_INVALID_WORK_GROUP_SIZE
Definition: pi.h:104
piEnqueueMemBufferCopy
pi_result piEnqueueMemBufferCopy(pi_queue, pi_mem, pi_mem, size_t, size_t, size_t, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1236
piEnqueueMemBufferRead
pi_result piEnqueueMemBufferRead(pi_queue Queue, pi_mem Src, pi_bool BlockingRead, size_t Offset, size_t Size, void *Dst, pi_uint32 NumEventsInWaitList, const pi_event *EventWaitList, pi_event *Event)
Definition: pi_esimd_emulator.cpp:1175
PI_IMAGE_CHANNEL_ORDER_ARGB
@ PI_IMAGE_CHANNEL_ORDER_ARGB
Definition: pi.h:454
_pi_image_desc
Definition: pi.h:881
PI_DEVICE_INFO_ENDIAN_LITTLE
@ PI_DEVICE_INFO_ENDIAN_LITTLE
Definition: pi.h:250
piQueueCreate
pi_result piQueueCreate(pi_context Context, pi_device Device, pi_queue_properties Properties, pi_queue *Queue)
Definition: pi_esimd_emulator.cpp:690
_pi_queue::CmQueuePtr
cm_support::CmQueue * CmQueuePtr
Definition: pi_esimd_emulator.hpp:88
piContextRelease
pi_result piContextRelease(pi_context Context)
Definition: pi_esimd_emulator.cpp:675
kernel_desc.hpp
_pi_event
PI Event mapping to CUevent.
Definition: pi_cuda.hpp:419
piQueueRelease
pi_result piQueueRelease(pi_queue Queue)
Definition: pi_esimd_emulator.cpp:728
DIE_NO_IMPLEMENTATION
#define DIE_NO_IMPLEMENTATION
Definition: pi_esimd_emulator.cpp:341
PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_INT
@ PI_DEVICE_INFO_NATIVE_VECTOR_WIDTH_INT
Definition: pi.h:214
PI_DEVICE_INFO_MAX_READ_IMAGE_ARGS
@ PI_DEVICE_INFO_MAX_READ_IMAGE_ARGS
Definition: pi.h:225
PI_OUT_OF_RESOURCES
@ PI_OUT_OF_RESOURCES
Definition: pi.h:99
piKernelCreate
pi_result piKernelCreate(pi_program, const char *, pi_kernel *)
Definition: pi_esimd_emulator.cpp:1035
piEventRelease
pi_result piEventRelease(pi_event Event)
Definition: pi_esimd_emulator.cpp:1122
piMemRetain
pi_result piMemRetain(pi_mem Mem)
Definition: pi_esimd_emulator.cpp:817
piKernelGetGroupInfo
pi_result piKernelGetGroupInfo(pi_kernel, pi_device, pi_kernel_group_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:1056
_pi_mem_info
_pi_mem_info
Definition: pi.h:1567
cm_fence_ptr
void(* cm_fence_ptr)(void)
Definition: esimd_emulator_functions_v1.h:40
pi_image_offset_struct
Definition: pi.h:818
piMemRelease
pi_result piMemRelease(pi_mem Mem)
Definition: pi_esimd_emulator.cpp:825
PI_INVALID_PLATFORM
@ PI_INVALID_PLATFORM
Definition: pi.h:89
PI_OUT_OF_HOST_MEMORY
@ PI_OUT_OF_HOST_MEMORY
Definition: pi.h:93
LambdaWrapper::LocalSize
const sycl::range< NDims > & LocalSize
Definition: pi_esimd_emulator.cpp:137
PI_DEVICE_INFO_VENDOR_ID
@ PI_DEVICE_INFO_VENDOR_ID
Definition: pi.h:188
piextEventCreateWithNativeHandle
pi_result piextEventCreateWithNativeHandle(pi_native_handle, pi_context, bool, pi_event *)
Creates PI event object from a native handle.
Definition: pi_esimd_emulator.cpp:1147
pi_image_region_struct
Definition: pi.h:827
piDeviceRetain
pi_result piDeviceRetain(pi_device Device)
Definition: pi_esimd_emulator.cpp:465
PI_DEVICE_INFO_EXTENSIONS
@ PI_DEVICE_INFO_EXTENSIONS
Definition: pi.h:268
piextUSMHostAlloc
pi_result piextUSMHostAlloc(void **, pi_context, pi_usm_mem_properties *, size_t, pi_uint32)
Allocates host memory accessible by the device.
Definition: pi_esimd_emulator.cpp:1409
piProgramRetain
pi_result piProgramRetain(pi_program)
Definition: pi_esimd_emulator.cpp:1022
cm_sbarrier_ptr
void(* cm_sbarrier_ptr)(uint32_t)
Definition: esimd_emulator_functions_v1.h:39
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_FLOAT
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_FLOAT
Definition: pi.h:205
_pi_usm_migration_flags
_pi_usm_migration_flags
Definition: pi.h:1588
common.hpp
piextContextCreateWithNativeHandle
pi_result piextContextCreateWithNativeHandle(pi_native_handle, pi_uint32, const pi_device *, bool, pi_context *)
Creates PI context object from a native handle.
Definition: pi_esimd_emulator.cpp:659
piEnqueueKernelLaunch
pi_result piEnqueueKernelLaunch(pi_queue Queue, pi_kernel Kernel, pi_uint32 WorkDim, const size_t *GlobalWorkOffset, const size_t *GlobalWorkSize, const size_t *LocalWorkSize, pi_uint32 NumEventsInWaitList, const pi_event *EventWaitList, pi_event *Event)
Definition: pi_esimd_emulator.cpp:1329
_pi_event::IsDummyEvent
bool IsDummyEvent
Definition: pi_esimd_emulator.hpp:149
nd_item.hpp
PI_DEVICE_INFO_TYPE
@ PI_DEVICE_INFO_TYPE
Definition: pi.h:187
piMemImageGetInfo
pi_result piMemImageGetInfo(pi_mem, pi_image_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:1267
PI_DEVICE_INFO_MAX_WORK_GROUP_SIZE
@ PI_DEVICE_INFO_MAX_WORK_GROUP_SIZE
Definition: pi.h:192
piPlatformsGet
pi_result piPlatformsGet(pi_uint32 NumEntries, pi_platform *Platforms, pi_uint32 *NumPlatforms)
Definition: pi_esimd_emulator.cpp:357
_pi_program_info
_pi_program_info
Definition: pi.h:311
_pi_profiling_info
_pi_profiling_info
Definition: pi.h:536
RangeBuilder< 3 >::create
static sycl::range< 3 > create(Gen G)
Definition: pi_esimd_emulator.cpp:177
CONTINUE_NO_IMPLEMENTATION
#define CONTINUE_NO_IMPLEMENTATION
Definition: pi_esimd_emulator.cpp:349
PI_DEVICE_INFO_IMAGE_SUPPORT
@ PI_DEVICE_INFO_IMAGE_SUPPORT
Definition: pi.h:224
piMemImageCreate
pi_result piMemImageCreate(pi_context Context, pi_mem_flags Flags, const pi_image_format *ImageFormat, const pi_image_desc *ImageDesc, void *HostPtr, pi_mem *RetImage)
Definition: pi_esimd_emulator.cpp:883
PI_DEVICE_INFO_MAX_PARAMETER_SIZE
@ PI_DEVICE_INFO_MAX_PARAMETER_SIZE
Definition: pi.h:235
libCMBatch
Definition: pi_esimd_emulator.cpp:219
_pi_kernel_group_info
_pi_kernel_group_info
Definition: pi.h:350
PI_DEVICE_INFO_SUB_GROUP_SIZES_INTEL
@ PI_DEVICE_INFO_SUB_GROUP_SIZES_INTEL
Definition: pi.h:282
piextQueueCreateWithNativeHandle
pi_result piextQueueCreateWithNativeHandle(pi_native_handle, pi_context, pi_queue *, bool)
Creates PI queue object from a native handle.
Definition: pi_esimd_emulator.cpp:753
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_HALF
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_HALF
Definition: pi.h:209
piEnqueueMemImageCopy
pi_result piEnqueueMemImageCopy(pi_queue, pi_mem, pi_mem, pi_image_offset, pi_image_offset, pi_image_region, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1311
piProgramRelease
pi_result piProgramRelease(pi_program)
Definition: pi_esimd_emulator.cpp:1024
piContextRetain
pi_result piContextRetain(pi_context Context)
Definition: pi_esimd_emulator.cpp:665
PI_DEVICE_INFO_MAX_CLOCK_FREQUENCY
@ PI_DEVICE_INFO_MAX_CLOCK_FREQUENCY
Definition: pi.h:221
_pi_mem::SurfaceIndex
int SurfaceIndex
Definition: pi_esimd_emulator.hpp:102
piEventSetStatus
pi_result piEventSetStatus(pi_event, pi_int32)
Definition: pi_esimd_emulator.cpp:1110
_pi_context::Device
pi_device Device
One-to-one mapping between Context and Device.
Definition: pi_esimd_emulator.hpp:75
PI_DEVICE_INFO_IMAGE2D_MAX_WIDTH
@ PI_DEVICE_INFO_IMAGE2D_MAX_WIDTH
Definition: pi.h:227
PI_DEVICE_INFO_AVAILABLE
@ PI_DEVICE_INFO_AVAILABLE
Definition: pi.h:251
pi_map_flags
pi_bitfield pi_map_flags
Definition: pi.h:557
piEventGetProfilingInfo
pi_result piEventGetProfilingInfo(pi_event Event, pi_profiling_info ParamName, size_t ParamValueSize, void *ParamValue, size_t *ParamValueSizeRet)
Definition: pi_esimd_emulator.cpp:1077
PI_DEVICE_INFO_IMAGE3D_MAX_HEIGHT
@ PI_DEVICE_INFO_IMAGE3D_MAX_HEIGHT
Definition: pi.h:230
_pi_event::OwnerQueue
cm_support::CmQueue * OwnerQueue
Definition: pi_esimd_emulator.hpp:147
piEnqueueMemImageWrite
pi_result piEnqueueMemImageWrite(pi_queue, pi_mem, pi_bool, pi_image_offset, pi_image_region, size_t, size_t, const void *, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1305
piQueueGetInfo
pi_result piQueueGetInfo(pi_queue, pi_queue_info, size_t, void *, size_t *)
Definition: pi_esimd_emulator.cpp:716
piSamplerRetain
pi_result piSamplerRetain(pi_sampler)
Definition: pi_esimd_emulator.cpp:1161
RangeBuilder< 2 >::create
static sycl::range< 2 > create(Gen G)
Definition: pi_esimd_emulator.cpp:172
piProgramLink
pi_result piProgramLink(pi_context, pi_uint32, const pi_device *, const char *, pi_uint32, const pi_program *, void(*)(pi_program, void *), void *, pi_program *)
Definition: pi_esimd_emulator.cpp:999
InvokeLambda
void InvokeLambda(void *Wrapper)
Definition: pi_esimd_emulator.cpp:185
LambdaWrapper::GlobalOffset
const sycl::id< NDims > & GlobalOffset
Definition: pi_esimd_emulator.cpp:139
pi_device_partition_property
intptr_t pi_device_partition_property
Definition: pi.h:610
PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_LONG
@ PI_DEVICE_INFO_PREFERRED_VECTOR_WIDTH_LONG
Definition: pi.h:203
ConvertPiImageFormatToCmFormat
cm_support::CM_SURFACE_FORMAT ConvertPiImageFormatToCmFormat(const pi_image_format *PiFormat)
Definition: pi_esimd_emulator.cpp:859
PI_INVALID_DEVICE
@ PI_INVALID_DEVICE
Definition: pi.h:90
piextKernelGetNativeHandle
pi_result piextKernelGetNativeHandle(pi_kernel, pi_native_handle *)
Gets the native handle of a PI kernel object.
Definition: pi_esimd_emulator.cpp:1394
piEnqueueMemImageFill
pi_result piEnqueueMemImageFill(pi_queue, pi_mem, const void *, const size_t *, const size_t *, pi_uint32, const pi_event *, pi_event *)
Definition: pi_esimd_emulator.cpp:1317
sycl_get_cm_buffer_params_ptr
void(* sycl_get_cm_buffer_params_ptr)(void *, char **, uint32_t *, std::mutex **)
Definition: esimd_emulator_functions_v1.h:47
pi_int32
int32_t pi_int32
Definition: pi.h:67
_pi_buffer::CmBufferPtr
cm_support::CmBuffer * CmBufferPtr
Definition: pi_esimd_emulator.hpp:125
_pi_context
PI context mapping to a CUDA context object.
Definition: pi_cuda.hpp:148
piEventRetain
pi_result piEventRetain(pi_event Event)
Definition: pi_esimd_emulator.cpp:1112
_pi_device
PI device mapping to a CUdevice.
Definition: pi_cuda.hpp:71
PI_DEVICE_INFO_PARTITION_PROPERTIES
@ PI_DEVICE_INFO_PARTITION_PROPERTIES
Definition: pi.h:273
_pi_kernel_sub_group_info
_pi_kernel_sub_group_info
Definition: pi.h:383
PI_PLATFORM_INFO_EXTENSIONS
@ PI_PLATFORM_INFO_EXTENSIONS
Definition: pi.h:130
PI_DEVICE_INFO_SUB_GROUP_INDEPENDENT_FORWARD_PROGRESS
@ PI_DEVICE_INFO_SUB_GROUP_INDEPENDENT_FORWARD_PROGRESS
Definition: pi.h:280
_pi_buffer::Size
size_t Size
Definition: pi_esimd_emulator.hpp:126