Classes | Namespaces | Macros | Functions
cxx11_tensor_builtins_sycl.cpp File Reference
#include "main.h"
#include <unsupported/Eigen/CXX11/Tensor>

Classes

struct  EqualAssignment
 
struct  PlusEqualAssignment
 

Namespaces

 SYCL
 

Macros

#define EIGEN_TEST_NO_LONGDOUBLE
 
#define EIGEN_TEST_NO_COMPLEX
 
#define EIGEN_DEFAULT_DENSE_INDEX_TYPE   int64_t
 
#define EIGEN_USE_SYCL
 
#define DECLARE_UNARY_STRUCT_NON_SYCL(FUNC)
 
#define DECLARE_BINARY_STRUCT_NON_SYCL(FUNC)
 
#define DECLARE_UNARY_STRUCT(FUNC)
 
#define RUN_UNARY_TEST(FUNC)    test_unary_builtins_for_scalar<DataType, DataLayout, Assignment, op_##FUNC>(sycl_device, tensor_range)
 
#define DECLARE_BINARY_STRUCT(FUNC)
 
#define DECLARE_BINARY_STRUCT_OP(NAME, OPERATOR)
 

Functions

template<typename T >
T SYCL::abs (T x)
 
template<>
Eigen::half SYCL::abs (Eigen::half x)
 
template<>
float SYCL::abs (float x)
 
template<>
double SYCL::abs (double x)
 
template<typename T >
T SYCL::square (T x)
 
template<typename T >
T SYCL::cube (T x)
 
template<typename T >
T SYCL::inverse (T x)
 
template<typename T >
T SYCL::cwiseMax (T x, T y)
 
template<>
Eigen::half SYCL::cwiseMax (Eigen::half x, Eigen::half y)
 
template<typename T >
T SYCL::cwiseMin (T x, T y)
 
template<>
Eigen::half SYCL::cwiseMin (Eigen::half x, Eigen::half y)
 
template<typename T >
T SYCL::sqrt (T x)
 
template<>
Eigen::half SYCL::sqrt (Eigen::half x)
 
template<typename T >
T SYCL::rsqrt (T x)
 
template<>
Eigen::half SYCL::rsqrt (Eigen::half x)
 
template<typename T >
T SYCL::tanh (T x)
 
template<>
Eigen::half SYCL::tanh (Eigen::half x)
 
template<typename T >
T SYCL::exp (T x)
 
template<>
Eigen::half SYCL::exp (Eigen::half x)
 
template<typename T >
T SYCL::expm1 (T x)
 
template<>
Eigen::half SYCL::expm1 (Eigen::half x)
 
template<typename T >
T SYCL::log (T x)
 
template<>
Eigen::half SYCL::log (Eigen::half x)
 
template<typename T >
T SYCL::ceil (T x)
 
template<>
Eigen::half SYCL::ceil (Eigen::half x)
 
template<typename T >
T SYCL::floor (T x)
 
template<>
Eigen::half SYCL::floor (Eigen::half x)
 
template<typename T >
T SYCL::round (T x)
 
template<>
Eigen::half SYCL::round (Eigen::half x)
 
template<typename T >
T SYCL::log1p (T x)
 
template<>
Eigen::half SYCL::log1p (Eigen::half x)
 
template<typename T >
T SYCL::sign (T x)
 
template<>
Eigen::half SYCL::sign (Eigen::half x)
 
template<typename T >
T SYCL::isnan (T x)
 
template<>
Eigen::half SYCL::isnan (Eigen::half x)
 
template<typename T >
T SYCL::isfinite (T x)
 
template<>
Eigen::half SYCL::isfinite (Eigen::half x)
 
template<typename T >
T SYCL::isinf (T x)
 
template<>
Eigen::half SYCL::isinf (Eigen::half x)
 
template<typename DataType , int DataLayout, typename Assignment , typename Operator >
void test_unary_builtins_for_scalar (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType , int DataLayout, typename Assignment >
void test_unary_builtins_for_assignement (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType , int DataLayout, typename Operator >
void test_unary_builtins_return_bool (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType , int DataLayout>
void test_unary_builtins (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType >
static void test_builtin_unary_sycl (const Eigen::SyclDevice &sycl_device)
 
template<typename DataType , int DataLayout, typename Operator >
void test_binary_builtins_func (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType , int DataLayout, typename Operator >
void test_binary_builtins_fixed_arg2 (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType , int DataLayout>
void test_binary_builtins (const Eigen::SyclDevice &sycl_device, const array< int64_t, 3 > &tensor_range)
 
template<typename DataType >
static void test_floating_builtin_binary_sycl (const Eigen::SyclDevice &sycl_device)
 
template<typename DataType >
static void test_integer_builtin_binary_sycl (const Eigen::SyclDevice &sycl_device)
 
 EIGEN_DECLARE_TEST (cxx11_tensor_builtins_sycl)
 

Macro Definition Documentation

◆ DECLARE_BINARY_STRUCT

#define DECLARE_BINARY_STRUCT (   FUNC)
Value:
struct op_##FUNC { \
template <typename T1, typename T2> \
auto operator()(const T1& x, const T2& y) -> decltype(cl::sycl::FUNC(x, y)) { \
return cl::sycl::FUNC(x, y); \
} \
template <typename T1, typename T2> \
auto operator()(const TensorMap<T1>& x, const TensorMap<T2>& y) -> decltype(x.FUNC(y)) { \
return x.FUNC(y); \
} \
};
Eigen::TensorMap
A tensor expression mapping an existing array of data.
Definition: TensorMap.h:33
y
Scalar * y
Definition: level1_cplx_impl.h:128
plotDoE.x
list x
Definition: plotDoE.py:28

◆ DECLARE_BINARY_STRUCT_NON_SYCL

#define DECLARE_BINARY_STRUCT_NON_SYCL (   FUNC)
Value:
struct op_##FUNC { \
template <typename T1, typename T2> \
auto operator()(const T1& x, const T2& y) { \
return SYCL::FUNC(x, y); \
} \
template <typename T1, typename T2> \
auto operator()(const TensorMap<T1>& x, const TensorMap<T2>& y) { \
return x.FUNC(y); \
} \
};

◆ DECLARE_BINARY_STRUCT_OP

#define DECLARE_BINARY_STRUCT_OP (   NAME,
  OPERATOR 
)
Value:
struct op_##NAME { \
template <typename T1, typename T2> \
auto operator()(const T1& x, const T2& y) -> decltype(x OPERATOR y) { \
return x OPERATOR y; \
} \
};

◆ DECLARE_UNARY_STRUCT

#define DECLARE_UNARY_STRUCT (   FUNC)
Value:
struct op_##FUNC { \
template <typename T> \
auto operator()(const T& x) -> decltype(SYCL::FUNC(x)) { \
return SYCL::FUNC(x); \
} \
template <typename T> \
auto operator()(const TensorMap<T>& x) -> decltype(x.FUNC()) { \
return x.FUNC(); \
} \
};

◆ DECLARE_UNARY_STRUCT_NON_SYCL

#define DECLARE_UNARY_STRUCT_NON_SYCL (   FUNC)
Value:
struct op_##FUNC { \
template <typename T> \
auto operator()(const T& x) { \
return SYCL::FUNC(x); \
} \
template <typename T> \
auto operator()(const TensorMap<T>& x) { \
return x.FUNC(); \
} \
};

◆ EIGEN_DEFAULT_DENSE_INDEX_TYPE

#define EIGEN_DEFAULT_DENSE_INDEX_TYPE   int64_t

◆ EIGEN_TEST_NO_COMPLEX

#define EIGEN_TEST_NO_COMPLEX

◆ EIGEN_TEST_NO_LONGDOUBLE

#define EIGEN_TEST_NO_LONGDOUBLE

◆ EIGEN_USE_SYCL

#define EIGEN_USE_SYCL

◆ RUN_UNARY_TEST

#define RUN_UNARY_TEST (   FUNC)     test_unary_builtins_for_scalar<DataType, DataLayout, Assignment, op_##FUNC>(sycl_device, tensor_range)

Function Documentation

◆ EIGEN_DECLARE_TEST()

EIGEN_DECLARE_TEST ( cxx11_tensor_builtins_sycl  )
499  {
500  for (const auto& device : Eigen::get_sycl_supported_devices()) {
501  QueueInterface queueInterface(device);
502  Eigen::SyclDevice sycl_device(&queueInterface);
503  CALL_SUBTEST_1(test_builtin_unary_sycl<half>(sycl_device));
504  CALL_SUBTEST_2(test_floating_builtin_binary_sycl<half>(sycl_device));
505  CALL_SUBTEST_3(test_builtin_unary_sycl<float>(sycl_device));
506  CALL_SUBTEST_4(test_floating_builtin_binary_sycl<float>(sycl_device));
507  CALL_SUBTEST_5(test_integer_builtin_binary_sycl<int>(sycl_device));
508  }
509 }
CALL_SUBTEST_3
#define CALL_SUBTEST_3(FUNC)
Definition: split_test_helper.h:16
CALL_SUBTEST_1
#define CALL_SUBTEST_1(FUNC)
Definition: split_test_helper.h:4
CALL_SUBTEST_5
#define CALL_SUBTEST_5(FUNC)
Definition: split_test_helper.h:28
CALL_SUBTEST_2
#define CALL_SUBTEST_2(FUNC)
Definition: split_test_helper.h:10
CALL_SUBTEST_4
#define CALL_SUBTEST_4(FUNC)
Definition: split_test_helper.h:22

References CALL_SUBTEST_1, CALL_SUBTEST_2, CALL_SUBTEST_3, CALL_SUBTEST_4, and CALL_SUBTEST_5.

◆ test_binary_builtins()

template<typename DataType , int DataLayout>
void test_binary_builtins ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
470  {
471  test_binary_builtins_func<DataType, DataLayout, op_cwiseMax>(sycl_device, tensor_range);
472  test_binary_builtins_func<DataType, DataLayout, op_cwiseMin>(sycl_device, tensor_range);
473  test_binary_builtins_func<DataType, DataLayout, op_plus>(sycl_device, tensor_range);
474  test_binary_builtins_func<DataType, DataLayout, op_minus>(sycl_device, tensor_range);
475  test_binary_builtins_func<DataType, DataLayout, op_times>(sycl_device, tensor_range);
476  test_binary_builtins_func<DataType, DataLayout, op_divide>(sycl_device, tensor_range);
477 }

◆ test_binary_builtins_fixed_arg2()

template<typename DataType , int DataLayout, typename Operator >
void test_binary_builtins_fixed_arg2 ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
421  {
422  /* out = op(in_1, 2) */
423  Operator op;
424  const DataType arg2(2);
425  Tensor<DataType, 3, DataLayout, int64_t> in_1(tensor_range);
426  Tensor<DataType, 3, DataLayout, int64_t> out(tensor_range);
427  in_1 = in_1.random();
428  Tensor<DataType, 3, DataLayout, int64_t> reference(out);
429  DataType* gpu_data_1 = static_cast<DataType*>(sycl_device.allocate(in_1.size() * sizeof(DataType)));
430  DataType* gpu_data_out = static_cast<DataType*>(sycl_device.allocate(out.size() * sizeof(DataType)));
431  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_1(gpu_data_1, tensor_range);
432  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_out(gpu_data_out, tensor_range);
433  sycl_device.memcpyHostToDevice(gpu_data_1, in_1.data(), (in_1.size()) * sizeof(DataType));
434  gpu_out.device(sycl_device) = op(gpu_1, arg2);
435  sycl_device.memcpyDeviceToHost(out.data(), gpu_data_out, (out.size()) * sizeof(DataType));
436  for (int64_t i = 0; i < out.size(); ++i) {
437  VERIFY_IS_APPROX(out(i), op(in_1(i), arg2));
438  }
439  sycl_device.deallocate(gpu_data_1);
440  sycl_device.deallocate(gpu_data_out);
441 }
i
int i
Definition: BiCGSTAB_step_by_step.cpp:9
Eigen::Tensor
The tensor class.
Definition: Tensor.h:68
VERIFY_IS_APPROX
#define VERIFY_IS_APPROX(a, b)
Definition: integer_types.cpp:13
op
char char * op
Definition: level2_impl.h:374
Eigen::numext::int64_t
std::int64_t int64_t
Definition: Meta.h:43
out
std::ofstream out("Result.txt")

References Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::data(), Eigen::TensorBase< Derived, AccessLevel >::device(), i, op, out(), Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::size(), and VERIFY_IS_APPROX.

◆ test_binary_builtins_func()

template<typename DataType , int DataLayout, typename Operator >
void test_binary_builtins_func ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
393  {
394  /* out = op(in_1, in_2) */
395  Operator op;
396  Tensor<DataType, 3, DataLayout, int64_t> in_1(tensor_range);
397  Tensor<DataType, 3, DataLayout, int64_t> in_2(tensor_range);
398  Tensor<DataType, 3, DataLayout, int64_t> out(tensor_range);
399  in_1 = in_1.random() + DataType(0.01);
400  in_2 = in_2.random() + DataType(0.01);
401  Tensor<DataType, 3, DataLayout, int64_t> reference(out);
402  DataType* gpu_data_1 = static_cast<DataType*>(sycl_device.allocate(in_1.size() * sizeof(DataType)));
403  DataType* gpu_data_2 = static_cast<DataType*>(sycl_device.allocate(in_2.size() * sizeof(DataType)));
404  DataType* gpu_data_out = static_cast<DataType*>(sycl_device.allocate(out.size() * sizeof(DataType)));
405  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_1(gpu_data_1, tensor_range);
406  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_2(gpu_data_2, tensor_range);
407  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_out(gpu_data_out, tensor_range);
408  sycl_device.memcpyHostToDevice(gpu_data_1, in_1.data(), (in_1.size()) * sizeof(DataType));
409  sycl_device.memcpyHostToDevice(gpu_data_2, in_2.data(), (in_2.size()) * sizeof(DataType));
410  gpu_out.device(sycl_device) = op(gpu_1, gpu_2);
411  sycl_device.memcpyDeviceToHost(out.data(), gpu_data_out, (out.size()) * sizeof(DataType));
412  for (int64_t i = 0; i < out.size(); ++i) {
413  VERIFY_IS_APPROX(out(i), op(in_1(i), in_2(i)));
414  }
415  sycl_device.deallocate(gpu_data_1);
416  sycl_device.deallocate(gpu_data_2);
417  sycl_device.deallocate(gpu_data_out);
418 }

References Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::data(), Eigen::TensorBase< Derived, AccessLevel >::device(), i, op, out(), Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::size(), and VERIFY_IS_APPROX.

◆ test_builtin_unary_sycl()

template<typename DataType >
static void test_builtin_unary_sycl ( const Eigen::SyclDevice &  sycl_device)
static
382  {
383  int64_t sizeDim1 = 10;
384  int64_t sizeDim2 = 10;
385  int64_t sizeDim3 = 10;
386  array<int64_t, 3> tensor_range = {{sizeDim1, sizeDim2, sizeDim3}};
387 
388  test_unary_builtins<DataType, RowMajor>(sycl_device, tensor_range);
389  test_unary_builtins<DataType, ColMajor>(sycl_device, tensor_range);
390 }
Eigen::array
std::array< T, N > array
Definition: EmulateArray.h:231

◆ test_floating_builtin_binary_sycl()

template<typename DataType >
static void test_floating_builtin_binary_sycl ( const Eigen::SyclDevice &  sycl_device)
static
480  {
481  int64_t sizeDim1 = 10;
482  int64_t sizeDim2 = 10;
483  int64_t sizeDim3 = 10;
484  array<int64_t, 3> tensor_range = {{sizeDim1, sizeDim2, sizeDim3}};
485  test_binary_builtins<DataType, RowMajor>(sycl_device, tensor_range);
486  test_binary_builtins<DataType, ColMajor>(sycl_device, tensor_range);
487 }

◆ test_integer_builtin_binary_sycl()

template<typename DataType >
static void test_integer_builtin_binary_sycl ( const Eigen::SyclDevice &  sycl_device)
static
490  {
491  int64_t sizeDim1 = 10;
492  int64_t sizeDim2 = 10;
493  int64_t sizeDim3 = 10;
494  array<int64_t, 3> tensor_range = {{sizeDim1, sizeDim2, sizeDim3}};
495  test_binary_builtins_fixed_arg2<DataType, RowMajor, op_modulo>(sycl_device, tensor_range);
496  test_binary_builtins_fixed_arg2<DataType, ColMajor, op_modulo>(sycl_device, tensor_range);
497 }

◆ test_unary_builtins()

template<typename DataType , int DataLayout>
void test_unary_builtins ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
373  {
374  test_unary_builtins_for_assignement<DataType, DataLayout, PlusEqualAssignment>(sycl_device, tensor_range);
375  test_unary_builtins_for_assignement<DataType, DataLayout, EqualAssignment>(sycl_device, tensor_range);
376  test_unary_builtins_return_bool<DataType, DataLayout, op_isnan>(sycl_device, tensor_range);
377  test_unary_builtins_return_bool<DataType, DataLayout, op_isfinite>(sycl_device, tensor_range);
378  test_unary_builtins_return_bool<DataType, DataLayout, op_isinf>(sycl_device, tensor_range);
379 }

◆ test_unary_builtins_for_assignement()

template<typename DataType , int DataLayout, typename Assignment >
void test_unary_builtins_for_assignement ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
331  {
332 #define RUN_UNARY_TEST(FUNC) \
333  test_unary_builtins_for_scalar<DataType, DataLayout, Assignment, op_##FUNC>(sycl_device, tensor_range)
334  RUN_UNARY_TEST(abs);
335  RUN_UNARY_TEST(sqrt);
336  RUN_UNARY_TEST(rsqrt);
337  RUN_UNARY_TEST(square);
338  RUN_UNARY_TEST(cube);
339  RUN_UNARY_TEST(inverse);
340  RUN_UNARY_TEST(tanh);
341  RUN_UNARY_TEST(exp);
342  RUN_UNARY_TEST(expm1);
343  RUN_UNARY_TEST(log);
344  RUN_UNARY_TEST(ceil);
345  RUN_UNARY_TEST(floor);
346  RUN_UNARY_TEST(round);
347  RUN_UNARY_TEST(log1p);
348  RUN_UNARY_TEST(sign);
349 }
abs
AnnoyingScalar abs(const AnnoyingScalar &x)
Definition: AnnoyingScalar.h:135
sqrt
AnnoyingScalar sqrt(const AnnoyingScalar &x)
Definition: AnnoyingScalar.h:134
RUN_UNARY_TEST
#define RUN_UNARY_TEST(FUNC)
inverse
void inverse(const MatrixType &m)
Definition: inverse.cpp:64
SYCL::log1p
Eigen::half log1p(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:167
SYCL::cube
T cube(T x)
Definition: cxx11_tensor_builtins_sycl.cpp:56
SYCL::floor
Eigen::half floor(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:149
SYCL::tanh
Eigen::half tanh(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:104
SYCL::round
Eigen::half round(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:158
SYCL::square
T square(T x)
Definition: cxx11_tensor_builtins_sycl.cpp:52
SYCL::sign
Eigen::half sign(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:176
SYCL::rsqrt
Eigen::half rsqrt(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:95
SYCL::exp
Eigen::half exp(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:113
SYCL::ceil
Eigen::half ceil(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:140
SYCL::expm1
Eigen::half expm1(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:122
SYCL::log
Eigen::half log(Eigen::half x)
Definition: cxx11_tensor_builtins_sycl.cpp:131

References abs(), SYCL::ceil(), SYCL::cube(), SYCL::exp(), SYCL::expm1(), SYCL::floor(), inverse(), SYCL::log(), SYCL::log1p(), SYCL::round(), SYCL::rsqrt(), RUN_UNARY_TEST, SYCL::sign(), sqrt(), SYCL::square(), and SYCL::tanh().

◆ test_unary_builtins_for_scalar()

template<typename DataType , int DataLayout, typename Assignment , typename Operator >
void test_unary_builtins_for_scalar ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
255  {
256  Operator op;
257  Assignment asgn;
258  {
259  /* Assignment(out, Operator(in)) */
260  Tensor<DataType, 3, DataLayout, int64_t> in(tensor_range);
261  Tensor<DataType, 3, DataLayout, int64_t> out(tensor_range);
262  in = in.random() + DataType(0.01);
263  out = out.random() + DataType(0.01);
264  Tensor<DataType, 3, DataLayout, int64_t> reference(out);
265  DataType* gpu_data = static_cast<DataType*>(sycl_device.allocate(in.size() * sizeof(DataType)));
266  DataType* gpu_data_out = static_cast<DataType*>(sycl_device.allocate(out.size() * sizeof(DataType)));
267  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu(gpu_data, tensor_range);
268  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_out(gpu_data_out, tensor_range);
269  sycl_device.memcpyHostToDevice(gpu_data, in.data(), (in.size()) * sizeof(DataType));
270  sycl_device.memcpyHostToDevice(gpu_data_out, out.data(), (out.size()) * sizeof(DataType));
271  auto device_expr = gpu_out.device(sycl_device);
272  asgn(device_expr, op(gpu));
273  sycl_device.memcpyDeviceToHost(out.data(), gpu_data_out, (out.size()) * sizeof(DataType));
274  for (int64_t i = 0; i < out.size(); ++i) {
275  DataType ver = reference(i);
276  asgn(ver, op(in(i)));
277  VERIFY_IS_APPROX(out(i), ver);
278  }
279  sycl_device.deallocate(gpu_data);
280  sycl_device.deallocate(gpu_data_out);
281  }
282  {
283  /* Assignment(out, Operator(out)) */
284  Tensor<DataType, 3, DataLayout, int64_t> out(tensor_range);
285  // Offset with 1 to avoid tiny output (< 1e-6) as they can easily fail.
286  out = out.random() + DataType(1);
287  Tensor<DataType, 3, DataLayout, int64_t> reference(out);
288  DataType* gpu_data_out = static_cast<DataType*>(sycl_device.allocate(out.size() * sizeof(DataType)));
289  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu_out(gpu_data_out, tensor_range);
290  sycl_device.memcpyHostToDevice(gpu_data_out, out.data(), (out.size()) * sizeof(DataType));
291  auto device_expr = gpu_out.device(sycl_device);
292  asgn(device_expr, op(gpu_out));
293  sycl_device.memcpyDeviceToHost(out.data(), gpu_data_out, (out.size()) * sizeof(DataType));
294  for (int64_t i = 0; i < out.size(); ++i) {
295  DataType ver = reference(i);
296  asgn(ver, op(reference(i)));
297  VERIFY_IS_APPROX(out(i), ver);
298  }
299  sycl_device.deallocate(gpu_data_out);
300  }
301 }
gpu
GpuHelper gpu
Definition: gpuhelper.cpp:18

References Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::data(), Eigen::TensorBase< Derived, AccessLevel >::device(), gpu, i, op, out(), Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::size(), and VERIFY_IS_APPROX.

◆ test_unary_builtins_return_bool()

template<typename DataType , int DataLayout, typename Operator >
void test_unary_builtins_return_bool ( const Eigen::SyclDevice &  sycl_device,
const array< int64_t, 3 > &  tensor_range 
)
352  {
353  /* out = op(in) */
354  Operator op;
355  Tensor<DataType, 3, DataLayout, int64_t> in(tensor_range);
356  Tensor<bool, 3, DataLayout, int64_t> out(tensor_range);
357  in = in.random() + DataType(0.01);
358  DataType* gpu_data = static_cast<DataType*>(sycl_device.allocate(in.size() * sizeof(DataType)));
359  bool* gpu_data_out = static_cast<bool*>(sycl_device.allocate(out.size() * sizeof(bool)));
360  TensorMap<Tensor<DataType, 3, DataLayout, int64_t>> gpu(gpu_data, tensor_range);
361  TensorMap<Tensor<bool, 3, DataLayout, int64_t>> gpu_out(gpu_data_out, tensor_range);
362  sycl_device.memcpyHostToDevice(gpu_data, in.data(), (in.size()) * sizeof(DataType));
363  gpu_out.device(sycl_device) = op(gpu);
364  sycl_device.memcpyDeviceToHost(out.data(), gpu_data_out, (out.size()) * sizeof(bool));
365  for (int64_t i = 0; i < out.size(); ++i) {
366  VERIFY_IS_EQUAL(out(i), op(in(i)));
367  }
368  sycl_device.deallocate(gpu_data);
369  sycl_device.deallocate(gpu_data_out);
370 }
VERIFY_IS_EQUAL
#define VERIFY_IS_EQUAL(a, b)
Definition: main.h:367

References Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::data(), Eigen::TensorBase< Derived, AccessLevel >::device(), gpu, i, op, out(), Eigen::Tensor< Scalar_, NumIndices_, Options_, IndexType_ >::size(), and VERIFY_IS_EQUAL.