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

Commit 6d3dd9d

Browse files
authored
[ESIMD]Test correctness of type conversion fix (#1170)
* Test correctness of type conversion fix * Add requirement of aspect-fp64 for the test
1 parent 44697ec commit 6d3dd9d

File tree

1 file changed

+83
-0
lines changed

1 file changed

+83
-0
lines changed
Lines changed: 83 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,83 @@
1+
//==--- double_conversion.cpp - DPC++ ESIMD on-device test ---------==//
2+
//
3+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4+
// See https://llvm.org/LICENSE.txt for license information.
5+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6+
//
7+
//===----------------------------------------------------------------------===//
8+
// This is basic test for testing conversion between double and unsigned int.
9+
10+
// REQUIRES: gpu, aspect-fp64
11+
// UNSUPPORTED: cuda || hip
12+
// RUN: %clangxx -fsycl %s -o %t.out
13+
// RUN: %GPU_RUN_PLACEHOLDER %t.out
14+
15+
#include <sycl/ext/intel/esimd.hpp>
16+
#include <sycl/sycl.hpp>
17+
18+
#include <iostream>
19+
#include <typeinfo>
20+
21+
namespace esimd = sycl::ext::intel::esimd;
22+
template <typename T> int test(sycl::queue queue, double TestValue) {
23+
double input = TestValue;
24+
T output = 0;
25+
const T expected =
26+
static_cast<T>(input); // exact values can be stored in double type, no
27+
// implementation-defined rounding here
28+
29+
// Call simd conversion constructor
30+
// Expectation: data not changed
31+
{
32+
sycl::range<1> range(1);
33+
sycl::buffer<double, 1> buffer_in(&input, range);
34+
sycl::buffer<T, 1> buffer_out(&output, range);
35+
36+
queue
37+
.submit([&](sycl::handler &cgh) {
38+
const auto in = buffer_in.get_access<sycl::access_mode::read,
39+
sycl::target::device>(cgh);
40+
const auto out =
41+
buffer_out.template get_access<sycl::access_mode::write,
42+
sycl::target::device>(cgh);
43+
44+
cgh.single_task([=]() SYCL_ESIMD_KERNEL {
45+
esimd::simd<double, 1> source;
46+
source.copy_from(in, 0);
47+
48+
esimd::simd<T, 1> destination(source);
49+
destination.copy_to(out, 0);
50+
});
51+
})
52+
.wait_and_throw();
53+
}
54+
55+
std::cout << " Expected value: " << expected << "\n";
56+
std::cout << " Retrieved value: " << output << "\n";
57+
58+
if (output == expected) {
59+
std::cout << "Test SUCCEED\n";
60+
return 0;
61+
} else {
62+
std::cout << "Test FAILED\n";
63+
return 1;
64+
}
65+
}
66+
67+
int main(int, char **) {
68+
sycl::queue queue;
69+
70+
auto dev = queue.get_device();
71+
std::cout << "Running on " << dev.get_info<sycl::info::device::name>()
72+
<< "\n";
73+
74+
int test_result = 0;
75+
76+
test_result |= test<unsigned int>(queue, 2147483647);
77+
test_result |= test<unsigned int>(queue, 2147483647UL + 1);
78+
test_result |= test<unsigned int>(queue, 4294967295);
79+
80+
test_result |= test<int>(queue, 2147483647);
81+
82+
return test_result;
83+
}

0 commit comments

Comments
 (0)