Skip to content

Commit 83e9287

Browse files
airliedairlied
committed
[SYCL] fix error building scheduler.cpp (v2.1)
On Fedora 28 with gcc version 8.2.1 20181105 (Red Hat 8.2.1-5) (GCC) building the SYCL stack gives: /home/airlied/devel/compute/intel/llvm/sycl/include/CL/sycl/detail/scheduler/scheduler.cpp: In member function ‘void cl::sycl::simple_scheduler::Node::addAccRequirement(cl::sycl::accessor<dataT, dimensions, accessMode, accessTarget, isPlaceholder>&&, int)’: /home/airlied/devel/compute/intel/llvm/sycl/include/CL/sycl/detail/scheduler/scheduler.cpp:66:48: error: expected ‘,’ or ‘;’ before ‘::’ token isPlaceholder>::__impl() As suggested Alexey Bader, remove the base class decleration. v2: add friend decls to all accessor subclasses, rename impl to avoid collision in the accessor class. v2.1: uggx impl is a secret detail inside SemaSYCL Signed-off-by: Dave Airlie <[email protected]> fix sched
1 parent 35cc013 commit 83e9287

File tree

3 files changed

+48
-31
lines changed

3 files changed

+48
-31
lines changed

‎clang/lib/Sema/SemaSYCL.cpp

Lines changed: 5 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -636,15 +636,15 @@ static void buildArgTys(ASTContext &Context, CXXRecordDecl *KernelObj,
636636

637637
CreateAndAddPrmDsc(Fld, PointerType);
638638

639-
FieldDecl *RangeFld = getFieldDeclByName(RecordDecl, {"__impl", "Range"});
639+
FieldDecl *RangeFld = getFieldDeclByName(RecordDecl, {"__implx", "Range"});
640640
assert(RangeFld &&
641641
"The accessor must contain the Range from the __impl field");
642642
CreateAndAddPrmDsc(RangeFld, RangeFld->getType());
643643

644644
FieldDecl *OffsetFld =
645-
getFieldDeclByName(RecordDecl, {"__impl", "Offset"});
645+
getFieldDeclByName(RecordDecl, {"__implx", "Offset"});
646646
assert(OffsetFld &&
647-
"The accessor must contain the Offset from the __impl field");
647+
"The accessor must contain the Offset from the __implx field");
648648
CreateAndAddPrmDsc(OffsetFld, OffsetFld->getType());
649649
} else if (Util::isSyclStreamType(ArgTy)) {
650650
// the parameter is a SYCL stream object
@@ -700,15 +700,15 @@ static void populateIntHeader(SYCLIntegrationHeader &H, const StringRef Name,
700700
getAccessTarget(AccTmplTy), Offset);
701701
// ... second descriptor (translated to range kernel parameter):
702702
FieldDecl *RngFld =
703-
getFieldDeclByName(AccTy, {"__impl", "Range"}, &Offset);
703+
getFieldDeclByName(AccTy, {"__implx", "Range"}, &Offset);
704704
uint64_t Sz = Ctx.getTypeSizeInChars(RngFld->getType()).getQuantity();
705705
H.addParamDesc(SYCLIntegrationHeader::kind_std_layout,
706706
static_cast<unsigned>(Sz), static_cast<unsigned>(Offset));
707707
// ... third descriptor (translated to id kernel parameter):
708708
// Get offset in bytes
709709
Offset = Layout.getFieldOffset(Fld->getFieldIndex()) / 8;
710710
FieldDecl *OffstFld =
711-
getFieldDeclByName(AccTy, {"__impl", "Offset"}, &Offset);
711+
getFieldDeclByName(AccTy, {"__implx", "Offset"}, &Offset);
712712
Sz = Ctx.getTypeSizeInChars(OffstFld->getType()).getQuantity();
713713
H.addParamDesc(SYCLIntegrationHeader::kind_std_layout,
714714
static_cast<unsigned>(Sz), static_cast<unsigned>(Offset));

‎sycl/include/CL/sycl/accessor.hpp

Lines changed: 37 additions & 14 deletions
Original file line numberDiff line numberDiff line change
@@ -355,6 +355,9 @@ SYCL_ACCESSOR_SUBCLASS(accessor_common, accessor_base, true /* always */) {
355355
template <int Dimensions = dimensions>
356356
typename std::enable_if<(Dimensions > 0), id<Dimensions>>::type
357357
get_offset() const { return this->__impl()->Offset; }
358+
359+
friend class ::cl::sycl::simple_scheduler::Node;
360+
friend class ::cl::sycl::simple_scheduler::Scheduler;
358361
};
359362

360363
SYCL_ACCESSOR_SUBCLASS(accessor_opdata_w, accessor_common,
@@ -366,6 +369,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_opdata_w, accessor_common,
366369
operator dataT &() const {
367370
return this->__impl()->Data[0];
368371
}
372+
friend class ::cl::sycl::simple_scheduler::Node;
373+
friend class ::cl::sycl::simple_scheduler::Scheduler;
369374
};
370375

371376
SYCL_ACCESSOR_SUBCLASS(accessor_subscript_wn, accessor_opdata_w,
@@ -387,6 +392,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_wn, accessor_opdata_w,
387392
return subscript_obj<dimensions, dataT, dimensions - 1, accessMode,
388393
accessTarget, isPlaceholder>(*this, ids);
389394
}
395+
friend class ::cl::sycl::simple_scheduler::Node;
396+
friend class ::cl::sycl::simple_scheduler::Scheduler;
390397
};
391398

392399
SYCL_ACCESSOR_SUBCLASS(accessor_subscript_w, accessor_subscript_wn,
@@ -408,6 +415,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_w, accessor_subscript_wn,
408415
dataT &operator[](size_t index) const {
409416
return this->__impl()->Data[index];
410417
}
418+
friend class ::cl::sycl::simple_scheduler::Node;
419+
friend class ::cl::sycl::simple_scheduler::Scheduler;
411420
};
412421

413422
SYCL_ACCESSOR_SUBCLASS(accessor_opdata_r, accessor_subscript_w,
@@ -416,6 +425,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_opdata_r, accessor_subscript_w,
416425
operator PureType() const {
417426
return this->__impl()->Data[0];
418427
}
428+
friend class ::cl::sycl::simple_scheduler::Node;
429+
friend class ::cl::sycl::simple_scheduler::Scheduler;
419430
};
420431

421432
SYCL_ACCESSOR_SUBCLASS(accessor_subscript_rn, accessor_opdata_r,
@@ -434,6 +445,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_rn, accessor_opdata_r,
434445
return subscript_obj<dimensions, dataT, dimensions - 1, accessMode,
435446
accessTarget, isPlaceholder>(*this, ids);
436447
}
448+
friend class ::cl::sycl::simple_scheduler::Node;
449+
friend class ::cl::sycl::simple_scheduler::Scheduler;
437450
};
438451

439452
SYCL_ACCESSOR_SUBCLASS(accessor_subscript_r, accessor_subscript_rn,
@@ -447,6 +460,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_r, accessor_subscript_rn,
447460
operator[](size_t index) const {
448461
return this->__impl()->Data[index];
449462
}
463+
friend class ::cl::sycl::simple_scheduler::Node;
464+
friend class ::cl::sycl::simple_scheduler::Scheduler;
450465
};
451466

452467
template <access::target accessTarget> struct getAddressSpace {
@@ -469,6 +484,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_atomic_eq0, accessor_subscript_r,
469484
return atomic<PureType, addressSpace>(
470485
multi_ptr<PureType, addressSpace>(&(this->__impl()->Data[0])));
471486
}
487+
friend class ::cl::sycl::simple_scheduler::Node;
488+
friend class ::cl::sycl::simple_scheduler::Scheduler;
472489
};
473490

474491
// Available when: accessMode == access::mode::atomic && dimensions > 0
@@ -483,6 +500,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_atomic_gt0,
483500
multi_ptr<PureType, addressSpace>(&(this->__impl()->Data[getOffsetForId(
484501
this->__impl()->Range, index, this->__impl()->Offset)])));
485502
}
503+
friend class ::cl::sycl::simple_scheduler::Node;
504+
friend class ::cl::sycl::simple_scheduler::Scheduler;
486505
};
487506

488507
// Available only when: accessMode == access::mode::atomic && dimensions == 1
@@ -496,6 +515,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_subscript_atomic_eq1,
496515
return atomic<PureType, addressSpace>(
497516
multi_ptr<PureType, addressSpace>(&(this->__impl()->Data[index])));
498517
}
518+
friend class ::cl::sycl::simple_scheduler::Node;
519+
friend class ::cl::sycl::simple_scheduler::Scheduler;
499520
};
500521

501522
// TODO:
@@ -535,6 +556,8 @@ SYCL_ACCESSOR_SUBCLASS(accessor_pointer, accessor_subscript_atomic_eq1, true) {
535556
get_pointer() const {
536557
return local_ptr<DataT>(this->__impl()->Data);
537558
}
559+
friend class ::cl::sycl::simple_scheduler::Node;
560+
friend class ::cl::sycl::simple_scheduler::Scheduler;
538561
};
539562

540563
} // namespace detail
@@ -557,13 +580,13 @@ class accessor
557580
// Make sure Impl field is the first in the class, so that it is
558581
// safe to reinterpret a pointer to accessor as a pointer to the
559582
// implementation.
560-
_ImplT __impl;
583+
_ImplT __implx;
561584

562585
void __init(_ValueType *Ptr, range<dimensions> Range,
563586
id<dimensions> Offset) {
564-
__impl.Data = Ptr;
565-
__impl.Range = Range;
566-
__impl.Offset = Offset;
587+
__implx.Data = Ptr;
588+
__implx.Range = Range;
589+
__implx.Offset = Offset;
567590
}
568591

569592
public:
@@ -593,7 +616,7 @@ class accessor
593616
AccessTarget == access::target::constant_buffer))) &&
594617
Dimensions == 0),
595618
buffer<DataT, 1>>::type &bufferRef)
596-
: __impl(detail::getSyclObjImpl(bufferRef)->BufPtr) {
619+
: __implx(detail::getSyclObjImpl(bufferRef)->BufPtr) {
597620
auto BufImpl = detail::getSyclObjImpl(bufferRef);
598621
if (AccessTarget == access::target::host_buffer) {
599622
if (BufImpl->OpenCLInterop) {
@@ -633,7 +656,7 @@ class accessor
633656
#ifdef __SYCL_DEVICE_ONLY__
634657
; // This ctor can't be used in device code, so no need to define it.
635658
#else // !__SYCL_DEVICE_ONLY__
636-
: __impl(detail::getSyclObjImpl(bufferRef)->BufPtr,
659+
: __implx(detail::getSyclObjImpl(bufferRef)->BufPtr,
637660
detail::getSyclObjImpl(bufferRef)->Range,
638661
&commandGroupHandlerRef) {
639662
auto BufImpl = detail::getSyclObjImpl(bufferRef);
@@ -643,7 +666,7 @@ class accessor
643666
"interoperability buffer");
644667
}
645668
commandGroupHandlerRef.AddBufDep<AccessMode, AccessTarget>(*BufImpl);
646-
__impl.m_Buf = BufImpl.get();
669+
__implx.m_Buf = BufImpl.get();
647670
}
648671
#endif // !__SYCL_DEVICE_ONLY__
649672

@@ -669,7 +692,7 @@ class accessor
669692
AccessTarget == access::target::constant_buffer))) &&
670693
Dimensions > 0),
671694
buffer<DataT, Dimensions>>::type &bufferRef)
672-
: __impl(detail::getSyclObjImpl(bufferRef)->BufPtr,
695+
: __implx(detail::getSyclObjImpl(bufferRef)->BufPtr,
673696
detail::getSyclObjImpl(bufferRef)->Range) {
674697
auto BufImpl = detail::getSyclObjImpl(bufferRef);
675698
if (AccessTarget == access::target::host_buffer) {
@@ -710,7 +733,7 @@ class accessor
710733
#ifdef __SYCL_DEVICE_ONLY__
711734
; // This ctor can't be used in device code, so no need to define it.
712735
#else
713-
: __impl(detail::getSyclObjImpl(bufferRef)->BufPtr,
736+
: __implx(detail::getSyclObjImpl(bufferRef)->BufPtr,
714737
detail::getSyclObjImpl(bufferRef)->Range,
715738
&commandGroupHandlerRef) {
716739
auto BufImpl = detail::getSyclObjImpl(bufferRef);
@@ -720,7 +743,7 @@ class accessor
720743
"interoperability buffer");
721744
}
722745
commandGroupHandlerRef.AddBufDep<AccessMode, AccessTarget>(*BufImpl);
723-
__impl.m_Buf = BufImpl.get();
746+
__implx.m_Buf = BufImpl.get();
724747
}
725748
#endif
726749

@@ -752,7 +775,7 @@ class accessor
752775
#ifdef __SYCL_DEVICE_ONLY__
753776
; // This ctor can't be used in device code, so no need to define it.
754777
#else // !__SYCL_DEVICE_ONLY__
755-
: __impl(detail::getSyclObjImpl(bufferRef)->BufPtr, Range, Offset) {
778+
: __implx(detail::getSyclObjImpl(bufferRef)->BufPtr, Range, Offset) {
756779
auto BufImpl = detail::getSyclObjImpl(bufferRef);
757780
if (AccessTarget == access::target::host_buffer) {
758781
if (BufImpl->OpenCLInterop) {
@@ -796,7 +819,7 @@ class accessor
796819
#ifdef __SYCL_DEVICE_ONLY__
797820
; // This ctor can't be used in device code, so no need to define it.
798821
#else // !__SYCL_DEVICE_ONLY__
799-
: __impl(detail::getSyclObjImpl(bufferRef)->BufPtr, Range,
822+
: __implx(detail::getSyclObjImpl(bufferRef)->BufPtr, Range,
800823
&commandGroupHandlerRef, Offset) {
801824
auto BufImpl = detail::getSyclObjImpl(bufferRef);
802825
if (BufImpl->OpenCLInterop && !BufImpl->isValidAccessToMem(accessMode)) {
@@ -805,7 +828,7 @@ class accessor
805828
"interoperability buffer");
806829
}
807830
commandGroupHandlerRef.AddBufDep<AccessMode, AccessTarget>(*BufImpl);
808-
__impl.m_Buf = BufImpl.get();
831+
__implx.m_Buf = BufImpl.get();
809832
}
810833
#endif // !__SYCL_DEVICE_ONLY__
811834

@@ -835,7 +858,7 @@ class accessor
835858
Dimensions > 0),
836859
range<Dimensions>>::type allocationSize,
837860
handler &commandGroupHandlerRef)
838-
: __impl(allocationSize, &commandGroupHandlerRef) {}
861+
: __implx(allocationSize, &commandGroupHandlerRef) {}
839862
};
840863

841864
} // namespace sycl

‎sycl/include/CL/sycl/detail/scheduler/scheduler.cpp

Lines changed: 6 additions & 12 deletions
Original file line numberDiff line numberDiff line change
@@ -61,9 +61,7 @@ void Node::addAccRequirement(
6161
accessor<dataT, dimensions, accessMode, accessTarget, isPlaceholder> &&Acc,
6262
int argIndex) {
6363
detail::buffer_impl<dataT, dimensions> *buf =
64-
Acc.template accessor_base<dataT, dimensions, accessMode, accessTarget,
65-
isPlaceholder>::__impl()
66-
->m_Buf;
64+
Acc.__impl()->m_Buf;
6765
addBufRequirement<accessMode, accessTarget, dataT, dimensions>(*buf);
6866
addInteropArg(nullptr, buf->get_size(), argIndex,
6967
getReqForBuffer(m_Bufs, *buf));
@@ -128,8 +126,7 @@ template <typename T, int Dimensions, access::mode mode, access::target tgt,
128126
access::placeholder isPlaceholder>
129127
void Node::addExplicitMemOp(
130128
accessor<T, Dimensions, mode, tgt, isPlaceholder> &Dest, T Src) {
131-
auto *DestBase = Dest.template accessor_base<T, Dimensions, mode, tgt,
132-
isPlaceholder>::__impl();
129+
auto *DestBase = Dest.__impl();
133130
assert(DestBase != nullptr &&
134131
"Accessor should have an initialized accessor_base");
135132
detail::buffer_impl<T, Dimensions> *Buf = DestBase->m_Buf;
@@ -153,13 +150,10 @@ template <typename T_src, int dim_src, access::mode mode_src,
153150
void Node::addExplicitMemOp(
154151
accessor<T_src, dim_src, mode_src, tgt_src, isPlaceholder_src> Src,
155152
accessor<T_dest, dim_dest, mode_dest, tgt_dest, isPlaceholder_dest> Dest) {
156-
auto *SrcBase = Src.template accessor_base<T_src, dim_src, mode_src, tgt_src,
157-
isPlaceholder_src>::__impl();
153+
auto *SrcBase = Src.__impl();
158154
assert(SrcBase != nullptr &&
159155
"Accessor should have an initialized accessor_base");
160-
auto *DestBase =
161-
Dest.template accessor_base<T_dest, dim_dest, mode_dest, tgt_dest,
162-
isPlaceholder_dest>::__impl();
156+
auto *DestBase = Dest.__impl();
163157
assert(DestBase != nullptr &&
164158
"Accessor should have an initialized accessor_base");
165159

@@ -191,8 +185,8 @@ template <typename T, int Dimensions, access::mode mode, access::target tgt,
191185
void Scheduler::updateHost(
192186
accessor<T, Dimensions, mode, tgt, isPlaceholder> &Acc,
193187
cl::sycl::event &Event) {
194-
auto *AccBase = Acc.template accessor_base<T, Dimensions, mode, tgt,
195-
isPlaceholder>::__impl();
188+
auto *AccBase = Acc.impl();
189+
196190
assert(AccBase != nullptr &&
197191
"Accessor should have an initialized accessor_base");
198192
detail::buffer_impl<T, Dimensions> *Buf = AccBase->m_Buf;

0 commit comments

Comments
 (0)