Skip to content
This repository was archived by the owner on Mar 28, 2023. It is now read-only.

[SYCL] Eliminate compilation warnings in reduction tests; disable 1 test #381

Closed
wants to merge 1 commit into from
Closed

[SYCL] Eliminate compilation warnings in reduction tests; disable 1 test #381

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
17 changes: 9 additions & 8 deletions SYCL/Reduction/reduction_big_data.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -4,7 +4,7 @@
// RUN: %CPU_RUN_PLACEHOLDER %t.out

// RUNx: %HOST_RUN_PLACEHOLDER %t.out
// TODO: Enable the test for HOST when it supports ONEAPI::reduce() and
// TODO: Enable the test for HOST when it supports ext::oneapi::reduce() and
// barrier()

// This test performs basic checks of parallel_for(nd_range, reduction, func)
Expand Down Expand Up @@ -64,12 +64,12 @@ void test(queue &Q, T Identity) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, Dim, access::mode::discard_write, access::target::global_buffer>
Out(OutBuf, CGH);
CGH.parallel_for<KernelName>(NDRange, ONEAPI::reduction(Out, Identity, BOp),
[=](nd_item<1> NDIt, auto &Sum) {
if (NDIt.get_global_linear_id() < NWorkItems)
Sum.combine(
In[NDIt.get_global_linear_id()]);
});
CGH.parallel_for<KernelName>(
NDRange, ext::oneapi::reduction(Out, Identity, BOp),
[=](nd_item<1> NDIt, auto &Sum) {
if (NDIt.get_global_linear_id() < NWorkItems)
Sum.combine(In[NDIt.get_global_linear_id()]);
});
});

// Check correctness.
Expand Down Expand Up @@ -99,7 +99,8 @@ template <class T> struct BigCustomVecPlus {

int main() {
queue Q;
test<class Test1, float, 0, ONEAPI::maximum<>>(Q, getMinimumFPValue<float>());
test<class Test1, float, 0, ext::oneapi::maximum<>>(
Q, getMinimumFPValue<float>());

using BCV = BigCustomVec<long long>;
test<class Test2, BCV, 1, BigCustomVecPlus<long long>>(Q, BCV(0));
Expand Down
18 changes: 9 additions & 9 deletions SYCL/Reduction/reduction_nd_N_vars.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -148,7 +148,7 @@ int testOne(queue &Q, T1 IdentityVal1, T1 InitVal1, BinaryOperation1 BOp1,
CGH.parallel_for<Name>(NDR, Redu1, Redu2, Redu3, Redu4, Lambda);
}).wait();
} else {
// Test ONEAPI reductions
// Test ext::oneapi reductions
Q.submit([&](handler &CGH) {
auto In1 = InBuf1.template get_access<access::mode::read>(CGH);
auto In2 = InBuf2.template get_access<access::mode::read>(CGH);
Expand All @@ -159,10 +159,10 @@ int testOne(queue &Q, T1 IdentityVal1, T1 InitVal1, BinaryOperation1 BOp1,
auto Out2 = OutBuf2.template get_access<Mode2>(CGH);
accessor<T3, 0, Mode3, access::target::global_buffer> Out3(OutBuf3, CGH);

auto Redu1 = ONEAPI::reduction(Out1, IdentityVal1, BOp1);
auto Redu2 = ONEAPI::reduction(Out2, IdentityVal2, BOp2);
auto Redu3 = ONEAPI::reduction(Out3, IdentityVal3, BOp3);
auto Redu4 = ONEAPI::reduction(Out4, IdentityVal4, BOp4);
auto Redu1 = ext::oneapi::reduction(Out1, IdentityVal1, BOp1);
auto Redu2 = ext::oneapi::reduction(Out2, IdentityVal2, BOp2);
auto Redu3 = ext::oneapi::reduction(Out3, IdentityVal3, BOp3);
auto Redu4 = ext::oneapi::reduction(Out4, IdentityVal4, BOp4);

auto Lambda = [=](nd_item<1> NDIt, auto &Sum1, auto &Sum2, auto &Sum3,
auto &Sum4) {
Expand Down Expand Up @@ -210,7 +210,7 @@ int testOne(queue &Q, T1 IdentityVal1, T1 InitVal1, BinaryOperation1 BOp1,
}

// Tests both implementations of reduction:
// sycl::reduction and sycl::ONEAPI::reduction
// sycl::reduction and sycl::ext::oneapi::reduction
template <class Name, typename T1, access::mode Mode1, typename T2,
access::mode Mode2, typename T3, access::mode Mode3, typename T4,
access::mode Mode4, class BinaryOperation1, class BinaryOperation2,
Expand All @@ -237,13 +237,13 @@ int testBoth(queue &Q, T1 IdentityVal1, T1 InitVal1, BinaryOperation1 BOp1,
int main() {
queue Q;
int Error = testBoth<class Case1, float, DW, int, RW, short, RW, int, RW>(
Q, 0, 1000, std::plus<float>{}, 0, 2000, std::plus<>{}, 0, 4000,
Q, 0, 1000, std::plus<>{}, 0, 2000, std::plus<>{}, 0, 4000,
std::bit_or<>{}, 0, 8000, std::bit_xor<>{}, usm::alloc::shared, 16, 16);

auto Add = [](auto x, auto y) { return (x + y); };
Error += testBoth<class Case2, float, RW, int, RW, short, DW, int, DW>(
Q, 0, 1000, std::plus<float>{}, 0, 2000, std::plus<>{}, 0, 4000, Add, 0,
8000, std::plus<>{}, usm::alloc::device, 5 * (256 + 1), 5);
Q, 0, 1000, std::plus<>{}, 0, 2000, std::plus<>{}, 0, 4000, Add, 0, 8000,
std::plus<>{}, usm::alloc::device, 5 * (256 + 1), 5);

if (!Error)
std::cout << "Test passed\n";
Expand Down
10 changes: 5 additions & 5 deletions SYCL/Reduction/reduction_nd_conditional.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -70,7 +70,7 @@ void test(queue &Q, T Identity, size_t WGSize, size_t NWItems) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, Dim, access::mode::discard_write, access::target::global_buffer>
Out(OutBuf, CGH);
auto Redu = ONEAPI::reduction(Out, Identity, BOp);
auto Redu = ext::oneapi::reduction(Out, Identity, BOp);

range<1> GlobalRange(NWItems);
range<1> LocalRange(WGSize);
Expand Down Expand Up @@ -100,10 +100,10 @@ void test(queue &Q, T Identity, size_t WGSize, size_t NWItems) {

int main() {
queue Q;
test<class A, int, 0, ONEAPI::plus<int>>(Q, 0, 2, 2);
test<class B, int, 1, ONEAPI::plus<int>>(Q, 0, 7, 7);
test<class C, int, 0, ONEAPI::plus<int>>(Q, 0, 2, 64);
test<class D, short, 1, ONEAPI::plus<short>>(Q, 0, 16, 256);
test<class A, int, 0, std::plus<>>(Q, 0, 2, 2);
test<class B, int, 1, std::plus<>>(Q, 0, 7, 7);
test<class C, int, 0, std::plus<>>(Q, 0, 2, 64);
test<class D, short, 1, std::plus<>>(Q, 0, 16, 256);

std::cout << "Test passed\n";
return 0;
Expand Down
26 changes: 13 additions & 13 deletions SYCL/Reduction/reduction_nd_ext_type.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -42,7 +42,7 @@ void test(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
Q.submit([&](handler &CGH) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, Dim, Mode, access::target::global_buffer> Out(OutBuf, CGH);
auto Redu = ONEAPI::reduction(Out, Identity, BOp);
auto Redu = ext::oneapi::reduction(Out, Identity, BOp);

CGH.parallel_for<Name>(NDRange, Redu, [=](nd_item<1> NDIt, auto &Sum) {
Sum.combine(In[NDIt.get_global_linear_id()]);
Expand Down Expand Up @@ -85,20 +85,20 @@ template <typename T> int runTests(const string_class &ExtensionName) {

testBoth<class A, T, 1, RW, std::multiplies<T>>(Q, 1, 77, 4, 4);

testBoth<class B1, T, 0, DW, ONEAPI::plus<T>>(Q, 0, 77, 4, 64);
testBoth<class B2, T, 1, RW, ONEAPI::plus<>>(Q, 0, 33, 3, 3 * 5);
testBoth<class B1, T, 0, DW, std::plus<T>>(Q, 0, 77, 4, 64);
testBoth<class B2, T, 1, RW, std::plus<>>(Q, 0, 33, 3, 3 * 5);

testBoth<class C1, T, 0, RW, ONEAPI::minimum<T>>(Q, getMaximumFPValue<T>(),
-10.0, 7, 7);
testBoth<class C2, T, 0, RW, ONEAPI::minimum<T>>(Q, getMaximumFPValue<T>(),
99.0, 7, 7);
testBoth<class C3, T, 1, DW, ONEAPI::minimum<>>(Q, getMaximumFPValue<T>(),
-99.0, 3, 3);
testBoth<class C1, T, 0, RW, ext::oneapi::minimum<T>>(
Q, getMaximumFPValue<T>(), -10.0, 7, 7);
testBoth<class C2, T, 0, RW, ext::oneapi::minimum<T>>(
Q, getMaximumFPValue<T>(), 99.0, 7, 7);
testBoth<class C3, T, 1, DW, ext::oneapi::minimum<>>(
Q, getMaximumFPValue<T>(), -99.0, 3, 3);

testBoth<class D1, T, 0, DW, ONEAPI::maximum<>>(Q, getMinimumFPValue<T>(),
99.0, 3, 3);
testBoth<class D2, T, 1, RW, ONEAPI::maximum<T>>(Q, getMinimumFPValue<T>(),
99.0, 7, 7 * 5);
testBoth<class D1, T, 0, DW, ext::oneapi::maximum<>>(
Q, getMinimumFPValue<T>(), 99.0, 3, 3);
testBoth<class D2, T, 1, RW, ext::oneapi::maximum<T>>(
Q, getMinimumFPValue<T>(), 99.0, 7, 7 * 5);
std::cout << "Test passed\n";
return 0;
}
6 changes: 3 additions & 3 deletions SYCL/Reduction/reduction_nd_lambda.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -13,14 +13,14 @@
using namespace cl::sycl;

// Note that this function is created only to test that if the accessor
// object passed to ONEAPI::reduction is destroyed right after
// ONEAPI::reduction creation, then the reduction still works properly,
// object passed to ext::oneapi::reduction is destroyed right after
// ext::oneapi::reduction creation, then the reduction still works properly,
// i.e. it holds a COPY of user's accessor.
template <typename T, typename BOpT>
auto createReduction(sycl::buffer<T, 1> Buffer, handler &CGH, T Identity,
BOpT BOp) {
auto Acc = Buffer.template get_access<access::mode::discard_write>(CGH);
return ONEAPI::reduction(Acc, Identity, BOp);
return ext::oneapi::reduction(Acc, Identity, BOp);
}

template <class KernelName, typename T, class BinaryOperation>
Expand Down
22 changes: 11 additions & 11 deletions SYCL/Reduction/reduction_nd_s0_dw.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -50,7 +50,7 @@ void test(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, 0, access::mode::discard_write, access::target::global_buffer>
Out(OutBuf, CGH);
auto Redu = ONEAPI::reduction(Out, Identity, BOp);
auto Redu = ext::oneapi::reduction(Out, Identity, BOp);

CGH.parallel_for<Name>(NDRange, Redu, [=](nd_item<1> NDIt, auto &Sum) {
Sum.combine(In[NDIt.get_global_linear_id()]);
Expand Down Expand Up @@ -81,24 +81,24 @@ int main() {
queue Q;

// Check some non power-of-two work-group sizes.
testBoth<class A1, int, ONEAPI::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, ONEAPI::plus<int>>(Q, 0, 99, 49, 49 * 5);
testBoth<class A1, int, std::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, std::plus<int>>(Q, 0, 99, 49, 49 * 5);

// Try some power-of-two work-group sizes.
testBoth<class B1, int, ONEAPI::plus<int>>(Q, 0, 99, 1, 32);
testBoth<class B1, int, std::plus<int>>(Q, 0, 99, 1, 32);
testBoth<class B2, int, std::multiplies<int>>(Q, 1, 99, 4, 32);
testBoth<class B4, int, ONEAPI::bit_xor<int>>(Q, 0, 99, 16, 256);
testBoth<class B5, int, ONEAPI::bit_and<int>>(Q, ~0, 99, 32, 256);
testBoth<class B6, int, ONEAPI::minimum<int>>(
testBoth<class B4, int, std::bit_xor<>>(Q, 0, 99, 16, 256);
testBoth<class B5, int, std::bit_and<>>(Q, ~0, 99, 32, 256);
testBoth<class B6, int, ext::oneapi::minimum<>>(
Q, (std::numeric_limits<int>::max)(), -99, 64, 256);
testBoth<class B7, int, ONEAPI::maximum<int>>(
testBoth<class B7, int, ext::oneapi::maximum<int>>(
Q, (std::numeric_limits<int>::min)(), 99, 128, 256);
testBoth<class B8, int, ONEAPI::plus<>>(Q, 0, 99, 256, 256);
testBoth<class B8, int, std::plus<>>(Q, 0, 99, 256, 256);

// Check with various types.
testBoth<class C1, float, std::multiplies<>>(Q, 1, 99, 8, 256);
testBoth<class C2, short, ONEAPI::minimum<>>(Q, 0x7fff, -99, 8, 256);
testBoth<class C3, unsigned char, ONEAPI::maximum<>>(Q, 0, 99, 8, 256);
testBoth<class C2, short, ext::oneapi::minimum<>>(Q, 0x7fff, -99, 8, 256);
testBoth<class C3, unsigned char, ext::oneapi::maximum<>>(Q, 0, 99, 8, 256);

// Check with CUSTOM type.
testBoth<class D1, CustomVec<long long>, CustomVecPlus<long long>>(
Expand Down
28 changes: 14 additions & 14 deletions SYCL/Reduction/reduction_nd_s0_rw.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -49,7 +49,7 @@ void test(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, 0, access::mode::read_write, access::target::global_buffer>
Out(OutBuf, CGH);
auto Redu = ONEAPI::reduction(Out, Identity, BOp);
auto Redu = ext::oneapi::reduction(Out, Identity, BOp);

CGH.parallel_for<Name>(NDRange, Redu, [=](nd_item<1> NDIt, auto &Sum) {
Sum.combine(In[NDIt.get_global_linear_id()]);
Expand Down Expand Up @@ -80,24 +80,24 @@ int main() {
queue Q;

// Check non power-of-two work-group sizes.
testBoth<class A1, int, ONEAPI::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, ONEAPI::plus<int>>(Q, 0, -99, 49, 49 * 5);
testBoth<class A1, int, std::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, std::plus<int>>(Q, 0, -99, 49, 49 * 5);

// Try some power-of-two work-group sizes.
testBoth<class B1, int, ONEAPI::plus<>>(Q, 0, 99, 2, 32);
testBoth<class B2, int, ONEAPI::plus<>>(Q, 0, 199, 32, 32);
testBoth<class B3, int, ONEAPI::plus<>>(Q, 0, 299, 128, 256);
testBoth<class B4, int, ONEAPI::plus<>>(Q, 0, 399, 256, 256);
testBoth<class B1, int, std::plus<>>(Q, 0, 99, 2, 32);
testBoth<class B2, int, std::plus<>>(Q, 0, 199, 32, 32);
testBoth<class B3, int, std::plus<>>(Q, 0, 299, 128, 256);
testBoth<class B4, int, std::plus<>>(Q, 0, 399, 256, 256);

// Check with various operations and types.
testBoth<class C1, int, std::multiplies<int>>(Q, 1, 2, 8, 256);
testBoth<class C2, float, std::multiplies<float>>(Q, 1, 1.2, 8, 32);
testBoth<class C3, short, ONEAPI::bit_or<>>(Q, 0, 0x3400, 4, 32);
testBoth<class C4, int, ONEAPI::bit_xor<int>>(Q, 0, 0x12340000, 4, 32);
testBoth<class C5, char, ONEAPI::bit_and<>>(Q, ~0, ~0, 4, 16);
testBoth<class C6, int, ONEAPI::minimum<int>>(
testBoth<class C1, int, std::multiplies<>>(Q, 1, 2, 8, 256);
testBoth<class C2, float, std::multiplies<>>(Q, 1, 1.2, 8, 32);
testBoth<class C3, short, std::bit_or<>>(Q, 0, 0x3400, 4, 32);
testBoth<class C4, int, std::bit_xor<>>(Q, 0, 0x12340000, 4, 32);
testBoth<class C5, char, std::bit_and<>>(Q, ~0, ~0, 4, 16);
testBoth<class C6, int, ext::oneapi::minimum<>>(
Q, (std::numeric_limits<int>::max)(), -99, 8, 256);
testBoth<class C7, int, ONEAPI::maximum<float>>(
testBoth<class C7, int, ext::oneapi::maximum<>>(
Q, (std::numeric_limits<int>::min)(), 99, 8, 256);

// Check with CUSTOM type.
Expand Down
28 changes: 14 additions & 14 deletions SYCL/Reduction/reduction_nd_s1_dw.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -47,7 +47,7 @@ void test(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, 1, access::mode::discard_write, access::target::global_buffer>
Out(OutBuf, CGH);
auto Redu = ONEAPI::reduction(Out, Identity, BOp);
auto Redu = ext::oneapi::reduction(Out, Identity, BOp);

CGH.parallel_for<Name>(NDRange, Redu, [=](nd_item<1> NDIt, auto &Sum) {
Sum.combine(In[NDIt.get_global_linear_id()]);
Expand Down Expand Up @@ -78,25 +78,25 @@ int main() {
queue Q;

// Check some non power-of-two work-group sizes.
testBoth<class A1, int, ONEAPI::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, ONEAPI::plus<int>>(Q, 0, 99, 49, 49 * 5);
testBoth<class A1, int, std::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, std::plus<int>>(Q, 0, 99, 49, 49 * 5);

// Try some power-of-two work-group sizes.
testBoth<class B1, int, ONEAPI::plus<int>>(Q, 0, 99, 1, 32);
testBoth<class B2, int, std::multiplies<int>>(Q, 1, 99, 4, 64);
testBoth<class B3, int, ONEAPI::bit_or<int>>(Q, 0, 99, 8, 128);
testBoth<class B4, int, ONEAPI::bit_xor<int>>(Q, 0, 99, 16, 256);
testBoth<class B5, int, ONEAPI::bit_and<int>>(Q, ~0, 99, 32, 256);
testBoth<class B6, int, ONEAPI::minimum<int>>(
testBoth<class B1, int, std::plus<>>(Q, 0, 99, 1, 32);
testBoth<class B2, int, std::multiplies<>>(Q, 1, 99, 4, 64);
testBoth<class B3, int, std::bit_or<>>(Q, 0, 99, 8, 128);
testBoth<class B4, int, std::bit_xor<>>(Q, 0, 99, 16, 256);
testBoth<class B5, int, std::bit_and<>>(Q, ~0, 99, 32, 256);
testBoth<class B6, int, ext::oneapi::minimum<>>(
Q, (std::numeric_limits<int>::max)(), -99, 64, 256);
testBoth<class B7, int, ONEAPI::maximum<int>>(
testBoth<class B7, int, ext::oneapi::maximum<>>(
Q, (std::numeric_limits<int>::min)(), 99, 128, 256);
testBoth<class B8, int, ONEAPI::plus<>>(Q, 0, 99, 256, 256);
testBoth<class B8, int, std::plus<>>(Q, 0, 99, 256, 256);

// Check with various types.
testBoth<class C1, float, std::multiplies<float>>(Q, 1, 99, 8, 256);
testBoth<class C2, short, ONEAPI::minimum<>>(Q, 0x7fff, -99, 8, 256);
testBoth<class C3, unsigned char, ONEAPI::maximum<>>(Q, 0, 99, 8, 256);
testBoth<class C1, float, std::multiplies<>>(Q, 1, 99, 8, 256);
testBoth<class C2, short, ext::oneapi::minimum<>>(Q, 0x7fff, -99, 8, 256);
testBoth<class C3, unsigned char, ext::oneapi::maximum<>>(Q, 0, 99, 8, 256);

// Check with CUSTOM type.
testBoth<class D1, CustomVec<long long>, CustomVecPlus<long long>>(
Expand Down
28 changes: 14 additions & 14 deletions SYCL/Reduction/reduction_nd_s1_rw.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -50,7 +50,7 @@ void test(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
accessor<T, 1, access::mode::read_write, access::target::global_buffer>
Out(OutBuf, CGH);
auto Redu = ONEAPI::reduction(Out, Identity, BOp);
auto Redu = ext::oneapi::reduction(Out, Identity, BOp);

CGH.parallel_for<Name>(NDRange, Redu, [=](nd_item<1> NDIt, auto &Sum) {
Sum.combine(In[NDIt.get_global_linear_id()]);
Expand Down Expand Up @@ -81,24 +81,24 @@ int main() {
queue Q;

// Check non power-of-two work-group sizes.
testBoth<class A1, int, ONEAPI::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, ONEAPI::plus<int>>(Q, 0, -99, 49, 49 * 5);
testBoth<class A1, int, std::plus<int>>(Q, 0, 99, 1, 7);
testBoth<class A2, int, std::plus<int>>(Q, 0, -99, 49, 49 * 5);

// Try some power-of-two work-group sizes.
testBoth<class B1, int, ONEAPI::plus<>>(Q, 0, 99, 2, 32);
testBoth<class B2, int, ONEAPI::plus<>>(Q, 0, 199, 32, 128);
testBoth<class B3, int, ONEAPI::plus<>>(Q, 0, 299, 128, 128);
testBoth<class B4, int, ONEAPI::plus<>>(Q, 0, 399, 256, 256);
testBoth<class B1, int, std::plus<>>(Q, 0, 99, 2, 32);
testBoth<class B2, int, std::plus<>>(Q, 0, 199, 32, 128);
testBoth<class B3, int, std::plus<>>(Q, 0, 299, 128, 128);
testBoth<class B4, int, std::plus<>>(Q, 0, 399, 256, 256);

// Check with various operations and types.
testBoth<class C1, int, std::multiplies<int>>(Q, 1, 2, 8, 256);
testBoth<class C2, float, std::multiplies<float>>(Q, 1, 1.2, 8, 16);
testBoth<class C3, short, ONEAPI::bit_or<>>(Q, 0, 0x3400, 4, 32);
testBoth<class C4, int, ONEAPI::bit_xor<int>>(Q, 0, 0x12340000, 2, 16);
testBoth<class C5, char, ONEAPI::bit_and<>>(Q, ~0, ~0, 4, 16);
testBoth<class C6, int, ONEAPI::minimum<int>>(
testBoth<class C1, int, std::multiplies<>>(Q, 1, 2, 8, 256);
testBoth<class C2, float, std::multiplies<>>(Q, 1, 1.2, 8, 16);
testBoth<class C3, short, std::bit_or<>>(Q, 0, 0x3400, 4, 32);
testBoth<class C4, int, std::bit_xor<>>(Q, 0, 0x12340000, 2, 16);
testBoth<class C5, char, std::bit_and<>>(Q, ~0, ~0, 4, 16);
testBoth<class C6, int, ext::oneapi::minimum<>>(
Q, (std::numeric_limits<int>::max)(), 99, 8, 256);
testBoth<class C7, int, ONEAPI::maximum<float>>(
testBoth<class C7, int, ext::oneapi::maximum<>>(
Q, (std::numeric_limits<int>::min)(), -99, 8, 256);

// Check with CUSTOM type.
Expand Down
16 changes: 8 additions & 8 deletions SYCL/Reduction/reduction_placeholder.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -4,7 +4,7 @@
// RUN: %ACC_RUN_PLACEHOLDER %t.out

// RUNx: %HOST_RUN_PLACEHOLDER %t.out
// TODO: Enable the test for HOST when it supports ONEAPI::reduce() and
// TODO: Enable the test for HOST when it supports ext::oneapi::reduce() and
// barrier()

// This test performs basic checks of parallel_for(nd_range, reduction, func)
Expand Down Expand Up @@ -39,7 +39,7 @@ void testOneCase(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
Q.submit([&](handler &CGH) {
auto In = InBuf.template get_access<access::mode::read>(CGH);
CGH.require(Out);
auto Redu = ONEAPI::reduction(Out, Identity, BinaryOperation());
auto Redu = ext::oneapi::reduction(Out, Identity, BinaryOperation());
range<1> GlobalRange(NWItems);
range<1> LocalRange(WGSize);
nd_range<1> NDRange(GlobalRange, LocalRange);
Expand Down Expand Up @@ -71,17 +71,17 @@ void test(queue &Q, T Identity, T Init, size_t WGSize, size_t NWItems) {
int main() {
queue Q;
// fast atomics and fast reduce
test<class AtomicReduce1, int, 1, ONEAPI::plus<>>(Q, 0, 77, 49, 49 * 5);
test<class AtomicReduce2, int, 0, ONEAPI::plus<int>>(Q, 0, -77, 8, 8);
test<class AtomicReduce1, int, 1, std::plus<>>(Q, 0, 77, 49, 49 * 5);
test<class AtomicReduce2, int, 0, std::plus<int>>(Q, 0, -77, 8, 8);

// fast atomics
test<class Atomic1, int, 0, ONEAPI::bit_or<int>>(Q, 0, 233, 7, 7 * 3);
test<class Atomic2, int, 1, ONEAPI::bit_or<int>>(Q, 0, 177, 4, 128);
test<class Atomic1, int, 0, std::bit_or<>>(Q, 0, 233, 7, 7 * 3);
test<class Atomic2, int, 1, std::bit_or<>>(Q, 0, 177, 4, 128);

// fast reduce
test<class Reduce1, float, 1, ONEAPI::minimum<float>>(
test<class Reduce1, float, 1, ext::oneapi::minimum<>>(
Q, getMaximumFPValue<float>(), -5.0, 5, 5 * 7);
test<class Reduce2, float, 0, ONEAPI::maximum<float>>(
test<class Reduce2, float, 0, ext::oneapi::maximum<>>(
Q, getMinimumFPValue<float>(), -5.0, 4, 128);

// generic algorithm
Expand Down
Loading