[SYCL] Implement SYCL-2020 reductions with read_write to reduction va…

[v-klochkov]

…riable

LIT tests: intel/llvm-test-suite#170

The patch has ABI non-breaking change requiring up-lifting the MINOR component
of the SYCL version.

This patch:

• adds SYCL-2020 property::reduction::initialize_to_identity
• implements support for SYCL-2020 reductions for which initialize_to_identity
  is NOT used (corresponds to read_write and USM ONEAPI::reductions).
  sycl::reduction re-uses ONEAPI::reduction implementation/classes and
  automatically creates placeholder accessors for sycl::reduction called
  with sycl::buffer argument.
• adds operator++ for reducer class defined in SYCL-2020
• fixes 2 errors in an ONEAPI::reduction used with placeholder accessors.

The attribute "initialize_to_identity" is NOT supported yet.
The multi-dimensional reductions are NOT supported yet.
Reductions accepting sycl::span are not supported yet.
Reductions cannot be yet used in parallel_for() accepting sycl::range

Signed-off-by: Vyacheslav N Klochkov vyacheslav.n.klochkov@intel.com

[Pennycook]

I might have missed it, but I don't see any support here for the new interface that accepts a sycl::span. I assume that's because we haven't implemented sycl::span yet, but I think it would be good to note that this feature isn't implemented in the commit message.

[v-klochkov]

Adding the new bool property property::reduction::initialize_to_identity
caused creation of new exported symbols, which is ABI non-breaking change
and thus requires up-lifting the MINOR component of the SYCL version - f22d47a

Adding @romanovvlad and @alexbatashev to review this aspect of the patch.

[alexbatashev]

Otherwise LGTM

[tomdeakin]

Having trouble using the version with exiting SYCL 1.2.1 and newer SYCL 2020 codes.

Similar errors occur for BabelStream main and sycl-2020 branches.
NB: the sycl-2020 branch needs source edits to misspell no_init as noinit and remove the use of sycl::property::reduction::initialize_to_identity.

I followed the instructions for building this LLVM with the Python scripts which worked. Exported the environment variables, and have the runtimes in place.

Build the BabelStream code as below:

$ make -f SYCL.make COMPILER=DPCPP SYCL_DPCPP_SYCLCXX=clang++ -B
clang++  -O3 --std=c++17  -DSYCL   main.cpp SYCLStream.cpp -o sycl-stream

The errors I get are below, which look like a load of missing libraries.

/tmp/SYCLStream-c82a44.o: In function `getDeviceList()':                                                                                                                                                    
SYCLStream.cpp:(.text+0x16): undefined reference to `cl::sycl::device::get_devices(cl::sycl::info::device_type)'                                                                                            
/tmp/SYCLStream-c82a44.o: In function `getDeviceName(int)':                                                                                                                                                 
SYCLStream.cpp:(.text+0x525): undefined reference to `cl::sycl::info::param_traits<cl::sycl::info::device, (cl::sycl::info::device)4139>::return_type cl::sycl::device::get_info<(cl::sycl::info::device)413
9>() const'                                                                                                                                                                                                 
/tmp/SYCLStream-c82a44.o: In function `getDeviceDriver(int)':                                                                                                                                               
SYCLStream.cpp:(.text+0x5f5): undefined reference to `cl::sycl::info::param_traits<cl::sycl::info::device, (cl::sycl::info::device)4141>::return_type cl::sycl::device::get_info<(cl::sycl::info::device)414
1>() const'                                                                                                                                                                                                 
/tmp/SYCLStream-c82a44.o: In function `SYCLStream<float>::SYCLStream(int, int)':                                                                                                                            
SYCLStream.cpp:(.text._ZN10SYCLStreamIfEC2Eii[_ZN10SYCLStreamIfEC5Eii]+0x88): undefined reference to `cl::sycl::device::is_cpu() const'                                                                     
SYCLStream.cpp:(.text._ZN10SYCLStreamIfEC2Eii[_ZN10SYCLStreamIfEC5Eii]+0x96): undefined reference to `cl::sycl::info::param_traits<cl::sycl::info::device, (cl::sycl::info::device)4098>::return_type cl::sy
cl::device::get_info<(cl::sycl::info::device)4098>() const'                                                                                                                                                 
SYCLStream.cpp:(.text._ZN10SYCLStreamIfEC2Eii[_ZN10SYCLStreamIfEC5Eii]+0xa7): undefined reference to `cl::sycl::info::param_traits<cl::sycl::info::device, (cl::sycl::info::device)4155>::return_type cl::sy
cl::device::get_info<(cl::sycl::info::device)4155>() const'                                                                                                                                                 
SYCLStream.cpp:(.text._ZN10SYCLStreamIfEC2Eii[_ZN10SYCLStreamIfEC5Eii]+0xb7): undefined reference to `cl::sycl::info::param_traits<cl::sycl::info::device, (cl::sycl::info::device)4098>::return_type cl::sy
cl::device::get_info<(cl::sycl::info::device)4098>() const'                                                                                                                                                 
SYCLStream.cpp:(.text._ZN10SYCLStreamIfEC2Eii[_ZN10SYCLStreamIfEC5Eii]+0xc9): undefined reference to `cl::sycl::info::param_traits<cl::sycl::info::device, (cl::sycl::info::device)4100>::return_type cl::sy
cl::device::get_info<(cl::sycl::info::device)4100>() const'                                                                                                                                                 
SYCLStream.cpp:(.text._ZN10SYCLStreamIfEC2Eii[_ZN10SYCLStreamIfEC5Eii]+0x2d1): undefined reference to `cl::sycl::queue::queue(cl::sycl::device const&, std::function<void (cl::sycl::exception_list)> const&
, cl::sycl::property_list const&)'                                                                                                                                                                          
/tmp/SYCLStream-c82a44.o: In function `SYCLStream<float>::copy()':                                                                                                                                          
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE4copyEv[_ZN10SYCLStreamIfE4copyEv]+0x53): undefined reference to `cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>, cl::sycl::detail::code_loca
tion const&)'
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE4copyEv[_ZN10SYCLStreamIfE4copyEv]+0x106): undefined reference to `cl::sycl::queue::wait_proxy(cl::sycl::detail::code_location const&)'
/tmp/SYCLStream-c82a44.o: In function `SYCLStream<float>::add()':
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE3addEv[_ZN10SYCLStreamIfE3addEv]+0x53): undefined reference to `cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>, cl::sycl::detail::code_locati
on const&)'
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE3addEv[_ZN10SYCLStreamIfE3addEv]+0x106): undefined reference to `cl::sycl::queue::wait_proxy(cl::sycl::detail::code_location const&)'
/tmp/SYCLStream-c82a44.o: In function `SYCLStream<float>::mul()':
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE3mulEv[_ZN10SYCLStreamIfE3mulEv]+0x65): undefined reference to `cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>, cl::sycl::detail::code_locati
on const&)'
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE3mulEv[_ZN10SYCLStreamIfE3mulEv]+0x118): undefined reference to `cl::sycl::queue::wait_proxy(cl::sycl::detail::code_location const&)'
/tmp/SYCLStream-c82a44.o: In function `SYCLStream<float>::triad()':
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE5triadEv[_ZN10SYCLStreamIfE5triadEv]+0x65): undefined reference to `cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>, cl::sycl::detail::code_lo
cation const&)'
SYCLStream.cpp:(.text._ZN10SYCLStreamIfE5triadEv[_ZN10SYCLStreamIfE5triadEv]+0x118): undefined reference to `cl::sycl::queue::wait_proxy(cl::sycl::detail::code_location const&)'
........ much more

[tomdeakin]

I was missing the hidden flags normally provided by the dpcpp wrapper: -fsycl -fsycl-unnamed-lambda.
Link issues go away with those flags.

It's a shame there is no range interface yet. I have to use nd_range and choose an arbitrary work-group size - can the implementation not just do that for me as a first attempt, and then optimise that choice later?

[v-klochkov]

Jenkins/Precommit testing reported only 1 test fail - SYCL/Reduction/reduction_reducer_op_eq.cpp
that test fails because of having such 2 lines:
using namespace sycl;
using namespace sycl::ONEAPI;
That is fixed in the corresponding patch for these changes: intel/llvm-test-suite#170

[v-klochkov]

@alexbatashev thank you for the review. I reverted the update of MINOR component of the VERSION.
Please replace 'change request' with approval if you ok with the changes.

The fail on only 1 LIT test (SYCL/Reduction/reduction_reducer_op_eq.cpp) is expected and is fixed by intel/llvm-test-suite#170

[v-klochkov]

BTW, I uploaded an additional NFC fix for failed LIT test: intel/llvm-test-suite#177
After it is merged I'll restart testing here and perhaps have clean testing results.