Skip to content

Commit 3a8ad28

Browse files
dklochkov-embdklochkov-emb
authored
[SYCL] Remove logical_or and logical_and from group algorithm tests (#18411)
PR fixes tests SYCL :: GroupAlgorithm/exclusive_scan_sycl2020.cpp SYCL :: GroupAlgorithm/inclusive_scan_sycl2020.cpp SYCL :: GroupAlgorithm/reduce_sycl2020.cpp due to becoming invalid in case of binary_op is logical_or and logical_and. Spec has _Mandates_ for each group algorithm which tells that binary_op: _binary_op(init, *first) must return a value of type T._ etc. but logical_or and logical_and in ABI breaking mode has return type _boolean_
1 parent 53ac927 commit 3a8ad28

File tree

4 files changed

+77
-30
lines changed

4 files changed

+77
-30
lines changed

‎sycl/test-e2e/GroupAlgorithm/exclusive_scan_sycl2020.cpp‎

Lines changed: 0 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -1,9 +1,6 @@
11
// RUN: %{build} -Wno-error=deprecated-declarations -fsycl-device-code-split=per_kernel -I . -o %t.out
22
// RUN: %{run} %t.out
33

4-
// XFAIL: preview-mode
5-
// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18390
6-
74
#include "../helpers.hpp"
85
#include "support.h"
96
#include <algorithm>
@@ -173,14 +170,9 @@ int main() {
173170
test<class KernelNameBitXorI>(input, sycl::bit_xor<int>(), 0);
174171
test<class KernelNameBitAndI>(input_small, sycl::bit_and<int>(), ~0);
175172

176-
test<class LogicalOrInt>(input, sycl::logical_or<int>(), 0);
177-
test<class LogicalAndInt>(input, sycl::logical_and<int>(), 1);
178-
179173
std::array<bool, N> bool_input = {};
180174
test<class LogicalOrBool>(bool_input, sycl::logical_or<bool>(), false);
181-
test<class LogicalOrVoid>(bool_input, sycl::logical_or<>(), false);
182175
test<class LogicalAndBool>(bool_input, sycl::logical_and<bool>(), true);
183-
test<class LogicalAndVoid>(bool_input, sycl::logical_and<>(), true);
184176

185177
std::array<int2, N> int2_input = {};
186178
std::iota(int2_input.begin(), int2_input.end(), 0);

‎sycl/test-e2e/GroupAlgorithm/inclusive_scan_sycl2020.cpp‎

Lines changed: 0 additions & 14 deletions
Original file line numberDiff line numberDiff line change
@@ -1,9 +1,6 @@
11
// RUN: %{build} -Wno-error=deprecated-declarations -fsycl-device-code-split=per_kernel -I . -o %t.out
22
// RUN: %{run} %t.out
33

4-
// XFAIL: preview-mode
5-
// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18390
6-
74
#include "../helpers.hpp"
85
#include "support.h"
96
#include <algorithm>
@@ -169,21 +166,10 @@ int main() {
169166
test<class KernelNameMultipliesI>(input_small, sycl::multiplies<int>(), 1);
170167
test<class KernelNameBitOrI>(input, sycl::bit_or<int>(), 0);
171168
test<class KernelNameBitXorI>(input, sycl::bit_xor<int>(), 0);
172-
test<class KernelNameBitAndI>(input_small, sycl::bit_and<int>(), ~0);
173-
174-
test<class LogicalOrInt>(input, sycl::logical_or<int>(), 0);
175-
test<class LogicalAndInt>(input, sycl::logical_and<int>(), 1);
176169

177170
std::array<bool, N> bool_input = {};
178171
test<class LogicalOrBool>(bool_input, sycl::logical_or<bool>(), false);
179-
test<class LogicalOrVoid>(bool_input, sycl::logical_or<>(), false);
180-
test<class LogicalAndBool>(bool_input, sycl::logical_and<bool>(), true);
181-
test<class LogicalAndVoid>(bool_input, sycl::logical_and<>(), true);
182-
183-
test<class LogicalOrBool>(bool_input, sycl::logical_or<bool>(), false);
184-
test<class LogicalOrVoid>(bool_input, sycl::logical_or<>(), false);
185172
test<class LogicalAndBool>(bool_input, sycl::logical_and<bool>(), true);
186-
test<class LogicalAndVoid>(bool_input, sycl::logical_and<>(), true);
187173

188174
std::array<int2, N> int2_input = {};
189175
std::iota(int2_input.begin(), int2_input.end(), 0);

‎sycl/test-e2e/GroupAlgorithm/reduce_sycl2020.cpp‎

Lines changed: 0 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -1,9 +1,6 @@
11
// RUN: %{build} -fsycl-device-code-split=per_kernel -I . -o %t.out
22
// RUN: %{run} %t.out
33

4-
// XFAIL: preview-mode
5-
// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18390
6-
74
#include "support.h"
85

96
#include <sycl/sub_group.hpp>
@@ -100,14 +97,9 @@ int main() {
10097
test<class KernelNameBitXorI>(input, sycl::bit_xor<int>(), 0);
10198
test<class KernelNameBitAndI>(input, sycl::bit_and<int>(), ~0);
10299

103-
test<class LogicalOrInt>(input, sycl::logical_or<int>(), 0);
104-
test<class LogicalAndInt>(input, sycl::logical_and<int>(), 1);
105-
106100
std::array<bool, N> bool_input = {};
107101
test<class LogicalOrBool>(bool_input, sycl::logical_or<bool>(), false);
108-
test<class LogicalOrVoid>(bool_input, sycl::logical_or<>(), false);
109102
test<class LogicalAndBool>(bool_input, sycl::logical_and<bool>(), true);
110-
test<class LogicalAndVoid>(bool_input, sycl::logical_and<>(), true);
111103

112104
std::array<int2, N> int2_input = {};
113105
std::iota(int2_input.begin(), int2_input.end(), 0);

‎sycl/test/group_algorithms/logical_or_and_group_algorithms.cpp‎

Lines changed: 77 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,77 @@
1+
// RUN: %clangxx -fsycl -Xclang -verify=expected -Xclang -verify-ignore-unexpected=note -fpreview-breaking-changes -fsyntax-only -fsycl-device-only -ferror-limit=0 %s
2+
3+
#include <sycl/group_algorithm.hpp>
4+
5+
using namespace sycl;
6+
7+
constexpr size_t N = 8;
8+
range<2> global{16, 32};
9+
range<2> local{4, 8};
10+
id<2> groupId{1, 2};
11+
group<2> g = detail::Builder::createGroup(global, local, groupId);
12+
int *rawIn = nullptr;
13+
int *rawOut = nullptr;
14+
15+
void ExclusiveScanOverGroup() {
16+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
17+
exclusive_scan_over_group(g, 0, sycl::logical_and<int>{});
18+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
19+
exclusive_scan_over_group(g, 0, 0, sycl::logical_and<int>{});
20+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
21+
exclusive_scan_over_group(g, 0, sycl::logical_or<int>{});
22+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
23+
exclusive_scan_over_group(g, 0, 0, sycl::logical_or<int>{});
24+
}
25+
26+
void JointExclusiveScan() {
27+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
28+
joint_exclusive_scan(g, rawIn, rawIn + N, rawOut, sycl::logical_and<int>{});
29+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
30+
joint_exclusive_scan(g, rawIn, rawIn + N, rawOut, sycl::logical_or<int>{});
31+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
32+
joint_exclusive_scan(g, rawIn, rawIn + N, rawOut, 0,
33+
sycl::logical_and<int>{});
34+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
35+
joint_exclusive_scan(g, rawIn, rawIn + N, rawOut, 0, sycl::logical_or<int>{});
36+
}
37+
38+
void InclusiveScanOverGroup() {
39+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
40+
inclusive_scan_over_group(g, 0, sycl::logical_and<int>{});
41+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
42+
inclusive_scan_over_group(g, 0, sycl::logical_and<int>{}, 0);
43+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
44+
inclusive_scan_over_group(g, 0, sycl::logical_or<int>{});
45+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
46+
inclusive_scan_over_group(g, 0, sycl::logical_or<int>{}, 0);
47+
}
48+
49+
void JointInclusiveScan() {
50+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
51+
joint_inclusive_scan(g, rawIn, rawIn + N, rawOut, sycl::logical_and<int>{});
52+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
53+
joint_inclusive_scan(g, rawIn, rawIn + N, rawOut, sycl::logical_or<int>{});
54+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
55+
joint_inclusive_scan(g, rawIn, rawIn + N, rawOut, sycl::logical_and<int>{},
56+
0);
57+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match scan accumulation type}}
58+
joint_inclusive_scan(g, rawIn, rawIn + N, rawOut, sycl::logical_or<int>{}, 0);
59+
}
60+
61+
void ReduceOverGroup() {
62+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match reduction accumulation type}}
63+
reduce_over_group(g, 0, sycl::logical_and<int>{});
64+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match reduction accumulation type}}
65+
reduce_over_group(g, 0, 0, sycl::logical_and<int>{});
66+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match reduction accumulation type}}
67+
reduce_over_group(g, 0, sycl::logical_or<int>{});
68+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match reduction accumulation type}}
69+
reduce_over_group(g, 0, 0, sycl::logical_or<int>{});
70+
}
71+
72+
void JointReduce() {
73+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match reduction accumulation type}}
74+
joint_reduce(g, rawIn, rawIn + N, 0, sycl::logical_and<int>{});
75+
// expected-error@sycl/group_algorithm.hpp:* {{Result type of binary_op must match reduction accumulation type}}
76+
joint_reduce(g, rawIn, rawIn + N, 0, sycl::logical_or<int>{});
77+
}

0 commit comments

Comments
 (0)