From f37a294c878bea15d059fc1c2ad88f1a0fab52be Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Mon, 25 May 2020 14:52:37 +0300 Subject: [PATCH 01/10] [SYCL] Introduce interop handle for host task Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 117 ++++++++++++++++++++++++ sycl/source/CMakeLists.txt | 1 + sycl/source/interop_handle.cpp | 28 ++++++ 3 files changed, 146 insertions(+) create mode 100644 sycl/include/CL/sycl/interop_handle.hpp create mode 100644 sycl/source/interop_handle.cpp diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp new file mode 100644 index 0000000000000..ba8704aa25b41 --- /dev/null +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -0,0 +1,117 @@ +//==------------ interop_handle.hpp --- SYCL interop handle ----------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +#pragma once + +#include +#include +#include +#include +#include + +#include + +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { + +namespace detail { +class AccessorBaseHost; +class ExecCGCommand; +class DispatchHostTask; +} // namespace detail + +template +class accessor; + +class interop_handle { +public: + /// Receives a SYCL accessor that has been defined is a requirement for the + /// command group, and returns the underlying OpenCL memory object that is + /// used by the SYCL runtime. If the accessor passed as parameter is not part + /// of the command group requirements (e.g. it is an unregistered placeholder + /// accessor), the exception `cl::sycl::invalid_object` is thrown + /// asynchronously. + template + typename std::enable_if::type + get_native_mem(const accessor &Acc) const { +#ifndef __SYCL_DEVICE_ONLY__ + // employ reinterpret_cast instead of static_cast due to cycle in includes + // involving CL/sycl/accessor.hpp + auto *AccBase = const_cast( + reinterpret_cast(&Acc)); + return getMemImpl(detail::getSyclObjImpl(*AccBase).get()); +#else + (void)Acc; + // we believe this won't be ever called on device side + return static_cast(0x0); +#endif + } + + template + typename std::enable_if::type + get_native_mem(const accessor &) const { + throw invalid_object_error("Getting memory object out of host accessor is " + "not allowed", + PI_INVALID_MEM_OBJECT); + } + + /// Returns an underlying OpenCL queue for the SYCL queue used to submit the + /// command group, or the fallback queue if this command-group is re-trying + /// execution on an OpenCL queue. The OpenCL command queue returned is + /// implementation-defined in cases where the SYCL queue maps to multiple + /// underlying OpenCL objects. It is responsibility of the SYCL runtime to + /// ensure the OpenCL queue returned is in a state that can be used to + /// dispatch work, and that other potential OpenCL command queues associated + /// with the same SYCL command queue are not executing commands while the host + /// task is executing. + cl_command_queue get_native_queue() const noexcept { return MQueue; } + + /// Returns an underlying OpenCL device associated with the SYCL queue used + /// to submit the command group, or the fallback queue if this command-group + /// is re-trying execution on an OpenCL queue. + cl_device_id get_native_device() const noexcept { return MDeviceId; } + + /// Returns an underlying OpenCL context associated with the SYCL queue used + /// to submit the command group, or the fallback queue if this command-group + /// is re-trying execution on an OpenCL queue. + cl_context get_native_context() const noexcept { return MContext; } + +private: + using ReqToMem = std::pair; + + template + friend class accessor; + friend class detail::ExecCGCommand; + friend class detail::DispatchHostTask; + +public: + // TODO set c-tor private + interop_handle(std::vector MemObjs, cl_command_queue Queue, + cl_device_id DeviceId, cl_context Context) + : MQueue(Queue), MDeviceId(DeviceId), MContext(Context), + MMemObjs(std::move(MemObjs)) {} + +private: + cl_mem getMemImpl(detail::Requirement *Req) const; + + cl_command_queue MQueue; + cl_device_id MDeviceId; + cl_context MContext; + std::vector MMemObjs; +}; + +} // namespace sycl +} // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index 17ff24db14b00..95501d4a616ab 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -150,6 +150,7 @@ set(SYCL_SOURCES "sampler.cpp" "stream.cpp" "spirv_ops.cpp" + "interop_handle.cpp" "$<$:detail/windows_pi.cpp>" "$<$,$>:detail/posix_pi.cpp>" ) diff --git a/sycl/source/interop_handle.cpp b/sycl/source/interop_handle.cpp new file mode 100644 index 0000000000000..c1df4993c700b --- /dev/null +++ b/sycl/source/interop_handle.cpp @@ -0,0 +1,28 @@ +//==------------ interop_handle.cpp --- SYCL interop handle ----------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +#include +#include + +#include + +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { + +cl_mem interop_handle::getMemImpl(detail::Requirement *Req) const { + auto Iter = std::find_if(std::begin(MMemObjs), std::end(MMemObjs), + [=](ReqToMem Elem) { return (Elem.first == Req); }); + + if (Iter == std::end(MMemObjs)) + throw("Invalid memory object used inside interop"); + + return detail::pi::cast(Iter->second); +} + +} // namespace sycl +} // __SYCL_INLINE_NAMESPACE(cl) From 26fe5e526448a176c98ef9582d74dfb26e02c7e6 Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Wed, 27 May 2020 11:51:58 +0300 Subject: [PATCH 02/10] [SYCL] Remove friend classes as unneeded for now. Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 6 ------ 1 file changed, 6 deletions(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index ba8704aa25b41..a30594ee0e359 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -91,12 +91,6 @@ class interop_handle { private: using ReqToMem = std::pair; - template - friend class accessor; - friend class detail::ExecCGCommand; - friend class detail::DispatchHostTask; - public: // TODO set c-tor private interop_handle(std::vector MemObjs, cl_command_queue Queue, From e4a35f89c13ce39738e283a122fe0faf5c22c5a8 Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Wed, 27 May 2020 14:46:40 +0300 Subject: [PATCH 03/10] [SYCL] Untie interop_handle from OpenCL. Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 90 ++++++++++++++++++------- sycl/source/interop_handle.cpp | 26 ++++++- 2 files changed, 89 insertions(+), 27 deletions(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index a30594ee0e359..f9084f5d1d9c3 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -8,11 +8,13 @@ #pragma once +#include #include #include #include #include #include +#include #include @@ -23,6 +25,9 @@ namespace detail { class AccessorBaseHost; class ExecCGCommand; class DispatchHostTask; +class queue_impl; +class device_impl; +class context_impl; } // namespace detail template - typename std::enable_if::type - get_native_mem(const accessor &Acc) const { + template + typename std::enable_if< + Target != access::target::host_buffer, + typename interop>::type>::type + get_native_mem(const accessor &Acc) const { #ifndef __SYCL_DEVICE_ONLY__ // employ reinterpret_cast instead of static_cast due to cycle in includes // involving CL/sycl/accessor.hpp auto *AccBase = const_cast( reinterpret_cast(&Acc)); - return getMemImpl(detail::getSyclObjImpl(*AccBase).get()); + return getMemImpl( + detail::getSyclObjImpl(*AccBase).get()); #else (void)Acc; // we believe this won't be ever called on device side - return static_cast(0x0); + return nullptr; #endif } - template - typename std::enable_if::type - get_native_mem(const accessor &) const { + template + typename std::enable_if< + Target == access::target::host_buffer, + typename interop>::type>::type + get_native_mem(const accessor &Acc) const { throw invalid_object_error("Getting memory object out of host accessor is " "not allowed", PI_INVALID_MEM_OBJECT); @@ -76,34 +84,68 @@ class interop_handle { /// dispatch work, and that other potential OpenCL command queues associated /// with the same SYCL command queue are not executing commands while the host /// task is executing. - cl_command_queue get_native_queue() const noexcept { return MQueue; } + template + auto + get_native_queue() const noexcept -> + typename interop::type { + return reinterpret_cast::type>( + getNativeQueue()); + } /// Returns an underlying OpenCL device associated with the SYCL queue used /// to submit the command group, or the fallback queue if this command-group /// is re-trying execution on an OpenCL queue. - cl_device_id get_native_device() const noexcept { return MDeviceId; } + template + auto get_native_device() const noexcept -> + typename interop::type + { + return reinterpret_cast::type>( + getNativeDevice()); + } /// Returns an underlying OpenCL context associated with the SYCL queue used /// to submit the command group, or the fallback queue if this command-group /// is re-trying execution on an OpenCL queue. - cl_context get_native_context() const noexcept { return MContext; } + template + auto get_native_context() const noexcept -> + typename interop::type + { + return reinterpret_cast::type>( + getNativeContext()); + } private: using ReqToMem = std::pair; public: // TODO set c-tor private - interop_handle(std::vector MemObjs, cl_command_queue Queue, - cl_device_id DeviceId, cl_context Context) - : MQueue(Queue), MDeviceId(DeviceId), MContext(Context), + interop_handle(std::vector MemObjs, + const std::shared_ptr &Queue, + const std::shared_ptr &Device, + const std::shared_ptr &Context) + : MQueue(Queue), MDevice(Device), MContext(Context), MMemObjs(std::move(MemObjs)) {} private: - cl_mem getMemImpl(detail::Requirement *Req) const; + template + auto getMemImpl(detail::Requirement *Req) const -> typename interop< + BackendName, accessor>::type { + return reinterpret_cast< + typename interop>::type>( + getNativeMem(Req)); + } + + pi_native_handle getNativeMem(detail::Requirement *Req) const; + pi_native_handle getNativeQueue() const; + pi_native_handle getNativeDevice() const; + pi_native_handle getNativeContext() const; + + std::shared_ptr MQueue; + std::shared_ptr MDevice; + std::shared_ptr MContext; - cl_command_queue MQueue; - cl_device_id MDeviceId; - cl_context MContext; std::vector MMemObjs; }; diff --git a/sycl/source/interop_handle.cpp b/sycl/source/interop_handle.cpp index c1df4993c700b..2d6f235c949d0 100644 --- a/sycl/source/interop_handle.cpp +++ b/sycl/source/interop_handle.cpp @@ -8,20 +8,40 @@ #include #include +#include +#include +#include #include __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { -cl_mem interop_handle::getMemImpl(detail::Requirement *Req) const { +pi_native_handle interop_handle::getNativeMem(detail::Requirement *Req) const { auto Iter = std::find_if(std::begin(MMemObjs), std::end(MMemObjs), [=](ReqToMem Elem) { return (Elem.first == Req); }); - if (Iter == std::end(MMemObjs)) + if (Iter == std::end(MMemObjs)) { throw("Invalid memory object used inside interop"); + } - return detail::pi::cast(Iter->second); + auto Plugin = MQueue->getPlugin(); + pi_native_handle Handle; + Plugin.call(Iter->second, + &Handle); + return Handle; +} + +pi_native_handle interop_handle::getNativeDevice() const { + return MDevice->getNative(); +} + +pi_native_handle interop_handle::getNativeContext() const { + return MContext->getNative(); +} + +pi_native_handle interop_handle::getNativeQueue() const { + return MQueue->getNative(); } } // namespace sycl From 6d8170accb7dee8b70246448250d959877c3307c Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Wed, 27 May 2020 14:54:22 +0300 Subject: [PATCH 04/10] [SYCL] Fix style issues Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 39 ++++++++++++------------- 1 file changed, 18 insertions(+), 21 deletions(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index f9084f5d1d9c3..edcb38e6908b9 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -8,8 +8,8 @@ #pragma once -#include #include +#include #include #include #include @@ -46,8 +46,8 @@ class interop_handle { access::mode Mode, access::target Target, access::placeholder IsPlh> typename std::enable_if< Target != access::target::host_buffer, - typename interop>::type>::type + typename interop>::type>::type get_native_mem(const accessor &Acc) const { #ifndef __SYCL_DEVICE_ONLY__ // employ reinterpret_cast instead of static_cast due to cycle in includes @@ -67,8 +67,8 @@ class interop_handle { access::mode Mode, access::target Target, access::placeholder IsPlh> typename std::enable_if< Target == access::target::host_buffer, - typename interop>::type>::type + typename interop>::type>::type get_native_mem(const accessor &Acc) const { throw invalid_object_error("Getting memory object out of host accessor is " "not allowed", @@ -85,10 +85,9 @@ class interop_handle { /// with the same SYCL command queue are not executing commands while the host /// task is executing. template - auto - get_native_queue() const noexcept -> + auto get_native_queue() const noexcept -> typename interop::type { - return reinterpret_cast::type>( + return reinterpret_cast::type>( getNativeQueue()); } @@ -97,9 +96,8 @@ class interop_handle { /// is re-trying execution on an OpenCL queue. template auto get_native_device() const noexcept -> - typename interop::type - { - return reinterpret_cast::type>( + typename interop::type { + return reinterpret_cast::type>( getNativeDevice()); } @@ -108,9 +106,8 @@ class interop_handle { /// is re-trying execution on an OpenCL queue. template auto get_native_context() const noexcept -> - typename interop::type - { - return reinterpret_cast::type>( + typename interop::type { + return reinterpret_cast::type>( getNativeContext()); } @@ -127,13 +124,13 @@ class interop_handle { MMemObjs(std::move(MemObjs)) {} private: - template - auto getMemImpl(detail::Requirement *Req) const -> typename interop< - BackendName, accessor>::type { - return reinterpret_cast< - typename interop>::type>( + template + auto getMemImpl(detail::Requirement *Req) const -> + typename interop>::type { + return reinterpret_cast>::type>( getNativeMem(Req)); } From 5f6ea1f1b4f0779404c54db8fcd267ca6409125a Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Fri, 29 May 2020 19:29:08 +0300 Subject: [PATCH 05/10] [SYCL] Address comments Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 2 +- sycl/source/CMakeLists.txt | 2 +- sycl/source/interop_handle.cpp | 4 +++- 3 files changed, 5 insertions(+), 3 deletions(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index edcb38e6908b9..3522ab48c4341 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -36,7 +36,7 @@ class accessor; class interop_handle { public: - /// Receives a SYCL accessor that has been defined is a requirement for the + /// Receives a SYCL accessor that has been defined as a requirement for the /// command group, and returns the underlying OpenCL memory object that is /// used by the SYCL runtime. If the accessor passed as parameter is not part /// of the command group requirements (e.g. it is an unregistered placeholder diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index 95501d4a616ab..49ea418a41cfe 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -143,6 +143,7 @@ set(SYCL_SOURCES "function_pointer.cpp" "half_type.cpp" "handler.cpp" + "interop_handle.cpp" "kernel.cpp" "platform.cpp" "program.cpp" @@ -150,7 +151,6 @@ set(SYCL_SOURCES "sampler.cpp" "stream.cpp" "spirv_ops.cpp" - "interop_handle.cpp" "$<$:detail/windows_pi.cpp>" "$<$,$>:detail/posix_pi.cpp>" ) diff --git a/sycl/source/interop_handle.cpp b/sycl/source/interop_handle.cpp index 2d6f235c949d0..2f52601e7abdb 100644 --- a/sycl/source/interop_handle.cpp +++ b/sycl/source/interop_handle.cpp @@ -7,6 +7,7 @@ //===----------------------------------------------------------------------===// #include +#include #include #include #include @@ -22,7 +23,8 @@ pi_native_handle interop_handle::getNativeMem(detail::Requirement *Req) const { [=](ReqToMem Elem) { return (Elem.first == Req); }); if (Iter == std::end(MMemObjs)) { - throw("Invalid memory object used inside interop"); + throw invalid_object_error("Invalid memory object used inside interop", + PI_INVALID_MEM_OBJECT); } auto Plugin = MQueue->getPlugin(); From a24cee7ad3dc9b6cf183e9b5b24b850385ad30f3 Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Mon, 1 Jun 2020 14:14:42 +0300 Subject: [PATCH 06/10] [SYCL] Use forward declaration instead of include. Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/detail/cg.hpp | 1 + sycl/include/CL/sycl/interop_handle.hpp | 11 +++++++---- 2 files changed, 8 insertions(+), 4 deletions(-) diff --git a/sycl/include/CL/sycl/detail/cg.hpp b/sycl/include/CL/sycl/detail/cg.hpp index 2ef0ff5170b74..3e446b7d7de93 100644 --- a/sycl/include/CL/sycl/detail/cg.hpp +++ b/sycl/include/CL/sycl/detail/cg.hpp @@ -18,6 +18,7 @@ #include #include #include +#include #include #include #include diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index 3522ab48c4341..9748964aae19f 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -14,7 +14,6 @@ #include #include #include -#include #include @@ -34,6 +33,10 @@ template class accessor; +class queue; +class device; +class context; + class interop_handle { public: /// Receives a SYCL accessor that has been defined as a requirement for the @@ -52,8 +55,8 @@ class interop_handle { #ifndef __SYCL_DEVICE_ONLY__ // employ reinterpret_cast instead of static_cast due to cycle in includes // involving CL/sycl/accessor.hpp - auto *AccBase = const_cast( - reinterpret_cast(&Acc)); + const auto *AccBase = + reinterpret_cast(&Acc); return getMemImpl( detail::getSyclObjImpl(*AccBase).get()); #else @@ -69,7 +72,7 @@ class interop_handle { Target == access::target::host_buffer, typename interop>::type>::type - get_native_mem(const accessor &Acc) const { + get_native_mem(const accessor &) const { throw invalid_object_error("Getting memory object out of host accessor is " "not allowed", PI_INVALID_MEM_OBJECT); From cb0925bb8d9fcde49d3348487042f208d71394e1 Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Mon, 1 Jun 2020 14:32:03 +0300 Subject: [PATCH 07/10] [SYCL][TEST] Add accesor.hpp include to interop_handle.hpp Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 1 + 1 file changed, 1 insertion(+) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index 9748964aae19f..85532bfef79c6 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -8,6 +8,7 @@ #pragma once +#include #include #include #include From 6cd8d33a96de626fb608c7ae7af6511677e2d2d1 Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Mon, 1 Jun 2020 15:35:58 +0300 Subject: [PATCH 08/10] [SYCL] Remove tesing include. Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 1 - 1 file changed, 1 deletion(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index 85532bfef79c6..9748964aae19f 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -8,7 +8,6 @@ #pragma once -#include #include #include #include From 605ea4e0e33dbc49e581739a8bd71eaad4891faf Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Tue, 2 Jun 2020 14:12:19 +0300 Subject: [PATCH 09/10] [SYCL] Employ static_cast for accessor Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 9 ++------- 1 file changed, 2 insertions(+), 7 deletions(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index 9748964aae19f..10e7d72a22d49 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -9,6 +9,7 @@ #pragma once #include +#include #include #include #include @@ -29,10 +30,6 @@ class device_impl; class context_impl; } // namespace detail -template -class accessor; - class queue; class device; class context; @@ -53,10 +50,8 @@ class interop_handle { accessor>::type>::type get_native_mem(const accessor &Acc) const { #ifndef __SYCL_DEVICE_ONLY__ - // employ reinterpret_cast instead of static_cast due to cycle in includes - // involving CL/sycl/accessor.hpp const auto *AccBase = - reinterpret_cast(&Acc); + static_cast(&Acc); return getMemImpl( detail::getSyclObjImpl(*AccBase).get()); #else From add913a9318fda05a49c8e1b30511afddc4506f6 Mon Sep 17 00:00:00 2001 From: Sergey Kanaev Date: Tue, 2 Jun 2020 14:47:07 +0300 Subject: [PATCH 10/10] [SYCL] Fix style issue Signed-off-by: Sergey Kanaev --- sycl/include/CL/sycl/interop_handle.hpp | 3 +-- 1 file changed, 1 insertion(+), 2 deletions(-) diff --git a/sycl/include/CL/sycl/interop_handle.hpp b/sycl/include/CL/sycl/interop_handle.hpp index 10e7d72a22d49..3296ab783bebf 100644 --- a/sycl/include/CL/sycl/interop_handle.hpp +++ b/sycl/include/CL/sycl/interop_handle.hpp @@ -50,8 +50,7 @@ class interop_handle { accessor>::type>::type get_native_mem(const accessor &Acc) const { #ifndef __SYCL_DEVICE_ONLY__ - const auto *AccBase = - static_cast(&Acc); + const auto *AccBase = static_cast(&Acc); return getMemImpl( detail::getSyclObjImpl(*AccBase).get()); #else