From 140709e4f844429556c5fcf5c822dab9b13cf118 Mon Sep 17 00:00:00 2001 From: Alexey Bader Date: Thu, 21 Feb 2019 18:19:19 +0300 Subject: [PATCH] [SYCL] Work around build issues with GCC 7+. Signed-off-by: Alexey Bader --- sycl/include/CL/sycl/accessor.hpp | 46 +++++++++---------- .../CL/sycl/detail/scheduler/scheduler.cpp | 17 ++----- 2 files changed, 28 insertions(+), 35 deletions(-) diff --git a/sycl/include/CL/sycl/accessor.hpp b/sycl/include/CL/sycl/accessor.hpp index 5c1fff316dc5a..307574e7c1f65 100644 --- a/sycl/include/CL/sycl/accessor.hpp +++ b/sycl/include/CL/sycl/accessor.hpp @@ -67,8 +67,8 @@ class subscript_objData[getOffsetForId( - accRef.__impl()->Range, ids, accRef.__impl()->Offset)]; + return accRef.__get_impl()->Data[getOffsetForId( + accRef.__get_impl()->Range, ids, accRef.__get_impl()->Offset)]; } }; @@ -89,8 +89,8 @@ class subscript_obj::type operator[](size_t index) { ids[accessorDim - 1] = index; - return accRef.__impl()->Data[getOffsetForId( - accRef.__impl()->Range, ids, accRef.__impl()->Offset)]; + return accRef.__get_impl()->Data[getOffsetForId( + accRef.__get_impl()->Range, ids, accRef.__get_impl()->Offset)]; } }; @@ -307,11 +307,11 @@ class accessor_base { using _ImplT = accessor_impl; - const _ImplT *__impl() const { + const _ImplT *__get_impl() const { return reinterpret_cast(this); } - _ImplT *__impl() { return reinterpret_cast<_ImplT *>(this); } + _ImplT *__get_impl() { return reinterpret_cast<_ImplT *>(this); } static_assert( std::is_same::type, @@ -347,15 +347,15 @@ SYCL_ACCESSOR_SUBCLASS(accessor_common, accessor_base, true /* always */) { size_t get_size() const { return this->get_count() * sizeof(dataT); } // Returns the number of accessed elements. - size_t get_count() const { return this->__impl()->get_count(); } + size_t get_count() const { return this->__get_impl()->get_count(); } template typename std::enable_if<(Dimensions > 0), range>::type - get_range() const { return this->__impl()->Range; } + get_range() const { return this->__get_impl()->Range; } template typename std::enable_if<(Dimensions > 0), id>::type - get_offset() const { return this->__impl()->Offset; } + get_offset() const { return this->__get_impl()->Offset; } }; SYCL_ACCESSOR_SUBCLASS(accessor_opdata_w, accessor_common, @@ -365,7 +365,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_opdata_w, accessor_common, accessMode == access::mode::discard_read_write) && dimensions == 0) { operator dataT &() const { - return this->__impl()->Data[0]; + return this->__get_impl()->Data[0]; } }; @@ -376,7 +376,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_wn, accessor_opdata_w, accessMode == access::mode::discard_read_write) && dimensions > 0) { dataT &operator[](id index) const { - return this->__impl()->Data[getOffsetForId( + return this->__get_impl()->Data[getOffsetForId( this->get_range(), index, this->get_offset())]; } @@ -407,7 +407,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_w, accessor_subscript_wn, getOffsetForId(this->get_range(), index, this->get_offset())); } dataT &operator[](size_t index) const { - return this->__impl()->Data[index]; + return this->__get_impl()->Data[index]; } }; @@ -415,7 +415,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_opdata_r, accessor_subscript_w, accessMode == access::mode::read && dimensions == 0) { using PureType = typename detail::remove_AS::type; operator PureType() const { - return this->__impl()->Data[0]; + return this->__get_impl()->Data[0]; } }; @@ -423,7 +423,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_rn, accessor_opdata_r, accessMode == access::mode::read && dimensions > 0) { typename detail::remove_AS::type operator[](id index) const { - return this->__impl()->Data[getOffsetForId( + return this->__get_impl()->Data[getOffsetForId( this->get_range(), index, this->get_offset())]; } @@ -446,7 +446,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_r, accessor_subscript_rn, } typename detail::remove_AS::type operator[](size_t index) const { - return this->__impl()->Data[index]; + return this->__get_impl()->Data[index]; } }; @@ -468,7 +468,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_atomic_eq0, accessor_subscript_r, getAddressSpace::value; operator atomic() const { return atomic( - multi_ptr(&(this->__impl()->Data[0]))); + multi_ptr(&(this->__get_impl()->Data[0]))); } }; @@ -481,8 +481,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_atomic_gt0, getAddressSpace::value; atomic operator[](id index) const { return atomic( - multi_ptr(&(this->__impl()->Data[getOffsetForId( - this->__impl()->Range, index, this->__impl()->Offset)]))); + multi_ptr(&(this->__get_impl()->Data[getOffsetForId( + this->__get_impl()->Range, index, this->__get_impl()->Offset)]))); } }; @@ -495,7 +495,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_atomic_eq1, getAddressSpace::value; atomic operator[](size_t index) const { return atomic( - multi_ptr(&(this->__impl()->Data[index]))); + multi_ptr(&(this->__get_impl()->Data[index]))); } }; @@ -510,7 +510,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_pointer, accessor_subscript_atomic_eq1, true) { typename std::enable_if<(AccessTarget == access::target::host_buffer), dataT *>::type get_pointer() const { - return this->__impl()->Data; + return this->__get_impl()->Data; } /* Available only when: accessTarget == access::target::global_buffer */ template ::type, @@ -518,7 +518,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_pointer, accessor_subscript_atomic_eq1, true) { typename std::enable_if<(AccessTarget == access::target::global_buffer), global_ptr>::type get_pointer() const { - return global_ptr(this->__impl()->Data); + return global_ptr(this->__get_impl()->Data); } /* Available only when: accessTarget == access::target::constant_buffer */ template ::type, @@ -526,7 +526,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_pointer, accessor_subscript_atomic_eq1, true) { typename std::enable_if<(AccessTarget == access::target::constant_buffer), constant_ptr>::type get_pointer() const { - return constant_ptr(this->__impl()->Data); + return constant_ptr(this->__get_impl()->Data); } /* Available only when: accessTarget == access::target::local */ template ::type, @@ -534,7 +534,7 @@ SYCL_ACCESSOR_SUBCLASS(accessor_pointer, accessor_subscript_atomic_eq1, true) { typename std::enable_if<(AccessTarget == access::target::local), local_ptr>::type get_pointer() const { - return local_ptr(this->__impl()->Data); + return local_ptr(this->__get_impl()->Data); } }; diff --git a/sycl/include/CL/sycl/detail/scheduler/scheduler.cpp b/sycl/include/CL/sycl/detail/scheduler/scheduler.cpp index d2ee4676af3e8..f197412210657 100644 --- a/sycl/include/CL/sycl/detail/scheduler/scheduler.cpp +++ b/sycl/include/CL/sycl/detail/scheduler/scheduler.cpp @@ -59,9 +59,7 @@ void Node::addAccRequirement( accessor &&Acc, int argIndex) { detail::buffer_impl> *buf = - Acc.template accessor_base::__impl() - ->m_Buf; + Acc.__get_impl()->m_Buf; addBufRequirement(*buf); addInteropArg(nullptr, buf->get_size(), argIndex, getReqForBuffer(m_Bufs, *buf)); @@ -126,8 +124,7 @@ template void Node::addExplicitMemOp( accessor &Dest, T Src) { - auto *DestBase = Dest.template accessor_base::__impl(); + auto *DestBase = Dest.__get_impl(); assert(DestBase != nullptr && "Accessor should have an initialized accessor_base"); detail::buffer_impl> *Buf = DestBase->m_Buf; @@ -151,13 +148,10 @@ template Src, accessor Dest) { - auto *SrcBase = Src.template accessor_base::__impl(); + auto *SrcBase = Src.__get_impl(); assert(SrcBase != nullptr && "Accessor should have an initialized accessor_base"); - auto *DestBase = - Dest.template accessor_base::__impl(); + auto *DestBase = Dest.__get_impl(); assert(DestBase != nullptr && "Accessor should have an initialized accessor_base"); @@ -191,8 +185,7 @@ template &Acc, cl::sycl::event &Event) { - auto *AccBase = Acc.template accessor_base::__impl(); + auto *AccBase = Acc.__get_impl(); assert(AccBase != nullptr && "Accessor should have an initialized accessor_base"); detail::buffer_impl> *Buf = AccBase->m_Buf;