-
Notifications
You must be signed in to change notification settings - Fork 80
Commit
This commit does not belong to any branch on this repository, and may belong to a fork outside of the repository.
Merge branch 'SYCL-2020' into cover_gap_accessor_user_defined_data_type
- Loading branch information
Showing
45 changed files
with
3,549 additions
and
792 deletions.
There are no files selected for viewing
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,193 @@ | ||
:sectnums: | ||
:xrefstyle: short | ||
|
||
= Test plan for bfloat16 | ||
|
||
This is a test plan for the APIs described in | ||
https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/supported/sycl_ext_oneapi_bfloat16.asciidoc[sycl_ext_oneapi_bfloat16] | ||
|
||
== Testing scope | ||
|
||
=== Device coverage | ||
|
||
All of the tests described below are performed only on the default device that | ||
is selected on the CTS command line. | ||
|
||
=== Type coverage | ||
|
||
All of the tests described below are performed using each of the following `typename T`: | ||
|
||
* `char` | ||
* `short` | ||
* `int` | ||
* `long long` | ||
* `std::size_t` | ||
* `bool` | ||
* `float` | ||
|
||
In addition, if the device has aspect `aspect::fp64`: | ||
|
||
* `double` | ||
|
||
In addition, if the device has `aspect::fp16`: | ||
|
||
* `sycl::half` | ||
|
||
Note that `T` must be convertible to `float` | ||
|
||
== Tests | ||
|
||
=== Size | ||
|
||
`bfloat16` must occupy 2 bytes of memory, check that `sizeof(bfloat16) == 2`. | ||
This check is required to calculate special values. | ||
|
||
=== Constructors | ||
|
||
`bfloat16()` + | ||
`bfloat16(const bfloat16 &)` + | ||
`~bfloat16()` | ||
|
||
Check that: | ||
|
||
* `std::is_default_constructible_v<bfloat16> == true`; | ||
* `std::is_copy_constructible_v<bfloat16> == true`; | ||
* `std::is_destructible_v<bfloat16> == true`. | ||
|
||
`bfloat16(const float &a)` + | ||
`bfloat16 &operator=(const float &a)` | ||
|
||
* Create a `float` variable `f` equal to 1; | ||
* Create a `bfloat16` variable `bf1` passing `f` to the constructor; | ||
* Create a `bfloat16` variable `bf2` by assinging it to `f`; | ||
* Verify that: | ||
** `bf1 == f` | ||
** `bf2 == f` | ||
|
||
`bfloat16(const sycl::half &a)` + | ||
`bfloat16 &operator=(const sycl::half &a)` | ||
|
||
Same as above, but with `sycl::half` instead of `float` | ||
|
||
=== Special values | ||
|
||
Use `uint*_t` variables representing `bfloat16` and `float` values in bitset format. | ||
|
||
==== Minimum positive normal value | ||
|
||
[source, c++] | ||
---- | ||
uint16_t bfloat16_bits = 0b0000000010000000; | ||
uint32_t float_bits = 0b00000000100000000000000000000000; | ||
bfloat16 bf_min = sycl::bit_cast<bfloat16>(bfloat16_bits); | ||
float float_min = sycl::bit_cast<float>(float_bits); | ||
---- | ||
|
||
Verify that the minimum values of `bfloat16` type is equal to the minimum value of `float` by using the `bfloat16` operator `==`. | ||
|
||
==== Zero | ||
|
||
[source, c++] | ||
---- | ||
uint16_t bfloat16_bits = 0b0000000000000000; | ||
uint32_t float_bits = 0b00000000000000000000000000000000; | ||
bfloat16 bf_zero = sycl::bit_cast<bfloat16>(bfloat16_bits); | ||
float float_zero = sycl::bit_cast<float>(float_bits); | ||
---- | ||
Verify that `bf_zero == float_zero`. | ||
|
||
==== NaN | ||
|
||
[source, c++] | ||
---- | ||
uint16_t bfloat16_bits[4] = { | ||
0b0111111111000001, // qNaN | ||
0b1111111111000001, | ||
0b0111111110000001, // sNaN | ||
0b1111111110000001}; | ||
bfloat16 bf_nan[4]; | ||
for (int i = 0; i < 4; i++) { | ||
bf_nan[i] = sycl::bit_cast<bfloat16>(bfloat16_bits[i]); | ||
} | ||
---- | ||
Verify that `std::isnan(bf_nan)` is `true` for all elements. | ||
|
||
==== Infinity | ||
|
||
[source, c++] | ||
---- | ||
uint16_t bfloat16_bits[2] = { | ||
0b0111111110000000, | ||
0b1111111110000000}; | ||
bfloat16 bf_inf_0 = sycl::bit_cast<bfloat16>(bfloat16_bits[0]); | ||
bfloat16 bf_inf_1 = sycl::bit_cast<bfloat16>(bfloat16_bits[1]); | ||
---- | ||
Verify that `std::isinf(bf_inf_0)` and `std::isinf(bf_inf_1)` are `true`. | ||
|
||
These tests will fail if the implementation does not use an 8 bit exponent and 7 bit significand for `bfloat16`. | ||
|
||
=== Conversion | ||
|
||
Check that: | ||
|
||
* `std::is_convertible_v<bfloat16, float> == true` | ||
* `std::is_convertible_v<bfloat16, sycl::half> == true` | ||
* `std::is_convertible_v<bfloat16, bool> == true` | ||
* `std::is_convertible_v<float, bfloat16> == true` | ||
* `std::is_convertible_v<sycl::half, bfloat16> == true` | ||
|
||
=== Operators | ||
|
||
`operator-(bfloat16 &bf)` | ||
|
||
Check that it constructs new instance of `bfloat16` class with negated value. | ||
Create `neg_bf` using this operator and verify: | ||
|
||
* `neg_bf == -bf` | ||
* `bf == -neg_bf` | ||
|
||
(Prefix) + | ||
`bfloat16 &operator++(bfloat16 &bf)` + | ||
`bfloat16 &operator--(bfloat16 &bf)` | ||
|
||
* Check if it adds/substracts 1 to the value of the object referenced by this `bf`. | ||
* Check that new value of the referenced object is equal to `(previous value +/- 1)`. | ||
* Check if it returns the copy of `bf`. | ||
* Check returned value type. | ||
|
||
(Postfix) + | ||
`bfloat16 operator++(bfloat16 &bf, int)` + | ||
`bfloat16 operator--(bfloat16 &bf, int)` | ||
|
||
Same as above, but check thar it returns value of `bf` before assignment instead of copy. | ||
|
||
OP is `+=`, `-=`, `*=`, `/=` + | ||
`bfloat16 &operatorOP(bfloat16 &lhs, const bfloat16 &rhs)` | ||
|
||
* Check results of arithmetic operations returned to initial `bfloat16` object. | ||
* Check returned value type. | ||
|
||
OP is `+`, `-`, `*`, `/` + | ||
`bfloat16 operatorOP(const bfloat16 &lhs, const bfloat16 &rhs)` | ||
|
||
* Check results of arithmetic operations. | ||
* Check returned value type. | ||
|
||
OP is `==`, `!=`, `<`, `>`, `+<=+`, `>=` + | ||
`bool operatorOP(const bfloat16 &lhs, const bfloat16 &rhs)` | ||
|
||
* Check results of equality and inequality between two `bfloat16` objects. | ||
* Check returned type is `bool`. | ||
|
||
OP is `==`, `!=`, `<`, `>`, `+<=+`, `>=` + | ||
`template <typename T>` + | ||
`bool operatorOP(const bfloat16 &lhs, const T &rhs)` + | ||
`template <typename T>` + | ||
`bool operatorOP(const T &lhs, const bfloat16 &rhs)` | ||
|
||
* Check results of equality and inequality between `bfloat16` and `T` objects. | ||
* Check returned type is `bool`. |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,130 @@ | ||
:sectnums: | ||
:xrefstyle: short | ||
|
||
= Test plan for Intel's Extensions for Device Information | ||
|
||
This is a test plan for the APIs described in | ||
https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/supported/sycl_ext_intel_device_info.md[sycl_ext_intel_device_info] | ||
|
||
== Testing scope | ||
|
||
=== Device coverage | ||
|
||
All of the tests described below are performed only on the default device that | ||
is selected on the CTS command line. | ||
|
||
=== Feature test macro | ||
|
||
All of the tests should use `#ifdef SYCL_EXT_INTEL_DEVICE_INFO` so they can be skipped | ||
if feature is not supported. | ||
|
||
== Tests | ||
|
||
Run the code in the sections below to check if device has aspect for this `info` and check type of `get_info()` with this `info`. | ||
|
||
=== Device ID | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_device_id)) { | ||
auto ID = dev.get_info<ext::intel::info::device::device_id>(); | ||
if (!std::is_same_v<decltype(ID), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Device UUID | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_device_info_uuid)) { | ||
auto UUID = dev.get_info<ext::intel::info::device::uuid>(); | ||
if (!std::is_same_v<decltype(UUID), std::array<unsigned char, 16>>) { /* test failed */ } | ||
} | ||
---- | ||
=== PCI Address | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_pci_address)) { | ||
auto BDF = dev.get_info<ext::intel::info::device::pci_address>(); | ||
if (!std::is_same_v<decltype(BDF), std::string>) { /* test failed */ } | ||
} | ||
---- | ||
=== Intel GPU Execution Unit SIMD Width | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_gpu_eu_simd_width)) { | ||
auto euSimdWidth = dev.get_info<ext::intel::info::device::gpu_eu_simd_width>(); | ||
if (!std::is_same_v<decltype(euSimdWidth), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Intel GPU Execution Unit Count | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_gpu_eu_count)) { | ||
auto euCount = dev.get_info<ext::intel::info::device::gpu_eu_count>(); | ||
if (!std::is_same_v<decltype(euCount), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Intel GPU Number of Slices | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_gpu_slices)) { | ||
auto slices = dev.get_info<ext::intel::info::device::gpu_slices>(); | ||
if (!std::is_same_v<decltype(slices), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Intel GPU Number of Subslices per Slice | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_gpu_subslices_per_slice)) { | ||
auto subslices = dev.get_info<ext::intel::info::device::gpu_subslices_per_slice>(); | ||
if (!std::is_same_v<decltype(subslices), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Intel GPU Number of Execution Units per Subslice | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_gpu_eu_count_per_subslice)) { | ||
auto euCount = dev.get_info<ext::intel::info::device::gpu_eu_count_per_subslice>(); | ||
if (!std::is_same_v<decltype(euCount), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Intel GPU Number of hardware threads per EU | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_gpu_hw_threads_per_eu)) { | ||
auto threadsCount = dev.get_info<ext::intel::info::device::gpu_hw_threads_per_eu>(); | ||
if (!std::is_same_v<decltype(threadsCount), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Maximum Memory Bandwidth | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_max_mem_bandwidth)) { | ||
auto maxBW = dev.get_info<ext::intel::info::device::max_mem_bandwidth>(); | ||
if (!std::is_same_v<decltype(maxBW), uint64_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Free Global Memory | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_free_memory)) { | ||
auto FreeMemory = dev.get_info<ext::intel::info::device::free_memory>(); | ||
if (!std::is_same_v<decltype(FreeMemory), uint64_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Memory Clock Rate | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_memory_clock_rate)) { | ||
auto MemoryClockRate = dev.get_info<ext::intel::info::device::memory_clock_rate>(); | ||
if (!std::is_same_v<decltype(MemoryClockRate), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
=== Memory Bus Width | ||
[source, c++] | ||
---- | ||
if (dev.has(aspect::ext_intel_memory_bus_width)) { | ||
auto MemoryBusWidth = dev.get_info<ext::intel::info::device::memory_bus_width>(); | ||
if (!std::is_same_v<decltype(MemoryBusWidth), uint32_t>) { /* test failed */ } | ||
} | ||
---- | ||
|
||
These tests will verify that `ext::intel::info::device` and `aspect` namespaces has these information descriptors. |
Oops, something went wrong.