|
1 |
| -// REQUIRES: aoc, accelerator |
2 |
| - |
3 |
| -// RUN: %clangxx -fsycl -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %S/Inputs/aot.cpp -o %t.out |
4 |
| -// RUN: env SYCL_DEVICE_TYPE=HOST %t.out |
5 |
| -// RUN: %ACC_RUN_PLACEHOLDER %t.out |
6 |
| - |
7 | 1 | //==----- accelerator.cpp - AOT compilation for fpga devices using aoc ------==//
|
8 | 2 | //
|
9 | 3 | // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
|
|
12 | 6 | //
|
13 | 7 | //===------------------------------------------------------------------------===//
|
14 | 8 |
|
15 |
| -#include <CL/sycl.hpp> |
16 |
| - |
17 |
| -#include <array> |
18 |
| -#include <iostream> |
19 |
| - |
20 |
| -constexpr cl::sycl::access::mode sycl_read = cl::sycl::access::mode::read; |
21 |
| -constexpr cl::sycl::access::mode sycl_write = cl::sycl::access::mode::write; |
22 |
| - |
23 |
| -template <typename T> |
24 |
| -class SimpleVadd; |
25 |
| - |
26 |
| -template <typename T, size_t N> |
27 |
| -void simple_vadd(const std::array<T, N>& VA, const std::array<T, N>& VB, |
28 |
| - std::array<T, N>& VC) { |
29 |
| - cl::sycl::queue deviceQueue([](cl::sycl::exception_list ExceptionList) { |
30 |
| - for (cl::sycl::exception_ptr_class ExceptionPtr : ExceptionList) { |
31 |
| - try { |
32 |
| - std::rethrow_exception(ExceptionPtr); |
33 |
| - } catch (cl::sycl::exception &E) { |
34 |
| - std::cerr << E.what(); |
35 |
| - } catch (...) { |
36 |
| - std::cerr << "Unknown async exception was caught." << std::endl; |
37 |
| - } |
38 |
| - } |
39 |
| - }); |
40 |
| - |
41 |
| - cl::sycl::range<1> numOfItems{N}; |
42 |
| - cl::sycl::buffer<T, 1> bufferA(VA.data(), numOfItems); |
43 |
| - cl::sycl::buffer<T, 1> bufferB(VB.data(), numOfItems); |
44 |
| - cl::sycl::buffer<T, 1> bufferC(VC.data(), numOfItems); |
45 |
| - |
46 |
| - deviceQueue.submit([&](cl::sycl::handler& cgh) { |
47 |
| - auto accessorA = bufferA.template get_access<sycl_read>(cgh); |
48 |
| - auto accessorB = bufferB.template get_access<sycl_read>(cgh); |
49 |
| - auto accessorC = bufferC.template get_access<sycl_write>(cgh); |
50 |
| - |
51 |
| - cgh.parallel_for<class SimpleVadd<T>>(numOfItems, |
52 |
| - [=](cl::sycl::id<1> wiID) { |
53 |
| - accessorC[wiID] = accessorA[wiID] + accessorB[wiID]; |
54 |
| - }); |
55 |
| - }); |
56 |
| - |
57 |
| - deviceQueue.wait_and_throw(); |
58 |
| -} |
| 9 | +// REQUIRES: aoc, accelerator |
59 | 10 |
|
60 |
| -int main() { |
61 |
| - const size_t array_size = 4; |
62 |
| - std::array<cl::sycl::cl_int, array_size> A = {{1, 2, 3, 4}}, |
63 |
| - B = {{1, 2, 3, 4}}, C; |
64 |
| - std::array<cl::sycl::cl_float, array_size> D = {{1.f, 2.f, 3.f, 4.f}}, |
65 |
| - E = {{1.f, 2.f, 3.f, 4.f}}, F; |
66 |
| - simple_vadd(A, B, C); |
67 |
| - simple_vadd(D, E, F); |
68 |
| - for (unsigned int i = 0; i < array_size; i++) { |
69 |
| - if (C[i] != A[i] + B[i]) { |
70 |
| - std::cout << "The results are incorrect (element " << i << " is " << C[i] |
71 |
| - << "!\n"; |
72 |
| - return 1; |
73 |
| - } |
74 |
| - if (F[i] != D[i] + E[i]) { |
75 |
| - std::cout << "The results are incorrect (element " << i << " is " << F[i] |
76 |
| - << "!\n"; |
77 |
| - return 1; |
78 |
| - } |
79 |
| - } |
80 |
| - std::cout << "The results are correct!\n"; |
81 |
| - return 0; |
82 |
| -} |
| 11 | +// RUN: %clangxx -fsycl -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %S/Inputs/aot.cpp -o %t.out |
| 12 | +// RUN: env SYCL_DEVICE_TYPE=HOST %t.out |
| 13 | +// RUN: %ACC_RUN_PLACEHOLDER %t.out |
0 commit comments