Skip to content

Commit d814b4a

Browse files
AlexeySachkovAlexeySachkov
authored
[SYCL] Add basic tests for virtual functions (#14209)
Test plan is available in #10540
1 parent accd0b5 commit d814b4a

File tree

6 files changed

+350
-0
lines changed

6 files changed

+350
-0
lines changed

‎sycl/test-e2e/VirtualFunctions/2/1/1/missing-overrides.cpp

Lines changed: 113 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,113 @@
1+
// UNSUPPORTED: cuda, hip, acc
2+
// FIXME: replace unsupported with an aspect check once we have it
3+
//
4+
// RUN: %{build} -o %t.out -Xclang -fsycl-allow-virtual-functions %helper-includes
5+
// RUN: %{run} %t.out
6+
7+
#include <sycl/detail/core.hpp>
8+
9+
#include "helpers.hpp"
10+
11+
#include <iostream>
12+
13+
namespace oneapi = sycl::ext::oneapi::experimental;
14+
15+
class Base {
16+
public:
17+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
18+
virtual void increment(int *) { /* do nothhing */
19+
}
20+
21+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
22+
virtual void multiply(int *) { /* do nothhing */
23+
}
24+
25+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
26+
virtual void substract(int *) { /* do nothhing */
27+
}
28+
};
29+
30+
class IncrementBy1 : public Base {
31+
public:
32+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
33+
void increment(int *Data) override { *Data += 1; }
34+
};
35+
36+
class IncrementBy1AndSubstractBy2 : public IncrementBy1 {
37+
public:
38+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
39+
void substract(int *Data) override { *Data -= 2; }
40+
};
41+
42+
class MultiplyBy2 : public Base {
43+
public:
44+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
45+
void multiply(int *Data) override { *Data *= 2; }
46+
};
47+
48+
class MultiplyBy2AndIncrementBy8 : public MultiplyBy2 {
49+
public:
50+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
51+
void increment(int *Data) override { *Data += 8; }
52+
};
53+
54+
class SubstractBy4 : public Base {
55+
public:
56+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
57+
void substract(int *Data) override { *Data -= 4; }
58+
};
59+
60+
class SubstractBy4AndMultiplyBy4 : public SubstractBy4 {
61+
public:
62+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
63+
void multiply(int *Data) override { *Data *= 4; }
64+
};
65+
66+
void applyOp(int *DataPtr, Base *ObjPtr) {
67+
ObjPtr->increment(DataPtr);
68+
ObjPtr->substract(DataPtr);
69+
ObjPtr->multiply(DataPtr);
70+
}
71+
72+
int main() try {
73+
using storage_t = obj_storage_t<IncrementBy1, IncrementBy1AndSubstractBy2,
74+
MultiplyBy2, MultiplyBy2AndIncrementBy8,
75+
SubstractBy4, SubstractBy4AndMultiplyBy4>;
76+
storage_t HostStorage;
77+
sycl::buffer<storage_t> DeviceStorage(sycl::range{1});
78+
79+
auto asyncHandler = [](sycl::exception_list list) {
80+
for (auto &e : list)
81+
std::rethrow_exception(e);
82+
};
83+
84+
sycl::queue q(asyncHandler);
85+
86+
constexpr oneapi::properties props{oneapi::calls_indirectly<>};
87+
for (unsigned TestCase = 0; TestCase < 6; ++TestCase) {
88+
int HostData = 42;
89+
int Data = HostData;
90+
sycl::buffer<int> DataStorage(&Data, sycl::range{1});
91+
92+
q.submit([&](sycl::handler &CGH) {
93+
sycl::accessor StorageAcc(DeviceStorage, CGH, sycl::write_only);
94+
sycl::accessor DataAcc(DataStorage, CGH, sycl::write_only);
95+
CGH.single_task(props, [=]() {
96+
auto *Ptr = StorageAcc[0].construct</* ret type = */ Base>(TestCase);
97+
applyOp(DataAcc.get_multi_ptr<sycl::access::decorated::no>().get(),
98+
Ptr);
99+
});
100+
});
101+
102+
Base *Ptr = HostStorage.construct</* ret type = */ Base>(TestCase);
103+
applyOp(&HostData, Ptr);
104+
105+
sycl::host_accessor HostAcc(DataStorage);
106+
assert(HostAcc[0] == HostData);
107+
}
108+
109+
return 0;
110+
} catch (sycl::exception &e) {
111+
std::cout << "Unexpected exception was thrown: " << e.what() << std::endl;
112+
return 1;
113+
}

‎sycl/test-e2e/VirtualFunctions/2/1/1/more-complex-hierarchy.cpp

Lines changed: 93 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,93 @@
1+
// UNSUPPORTED: cuda, hip, acc
2+
// FIXME: replace unsupported with an aspect check once we have it
3+
//
4+
// RUN: %{build} -o %t.out -Xclang -fsycl-allow-virtual-functions %helper-includes
5+
// RUN: %{run} %t.out
6+
7+
#include <sycl/detail/core.hpp>
8+
9+
#include "helpers.hpp"
10+
11+
#include <iostream>
12+
13+
namespace oneapi = sycl::ext::oneapi::experimental;
14+
15+
class AbstractOp {
16+
public:
17+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
18+
virtual void applyOp(int *) = 0;
19+
};
20+
21+
class IncrementOp : public AbstractOp {
22+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
23+
void applyOp(int *Data) final override { increment(Data); }
24+
25+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
26+
virtual void increment(int *) = 0;
27+
};
28+
29+
class IncrementBy1 : public IncrementOp {
30+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
31+
void increment(int *Data) override { *Data += 1; }
32+
};
33+
34+
class IncrementBy2 : public IncrementOp {
35+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
36+
void increment(int *Data) override { *Data += 2; }
37+
};
38+
39+
class IncrementBy4 : public IncrementOp {
40+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
41+
void increment(int *Data) override { *Data += 4; }
42+
};
43+
44+
class IncrementBy8 : public IncrementOp {
45+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
46+
void increment(int *Data) override { *Data += 8; }
47+
};
48+
49+
void applyOp(int *Data, AbstractOp *Obj) { Obj->applyOp(Data); }
50+
51+
int main() try {
52+
using storage_t =
53+
obj_storage_t<IncrementBy1, IncrementBy2, IncrementBy4, IncrementBy8>;
54+
55+
storage_t HostStorage;
56+
sycl::buffer<storage_t> DeviceStorage(sycl::range{1});
57+
58+
auto asyncHandler = [](sycl::exception_list list) {
59+
for (auto &e : list)
60+
std::rethrow_exception(e);
61+
};
62+
63+
sycl::queue q(asyncHandler);
64+
65+
constexpr oneapi::properties props{oneapi::calls_indirectly<>};
66+
for (unsigned TestCase = 0; TestCase < 4; ++TestCase) {
67+
int HostData = 42;
68+
int Data = HostData;
69+
sycl::buffer<int> DataStorage(&Data, sycl::range{1});
70+
71+
q.submit([&](sycl::handler &CGH) {
72+
sycl::accessor StorageAcc(DeviceStorage, CGH, sycl::write_only);
73+
sycl::accessor DataAcc(DataStorage, CGH, sycl::write_only);
74+
CGH.single_task(props, [=]() {
75+
auto *Ptr =
76+
StorageAcc[0].construct</* ret type = */ AbstractOp>(TestCase);
77+
applyOp(DataAcc.get_multi_ptr<sycl::access::decorated::no>().get(),
78+
Ptr);
79+
});
80+
});
81+
82+
auto *Ptr = HostStorage.construct</* ret type = */ AbstractOp>(TestCase);
83+
Ptr->applyOp(&HostData);
84+
85+
sycl::host_accessor HostAcc(DataStorage);
86+
assert(HostAcc[0] == HostData);
87+
}
88+
89+
return 0;
90+
} catch (sycl::exception &e) {
91+
std::cout << "Unexpected exception was thrown: " << e.what() << std::endl;
92+
return 1;
93+
}

‎sycl/test-e2e/VirtualFunctions/2/1/1/simple-hierarchy.cpp

Lines changed: 78 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,78 @@
1+
// UNSUPPORTED: cuda, hip, acc
2+
// FIXME: replace unsupported with an aspect check once we have it
3+
//
4+
// RUN: %{build} -o %t.out -Xclang -fsycl-allow-virtual-functions %helper-includes
5+
// RUN: %{run} %t.out
6+
7+
#include <sycl/detail/core.hpp>
8+
9+
#include "helpers.hpp"
10+
11+
#include <iostream>
12+
13+
namespace oneapi = sycl::ext::oneapi::experimental;
14+
15+
class BaseIncrement {
16+
public:
17+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
18+
virtual void increment(int *Data) { *Data += 1; }
19+
};
20+
21+
class IncrementBy2 : public BaseIncrement {
22+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
23+
void increment(int *Data) override { *Data += 2; }
24+
};
25+
26+
class IncrementBy4 : public BaseIncrement {
27+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
28+
void increment(int *Data) override { *Data += 4; }
29+
};
30+
31+
class IncrementBy8 : public BaseIncrement {
32+
SYCL_EXT_ONEAPI_FUNCTION_PROPERTY(oneapi::indirectly_callable<>)
33+
void increment(int *Data) override { *Data += 8; }
34+
};
35+
36+
int main() try {
37+
using storage_t =
38+
obj_storage_t<BaseIncrement, IncrementBy2, IncrementBy4, IncrementBy8>;
39+
40+
storage_t HostStorage;
41+
sycl::buffer<storage_t> DeviceStorage(sycl::range{1});
42+
43+
auto asyncHandler = [](sycl::exception_list list) {
44+
for (auto &e : list)
45+
std::rethrow_exception(e);
46+
};
47+
48+
sycl::queue q(asyncHandler);
49+
50+
constexpr oneapi::properties props{oneapi::calls_indirectly<>};
51+
for (unsigned TestCase = 0; TestCase < 4; ++TestCase) {
52+
int HostData = 42;
53+
int Data = HostData;
54+
sycl::buffer<int> DataStorage(&Data, sycl::range{1});
55+
56+
q.submit([&](sycl::handler &CGH) {
57+
sycl::accessor StorageAcc(DeviceStorage, CGH, sycl::write_only);
58+
sycl::accessor DataAcc(DataStorage, CGH, sycl::write_only);
59+
CGH.single_task(props, [=]() {
60+
auto *Ptr =
61+
StorageAcc[0].construct</* ret type = */ BaseIncrement>(TestCase);
62+
Ptr->increment(
63+
DataAcc.get_multi_ptr<sycl::access::decorated::no>().get());
64+
});
65+
});
66+
67+
auto *Ptr = HostStorage.construct</* ret type = */ BaseIncrement>(TestCase);
68+
Ptr->increment(&HostData);
69+
70+
sycl::host_accessor HostAcc(DataStorage);
71+
assert(HostAcc[0] == HostData);
72+
}
73+
74+
return 0;
75+
} catch (sycl::exception &e) {
76+
std::cout << "Unexpected exception was thrown: " << e.what() << std::endl;
77+
return 1;
78+
}

‎sycl/test-e2e/VirtualFunctions/README.md

Lines changed: 7 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,7 @@
1+
# E2E tests for `sycl_ext_oneapi_virtual_functions` extension
2+
3+
Note about naming convention and files organization for this folder: the tests,
4+
files and directories are named and organized in a way that resembles their
5+
description in the corresponding test plan document: link to be inserted here
6+
later, but for now look into
7+
[intel/llvm#10540](https://github.com/intel/llvm/pull/10540) PR.

‎sycl/test-e2e/VirtualFunctions/helpers.hpp

Lines changed: 53 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,53 @@
1+
#include <algorithm>
2+
#include <type_traits>
3+
4+
// TODO: strictly speaking, selecting a max alignment here may not be always
5+
// valid, but for test cases that we have now we expect alignment of all types
6+
// to be the same.
7+
// std::aligned_storage uses double under the hood which prevents us from
8+
// using it on some HW. Therefore we use a custom implementation.
9+
template <typename... T> struct aligned_storage {
10+
static constexpr size_t Len = std::max({sizeof(T)...});
11+
static constexpr size_t Align = std::max({alignof(T)...});
12+
13+
struct type {
14+
alignas(Align) unsigned char data[Len];
15+
};
16+
};
17+
18+
// Helper data structure that automatically creates a right (in terms of size
19+
// and alignment) storage to accomodate a value of any of types T...
20+
template <typename... T> struct obj_storage_t {
21+
static_assert(std::max({alignof(T)...}) == std::min({alignof(T)...}),
22+
"Unsupported alignment of input types");
23+
using type = typename aligned_storage<T...>::type;
24+
static constexpr size_t size = std::max({sizeof(T)...});
25+
26+
type storage;
27+
28+
template <typename RetT> RetT *construct(const unsigned int TypeIndex) {
29+
if (TypeIndex >= sizeof...(T)) {
30+
#ifndef __SYCL_DEVICE_ONLY__
31+
assert(false && "Type index is invalid");
32+
#endif
33+
return nullptr;
34+
}
35+
36+
return constructHelper<RetT, T...>(TypeIndex, 0);
37+
}
38+
39+
private:
40+
template <typename RetT> RetT *constructHelper(const int, const int) {
41+
// Won't be ever called, but required to compile
42+
return nullptr;
43+
}
44+
45+
template <typename RetT, typename Type, typename... Rest>
46+
RetT *constructHelper(const int TargetIndex, const int CurIndex) {
47+
if (TargetIndex != CurIndex)
48+
return constructHelper<RetT, Rest...>(TargetIndex, CurIndex + 1);
49+
50+
RetT *Ptr = new (reinterpret_cast<Type *>(&storage)) Type;
51+
return Ptr;
52+
}
53+
};

‎sycl/test-e2e/VirtualFunctions/lit.local.cfg

Lines changed: 6 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,6 @@
1+
import os
2+
3+
# Tests are sharing some common header, but we don't won't to use relative
4+
# paths like "../../../helper.hpp" in them, so let's just register a
5+
# substitution to add directory with helper headers into include search path
6+
config.substitutions.append(("%helper-includes", "-I {}".format(os.path.dirname(os.path.abspath(__file__)))))

0 commit comments

Comments
 (0)