Skip to content

Commit 4dca823

Browse files
aelovikov-intelaelovikov-intel
authored
[SYCL] Remove ESIMD Emulator (#13295)
1 parent f5178eb commit 4dca823

20 files changed

+18
-172
lines changed

‎sycl/doc/GetStartedGuide.md

Lines changed: 0 additions & 27 deletions
Original file line numberDiff line numberDiff line change
@@ -12,7 +12,6 @@ and a wide range of compute accelerators such as GPU and FPGA.
1212
* [Build DPC++ toolchain with support for NVIDIA CUDA](#build-dpc-toolchain-with-support-for-nvidia-cuda)
1313
* [Build DPC++ toolchain with support for HIP AMD](#build-dpc-toolchain-with-support-for-hip-amd)
1414
* [Build DPC++ toolchain with support for HIP NVIDIA](#build-dpc-toolchain-with-support-for-hip-nvidia)
15-
* [Build DPC++ toolchain with support for ESIMD CPU Emulation](#build-dpc-toolchain-with-support-for-esimd-cpu-emulation)
1615
* [Build DPC++ toolchain with support for runtime kernel fusion](#build-dpc-toolchain-with-support-for-runtime-kernel-fusion)
1716
* [Build DPC++ toolchain with a custom Unified Runtime](#build-dpc-toolchain-with-a-custom-unified-runtime)
1817
* [Build Doxygen documentation](#build-doxygen-documentation)
@@ -302,32 +301,6 @@ as well as the CUDA Runtime API to be installed, see
302301
Currently, this has only been tried on Linux, with ROCm 4.2.0 or 4.3.0, with
303302
CUDA 11, and using a GeForce 1060 device.
304303

305-
### Build DPC++ toolchain with support for ESIMD CPU Emulation
306-
307-
There is experimental support for DPC++ for using ESIMD CPU Emulation.
308-
309-
This feature supports ESIMD CPU Emulation using CM_EMU library
310-
[CM Emulation project](https://github.com/intel/cm-cpu-emulation). The library
311-
package will be generated from source codes downloaded from its open source
312-
project and installed in your deploy directory during toolchain build.
313-
314-
To enable support for ESIMD CPU emulation, follow the instructions for the Linux
315-
DPC++ toolchain, but add the `--enable-esimd-emulator`.
316-
317-
Enabling this flag requires following packages installed.
318-
319-
* Ubuntu 22.04
320-
* libva-dev / 2.7.0-2
321-
* libffi-dev / 3.3-4
322-
* libtool
323-
* RHEL 8.\*
324-
* libffi
325-
* libffi-devel
326-
* libva
327-
* libva-devel
328-
329-
Currently, this feature was tested and verified on Ubuntu 22.04 environment.
330-
331304
### Build DPC++ toolchain with support for runtime kernel fusion
332305

333306
Support for the experimental SYCL extension for user-driven kernel fusion at

‎sycl/include/sycl/backend_types.hpp

Lines changed: 2 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -21,8 +21,8 @@ enum class backend : char {
2121
ext_oneapi_level_zero = 2,
2222
ext_oneapi_cuda = 3,
2323
all = 4,
24-
ext_intel_esimd_emulator __SYCL_DEPRECATED(
25-
"esimd emulator is no longer supported") = 5,
24+
// No support anymore:
25+
// ext_intel_esimd_emulator = 5,
2626
ext_oneapi_hip = 6,
2727
ext_oneapi_native_cpu = 7,
2828
};
@@ -50,9 +50,6 @@ inline std::ostream &operator<<(std::ostream &Out, backend be) {
5050
case backend::ext_oneapi_cuda:
5151
Out << "ext_oneapi_cuda";
5252
break;
53-
case backend::ext_intel_esimd_emulator:
54-
Out << "ext_intel_esimd_emulator";
55-
break;
5653
case backend::ext_oneapi_hip:
5754
Out << "ext_oneapi_hip";
5855
break;
@@ -76,8 +73,6 @@ inline std::string_view get_backend_name_no_vendor(backend Backend) {
7673
return "level_zero";
7774
case backend::ext_oneapi_cuda:
7875
return "cuda";
79-
case backend::ext_intel_esimd_emulator:
80-
return "esimd_emulator";
8176
case backend::ext_oneapi_hip:
8277
return "hip";
8378
case backend::ext_oneapi_native_cpu:

‎sycl/include/sycl/detail/pi.h

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -312,7 +312,8 @@ typedef enum {
312312
PI_EXT_PLATFORM_BACKEND_OPENCL = 2, ///< The backend is OpenCL
313313
PI_EXT_PLATFORM_BACKEND_CUDA = 3, ///< The backend is CUDA
314314
PI_EXT_PLATFORM_BACKEND_HIP = 4, ///< The backend is HIP
315-
PI_EXT_PLATFORM_BACKEND_ESIMD = 5, ///< The backend is ESIMD
315+
// Not supported anymore:
316+
// PI_EXT_PLATFORM_BACKEND_ESIMD = 5,
316317
PI_EXT_PLATFORM_BACKEND_NATIVE_CPU = 6, ///< The backend is NATIVE_CPU
317318
} _pi_platform_backend;
318319

‎sycl/include/sycl/detail/pi.hpp

Lines changed: 0 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -69,15 +69,13 @@ bool trace(TraceLevel level);
6969
#define __SYCL_OPENCL_PLUGIN_NAME "pi_opencl.dll"
7070
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME "pi_level_zero.dll"
7171
#define __SYCL_CUDA_PLUGIN_NAME "pi_cuda.dll"
72-
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME "pi_esimd_emulator.dll"
7372
#define __SYCL_HIP_PLUGIN_NAME "pi_hip.dll"
7473
#define __SYCL_UR_PLUGIN_NAME "pi_unified_runtime.dll"
7574
#define __SYCL_NATIVE_CPU_PLUGIN_NAME "pi_native_cpu.dll"
7675
#else
7776
#define __SYCL_OPENCL_PLUGIN_NAME "libpi_opencl.dll"
7877
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME "libpi_level_zero.dll"
7978
#define __SYCL_CUDA_PLUGIN_NAME "libpi_cuda.dll"
80-
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME "libpi_esimd_emulator.dll"
8179
#define __SYCL_HIP_PLUGIN_NAME "libpi_hip.dll"
8280
#define __SYCL_UR_PLUGIN_NAME "libpi_unified_runtime.dll"
8381
#define __SYCL_NATIVE_CPU_PLUGIN_NAME "libpi_native_cpu.dll"
@@ -86,15 +84,13 @@ bool trace(TraceLevel level);
8684
#define __SYCL_OPENCL_PLUGIN_NAME "libpi_opencl.so"
8785
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME "libpi_level_zero.so"
8886
#define __SYCL_CUDA_PLUGIN_NAME "libpi_cuda.so"
89-
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME "libpi_esimd_emulator.so"
9087
#define __SYCL_HIP_PLUGIN_NAME "libpi_hip.so"
9188
#define __SYCL_UR_PLUGIN_NAME "libpi_unified_runtime.so"
9289
#define __SYCL_NATIVE_CPU_PLUGIN_NAME "libpi_native_cpu.so"
9390
#elif defined(__SYCL_RT_OS_DARWIN)
9491
#define __SYCL_OPENCL_PLUGIN_NAME "libpi_opencl.dylib"
9592
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME "libpi_level_zero.dylib"
9693
#define __SYCL_CUDA_PLUGIN_NAME "libpi_cuda.dylib"
97-
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME "libpi_esimd_emulator.dylib"
9894
#define __SYCL_HIP_PLUGIN_NAME "libpi_hip.dylib"
9995
#define __SYCL_UR_PLUGIN_NAME "libpi_unified_runtime.dylib"
10096
#define __SYCL_NATIVE_CPU_PLUGIN_NAME "libpi_native_cpu.dylib"

‎sycl/pi_win_proxy_loader/pi_win_proxy_loader.cpp

Lines changed: 0 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -89,15 +89,13 @@ std::wstring getCurrentDSODir() {
8989
#define __SYCL_OPENCL_PLUGIN_NAME "pi_opencl.dll"
9090
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME "pi_level_zero.dll"
9191
#define __SYCL_CUDA_PLUGIN_NAME "pi_cuda.dll"
92-
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME "pi_esimd_emulator.dll"
9392
#define __SYCL_HIP_PLUGIN_NAME "pi_hip.dll"
9493
#define __SYCL_UNIFIED_RUNTIME_PLUGIN_NAME "pi_unified_runtime.dll"
9594
#define __SYCL_NATIVE_CPU_PLUGIN_NAME "pi_native_cpu.dll"
9695
#else // llvm-mingw
9796
#define __SYCL_OPENCL_PLUGIN_NAME "libpi_opencl.dll"
9897
#define __SYCL_LEVEL_ZERO_PLUGIN_NAME "libpi_level_zero.dll"
9998
#define __SYCL_CUDA_PLUGIN_NAME "libpi_cuda.dll"
100-
#define __SYCL_ESIMD_EMULATOR_PLUGIN_NAME "libpi_esimd_emulator.dll"
10199
#define __SYCL_HIP_PLUGIN_NAME "libpi_hip.dll"
102100
#define __SYCL_UNIFIED_RUNTIME_PLUGIN_NAME "libpi_unified_runtime.dll"
103101
#define __SYCL_NATIVE_CPU_PLUGIN_NAME "libpi_native_cpu.dll"
@@ -147,7 +145,6 @@ void preloadLibraries() {
147145
loadPlugin(__SYCL_OPENCL_PLUGIN_NAME);
148146
loadPlugin(__SYCL_LEVEL_ZERO_PLUGIN_NAME);
149147
loadPlugin(__SYCL_CUDA_PLUGIN_NAME);
150-
loadPlugin(__SYCL_ESIMD_EMULATOR_PLUGIN_NAME);
151148
loadPlugin(__SYCL_HIP_PLUGIN_NAME);
152149
loadPlugin(__SYCL_UNIFIED_RUNTIME_PLUGIN_NAME);
153150
loadPlugin(__SYCL_NATIVE_CPU_PLUGIN_NAME);

‎sycl/source/CMakeLists.txt

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -261,7 +261,6 @@ set(SYCL_COMMON_SOURCES
261261
"sampler.cpp"
262262
"stream.cpp"
263263
"spirv_ops.cpp"
264-
"esimd_emulator_device_interface.cpp"
265264
"$<$<PLATFORM_ID:Windows>:detail/windows_pi.cpp>"
266265
"$<$<OR:$<PLATFORM_ID:Linux>,$<PLATFORM_ID:Darwin>>:detail/posix_pi.cpp>"
267266
)

‎sycl/source/backend.cpp

Lines changed: 0 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -58,8 +58,6 @@ backend convertBackend(pi_platform_backend PiBackend) {
5858
return backend::ext_oneapi_cuda;
5959
case PI_EXT_PLATFORM_BACKEND_HIP:
6060
return backend::ext_oneapi_hip;
61-
case PI_EXT_PLATFORM_BACKEND_ESIMD:
62-
return backend::ext_intel_esimd_emulator;
6361
case PI_EXT_PLATFORM_BACKEND_NATIVE_CPU:
6462
return backend::ext_oneapi_native_cpu;
6563
}

‎sycl/source/detail/config.cpp

Lines changed: 2 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -163,17 +163,15 @@ void dumpConfig() {
163163

164164
// Array is used by SYCL_DEVICE_FILTER and SYCL_DEVICE_ALLOWLIST and
165165
// ONEAPI_DEVICE_SELECTOR
166-
// TODO: Remove esimd_emulator in the next ABI breaking window.
167166
// TODO: host device type will be removed once sycl_ext_oneapi_filter_selector
168167
// is removed.
169-
const std::array<std::pair<std::string, backend>, 8> &getSyclBeMap() {
170-
static const std::array<std::pair<std::string, backend>, 8> SyclBeMap = {
168+
const std::array<std::pair<std::string, backend>, 7> &getSyclBeMap() {
169+
static const std::array<std::pair<std::string, backend>, 7> SyclBeMap = {
171170
{{"host", backend::host},
172171
{"opencl", backend::opencl},
173172
{"level_zero", backend::ext_oneapi_level_zero},
174173
{"cuda", backend::ext_oneapi_cuda},
175174
{"hip", backend::ext_oneapi_hip},
176-
{"esimd_emulator", backend::ext_intel_esimd_emulator},
177175
{"native_cpu", backend::ext_oneapi_native_cpu},
178176
{"*", backend::all}}};
179177
return SyclBeMap;

‎sycl/source/detail/config.hpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -251,7 +251,7 @@ getSyclDeviceTypeMap() {
251251

252252
// Array is used by SYCL_DEVICE_FILTER and SYCL_DEVICE_ALLOWLIST and
253253
// ONEAPI_DEVICE_SELECTOR
254-
const std::array<std::pair<std::string, backend>, 8> &getSyclBeMap();
254+
const std::array<std::pair<std::string, backend>, 7> &getSyclBeMap();
255255

256256
// ---------------------------------------
257257
// ONEAPI_DEVICE_SELECTOR support

‎sycl/source/detail/device_impl.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -363,7 +363,7 @@ bool device_impl::has(aspect Aspect) const {
363363
return is_accelerator();
364364
case aspect::custom:
365365
return false;
366-
// TODO: Implement this for FPGA and ESIMD emulators.
366+
// TODO: Implement this for FPGA emulator.
367367
case aspect::emulated:
368368
return false;
369369
case aspect::host_debuggable:

‎sycl/source/detail/pi.cpp

Lines changed: 0 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -68,9 +68,6 @@ template <sycl::backend BE> void *getPluginOpaqueData(void *OpaqueDataParam) {
6868
return ReturnOpaqueData;
6969
}
7070

71-
template __SYCL_EXPORT void *
72-
getPluginOpaqueData<sycl::backend::ext_intel_esimd_emulator>(void *);
73-
7471
namespace pi {
7572

7673
static void initializePlugins(std::vector<PluginPtr> &Plugins);
@@ -514,8 +511,6 @@ template <backend BE> const PluginPtr &getPlugin() {
514511
template __SYCL_EXPORT const PluginPtr &getPlugin<backend::opencl>();
515512
template __SYCL_EXPORT const PluginPtr &
516513
getPlugin<backend::ext_oneapi_level_zero>();
517-
template __SYCL_EXPORT const PluginPtr &
518-
getPlugin<backend::ext_intel_esimd_emulator>();
519514
template __SYCL_EXPORT const PluginPtr &getPlugin<backend::ext_oneapi_cuda>();
520515
template __SYCL_EXPORT const PluginPtr &getPlugin<backend::ext_oneapi_hip>();
521516

‎sycl/source/detail/scheduler/commands.cpp

Lines changed: 3 additions & 16 deletions
Original file line numberDiff line numberDiff line change
@@ -2992,8 +2992,7 @@ pi_int32 ExecCGCommand::enqueueImpQueue() {
29922992
NDRDescT &NDRDesc = ExecKernel->MNDRDesc;
29932993
std::vector<ArgDesc> &Args = ExecKernel->MArgs;
29942994

2995-
if (MQueue->is_host() || (MQueue->getDeviceImplPtr()->getBackend() ==
2996-
backend::ext_intel_esimd_emulator)) {
2995+
if (MQueue->is_host()) {
29972996
for (ArgDesc &Arg : Args)
29982997
if (kernel_param_kind_t::kind_accessor == Arg.MType) {
29992998
Requirement *Req = (Requirement *)(Arg.MPtr);
@@ -3006,20 +3005,8 @@ pi_int32 ExecCGCommand::enqueueImpQueue() {
30063005
Plugin->call<PiApiKind::piEventsWait>(RawEvents.size(), &RawEvents[0]);
30073006
}
30083007

3009-
if (MQueue->is_host()) {
3010-
ExecKernel->MHostKernel->call(NDRDesc,
3011-
getEvent()->getHostProfilingInfo());
3012-
} else {
3013-
assert(MQueue->getDeviceImplPtr()->getBackend() ==
3014-
backend::ext_intel_esimd_emulator);
3015-
if (MEvent != nullptr)
3016-
MEvent->setHostEnqueueTime();
3017-
MQueue->getPlugin()->call<PiApiKind::piEnqueueKernelLaunch>(
3018-
nullptr,
3019-
reinterpret_cast<pi_kernel>(ExecKernel->MHostKernel->getPtr()),
3020-
NDRDesc.Dims, &NDRDesc.GlobalOffset[0], &NDRDesc.GlobalSize[0],
3021-
&NDRDesc.LocalSize[0], 0, nullptr, nullptr);
3022-
}
3008+
ExecKernel->MHostKernel->call(NDRDesc,
3009+
getEvent()->getHostProfilingInfo());
30233010
return PI_SUCCESS;
30243011
}
30253012

‎sycl/source/device_selector.cpp

Lines changed: 0 additions & 11 deletions
Original file line numberDiff line numberDiff line change
@@ -179,13 +179,6 @@ __SYCL_EXPORT int default_selector_v(const device &dev) {
179179
// The default selector doesn't reject any devices.
180180
int Score = 0;
181181

182-
// we give the esimd_emulator device a score of zero to prevent it from being
183-
// chosen among other devices. The same thing is done for gpu_selector_v
184-
// below.
185-
if (dev.get_backend() == backend::ext_intel_esimd_emulator) {
186-
return 0;
187-
}
188-
189182
traceDeviceSelector("info::device_type::automatic");
190183

191184
if (dev.is_gpu())
@@ -209,10 +202,6 @@ __SYCL_EXPORT int default_selector_v(const device &dev) {
209202
__SYCL_EXPORT int gpu_selector_v(const device &dev) {
210203
int Score = detail::REJECT_DEVICE_SCORE;
211204

212-
if (dev.get_backend() == backend::ext_intel_esimd_emulator) {
213-
return 0;
214-
}
215-
216205
traceDeviceSelector("info::device_type::gpu");
217206
if (dev.is_gpu()) {
218207
Score = 1000;

‎sycl/source/esimd_emulator_device_interface.cpp

Lines changed: 0 additions & 38 deletions
This file was deleted.

‎sycl/source/handler.cpp

Lines changed: 4 additions & 34 deletions
Original file line numberDiff line numberDiff line change
@@ -279,40 +279,10 @@ event handler::finalize() {
279279
: nullptr);
280280
Result = PI_SUCCESS;
281281
} else {
282-
if (MQueue->getDeviceImplPtr()->getBackend() ==
283-
backend::ext_intel_esimd_emulator) {
284-
// Capture the host timestamp for profiling (queue time)
285-
if (NewEvent != nullptr)
286-
NewEvent->setHostEnqueueTime();
287-
[&](auto... Args) {
288-
if (MImpl->MKernelIsCooperative) {
289-
MQueue->getPlugin()
290-
->call<
291-
detail::PiApiKind::piextEnqueueCooperativeKernelLaunch>(
292-
Args...);
293-
} else {
294-
MQueue->getPlugin()
295-
->call<detail::PiApiKind::piEnqueueKernelLaunch>(Args...);
296-
}
297-
}(/* queue */
298-
nullptr,
299-
/* kernel */
300-
reinterpret_cast<pi_kernel>(MHostKernel->getPtr()),
301-
/* work_dim */
302-
MNDRDesc.Dims,
303-
/* global_work_offset */ &MNDRDesc.GlobalOffset[0],
304-
/* global_work_size */ &MNDRDesc.GlobalSize[0],
305-
/* local_work_size */ &MNDRDesc.LocalSize[0],
306-
/* num_events_in_wait_list */ 0,
307-
/* event_wait_list */ nullptr,
308-
/* event */ nullptr);
309-
Result = PI_SUCCESS;
310-
} else {
311-
Result = enqueueImpKernel(
312-
MQueue, MNDRDesc, MArgs, KernelBundleImpPtr, MKernel,
313-
MKernelName.c_str(), RawEvents, NewEvent, nullptr,
314-
MImpl->MKernelCacheConfig, MImpl->MKernelIsCooperative);
315-
}
282+
Result = enqueueImpKernel(
283+
MQueue, MNDRDesc, MArgs, KernelBundleImpPtr, MKernel,
284+
MKernelName.c_str(), RawEvents, NewEvent, nullptr,
285+
MImpl->MKernelCacheConfig, MImpl->MKernelIsCooperative);
316286
}
317287
#ifdef XPTI_ENABLE_INSTRUMENTATION
318288
// Emit signal only when event is created

‎sycl/source/kernel_bundle.cpp

Lines changed: 0 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -312,11 +312,6 @@ bool is_compatible(const std::vector<kernel_id> &KernelIDs, const device &Dev) {
312312
const detail::RTDeviceBinaryImage &Img) {
313313
const char *Target = Img.getRawData().DeviceTargetSpec;
314314
auto BE = Dev.get_backend();
315-
// ESIMD emulator backend is only compatible with esimd kernels.
316-
if (BE == sycl::backend::ext_intel_esimd_emulator) {
317-
pi_device_binary_property Prop = Img.getProperty("isEsimdImage");
318-
return (Prop && (detail::DeviceBinaryProperty(Prop).asUint32() != 0));
319-
}
320315
if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64) == 0) {
321316
return (BE == sycl::backend::opencl ||
322317
BE == sycl::backend::ext_oneapi_level_zero);

‎sycl/test/abi/sycl_symbols_linux.dump

Lines changed: 0 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -3359,7 +3359,6 @@ _ZN4sycl3_V16detail18stringifyErrorCodeEi
33593359
_ZN4sycl3_V16detail19convertChannelOrderE23_pi_image_channel_order
33603360
_ZN4sycl3_V16detail19convertChannelOrderENS0_19image_channel_orderE
33613361
_ZN4sycl3_V16detail19getImageElementSizeEhNS0_18image_channel_typeE
3362-
_ZN4sycl3_V16detail19getPluginOpaqueDataILNS0_7backendE5EEEPvS4_
33633362
_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_get_kernelERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE
33643363
_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_has_kernelERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE
33653364
_ZN4sycl3_V16detail19kernel_bundle_plain32set_specialization_constant_implEPKcPvm
@@ -3384,7 +3383,6 @@ _ZN4sycl3_V16detail22has_kernel_bundle_implERKNS0_7contextERKSt6vectorINS0_6devi
33843383
_ZN4sycl3_V16detail22reduGetPreferredWGSizeERSt10shared_ptrINS1_10queue_implEEm
33853384
_ZN4sycl3_V16detail22removeDuplicateDevicesERKSt6vectorINS0_6deviceESaIS3_EE
33863385
_ZN4sycl3_V16detail23constructorNotificationEPvS2_NS0_6access6targetENS3_4modeERKNS1_13code_locationE
3387-
_ZN4sycl3_V16detail23getESIMDDeviceInterfaceEv
33883386
_ZN4sycl3_V16detail24find_device_intersectionERKSt6vectorINS0_13kernel_bundleILNS0_12bundle_stateE1EEESaIS5_EE
33893387
_ZN4sycl3_V16detail26isDeviceGlobalUsedInKernelEPKv
33903388
_ZN4sycl3_V16detail27getPixelCoordLinearFiltModeENS0_3vecIfLi4EEENS0_15addressing_modeENS0_5rangeILi3EEERS3_
@@ -3400,7 +3398,6 @@ _ZN4sycl3_V16detail2pi9assertionEbPKc
34003398
_ZN4sycl3_V16detail2pi9getPluginILNS0_7backendE1EEERKSt10shared_ptrINS1_6pluginEEv
34013399
_ZN4sycl3_V16detail2pi9getPluginILNS0_7backendE2EEERKSt10shared_ptrINS1_6pluginEEv
34023400
_ZN4sycl3_V16detail2pi9getPluginILNS0_7backendE3EEERKSt10shared_ptrINS1_6pluginEEv
3403-
_ZN4sycl3_V16detail2pi9getPluginILNS0_7backendE5EEERKSt10shared_ptrINS1_6pluginEEv
34043401
_ZN4sycl3_V16detail2pi9getPluginILNS0_7backendE6EEERKSt10shared_ptrINS1_6pluginEEv
34053402
_ZN4sycl3_V16detail30UnsampledImageAccessorBaseHost10getAccDataEv
34063403
_ZN4sycl3_V16detail30UnsampledImageAccessorBaseHost6getPtrEv

0 commit comments

Comments
 (0)