This repository was archived by the owner on Mar 28, 2023. It is now read-only.
[SYCL] Test corrections after moving bfloat16 support out of experimental status. #1129
There are no files selected for viewing
This file contains hidden or 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
This file contains hidden or 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,70 @@ | ||
| // UNSUPPORTED: hip | ||
| // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -fsycl-device-code-split=per_kernel %s -o %t.out | ||
| // Currently the feature is supported only on CPU and GPU, natively or by | ||
| // software emulation. | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
| // RUNx: %ACC_RUN_PLACEHOLDER %t.out | ||
|
|
||
| //==---------- bfloat16_conversions.cpp - SYCL bfloat16 type test ---------==// | ||
| // | ||
| // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. | ||
| // See https://llvm.org/LICENSE.txt for license information. | ||
| // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception | ||
| // | ||
| //===---------------------------------------------------------------------===// | ||
|
|
||
| #include <iostream> | ||
| #include <sycl/sycl.hpp> | ||
|
|
||
| using namespace sycl; | ||
|
|
||
| template <typename T> T calculate(T a, T b) { | ||
| sycl::ext::oneapi::bfloat16 x = -a; | ||
| sycl::ext::oneapi::bfloat16 y = b; | ||
| sycl::ext::oneapi::bfloat16 z = x + y; | ||
| T result = z; | ||
| return result; | ||
| } | ||
|
|
||
| template <typename T> int test_device(queue Q) { | ||
| T data[3] = {-7.0f, 8.1f, 0.0f}; | ||
|
|
||
| buffer<T, 1> buf{data, 3}; | ||
| Q.submit([&](handler &cgh) { | ||
| accessor numbers{buf, cgh, read_write}; | ||
| cgh.single_task([=]() { numbers[2] = calculate(numbers[0], numbers[1]); }); | ||
| }); | ||
|
|
||
| host_accessor hostOutAcc{buf, read_only}; | ||
| std::cout << "Device Result = " << hostOutAcc[2] << std::endl; | ||
| if (hostOutAcc[2] == 15.125f) | ||
| return 0; | ||
| return 1; | ||
| } | ||
|
|
||
| template <typename T> int test_host() { | ||
| T a{-5.6f}; | ||
| T b{-1.1f}; | ||
| T result = calculate(a, b); | ||
| std::cout << "Host Result = " << result << std::endl; | ||
| if (result == 4.5f) | ||
| return 0; | ||
| return 1; | ||
| } | ||
|
|
||
| int main() { | ||
| queue Q; | ||
| int result; | ||
| result = test_host<sycl::half>(); | ||
| result |= test_host<float>(); | ||
| if (Q.get_device().has(aspect::fp16)) | ||
| result |= test_device<sycl::half>(Q); | ||
| result |= test_device<float>(Q); | ||
| if (result) | ||
| std::cout << "FAIL\n"; | ||
| else | ||
| std::cout << "PASS\n"; | ||
|
|
||
| return result; | ||
| } |
This file contains hidden or 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,83 @@ | ||
| /// | ||
| /// Check if bfloat16 example works using fallback libraries | ||
| /// | ||
|
|
||
| // REQUIRES: opencl-aot, ocloc, cpu, gpu-intel-gen9 | ||
| // UNSUPPORTED: cuda | ||
| // CUDA is not compatible with SPIR. | ||
|
|
||
| // RUN: %clangxx -fsycl %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64 %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64_gen -Xsycl-target-backend "-device gen9" %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64_gen -Xsycl-target-backend "-device *" %s -o %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64,spir64_gen -Xsycl-target-backend=spir64_gen "-device gen9" %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64_x86_64,spir64_gen -Xsycl-target-backend=spir64_gen "-device gen9" %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
| // RUN: %GPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
|
There was a problem hiding this comment. I think we can add the following test cases: These are interesting because they compile for PVC (native bfloat support) and run on CPU (fallback support). |
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64,spir64_gen -Xsycl-target-backend=spir64_gen "-device pvc" %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| // RUN: %clangxx -fsycl -fsycl-targets=spir64_x86_64,spir64_gen -Xsycl-target-backend=spir64_gen "-device pvc" %s -o %t.out | ||
| // RUN: %CPU_RUN_PLACEHOLDER %t.out | ||
|
|
||
| #include <sycl/sycl.hpp> | ||
|
|
||
| using namespace sycl; | ||
| using sycl::ext::oneapi::bfloat16; | ||
|
|
||
| float foo(float a, float b) { | ||
| // Convert from float to bfloat16. | ||
| bfloat16 A{a}; | ||
| bfloat16 B{b}; | ||
|
|
||
| // Convert A and B from bfloat16 to float, do addition on floating-point | ||
| // numbers, then convert the result to bfloat16 and store it in C. | ||
| bfloat16 C = A + B; | ||
|
|
||
| // Return the result converted from bfloat16 to float. | ||
| return C; | ||
| } | ||
|
|
||
| int main(int argc, char *argv[]) { | ||
| float data[3] = {7.0f, 8.1f, 0.0f}; | ||
|
|
||
| float result_host = foo(7.0f, 8.1f); | ||
| std::cout << "CPU Result = " << result_host << std::endl; | ||
| if (std::abs(15.1f - result_host) > 0.1f) { | ||
| std::cout << "Test failed. Expected CPU Result ~= 15.1" << std::endl; | ||
| return 1; | ||
| } | ||
|
|
||
| queue deviceQueue; | ||
| buffer<float, 1> buf{data, 3}; | ||
|
|
||
| deviceQueue.submit([&](handler &cgh) { | ||
| accessor numbers{buf, cgh, read_write}; | ||
| cgh.single_task([=]() { numbers[2] = foo(numbers[0], numbers[1]); }); | ||
| }); | ||
|
|
||
| host_accessor hostOutAcc{buf, read_only}; | ||
| float result_device = hostOutAcc[2]; | ||
| std::cout << "GPU Result = " << result_device << std::endl; | ||
| if (std::abs(result_host - result_device) > 0.1f) { | ||
| std::cout << "Test failed. CPU Result !~= GPU result" << std::endl; | ||
| return 1; | ||
| } | ||
|
|
||
| return 0; | ||
| } | ||
This file contains hidden or 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
This file contains hidden or 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
This file contains hidden or 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 |
|---|---|---|
|
|
@@ -12,13 +12,4 @@ | |
|
|
||
| #include "bfloat16_type.hpp" | ||
|
|
||
| int main() { return run_tests(); } | ||
Oops, something went wrong.
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
In Line 225, should it be ext_oneapi_bfloat16_math_functions?
Thanks
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Yes, it should be ext_oneapi_bfloat16_math_functions. Made the change.