Skip to content

Commit 8b53027

Browse files
yuriykochbb-sycl
yuriykoch
authored and
bb-sycl
committed
[SYCL][ESIMD] Provide the for_all_combinations utility (intel#721)
[SYCL][ESIMD] Provide the for_all_combinations utility Provides a generic way to represent a combinations of types and compile-time values to cover. Supports any number of type and value packs to iterate over Adds unnamed type pack support By doing so we avoid additional restriction on architecture of the test case functors and on logging architecture, as we don't need to construct string description for every type/value on the top level. We are still free to use named type packs, giving us a possibility to have any combination of named packs and unnamed packs, with value packs supported in both variations. Tests are updated as there is no more for_all_types_and_dims function Signed-off-by: Kochetkov, Yuriy <[email protected]>
1 parent a85558e commit 8b53027

File tree

8 files changed

+383
-7
lines changed

8 files changed

+383
-7
lines changed

‎SYCL/ESIMD/api/functional/ctors/ctor_array_core.cpp

Lines changed: 150 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -25,6 +25,149 @@
2525
#include "ctor_array.hpp"
2626

2727
using namespace esimd_test::api::functional;
28+
<<<<<<< HEAD
29+
=======
30+
using namespace sycl::ext::intel::experimental::esimd;
31+
32+
// Descriptor class for the case of calling constructor in initializer context.
33+
struct initializer {
34+
static std::string get_description() { return "initializer"; }
35+
36+
template <typename DataT, int NumElems>
37+
static void call_simd_ctor(DataT (&&ref_data)[NumElems], DataT *const out) {
38+
static_assert(
39+
type_traits::is_nonconst_rvalue_reference_v<decltype(ref_data)>,
40+
"Provided input data is not nonconst rvalue reference");
41+
const auto simd_by_init = simd<DataT, NumElems>(std::move(ref_data));
42+
simd_by_init.copy_to(out);
43+
}
44+
};
45+
46+
// Descriptor class for the case of calling constructor in variable declaration
47+
// context.
48+
struct var_decl {
49+
static std::string get_description() { return "variable declaration"; }
50+
51+
template <typename DataT, int NumElems>
52+
static void call_simd_ctor(DataT (&&ref_data)[NumElems], DataT *const out) {
53+
static_assert(
54+
type_traits::is_nonconst_rvalue_reference_v<decltype(ref_data)>,
55+
"Provided input data is not nonconst rvalue reference");
56+
const simd<DataT, NumElems> simd_by_var_decl(std::move(ref_data));
57+
simd_by_var_decl.copy_to(out);
58+
}
59+
};
60+
61+
// Descriptor class for the case of calling constructor in rvalue in an
62+
// expression context.
63+
struct rval_in_expr {
64+
static std::string get_description() { return "rvalue in an expression"; }
65+
66+
template <typename DataT, int NumElems>
67+
static void call_simd_ctor(DataT (&&ref_data)[NumElems], DataT *const out) {
68+
static_assert(
69+
type_traits::is_nonconst_rvalue_reference_v<decltype(ref_data)>,
70+
"Provided input data is not nonconst rvalue reference");
71+
simd<DataT, NumElems> simd_by_rval;
72+
simd_by_rval = simd<DataT, NumElems>(std::move(ref_data));
73+
simd_by_rval.copy_to(out);
74+
}
75+
};
76+
77+
// Descriptor class for the case of calling constructor in const reference
78+
// context.
79+
class const_ref {
80+
public:
81+
static std::string get_description() { return "const reference"; }
82+
83+
template <typename DataT, int NumElems>
84+
static void call_simd_ctor(DataT (&&ref_data)[NumElems], DataT *const out) {
85+
static_assert(
86+
type_traits::is_nonconst_rvalue_reference_v<decltype(ref_data)>,
87+
"Provided input data is not nonconst rvalue reference");
88+
call_simd_by_const_ref<DataT, NumElems>(
89+
simd<DataT, NumElems>(std::move(ref_data)), out);
90+
}
91+
92+
private:
93+
template <typename DataT, int NumElems>
94+
static void
95+
call_simd_by_const_ref(const simd<DataT, NumElems> &simd_by_const_ref,
96+
DataT *out) {
97+
simd_by_const_ref.copy_to(out);
98+
}
99+
};
100+
101+
// The main test routine.
102+
// Using functor class to be able to iterate over the pre-defined data types.
103+
template <typename DataT, typename DimT, typename TestCaseT> class run_test {
104+
static constexpr int NumElems = DimT::value;
105+
106+
public:
107+
bool operator()(sycl::queue &queue, const std::string &data_type) {
108+
109+
bool passed = true;
110+
const std::vector<DataT> ref_data = generate_ref_data<DataT, NumElems>();
111+
112+
// If current number of elements is equal to one, then run test with each
113+
// one value from reference data.
114+
// If current number of elements is greater than one, then run tests with
115+
// whole reference data.
116+
if constexpr (NumElems == 1) {
117+
for (size_t i = 0; i < ref_data.size(); ++i) {
118+
passed &= run_verification(queue, {ref_data[i]}, data_type);
119+
}
120+
} else {
121+
passed &= run_verification(queue, ref_data, data_type);
122+
}
123+
return passed;
124+
}
125+
126+
private:
127+
bool run_verification(sycl::queue &queue, const std::vector<DataT> &ref_data,
128+
const std::string &data_type) {
129+
assert(ref_data.size() == NumElems &&
130+
"Reference data size is not equal to the simd vector length.");
131+
132+
bool passed = true;
133+
134+
shared_allocator<DataT> allocator(queue);
135+
shared_vector<DataT> result(NumElems, allocator);
136+
shared_vector<DataT> shared_ref_data(ref_data.begin(), ref_data.end(),
137+
allocator);
138+
139+
queue.submit([&](sycl::handler &cgh) {
140+
const DataT *const ref = shared_ref_data.data();
141+
DataT *const out = result.data();
142+
143+
cgh.single_task<ctors::Kernel<DataT, NumElems, TestCaseT>>(
144+
[=]() SYCL_ESIMD_KERNEL {
145+
DataT ref_on_dev[NumElems];
146+
for (size_t i = 0; i < NumElems; ++i) {
147+
ref_on_dev[i] = ref[i];
148+
}
149+
150+
TestCaseT::template call_simd_ctor<DataT, NumElems>(
151+
std::move(ref_on_dev), out);
152+
});
153+
});
154+
queue.wait_and_throw();
155+
156+
for (size_t i = 0; i < result.size(); ++i) {
157+
if (!are_bitwise_equal(ref_data[i], result[i])) {
158+
passed = false;
159+
160+
const auto description =
161+
ctors::TestDescription<DataT, NumElems, TestCaseT>(
162+
i, result[i], ref_data[i], data_type);
163+
log::fail(description);
164+
}
165+
}
166+
167+
return passed;
168+
}
169+
};
170+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721))
28171

29172
int main(int, char **) {
30173
sycl::queue queue(esimd_test::ESIMDSelector{},
@@ -41,9 +184,16 @@ int main(int, char **) {
41184
=======
42185
const auto types = get_tested_types<tested_types::all>();
43186
const auto dims = get_all_dimensions();
187+
<<<<<<< HEAD
44188
>>>>>>> 7ff842b8f ([SYCL][ESIMD] Enable verification for 32 simd vector length (#758))
45189

46190
passed &= for_all_combinations<ctors::run_test>(types, dims, contexts, queue);
191+
=======
192+
const auto contexts = unnamed_type_pack<initializer, var_decl, rval_in_expr,
193+
const_ref>::generate();
194+
195+
passed &= for_all_combinations<run_test>(types, dims, contexts, queue);
196+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721))
47197

48198
std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
49199
return passed ? 0 : 1;

‎SYCL/ESIMD/api/functional/ctors/ctor_copy.hpp

Lines changed: 19 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -162,4 +162,23 @@ template <typename DataT, typename DimT, typename TestCaseT> class run_test {
162162
}
163163
};
164164

165+
<<<<<<< HEAD:SYCL/ESIMD/api/functional/ctors/ctor_copy.hpp
165166
} // namespace esimd_test::api::functional::ctors
167+
=======
168+
int main(int, char **) {
169+
sycl::queue queue(esimd_test::ESIMDSelector{},
170+
esimd_test::createExceptionHandler());
171+
172+
bool passed = true;
173+
174+
const auto types = get_tested_types<tested_types::all>();
175+
const auto dims = get_all_dimensions();
176+
const auto contexts = unnamed_type_pack<initializer, var_decl, rval_in_expr,
177+
const_ref>::generate();
178+
179+
passed &= for_all_combinations<run_test>(types, dims, contexts, queue);
180+
181+
std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
182+
return passed ? 0 : 1;
183+
}
184+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721)):SYCL/ESIMD/api/functional/ctors/ctor_copy.cpp

‎SYCL/ESIMD/api/functional/ctors/ctor_default.hpp

Lines changed: 19 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -108,4 +108,23 @@ template <typename DataT, typename DimT, typename TestCaseT> struct run_test {
108108
}
109109
};
110110

111+
<<<<<<< HEAD:SYCL/ESIMD/api/functional/ctors/ctor_default.hpp
111112
} // namespace esimd_test::api::functional::ctors
113+
=======
114+
int main(int, char **) {
115+
sycl::queue queue(esimd_test::ESIMDSelector{},
116+
esimd_test::createExceptionHandler());
117+
118+
bool passed = true;
119+
120+
const auto types = get_tested_types<tested_types::all>();
121+
const auto dims = get_all_dimensions();
122+
const auto contexts = unnamed_type_pack<initializer, var_decl, rval_in_expr,
123+
const_ref>::generate();
124+
125+
passed &= for_all_combinations<run_test>(types, dims, contexts, queue);
126+
127+
std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
128+
return passed ? 0 : 1;
129+
}
130+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721)):SYCL/ESIMD/api/functional/ctors/ctor_default.cpp

‎SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_fp_extra.cpp

Lines changed: 25 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -27,39 +27,62 @@
2727
#include "ctor_fill.hpp"
2828

2929
using namespace esimd_test::api::functional;
30+
using init_val = ctors::init_val;
3031

3132
int main(int, char **) {
3233
sycl::queue queue(esimd_test::ESIMDSelector{},
3334
esimd_test::createExceptionHandler());
3435

3536
bool passed = true;
3637

38+
<<<<<<< HEAD:SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_fp_extra.cpp
3739
const auto types = get_tested_types<tested_types::fp_extra>();
3840
const auto single_dim = get_dimensions<8>();
3941
const auto context = unnamed_type_pack<ctors::var_decl>::generate();
42+
=======
43+
// Using single dimension and context to verify the accuracy of operations
44+
// with floating point data types
45+
const auto types = get_tested_types<tested_types::fp>();
46+
const auto dims = get_dimensions<8>();
47+
const auto contexts = unnamed_type_pack<ctors::var_decl>::generate();
48+
49+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721)):SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_core.cpp
4050
// Run for specific combinations of types, base and step values and vector
4151
// length.
42-
// The first init_val value it's a base value and the second init_val value
43-
// it's a step value.
4452
#ifdef SIMD_RUN_TEST_WITH_DENORM_INIT_VAL_AND_ULP_STEP
4553
{
54+
<<<<<<< HEAD:SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_fp_extra.cpp
4655
const auto base_values =
4756
ctors::get_init_values_pack<ctors::init_val::denorm>();
4857
const auto step_values =
4958
ctors::get_init_values_pack<ctors::init_val::ulp>();
5059
passed &= for_all_combinations<ctors::run_test>(
5160
types, single_dim, context, base_values, step_values, queue);
61+
=======
62+
const auto base_values = ctors::get_init_values_pack<init_val::denorm>();
63+
const auto step_values = ctors::get_init_values_pack<init_val::ulp>();
64+
passed &= for_all_combinations<ctors::run_test>(
65+
types, dims, contexts, base_values, step_values, queue);
66+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721)):SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_core.cpp
5267
}
5368
#endif
5469
{
5570
const auto base_values =
71+
<<<<<<< HEAD:SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_fp_extra.cpp
5672
ctors::get_init_values_pack<ctors::init_val::inexact,
5773
ctors::init_val::min>();
5874
const auto step_values =
5975
ctors::get_init_values_pack<ctors::init_val::ulp,
6076
ctors::init_val::ulp_half>();
6177
passed &= for_all_combinations<ctors::run_test>(
6278
types, single_dim, context, base_values, step_values, queue);
79+
=======
80+
ctors::get_init_values_pack<init_val::inexact, init_val::min>();
81+
const auto step_values =
82+
ctors::get_init_values_pack<init_val::ulp, init_val::ulp_half>();
83+
passed &= for_all_combinations<ctors::run_test>(
84+
types, dims, contexts, base_values, step_values, queue);
85+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721)):SYCL/ESIMD/api/functional/ctors/ctor_fill_accuracy_core.cpp
6386
}
6487

6588
std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");

‎SYCL/ESIMD/api/functional/ctors/ctor_fill_core.cpp

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -36,6 +36,9 @@ int main(int, char **) {
3636
// boolean flag.
3737

3838
<<<<<<< HEAD
39+
<<<<<<< HEAD
40+
=======
41+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721))
3942
{
4043
// Validate basic functionality works for every invocation context
4144
const auto types = named_type_pack<char, int>::generate("char", "int");
@@ -61,7 +64,11 @@ int main(int, char **) {
6164
}
6265
{
6366
// Validate basic functionality works for every type
67+
<<<<<<< HEAD
6468
const auto types = get_tested_types<tested_types::core>();
69+
=======
70+
const auto types = get_tested_types<tested_types::all>();
71+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721))
6572
const auto dims = get_all_dimensions();
6673
const auto contexts = unnamed_type_pack<ctors::var_decl>::generate();
6774
{
@@ -145,6 +152,7 @@ int main(int, char **) {
145152
}
146153
}
147154
}
155+
<<<<<<< HEAD
148156
=======
149157
const auto two_dims = values_pack<1, 8>();
150158
const auto char_int_types = named_type_pack<char, int>({"char", "int"});
@@ -257,6 +265,8 @@ int main(int, char **) {
257265
ctors::init_val::nan>(queue, single_dim,
258266
fp_types);
259267
>>>>>>> 7ff842b8f ([SYCL][ESIMD] Enable verification for 32 simd vector length (#758))
268+
=======
269+
>>>>>>> 6870ea3ee ([SYCL][ESIMD] Provide the for_all_combinations utility (#721))
260270

261271
std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
262272
return passed ? 0 : 1;

‎SYCL/ESIMD/api/functional/ctors/ctor_move.cpp

Lines changed: 5 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -91,7 +91,8 @@ class const_ref {
9191
// The core test functionality.
9292
// Runs a TestCaseT, specific for each C++ context, for a simd<DataT,NumElems>
9393
// instance
94-
template <typename DataT, int NumElems, typename TestCaseT> class run_test {
94+
template <typename DataT, typename DimT, typename TestCaseT> class run_test {
95+
static constexpr int NumElems = DimT::value;
9596
using KernelName = ctors::Kernel<DataT, NumElems, TestCaseT>;
9697

9798
public:
@@ -217,15 +218,14 @@ int main(int, char **) {
217218
bool passed = true;
218219
const auto types = get_tested_types<tested_types::all>();
219220
const auto dims = get_all_dimensions();
221+
const auto contexts = unnamed_type_pack<initializer, var_decl, rval_in_expr,
222+
const_ref>::generate();
220223

221224
sycl::queue queue(esimd_test::ESIMDSelector{},
222225
esimd_test::createExceptionHandler());
223226

224227
// Run for all combinations possible
225-
passed &= for_all_types_and_dims<run_test, initializer>(types, dims, queue);
226-
passed &= for_all_types_and_dims<run_test, var_decl>(types, dims, queue);
227-
passed &= for_all_types_and_dims<run_test, rval_in_expr>(types, dims, queue);
228-
passed &= for_all_types_and_dims<run_test, const_ref>(types, dims, queue);
228+
passed &= for_all_combinations<run_test>(types, dims, contexts, queue);
229229

230230
std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
231231
return passed ? 0 : 1;

0 commit comments

Comments
 (0)