From 3fe89b994396c37fae7dd468984a462a0575ed11 Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Wed, 15 Apr 2020 14:41:37 +0300 Subject: [PATCH 1/8] Warnings handeling Signed-off-by: Aleksander Fadeev Werror is added Signed-off-by: Aleksander Fadeev Path fix Signed-off-by: Aleksander Fadeev __SYCL_UNUSED_ARUMENT__ Signed-off-by: Aleksander Fadeev __SYCL_UNUSED_ARGUMENT__ Signed-off-by: Aleksander Fadeev path fix Signed-off-by: Aleksander Fadeev Fix Signed-off-by: Aleksander Fadeev usm_allocator.hpp fix Signed-off-by: Aleksander Fadeev test setup Signed-off-by: Aleksander Fadeev warnings.cpp path correction Signed-off-by: Aleksander Fadeev Formatting Signed-off-by: Aleksander Fadeev suppres depricate warning Signed-off-by: Aleksander Fadeev Path creating Signed-off-by: Aleksander Fadeev warnings.cpp fix Signed-off-by: Aleksander Fadeev defines.hpp revert Signed-off-by: Aleksander Fadeev ordered_queue.cpp fix Signed-off-by: Aleksander Fadeev Path improvement Signed-off-by: Aleksander Fadeev Formatting 3 Signed-off-by: Aleksander Fadeev pragma fix Signed-off-by: Aleksander Fadeev FIX rebase Signed-off-by: Aleksander Fadeev Formatting3 Signed-off-by: Aleksander Fadeev orderes_queue.hpp fix Signed-off-by: Aleksander Fadeev Formatting4 Signed-off-by: Aleksander Fadeev --- sycl/include/CL/__spirv/spirv_ops.hpp | 12 ++-- sycl/include/CL/sycl/accessor.hpp | 15 +++- sycl/include/CL/sycl/atomic.hpp | 24 +++---- sycl/include/CL/sycl/buffer.hpp | 3 +- .../CL/sycl/detail/aligned_allocator.hpp | 6 +- sycl/include/CL/sycl/detail/array.hpp | 1 + sycl/include/CL/sycl/detail/buffer_impl.hpp | 3 +- sycl/include/CL/sycl/detail/cg.hpp | 4 +- .../CL/sycl/detail/image_accessor_util.hpp | 3 +- sycl/include/CL/sycl/detail/image_impl.hpp | 4 +- sycl/include/CL/sycl/detail/kernel_desc.hpp | 2 +- sycl/include/CL/sycl/detail/pi.hpp | 4 +- sycl/include/CL/sycl/detail/stream_impl.hpp | 9 ++- .../CL/sycl/detail/sycl_mem_obj_allocator.hpp | 2 +- .../include/CL/sycl/detail/sycl_mem_obj_t.hpp | 4 ++ sycl/include/CL/sycl/group.hpp | 2 + sycl/include/CL/sycl/handler.hpp | 28 ++++++-- sycl/include/CL/sycl/intel/functional.hpp | 2 +- .../include/CL/sycl/intel/group_algorithm.hpp | 69 ++++++++++++++++--- sycl/include/CL/sycl/intel/pipes.hpp | 8 +++ sycl/include/CL/sycl/intel/reduction.hpp | 8 +-- sycl/include/CL/sycl/intel/sub_group_host.hpp | 43 ++++++------ sycl/include/CL/sycl/multi_ptr.hpp | 24 +++---- sycl/include/CL/sycl/nd_item.hpp | 1 + sycl/include/CL/sycl/program.hpp | 2 + sycl/include/CL/sycl/sampler.hpp | 1 + sycl/include/CL/sycl/stream.hpp | 1 + sycl/include/CL/sycl/types.hpp | 10 +-- sycl/include/CL/sycl/usm/usm_allocator.hpp | 10 +-- sycl/test/CMakeLists.txt | 1 + sycl/test/lit.cfg.py | 1 + sycl/test/lit.site.cfg.py.in | 1 + sycl/test/warnings/warnings.cpp | 2 +- 33 files changed, 203 insertions(+), 107 deletions(-) diff --git a/sycl/include/CL/__spirv/spirv_ops.hpp b/sycl/include/CL/__spirv/spirv_ops.hpp index b93b72d9cf47b..aa7696d45e321 100644 --- a/sycl/include/CL/__spirv/spirv_ops.hpp +++ b/sycl/include/CL/__spirv/spirv_ops.hpp @@ -235,9 +235,9 @@ __spirv_ocl_prefetch(const __attribute__((opencl_global)) char *Ptr, template __SYCL_CONVERGENT__ extern __ocl_event_t -OpGroupAsyncCopyGlobalToLocal(__spv::Scope::Flag Execution, dataT *Dest, - dataT *Src, size_t NumElements, size_t Stride, - __ocl_event_t E) noexcept { +OpGroupAsyncCopyGlobalToLocal(__spv::Scope::Flag, dataT *Dest, dataT *Src, + size_t NumElements, size_t Stride, + __ocl_event_t) noexcept { for (size_t i = 0; i < NumElements; i++) { Dest[i] = Src[i * Stride]; } @@ -247,9 +247,9 @@ OpGroupAsyncCopyGlobalToLocal(__spv::Scope::Flag Execution, dataT *Dest, template __SYCL_CONVERGENT__ extern __ocl_event_t -OpGroupAsyncCopyLocalToGlobal(__spv::Scope::Flag Execution, dataT *Dest, - dataT *Src, size_t NumElements, size_t Stride, - __ocl_event_t E) noexcept { +OpGroupAsyncCopyLocalToGlobal(__spv::Scope::Flag, dataT *Dest, dataT *Src, + size_t NumElements, size_t Stride, + __ocl_event_t) noexcept { for (size_t i = 0; i < NumElements; i++) { Dest[i * Stride] = Src[i]; } diff --git a/sycl/include/CL/sycl/accessor.hpp b/sycl/include/CL/sycl/accessor.hpp index 97feaa283182c..cbc13528a1e41 100644 --- a/sycl/include/CL/sycl/accessor.hpp +++ b/sycl/include/CL/sycl/accessor.hpp @@ -387,6 +387,8 @@ class image_accessor image_accessor(image &ImageRef, int ImageElementSize) #ifdef __SYCL_DEVICE_ONLY__ { + (void)ImageRef; + (void)ImageElementSize; // No implementation needed for device. The constructor is only called by // host. } @@ -415,6 +417,9 @@ class image_accessor handler &CommandGroupHandlerRef, int ImageElementSize) #ifdef __SYCL_DEVICE_ONLY__ { + (void)ImageRef; + (void)CommandGroupHandlerRef; + (void)ImageElementSize; // No implementation needed for device. The constructor is only called by // host. } @@ -784,6 +789,7 @@ class accessor : handler &CommandGroupHandler) #ifdef __SYCL_DEVICE_ONLY__ : impl(id(), range<1>{1}, BufferRef.get_range()) { + (void)CommandGroupHandler; } #else : AccessorBaseHost( @@ -824,6 +830,7 @@ class accessor : handler &CommandGroupHandler) #ifdef __SYCL_DEVICE_ONLY__ : impl(id(), BufferRef.get_range(), BufferRef.get_range()) { + (void)CommandGroupHandler; } #else : AccessorBaseHost( @@ -867,6 +874,7 @@ class accessor : id AccessOffset = {}) #ifdef __SYCL_DEVICE_ONLY__ : impl(AccessOffset, AccessRange, BufferRef.get_range()) { + (void)CommandGroupHandler; } #else : AccessorBaseHost(detail::convertToArrayOfN<3, 0>(AccessOffset), @@ -1033,7 +1041,7 @@ class accessor &getSize() const { return impl.MemRange; } void __init(ConcreteASPtrType Ptr, range AccessRange, - range MemRange, id Offset) { + range, id) { MData = Ptr; for (int I = 0; I < AdjustedDim; ++I) getSize()[I] = AccessRange[I]; @@ -1075,7 +1083,7 @@ class accessor> - accessor(handler &CommandGroupHandler) + accessor(handler &) #ifdef __SYCL_DEVICE_ONLY__ : impl(range{1}) { } @@ -1085,7 +1093,7 @@ class accessor 0)>> - accessor(range AllocationSize, handler &CommandGroupHandler) + accessor(range AllocationSize, handler &) #ifdef __SYCL_DEVICE_ONLY__ : impl(AllocationSize) { } @@ -1271,6 +1279,7 @@ struct hash -void __spirv_AtomicStore(std::atomic *Ptr, __spv::Scope::Flag S, +void __spirv_AtomicStore(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { Ptr->store(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -T __spirv_AtomicLoad(const std::atomic *Ptr, __spv::Scope::Flag S, +T __spirv_AtomicLoad(const std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS) { return Ptr->load(::cl::sycl::detail::getStdMemoryOrder(MS)); } template -T __spirv_AtomicExchange(std::atomic *Ptr, __spv::Scope::Flag S, +T __spirv_AtomicExchange(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { return Ptr->exchange(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -extern T __spirv_AtomicIAdd(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicIAdd(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { return Ptr->fetch_add(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -extern T __spirv_AtomicISub(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicISub(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { return Ptr->fetch_sub(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -extern T __spirv_AtomicAnd(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicAnd(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { return Ptr->fetch_and(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -extern T __spirv_AtomicOr(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicOr(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { return Ptr->fetch_or(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -extern T __spirv_AtomicXor(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicXor(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { return Ptr->fetch_xor(V, ::cl::sycl::detail::getStdMemoryOrder(MS)); } template -extern T __spirv_AtomicMin(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicMin(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { std::memory_order MemoryOrder = ::cl::sycl::detail::getStdMemoryOrder(MS); T Val = Ptr->load(MemoryOrder); @@ -147,7 +147,7 @@ extern T __spirv_AtomicMin(std::atomic *Ptr, __spv::Scope::Flag S, } template -extern T __spirv_AtomicMax(std::atomic *Ptr, __spv::Scope::Flag S, +extern T __spirv_AtomicMax(std::atomic *Ptr, __spv::Scope::Flag, __spv::MemorySemanticsMask::Flag MS, T V) { std::memory_order MemoryOrder = ::cl::sycl::detail::getStdMemoryOrder(MS); T Val = Ptr->load(MemoryOrder); diff --git a/sycl/include/CL/sycl/buffer.hpp b/sycl/include/CL/sycl/buffer.hpp index 0661aada4a4a6..407bf25a8610a 100644 --- a/sycl/include/CL/sycl/buffer.hpp +++ b/sycl/include/CL/sycl/buffer.hpp @@ -337,8 +337,7 @@ class buffer { return outOfBounds; } - bool isContiguousRegion(const id<1> &offset, const range<1> &newRange, - const range<1> &parentRange) { + bool isContiguousRegion(const id<1> &, const range<1> &, const range<1> &) { // 1D sub buffer always has contiguous region return true; } diff --git a/sycl/include/CL/sycl/detail/aligned_allocator.hpp b/sycl/include/CL/sycl/detail/aligned_allocator.hpp index e4fe03328d6e8..624d435d5c6d1 100644 --- a/sycl/include/CL/sycl/detail/aligned_allocator.hpp +++ b/sycl/include/CL/sycl/detail/aligned_allocator.hpp @@ -63,13 +63,13 @@ template class aligned_allocator { } // Release allocated memory - void deallocate(pointer Ptr, size_t size) { + void deallocate(pointer Ptr, size_t) { if (Ptr) detail::OSUtil::alignedFree(Ptr); } - bool operator==(const aligned_allocator&) { return true; } - bool operator!=(const aligned_allocator& rhs) { return false; } + bool operator==(const aligned_allocator &) { return true; } + bool operator!=(const aligned_allocator &) { return false; } void setAlignment(size_t Alignment) { MAlignment = Alignment; } diff --git a/sycl/include/CL/sycl/detail/array.hpp b/sycl/include/CL/sycl/detail/array.hpp index d4aada6d7dd86..668f5b2942e9c 100644 --- a/sycl/include/CL/sycl/detail/array.hpp +++ b/sycl/include/CL/sycl/detail/array.hpp @@ -115,6 +115,7 @@ template class array { PI_INVALID_VALUE); } #endif + (void)dimension; } }; diff --git a/sycl/include/CL/sycl/detail/buffer_impl.hpp b/sycl/include/CL/sycl/detail/buffer_impl.hpp index a39fcdb7ed65a..3052cd768372d 100644 --- a/sycl/include/CL/sycl/detail/buffer_impl.hpp +++ b/sycl/include/CL/sycl/detail/buffer_impl.hpp @@ -43,8 +43,7 @@ class __SYCL_EXPORT buffer_impl final : public SYCLMemObjT { using typename BaseT::MemObjType; public: - buffer_impl(size_t SizeInBytes, size_t RequiredAlign, - const property_list &Props, + buffer_impl(size_t SizeInBytes, size_t, const property_list &Props, unique_ptr_class Allocator) : BaseT(SizeInBytes, Props, std::move(Allocator)) {} diff --git a/sycl/include/CL/sycl/detail/cg.hpp b/sycl/include/CL/sycl/detail/cg.hpp index 26ee7b6f89837..2ef0ff5170b74 100644 --- a/sycl/include/CL/sycl/detail/cg.hpp +++ b/sycl/include/CL/sycl/detail/cg.hpp @@ -251,7 +251,7 @@ class HostKernel : public HostKernelBase { // If local size for host is not set explicitly, let's adjust it to 1, // so nd_range_error for zero local size is not thrown. if (AdjustedRange.LocalSize[0] == 0) - for (int I = 0; I < AdjustedRange.Dims; ++I) + for (size_t I = 0; I < AdjustedRange.Dims; ++I) AdjustedRange.LocalSize[I] = 1; if (HPI) HPI->start(); @@ -264,7 +264,7 @@ class HostKernel : public HostKernelBase { template typename std::enable_if::value>::type - runOnHost(const NDRDescT &NDRDesc) { + runOnHost(const NDRDescT &) { MKernel(); } diff --git a/sycl/include/CL/sycl/detail/image_accessor_util.hpp b/sycl/include/CL/sycl/detail/image_accessor_util.hpp index b7ac161e7358a..63c805ca7a833 100644 --- a/sycl/include/CL/sycl/detail/image_accessor_util.hpp +++ b/sycl/include/CL/sycl/detail/image_accessor_util.hpp @@ -78,8 +78,7 @@ convertToFloat4(vec Coords) { // ptr. template detail::enable_if_t::value, size_t> -getImageOffset(const T &Coords, const id<3> ImgPitch, - const uint8_t ElementSize) { +getImageOffset(const T &Coords, const id<3>, const uint8_t ElementSize) { return Coords * ElementSize; } diff --git a/sycl/include/CL/sycl/detail/image_impl.hpp b/sycl/include/CL/sycl/detail/image_impl.hpp index 2ff2f67b12902..a8ad3e452357f 100644 --- a/sycl/include/CL/sycl/detail/image_impl.hpp +++ b/sycl/include/CL/sycl/detail/image_impl.hpp @@ -242,9 +242,7 @@ class __SYCL_EXPORT image_impl final : public SYCLMemObjT { }); } - template bool checkAnyImpl(T Value) { - return false; - } + template bool checkAnyImpl(T) { return false; } template bool checkAnyImpl(ValT Value, VarT Variant, Args... Arguments) { diff --git a/sycl/include/CL/sycl/detail/kernel_desc.hpp b/sycl/include/CL/sycl/detail/kernel_desc.hpp index 7f1bcf496a55e..120fb9dc4e96c 100644 --- a/sycl/include/CL/sycl/detail/kernel_desc.hpp +++ b/sycl/include/CL/sycl/detail/kernel_desc.hpp @@ -52,7 +52,7 @@ template struct SpecConstantInfo { #ifndef __SYCL_UNNAMED_LAMBDA__ template struct KernelInfo { static constexpr unsigned getNumParams() { return 0; } - static const kernel_param_desc_t &getParamDesc(int Idx) { + static const kernel_param_desc_t &getParamDesc(int) { static kernel_param_desc_t Dummy; return Dummy; } diff --git a/sycl/include/CL/sycl/detail/pi.hpp b/sycl/include/CL/sycl/detail/pi.hpp index d94ccaa3960fb..113255d60beba 100644 --- a/sycl/include/CL/sycl/detail/pi.hpp +++ b/sycl/include/CL/sycl/detail/pi.hpp @@ -328,12 +328,12 @@ template inline To cast(From value) { } // These conversions should use PI interop API. -template <> inline pi::PiProgram cast(cl_program interop) { +template <> inline pi::PiProgram cast(cl_program) { RT::assertion(false, "pi::cast -> use piextProgramFromNative"); return {}; } -template <> inline pi::PiDevice cast(cl_device_id interop) { +template <> inline pi::PiDevice cast(cl_device_id) { RT::assertion(false, "pi::cast -> use piextDeviceFromNative"); return {}; } diff --git a/sycl/include/CL/sycl/detail/stream_impl.hpp b/sycl/include/CL/sycl/detail/stream_impl.hpp index dfa416d4da5a4..e320e0e8455b9 100644 --- a/sycl/include/CL/sycl/detail/stream_impl.hpp +++ b/sycl/include/CL/sycl/detail/stream_impl.hpp @@ -399,9 +399,8 @@ checkForInfNan(char *Buf, T Val) { // Returns number of symbols written to the buffer template -inline EnableIfFP ScalarToStr(const T &Val, char *Buf, - unsigned Flags, int Width, - int Precision = -1) { +inline EnableIfFP +ScalarToStr(const T &Val, char *Buf, unsigned Flags, int, int Precision = -1) { unsigned Offset = checkForInfNan(Buf, Val); if (Offset) return Offset; @@ -428,8 +427,8 @@ inline EnableIfFP ScalarToStr(const T &Val, char *Buf, // Returns number of symbols written to the buffer template inline typename std::enable_if::value, unsigned>::type -ScalarToStr(const T &Val, char *Buf, unsigned Flags, int Width, - int Precision = -1) { +ScalarToStr(const T &Val, char *Buf, unsigned Flags, int, int Precision = -1) { + (void)Precision; int Base = 10; // append base manipulator diff --git a/sycl/include/CL/sycl/detail/sycl_mem_obj_allocator.hpp b/sycl/include/CL/sycl/detail/sycl_mem_obj_allocator.hpp index 512e105b60bf6..827a54ff9f6af 100644 --- a/sycl/include/CL/sycl/detail/sycl_mem_obj_allocator.hpp +++ b/sycl/include/CL/sycl/detail/sycl_mem_obj_allocator.hpp @@ -74,7 +74,7 @@ class SYCLMemObjAllocatorHolder : public SYCLMemObjAllocator { private: template - EnableIfNonDefaultAllocator setAlignImpl(std::size_t RequiredAlign) { + EnableIfNonDefaultAllocator setAlignImpl(std::size_t) { // Do nothing in case of user's allocator. } diff --git a/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp b/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp index 21b4ff925deb2..4043b2ec28b1e 100644 --- a/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp +++ b/sycl/include/CL/sycl/detail/sycl_mem_obj_t.hpp @@ -263,6 +263,10 @@ class __SYCL_EXPORT SYCLMemObjT : public SYCLMemObjI { void *allocateMem(ContextImplPtr Context, bool InitFromUserData, void *HostPtr, RT::PiEvent &InteropEvent) override { + (void)Context; + (void)InitFromUserData; + (void)HostPtr; + (void)InteropEvent; throw runtime_error("Not implemented", PI_INVALID_OPERATION); } diff --git a/sycl/include/CL/sycl/group.hpp b/sycl/include/CL/sycl/group.hpp index dd99037e0a3d7..3dfb4fdee0789 100644 --- a/sycl/include/CL/sycl/group.hpp +++ b/sycl/include/CL/sycl/group.hpp @@ -56,6 +56,7 @@ template class private_memory { // in the group: Val.reset(new T[G.get_local_range().size()]); #endif // __SYCL_DEVICE_ONLY__ + (void)G; } // Access the instance for the current work-item @@ -66,6 +67,7 @@ template class private_memory { size_t Ind = Id.get_physical_local().get_linear_id(); return Val.get()[Ind]; #else + (void)Id; return Val; #endif // __SYCL_DEVICE_ONLY__ } diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp index cd8f4112b33c8..c65eea75bb7ba 100644 --- a/sycl/include/CL/sycl/handler.hpp +++ b/sycl/include/CL/sycl/handler.hpp @@ -312,7 +312,7 @@ class __SYCL_EXPORT handler { setArgsHelper(++ArgIndex, std::move(Args)...); } - void setArgsHelper(int ArgIndex) {} + void setArgsHelper(int) {} // setArgHelper for local accessor argument. template getDelinearizedIndex(const range<1> Range, const size_t Index) { + static id<1> getDelinearizedIndex(const range<1>, const size_t Index) { return {Index}; } @@ -718,7 +718,7 @@ class __SYCL_EXPORT handler { /// /// \param Event is a valid SYCL event to wait on. void depends_on(event Event) { - MEvents.push_back(std::move(detail::getSyclObjImpl(Event))); + MEvents.push_back(detail::getSyclObjImpl(Event)); } /// Registers event dependencies on this command group. @@ -726,7 +726,7 @@ class __SYCL_EXPORT handler { /// \param Events is a vector of valid SYCL events to wait on. void depends_on(vector_class Events) { for (event &Event : Events) { - MEvents.push_back(std::move(detail::getSyclObjImpl(Event))); + MEvents.push_back(detail::getSyclObjImpl(Event)); } } @@ -789,6 +789,7 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)NumWorkItems; kernel_parallel_for(KernelFunc); #else MNDRDesc.set(std::move(NumWorkItems)); @@ -845,6 +846,8 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)NumWorkItems; + (void)WorkItemOffset; kernel_parallel_for(KernelFunc); #else MNDRDesc.set(std::move(NumWorkItems), std::move(WorkItemOffset)); @@ -872,6 +875,7 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)ExecutionRange; kernel_parallel_for_nd_range(KernelFunc); #else MNDRDesc.set(std::move(ExecutionRange)); @@ -1040,6 +1044,7 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)NumWorkGroups; kernel_parallel_for_work_group(KernelFunc); #else MNDRDesc.setNumWorkGroups(NumWorkGroups); @@ -1069,6 +1074,8 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)NumWorkGroups; + (void)WorkGroupSize; kernel_parallel_for_work_group(KernelFunc); #else MNDRDesc.set(nd_range(NumWorkGroups * WorkGroupSize, WorkGroupSize)); @@ -1157,6 +1164,7 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)Kernel; kernel_single_task(KernelFunc); #else MNDRDesc.set(range<1>{1}); @@ -1193,6 +1201,8 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)Kernel; + (void)NumWorkItems; kernel_parallel_for(KernelFunc); #else MNDRDesc.set(std::move(NumWorkItems)); @@ -1222,6 +1232,9 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)Kernel; + (void)NumWorkItems; + (void)WorkItemOffset; kernel_parallel_for(KernelFunc); #else MNDRDesc.set(std::move(NumWorkItems), std::move(WorkItemOffset)); @@ -1251,6 +1264,8 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)Kernel; + (void)NDRange; kernel_parallel_for_nd_range(KernelFunc); #else MNDRDesc.set(std::move(NDRange)); @@ -1284,6 +1299,8 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)Kernel; + (void)NumWorkGroups; kernel_parallel_for_work_group(KernelFunc); #else MNDRDesc.setNumWorkGroups(NumWorkGroups); @@ -1317,6 +1334,9 @@ class __SYCL_EXPORT handler { using NameT = typename detail::get_kernel_name_t::name; #ifdef __SYCL_DEVICE_ONLY__ + (void)Kernel; + (void)NumWorkGroups; + (void)WorkGroupSize; kernel_parallel_for_work_group(KernelFunc); #else MNDRDesc.set(nd_range(NumWorkGroups * WorkGroupSize, WorkGroupSize)); diff --git a/sycl/include/CL/sycl/intel/functional.hpp b/sycl/include/CL/sycl/intel/functional.hpp index f54e3777d3375..ee4ed21b33ffd 100644 --- a/sycl/include/CL/sycl/intel/functional.hpp +++ b/sycl/include/CL/sycl/intel/functional.hpp @@ -85,7 +85,7 @@ struct GroupOpTag::value>> { #define __SYCL_CALC_OVERLOAD(GroupTag, SPIRVOperation, BinaryOperation) \ template \ - static T calc(GroupTag, T x, BinaryOperation op) { \ + static T calc(GroupTag, T x, BinaryOperation) { \ using OCLT = detail::ConvertToOpenCLType_t; \ OCLT Arg = x; \ OCLT Ret = \ diff --git a/sycl/include/CL/sycl/intel/group_algorithm.hpp b/sycl/include/CL/sycl/intel/group_algorithm.hpp index 5a11d23812cd7..5142ab226a8b2 100644 --- a/sycl/include/CL/sycl/intel/group_algorithm.hpp +++ b/sycl/include/CL/sycl/intel/group_algorithm.hpp @@ -42,7 +42,7 @@ typename Group::linear_id_type get_local_linear_id(Group g); #ifdef __SYCL_DEVICE_ONLY__ #define __SYCL_GROUP_GET_LOCAL_LINEAR_ID(D) \ template <> \ - group::linear_id_type get_local_linear_id>(group g) { \ + group::linear_id_type get_local_linear_id>(group) { \ nd_item it = cl::sycl::detail::Builder::getNDItem(); \ return it.get_local_linear_id(); \ } @@ -60,7 +60,7 @@ get_local_linear_id(intel::sub_group g) { template id linear_id_to_id(range, size_t linear_id); -template <> inline id<1> linear_id_to_id(range<1> r, size_t linear_id) { +template <> inline id<1> linear_id_to_id(range<1>, size_t linear_id) { return id<1>(linear_id); } template <> inline id<2> linear_id_to_id(range<2> r, size_t linear_id) { @@ -115,6 +115,10 @@ Function for_each(Group g, Ptr first, Ptr last, Function f) { } return f; #else + (void)g; + (void)first; + (void)last; + (void)f; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -136,13 +140,14 @@ template using EnableIfIsPointer = cl::sycl::detail::enable_if_t::value, T>; -template bool all_of(Group g, bool pred) { +template bool all_of(Group, bool pred) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " "intel::sub_group class."); #ifdef __SYCL_DEVICE_ONLY__ return sycl::detail::spirv::GroupAll(pred); #else + (void)pred; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -169,18 +174,23 @@ EnableIfIsPointer all_of(Group g, Ptr first, Ptr last, [&](const typename Ptr::element_type &x) { partial &= pred(x); }); return all_of(g, partial); #else + (void)g; + (void)first; + (void)last; + (void)pred; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif } -template bool any_of(Group g, bool pred) { +template bool any_of(Group, bool pred) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " "intel::sub_group class."); #ifdef __SYCL_DEVICE_ONLY__ return sycl::detail::spirv::GroupAny(pred); #else + (void)pred; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -207,18 +217,23 @@ EnableIfIsPointer any_of(Group g, Ptr first, Ptr last, [&](const typename Ptr::element_type &x) { partial |= pred(x); }); return any_of(g, partial); #else + (void)g; + (void)first; + (void)last; + (void)pred; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif } -template bool none_of(Group g, bool pred) { +template bool none_of(Group, bool pred) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " "intel::sub_group class."); #ifdef __SYCL_DEVICE_ONLY__ return sycl::detail::spirv::GroupAll(!pred); #else + (void)pred; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -241,13 +256,17 @@ EnableIfIsPointer none_of(Group g, Ptr first, Ptr last, "intel::sub_group class."); return !any_of(g, first, last, pred); #else + (void)g; + (void)first; + (void)last; + (void)pred; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif } template -EnableIfIsScalarArithmetic broadcast(Group g, T x, +EnableIfIsScalarArithmetic broadcast(Group, T x, typename Group::id_type local_id) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " @@ -255,6 +274,8 @@ EnableIfIsScalarArithmetic broadcast(Group g, T x, #ifdef __SYCL_DEVICE_ONLY__ return sycl::detail::spirv::GroupBroadcast(x, local_id); #else + (void)x; + (void)local_id; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -273,6 +294,9 @@ EnableIfIsVectorArithmetic broadcast(Group g, T x, } return result; #else + (void)g; + (void)x; + (void)local_id; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -289,6 +313,9 @@ broadcast(Group g, T x, typename Group::linear_id_type linear_local_id) { g, x, sycl::detail::linear_id_to_id(g.get_local_range(), linear_local_id)); #else + (void)g; + (void)x; + (void)linear_local_id; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -307,6 +334,9 @@ broadcast(Group g, T x, typename Group::linear_id_type linear_local_id) { } return result; #else + (void)g; + (void)x; + (void)linear_local_id; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -320,6 +350,8 @@ EnableIfIsScalarArithmetic broadcast(Group g, T x) { #ifdef __SYCL_DEVICE_ONLY__ return broadcast(g, x, 0); #else + (void)g; + (void)x; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -337,13 +369,15 @@ EnableIfIsVectorArithmetic broadcast(Group g, T x) { } return result; #else + (void)g; + (void)x; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif } template -EnableIfIsScalarArithmetic reduce(Group g, T x, BinaryOperation binary_op) { +EnableIfIsScalarArithmetic reduce(Group, T x, BinaryOperation binary_op) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " "intel::sub_group class."); @@ -397,6 +431,7 @@ EnableIfIsScalarArithmetic reduce(Group g, V x, T init, #ifdef __SYCL_DEVICE_ONLY__ return binary_op(init, reduce(g, x, binary_op)); #else + (void)g; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -422,6 +457,7 @@ EnableIfIsVectorArithmetic reduce(Group g, V x, T init, } return result; #else + (void)g; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -450,6 +486,9 @@ reduce(Group g, Ptr first, Ptr last, BinaryOperation binary_op) { }); return reduce(g, partial, binary_op); #else + (void)g; + (void)last; + (void)binary_op; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -476,13 +515,15 @@ EnableIfIsPointer reduce(Group g, Ptr first, Ptr last, T init, }); return reduce(g, partial, init, binary_op); #else + (void)g; + (void)last; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif } template -EnableIfIsScalarArithmetic exclusive_scan(Group g, T x, +EnableIfIsScalarArithmetic exclusive_scan(Group, T x, BinaryOperation binary_op) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " @@ -565,6 +606,7 @@ EnableIfIsScalarArithmetic exclusive_scan(Group g, V x, T init, } return scan; #else + (void)g; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -607,6 +649,10 @@ exclusive_scan(Group g, InPtr first, InPtr last, OutPtr result, T init, } return result + N; #else + (void)g; + (void)last; + (void)result; + (void)init; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -651,7 +697,7 @@ EnableIfIsVectorArithmetic inclusive_scan(Group g, T x, } template -EnableIfIsScalarArithmetic inclusive_scan(Group g, T x, +EnableIfIsScalarArithmetic inclusive_scan(Group, T x, BinaryOperation binary_op) { static_assert(sycl::detail::is_generic_group::value, "Group algorithms only support the sycl::group and " @@ -688,6 +734,7 @@ inclusive_scan(Group g, V x, BinaryOperation binary_op, T init) { } return inclusive_scan(g, x, binary_op); #else + (void)g; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -749,6 +796,9 @@ inclusive_scan(Group g, InPtr first, InPtr last, OutPtr result, } return result + N; #else + (void)g; + (void)last; + (void)result; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif @@ -780,6 +830,7 @@ template bool leader(Group g) { sycl::detail::get_local_linear_id(g); return (linear_id == 0); #else + (void)g; throw runtime_error("Group algorithms are not supported on host device.", PI_INVALID_DEVICE); #endif diff --git a/sycl/include/CL/sycl/intel/pipes.hpp b/sycl/include/CL/sycl/intel/pipes.hpp index 9f8de0e0e4efe..8396bc1e215fc 100644 --- a/sycl/include/CL/sycl/intel/pipes.hpp +++ b/sycl/include/CL/sycl/intel/pipes.hpp @@ -30,6 +30,7 @@ template class pipe { __spirv_ReadPipe(RPipe, &TempData, m_Size, m_Alignment)); return TempData; #else + (void)Success; assert(!"Pipes are not supported on a host device!"); #endif // __SYCL_DEVICE_ONLY__ } @@ -43,6 +44,8 @@ template class pipe { Success = !static_cast( __spirv_WritePipe(WPipe, &Data, m_Size, m_Alignment)); #else + (void)Success; + (void)Data; assert(!"Pipes are not supported on a host device!"); #endif // __SYCL_DEVICE_ONLY__ } @@ -70,6 +73,7 @@ template class pipe { __spirv_CreatePipeFromPipeStorage_write(&m_Storage); __spirv_WritePipeBlockingINTEL(WPipe, &Data, m_Size, m_Alignment); #else + (void)Data; assert(!"Pipes are not supported on a host device!"); #endif // __SYCL_DEVICE_ONLY__ } @@ -119,6 +123,7 @@ class kernel_readable_io_pipe { __spirv_ReadPipe(RPipe, &TempData, m_Size, m_Alignment)); return TempData; #else + (void)Success; assert(!"Pipes are not supported on a host device!"); #endif // __SYCL_DEVICE_ONLY__ } @@ -162,6 +167,8 @@ class kernel_writeable_io_pipe { Success = !static_cast( __spirv_WritePipe(WPipe, &Data, m_Size, m_Alignment)); #else + (void)Data; + (void)Success; assert(!"Pipes are not supported on a host device!"); #endif // __SYCL_DEVICE_ONLY__ } @@ -175,6 +182,7 @@ class kernel_writeable_io_pipe { __spirv_CreatePipeFromPipeStorage_write(&m_Storage); __spirv_WritePipeBlockingINTEL(WPipe, &Data, m_Size, m_Alignment); #else + (void)Data; assert(!"Pipes are not supported on a host device!"); #endif // __SYCL_DEVICE_ONLY__ } diff --git a/sycl/include/CL/sycl/intel/reduction.hpp b/sycl/include/CL/sycl/intel/reduction.hpp index 4b9bec9829c81..41ee8dcfcfb65 100644 --- a/sycl/include/CL/sycl/intel/reduction.hpp +++ b/sycl/include/CL/sycl/intel/reduction.hpp @@ -177,7 +177,7 @@ class reducer::value>> { public: reducer() : MValue(getIdentity()) {} - reducer(const T &Identity) : MValue(getIdentity()) {} + reducer(const T &) : MValue(getIdentity()) {} void combine(const T &Partial) { BinaryOperation BOp; @@ -594,7 +594,7 @@ struct get_reduction_aux_2nd_kernel_name_t { template enable_if_t reduCGFunc(handler &CGH, KernelType KernelFunc, const nd_range &Range, - Reduction &Redu, typename Reduction::rw_accessor_type &Out) { + Reduction &, typename Reduction::rw_accessor_type &Out) { size_t NWorkItems = Range.get_global_range().size(); size_t WGSize = Range.get_local_range().size(); @@ -1095,7 +1095,7 @@ template detail::reduction_impl reduction(accessor &Acc, - const T &Identity, BinaryOperation Combiner) { + const T &Identity, BinaryOperation) { // The Combiner argument was needed only to define the BinaryOperation param. return detail::reduction_impl( Acc, Identity); @@ -1112,7 +1112,7 @@ detail::enable_if_t< detail::IsKnownIdentityOp::value, detail::reduction_impl> reduction(accessor &Acc, - BinaryOperation Combiner) { + BinaryOperation) { // The Combiner argument was needed only to define the BinaryOperation param. return detail::reduction_impl( Acc); diff --git a/sycl/include/CL/sycl/intel/sub_group_host.hpp b/sycl/include/CL/sycl/intel/sub_group_host.hpp index d805c42b8869c..0c5762462e1ff 100644 --- a/sycl/include/CL/sycl/intel/sub_group_host.hpp +++ b/sycl/include/CL/sycl/intel/sub_group_host.hpp @@ -59,55 +59,55 @@ struct sub_group { /* --- vote / ballot functions --- */ - bool any(bool predicate) const { + bool any(bool) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } - bool all(bool predicate) const { + bool all(bool) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } /* --- collectives --- */ - template T broadcast(T x, id<1> local_id) const { + template T broadcast(T, id<1>) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - T reduce(T x, BinaryOperation op) const { + T reduce(T, BinaryOperation) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - T reduce(T x, T init, BinaryOperation op) const { + T reduce(T, T, BinaryOperation) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - T exclusive_scan(T x, BinaryOperation op) const { + T exclusive_scan(T, BinaryOperation) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - T exclusive_scan(T x, T init, BinaryOperation op) const { + T exclusive_scan(T, T, BinaryOperation) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - T inclusive_scan(T x, BinaryOperation op) const { + T inclusive_scan(T, BinaryOperation) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - T inclusive_scan(T x, BinaryOperation op, T init) const { + T inclusive_scan(T, BinaryOperation, T) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } @@ -115,38 +115,36 @@ struct sub_group { /* --- one - input shuffles --- */ /* indices in [0 , sub - group size ) */ - template T shuffle(T x, id<1> local_id) const { + template T shuffle(T, id<1>) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } - template T shuffle_down(T x, uint32_t delta) const { + template T shuffle_down(T, uint32_t) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } - template T shuffle_up(T x, uint32_t delta) const { + template T shuffle_up(T, uint32_t) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } - template T shuffle_xor(T x, id<1> value) const { + template T shuffle_xor(T, id<1>) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } /* --- two - input shuffles --- */ /* indices in [0 , 2* sub - group size ) */ - template T shuffle(T x, T y, id<1> local_id) const { + template T shuffle(T, T, id<1>) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } - template - T shuffle_down(T current, T next, uint32_t delta) const { + template T shuffle_down(T, T, uint32_t) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } - template - T shuffle_up(T previous, T current, uint32_t delta) const { + template T shuffle_up(T, T, uint32_t) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } @@ -154,25 +152,25 @@ struct sub_group { /* --- sub - group load / stores --- */ /* these can map to SIMD or block read / write hardware where available */ template - T load(const multi_ptr src) const { + T load(const multi_ptr) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - vec load(const multi_ptr src) const { + vec load(const multi_ptr) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - void store(multi_ptr dst, const T &x) const { + void store(multi_ptr, const T &) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } template - void store(multi_ptr dst, const vec &x) const { + void store(multi_ptr, const vec &) const { throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } @@ -180,6 +178,7 @@ struct sub_group { /* --- synchronization functions --- */ void barrier(access::fence_space accessSpace = access::fence_space::global_and_local) const { + (void)accessSpace; throw runtime_error("Subgroups are not supported on host device. ", PI_INVALID_DEVICE); } diff --git a/sycl/include/CL/sycl/multi_ptr.hpp b/sycl/include/CL/sycl/multi_ptr.hpp index dbdc0d045bca6..9dc45483d61a6 100644 --- a/sycl/include/CL/sycl/multi_ptr.hpp +++ b/sycl/include/CL/sycl/multi_ptr.hpp @@ -597,62 +597,62 @@ bool operator>=(const multi_ptr &lhs, } template -bool operator!=(const multi_ptr &lhs, std::nullptr_t rhs) { +bool operator!=(const multi_ptr &lhs, std::nullptr_t) { return lhs.get() != nullptr; } template -bool operator!=(std::nullptr_t lhs, const multi_ptr &rhs) { +bool operator!=(std::nullptr_t, const multi_ptr &rhs) { return rhs.get() != nullptr; } template -bool operator==(const multi_ptr &lhs, std::nullptr_t rhs) { +bool operator==(const multi_ptr &lhs, std::nullptr_t) { return lhs.get() == nullptr; } template -bool operator==(std::nullptr_t lhs, const multi_ptr &rhs) { +bool operator==(std::nullptr_t, const multi_ptr &rhs) { return rhs.get() == nullptr; } template -bool operator>(const multi_ptr &lhs, std::nullptr_t rhs) { +bool operator>(const multi_ptr &lhs, std::nullptr_t) { return lhs.get() != nullptr; } template -bool operator>(std::nullptr_t lhs, const multi_ptr &rhs) { +bool operator>(std::nullptr_t, const multi_ptr &) { return false; } template -bool operator<(const multi_ptr &lhs, std::nullptr_t rhs) { +bool operator<(const multi_ptr &, std::nullptr_t) { return false; } template -bool operator<(std::nullptr_t lhs, const multi_ptr &rhs) { +bool operator<(std::nullptr_t, const multi_ptr &rhs) { return rhs.get() != nullptr; } template -bool operator>=(const multi_ptr &lhs, std::nullptr_t rhs) { +bool operator>=(const multi_ptr &, std::nullptr_t) { return true; } template -bool operator>=(std::nullptr_t lhs, const multi_ptr &rhs) { +bool operator>=(std::nullptr_t, const multi_ptr &rhs) { return rhs.get() == nullptr; } template -bool operator<=(const multi_ptr &lhs, std::nullptr_t rhs) { +bool operator<=(const multi_ptr &lhs, std::nullptr_t) { return lhs.get() == nullptr; } template -bool operator<=(std::nullptr_t lhs, const multi_ptr &rhs) { +bool operator<=(std::nullptr_t, const multi_ptr &rhs) { return rhs.get() == nullptr; } diff --git a/sycl/include/CL/sycl/nd_item.hpp b/sycl/include/CL/sycl/nd_item.hpp index 4ac682ea6e2b9..45e474621434c 100644 --- a/sycl/include/CL/sycl/nd_item.hpp +++ b/sycl/include/CL/sycl/nd_item.hpp @@ -122,6 +122,7 @@ template class nd_item { accessMode == access::mode::read_write, access::fence_space>::type accessSpace = access::fence_space::global_and_local) const { + (void)accessSpace; Group.mem_fence(); } diff --git a/sycl/include/CL/sycl/program.hpp b/sycl/include/CL/sycl/program.hpp index 9e3d2e5714086..eaf885984d9f3 100644 --- a/sycl/include/CL/sycl/program.hpp +++ b/sycl/include/CL/sycl/program.hpp @@ -307,6 +307,8 @@ class __SYCL_EXPORT program { std::is_floating_point::value, "unsupported specialization constant type"); #ifdef __SYCL_DEVICE_ONLY__ + (void)Cst; + (void)Name; return experimental::spec_constant(); #else set_spec_constant_impl(Name, &Cst, sizeof(T)); diff --git a/sycl/include/CL/sycl/sampler.hpp b/sycl/include/CL/sycl/sampler.hpp index 2f099aae25ce3..e7a72d3d71a74 100644 --- a/sycl/include/CL/sycl/sampler.hpp +++ b/sycl/include/CL/sycl/sampler.hpp @@ -91,6 +91,7 @@ namespace std { template <> struct hash { size_t operator()(const cl::sycl::sampler &s) const { #ifdef __SYCL_DEVICE_ONLY__ + (void)s; return 0; #else return hash>()( diff --git a/sycl/include/CL/sycl/stream.hpp b/sycl/include/CL/sycl/stream.hpp index 4c525b5a9b962..4fa17cdf2bd43 100644 --- a/sycl/include/CL/sycl/stream.hpp +++ b/sycl/include/CL/sycl/stream.hpp @@ -483,6 +483,7 @@ namespace std { template <> struct hash { size_t operator()(const cl::sycl::stream &S) const { #ifdef __SYCL_DEVICE_ONLY__ + (void)S; return 0; #else return hash>()( diff --git a/sycl/include/CL/sycl/types.hpp b/sycl/include/CL/sycl/types.hpp index b94c183e01e41..3e08af085bae1 100644 --- a/sycl/include/CL/sycl/types.hpp +++ b/sycl/include/CL/sycl/types.hpp @@ -117,8 +117,8 @@ using rel_t = typename std::conditional< template class GetOp { public: using DataT = T; - DataT getValue(size_t Index) const { return 0; } - DataT operator()(DataT LHS, DataT Rhs) { return 0; } + DataT getValue(size_t) const { return 0; } + DataT operator()(DataT, DataT) { return 0; } }; // Special type for working SwizzleOp with scalars, stores a scalar and gives @@ -128,7 +128,7 @@ template class GetScalarOp { public: using DataT = T; GetScalarOp(DataT Data) : m_Data(Data) {} - DataT getValue(size_t Index) const { return m_Data; } + DataT getValue(size_t) const { return m_Data; } private: DataT m_Data; @@ -1216,13 +1216,13 @@ template class vec { template ::type> - void setValue(int Index, const DataT &Value, float) { + void setValue(int, const DataT &Value, float) { m_Data = Value; } template ::type> - DataT getValue(int Index, float) const { + DataT getValue(int, float) const { return m_Data; } diff --git a/sycl/include/CL/sycl/usm/usm_allocator.hpp b/sycl/include/CL/sycl/usm/usm_allocator.hpp index 1ee4d6ae44e66..f821175205c03 100644 --- a/sycl/include/CL/sycl/usm/usm_allocator.hpp +++ b/sycl/include/CL/sycl/usm/usm_allocator.hpp @@ -65,7 +65,7 @@ class usm_allocator { template < usm::alloc AllocT = AllocKind, typename std::enable_if::type = 0> - void construct(pointer Ptr, const_reference Val) { + void construct(pointer, const_reference) { throw feature_not_supported( "Device pointers do not support construct on host", PI_INVALID_OPERATION); @@ -86,7 +86,7 @@ class usm_allocator { template < usm::alloc AllocT = AllocKind, typename std::enable_if::type = 0> - void destroy(pointer Ptr) { + void destroy(pointer) { // This method must be a NOP for device pointers. } @@ -104,7 +104,7 @@ class usm_allocator { template < usm::alloc AllocT = AllocKind, typename std::enable_if::type = 0> - pointer address(reference Val) const { + pointer address(reference) const { throw feature_not_supported( "Device pointers do not support address on host", PI_INVALID_OPERATION); } @@ -119,7 +119,7 @@ class usm_allocator { template < usm::alloc AllocT = AllocKind, typename std::enable_if::type = 0> - const_pointer address(const_reference Val) const { + const_pointer address(const_reference) const { throw feature_not_supported( "Device pointers do not support address on host", PI_INVALID_OPERATION); } @@ -142,7 +142,7 @@ class usm_allocator { /// /// \param Ptr is a pointer to memory being deallocated. /// \param Size is a number of elements previously passed to allocate. - void deallocate(pointer Ptr, size_t Size) { + void deallocate(pointer Ptr, size_t) { if (Ptr) { free(Ptr, MContext); } diff --git a/sycl/test/CMakeLists.txt b/sycl/test/CMakeLists.txt index ffbbbe6e38be3..9f53afd5b8af4 100644 --- a/sycl/test/CMakeLists.txt +++ b/sycl/test/CMakeLists.txt @@ -6,6 +6,7 @@ set(SYCL_INCLUDE "${SYCL_INCLUDE_BUILD_DIR}") set(SYCL_TOOLS_SRC_DIR "${PROJECT_SOURCE_DIR}/tools/") set(LLVM_BUILD_BINARY_DIRS "${LLVM_BINARY_DIR}/bin/") set(LLVM_BUILD_LIBRARY_DIRS "${LLVM_BINARY_DIR}/lib/") +set(SOURCE_SYCL_INCLUDE "${sycl_inc_dir}") set(RT_TEST_ARGS ${RT_TEST_ARGS} "-v") set(DEPLOY_RT_TEST_ARGS ${DEPLOY_RT_TEST_ARGS} "-v -D SYCL_TOOLS_DIR=${CMAKE_INSTALL_PREFIX}/bin -D SYCL_LIBS_DIR=${CMAKE_INSTALL_PREFIX}/lib${LLVM_LIBDIR_SUFFIX} -D SYCL_INCLUDE=${SYCL_INCLUDE_DEPLOY_DIR}") diff --git a/sycl/test/lit.cfg.py b/sycl/test/lit.cfg.py index 4baaeac0f1dd3..2b877faf483ee 100644 --- a/sycl/test/lit.cfg.py +++ b/sycl/test/lit.cfg.py @@ -59,6 +59,7 @@ llvm_config.with_environment('PATH', config.sycl_tools_dir, append_path=True) config.substitutions.append( ('%threads_lib', config.sycl_threads_lib) ) +config.substitutions.append( ('%source_sycl_include', config.source_sycl_include) ) config.substitutions.append( ('%sycl_libs_dir', config.sycl_libs_dir ) ) config.substitutions.append( ('%sycl_include', config.sycl_include ) ) config.substitutions.append( ('%sycl_source_dir', config.sycl_source_dir) ) diff --git a/sycl/test/lit.site.cfg.py.in b/sycl/test/lit.site.cfg.py.in index a8d456eddd3b5..784784f3b762f 100644 --- a/sycl/test/lit.site.cfg.py.in +++ b/sycl/test/lit.site.cfg.py.in @@ -6,6 +6,7 @@ config.llvm_tools_dir = "@LLVM_TOOLS_DIR@" config.lit_tools_dir = "@LLVM_LIT_TOOLS_DIR@" config.sycl_tools_dir = lit_config.params.get('SYCL_TOOLS_DIR', "@LLVM_TOOLS_DIR@") config.sycl_include = lit_config.params.get('SYCL_INCLUDE', "@SYCL_INCLUDE@") +config.source_sycl_include = lit_config.params.get('SOURCE_SYCL_INCLUDE', "@SOURCE_SYCL_INCLUDE@") config.sycl_obj_root = "@SYCL_BINARY_DIR@" config.sycl_source_dir = "@SYCL_SOURCE_DIR@/source" config.sycl_libs_dir = lit_config.params.get('SYCL_LIBS_DIR', "@LLVM_LIBS_DIR@") diff --git a/sycl/test/warnings/warnings.cpp b/sycl/test/warnings/warnings.cpp index ea2409f828f89..c22d61c6b5296 100644 --- a/sycl/test/warnings/warnings.cpp +++ b/sycl/test/warnings/warnings.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -Wall -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version -fsycl %s -o %t.out +// RUN: %clangxx -fsycl -I %source_sycl_include -Wall -Wextra -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out #include From 191be27aa2986094894ba4bbbffac4c18ee1439f Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Tue, 19 May 2020 12:27:40 +0300 Subject: [PATCH 2/8] warnings.cpp fix Signed-off-by: Aleksander Fadeev --- sycl/test/warnings/warnings.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/warnings/warnings.cpp b/sycl/test/warnings/warnings.cpp index c22d61c6b5296..5a6d91708c79f 100644 --- a/sycl/test/warnings/warnings.cpp +++ b/sycl/test/warnings/warnings.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -I %source_sycl_include -Wall -Wextra -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out +// RUN: %clangxx -fsycl -I %source_sycl_include -Wall -Wextra -Wno-ignored-attributes -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out #include From 0a6b77df400bc7a651040721b39f33dfe3fae615 Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Tue, 19 May 2020 13:43:23 +0300 Subject: [PATCH 3/8] Warnings fix Signed-off-by: Aleksander Fadeev --- sycl/include/CL/sycl/intel/reduction.hpp | 2 ++ sycl/include/CL/sycl/types.hpp | 8 ++++---- 2 files changed, 6 insertions(+), 4 deletions(-) diff --git a/sycl/include/CL/sycl/intel/reduction.hpp b/sycl/include/CL/sycl/intel/reduction.hpp index 41ee8dcfcfb65..e1b4229dab824 100644 --- a/sycl/include/CL/sycl/intel/reduction.hpp +++ b/sycl/include/CL/sycl/intel/reduction.hpp @@ -1128,6 +1128,7 @@ detail::reduction_impl reduction(T *VarPtr, const T &Identity, BinaryOperation Combiner) { // The Combiner argument was needed only to define the BinaryOperation param. + (void)Combiner; return detail::reduction_impl(VarPtr, Identity); @@ -1146,6 +1147,7 @@ detail::enable_if_t::value, access::placeholder::true_t>> reduction(T *VarPtr, BinaryOperation Combiner) { // The Combiner argument was needed only to define the BinaryOperation param. + (void)Combiner; return detail::reduction_impl(VarPtr); diff --git a/sycl/include/CL/sycl/types.hpp b/sycl/include/CL/sycl/types.hpp index 3e08af085bae1..e7605d6a5cf19 100644 --- a/sycl/include/CL/sycl/types.hpp +++ b/sycl/include/CL/sycl/types.hpp @@ -211,8 +211,8 @@ using is_uint_to_uint = template using is_sint_to_from_uint = std::integral_constant< - bool, is_sugeninteger::value && is_sigeninteger::value || - is_sigeninteger::value && is_sugeninteger::value>; + bool, (is_sugeninteger::value && is_sigeninteger::value) || + (is_sigeninteger::value && is_sugeninteger::value)>; template using is_sint_to_float = @@ -486,8 +486,8 @@ __SYCL_GENERATE_CONVERT_IMPL_FOR_ROUNDING_MODE(rtn, Rtn) template detail::enable_if_t< - (!is_standard_type::value && !is_standard_type::value || - !is_standard_type::value && !is_standard_type::value) && + ((!is_standard_type::value && !is_standard_type::value) || + (!is_standard_type::value && !is_standard_type::value)) && !std::is_same::value, R> convertImpl(T Value) { From ac59ae0d8427d2b04ffd7859e4606c225d7a3e01 Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Tue, 19 May 2020 15:12:09 +0300 Subject: [PATCH 4/8] Warnings fix 2 Signed-off-by: Aleksander Fadeev --- sycl/test/warnings/warnings.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/warnings/warnings.cpp b/sycl/test/warnings/warnings.cpp index 5a6d91708c79f..57483e33db751 100644 --- a/sycl/test/warnings/warnings.cpp +++ b/sycl/test/warnings/warnings.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -I %source_sycl_include -Wall -Wextra -Wno-ignored-attributes -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out +// RUN: %clangxx -fsycl -I %source_sycl_include -Wall -Wextra -Wno-ignored-attributes -Wno-deprecated-declarations -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out #include From 81d2657d0e468c60cc15f5598cf1d94a59cdb3da Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Wed, 20 May 2020 13:54:11 +0300 Subject: [PATCH 5/8] rebase fix Signed-off-by: Aleksander Fadeev --- sycl/include/CL/sycl/intel/reduction.hpp | 1 + 1 file changed, 1 insertion(+) diff --git a/sycl/include/CL/sycl/intel/reduction.hpp b/sycl/include/CL/sycl/intel/reduction.hpp index e1b4229dab824..19fdac7e1b410 100644 --- a/sycl/include/CL/sycl/intel/reduction.hpp +++ b/sycl/include/CL/sycl/intel/reduction.hpp @@ -394,6 +394,7 @@ class reduction_impl { enable_if_t::value> * = nullptr> reduction_impl(accessor_type &Acc, const T &Identity) : MAcc(Acc), MIdentity(getIdentity()) { + (void)Identity; assert(Acc.get_count() == 1 && "Only scalar/1-element reductions are supported now."); // For now the implementation ignores the identity value given by user From 0ece5d976f5b2e2cffe4a50477244a900967272f Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Thu, 21 May 2020 10:05:27 +0300 Subject: [PATCH 6/8] warnings.cpp RUN fix Signed-off-by: Aleksander Fadeev --- sycl/test/warnings/warnings.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/test/warnings/warnings.cpp b/sycl/test/warnings/warnings.cpp index 57483e33db751..438e1ae9a3960 100644 --- a/sycl/test/warnings/warnings.cpp +++ b/sycl/test/warnings/warnings.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -I %source_sycl_include -Wall -Wextra -Wno-ignored-attributes -Wno-deprecated-declarations -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out +// RUN: %clangxx -fsycl --no-system-header-prefix=CL/ -Wall -Wextra -Wno-ignored-attributes -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out #include @@ -7,7 +7,7 @@ using namespace cl::sycl; int main(void) { // add a very simple kernel to see if compilation succeeds with -Werror int data1[10] = {-1, -1, -1, -1, -1, -1, -1, -1, -1, -1}; - + exit(1); buffer B(data1, range<1>(10), {property::buffer::use_host_ptr()}); queue Q; Q.submit([&](handler &CGH) { From a3eb3d80d64bf7eae80a5b4810f58843273441b2 Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Thu, 21 May 2020 10:31:17 +0300 Subject: [PATCH 7/8] lit.cfg.py fix Signed-off-by: Aleksander Fadeev --- sycl/test/CMakeLists.txt | 1 - sycl/test/lit.cfg.py | 1 - sycl/test/lit.site.cfg.py.in | 1 - 3 files changed, 3 deletions(-) diff --git a/sycl/test/CMakeLists.txt b/sycl/test/CMakeLists.txt index 9f53afd5b8af4..ffbbbe6e38be3 100644 --- a/sycl/test/CMakeLists.txt +++ b/sycl/test/CMakeLists.txt @@ -6,7 +6,6 @@ set(SYCL_INCLUDE "${SYCL_INCLUDE_BUILD_DIR}") set(SYCL_TOOLS_SRC_DIR "${PROJECT_SOURCE_DIR}/tools/") set(LLVM_BUILD_BINARY_DIRS "${LLVM_BINARY_DIR}/bin/") set(LLVM_BUILD_LIBRARY_DIRS "${LLVM_BINARY_DIR}/lib/") -set(SOURCE_SYCL_INCLUDE "${sycl_inc_dir}") set(RT_TEST_ARGS ${RT_TEST_ARGS} "-v") set(DEPLOY_RT_TEST_ARGS ${DEPLOY_RT_TEST_ARGS} "-v -D SYCL_TOOLS_DIR=${CMAKE_INSTALL_PREFIX}/bin -D SYCL_LIBS_DIR=${CMAKE_INSTALL_PREFIX}/lib${LLVM_LIBDIR_SUFFIX} -D SYCL_INCLUDE=${SYCL_INCLUDE_DEPLOY_DIR}") diff --git a/sycl/test/lit.cfg.py b/sycl/test/lit.cfg.py index 2b877faf483ee..4baaeac0f1dd3 100644 --- a/sycl/test/lit.cfg.py +++ b/sycl/test/lit.cfg.py @@ -59,7 +59,6 @@ llvm_config.with_environment('PATH', config.sycl_tools_dir, append_path=True) config.substitutions.append( ('%threads_lib', config.sycl_threads_lib) ) -config.substitutions.append( ('%source_sycl_include', config.source_sycl_include) ) config.substitutions.append( ('%sycl_libs_dir', config.sycl_libs_dir ) ) config.substitutions.append( ('%sycl_include', config.sycl_include ) ) config.substitutions.append( ('%sycl_source_dir', config.sycl_source_dir) ) diff --git a/sycl/test/lit.site.cfg.py.in b/sycl/test/lit.site.cfg.py.in index 784784f3b762f..a8d456eddd3b5 100644 --- a/sycl/test/lit.site.cfg.py.in +++ b/sycl/test/lit.site.cfg.py.in @@ -6,7 +6,6 @@ config.llvm_tools_dir = "@LLVM_TOOLS_DIR@" config.lit_tools_dir = "@LLVM_LIT_TOOLS_DIR@" config.sycl_tools_dir = lit_config.params.get('SYCL_TOOLS_DIR', "@LLVM_TOOLS_DIR@") config.sycl_include = lit_config.params.get('SYCL_INCLUDE', "@SYCL_INCLUDE@") -config.source_sycl_include = lit_config.params.get('SOURCE_SYCL_INCLUDE', "@SOURCE_SYCL_INCLUDE@") config.sycl_obj_root = "@SYCL_BINARY_DIR@" config.sycl_source_dir = "@SYCL_SOURCE_DIR@/source" config.sycl_libs_dir = lit_config.params.get('SYCL_LIBS_DIR', "@LLVM_LIBS_DIR@") From 735e7b8c79a0638f57656c518ff6000c514702b4 Mon Sep 17 00:00:00 2001 From: Aleksander Fadeev Date: Thu, 21 May 2020 15:41:42 +0300 Subject: [PATCH 8/8] warnings.cpp fix 2 Signed-off-by: Aleksander Fadeev --- sycl/test/warnings/warnings.cpp | 3 +-- 1 file changed, 1 insertion(+), 2 deletions(-) diff --git a/sycl/test/warnings/warnings.cpp b/sycl/test/warnings/warnings.cpp index 438e1ae9a3960..1d81132aa3edf 100644 --- a/sycl/test/warnings/warnings.cpp +++ b/sycl/test/warnings/warnings.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl --no-system-header-prefix=CL/ -Wall -Wextra -Wno-ignored-attributes -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out +// RUN: %clangxx -fsycl --no-system-header-prefix=CL/ -Wall -Wextra -Wno-ignored-attributes -Wno-deprecated-declarations -Wpessimizing-move -Wunused-variable -Wmismatched-tags -Wunneeded-internal-declaration -Werror -Wno-unknown-cuda-version %s -o %t.out #include @@ -7,7 +7,6 @@ using namespace cl::sycl; int main(void) { // add a very simple kernel to see if compilation succeeds with -Werror int data1[10] = {-1, -1, -1, -1, -1, -1, -1, -1, -1, -1}; - exit(1); buffer B(data1, range<1>(10), {property::buffer::use_host_ptr()}); queue Q; Q.submit([&](handler &CGH) {