-
Notifications
You must be signed in to change notification settings - Fork 75
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
Test for [[sycl::reqd_work_group_size]]
exception when mismatched nd_range
was given
#285
Merged
bader
merged 18 commits into
KhronosGroup:SYCL-2020
from
andreyromanof:optional_kernel_features/missmatched_nd_range_exception
Mar 10, 2022
Merged
Changes from 11 commits
Commits
Show all changes
18 commits
Select commit
Hold shift + click to select a range
3822a72
Test for `[[sycl::reqd_work_group_size]]` exception when mismatched `…
b03925a
Add `optional_kernel_features` to filter for `computecpp` and `hipsycl`
d50c118
Fix typo in file name
ee28a82
Fix typo in code
a02a67f
Sort lines
88bda24
Change `item` to `nd_item` in kernels args
373ae90
Add kernel names
9871be2
Replace `auto` to `size_to` for `max_wg_size`
bb33e39
Return computecpp to the filter
36a27b4
Apply comments
78147fe
Remove duplicate of the test tag from the test name
e09a86b
Merge branch 'SYCL-2020' of https://github.com/KhronosGroup/SYCL-CTS …
8d399e3
Add compile time disable for hipSYCL
12d9404
Temorarly remove of SECTION
9bb4f1a
Revert "Temorarly remove of SECTION"
329bbbb
Merge branch 'SYCL-2020' of https://github.com/KhronosGroup/SYCL-CTS …
b790797
Merge branch 'SYCL-2020' of https://github.com/KhronosGroup/SYCL-CTS …
951f4ee
Disable mismathched nd range test for ComputeCpp
File filter
Filter by extension
Conversations
Failed to load comments.
Jump to
Jump to file
Failed to load files.
Diff view
Diff view
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
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 |
---|---|---|
|
@@ -22,6 +22,7 @@ math_builtin_api | |
multi_ptr | ||
nd_item | ||
nd_range | ||
optional_kernel_features | ||
queue | ||
range | ||
reduction | ||
|
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,3 @@ | ||
file(GLOB test_cases_list *.cpp) | ||
|
||
add_cts_test(${test_cases_list}) |
104 changes: 104 additions & 0 deletions
104
tests/optional_kernel_features/kernel_features_mismatched_nd_range_exception.cpp
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,104 @@ | ||
/******************************************************************************* | ||
// | ||
// SYCL 2020 Conformance Test Suite | ||
// | ||
// Provides tests for the exception that is thrown by | ||
// [[sycl::reqd_work_group_size]] attribute when use nd_range of wrong size in | ||
// kernel. | ||
// | ||
*******************************************************************************/ | ||
|
||
#include "../common/common.h" | ||
#include "catch2/catch_template_test_macros.hpp" | ||
|
||
namespace kernel_features_mismatched_nd_range_exception { | ||
using namespace sycl_cts; | ||
|
||
struct kernel_separate_lambda; | ||
struct kernel_functor; | ||
struct kernel_submission_call; | ||
|
||
// Define required size of work group for attribute | ||
constexpr int testing_wg_size = 1; | ||
|
||
class Functor { | ||
public: | ||
[[sycl::reqd_work_group_size(testing_wg_size)]] void operator()( | ||
sycl::nd_item<1>) const {} | ||
}; | ||
|
||
TEST_CASE( | ||
"Submitting a kernel with an nd-range that does not match." | ||
"[[sycl::reqd_work_group_size]] throws an exception", | ||
"[kernel_features]") { | ||
auto queue = util::get_cts_object::queue(); | ||
const size_t max_wg_size = | ||
queue.get_device().get_info<sycl::info::device::max_work_group_size>(); | ||
if (max_wg_size < testing_wg_size) { | ||
WARN("Device supported work group size too small. Skipping.."); | ||
return; | ||
} | ||
|
||
const bool is_exception_expected = true; | ||
sycl::errc errc_expected = sycl::errc::nd_range; | ||
|
||
const auto separate_lambda = [=](sycl::nd_item<1>) | ||
[[sycl::reqd_work_group_size(testing_wg_size)]]{}; | ||
|
||
// Create nd_range that have to cause an exception | ||
sycl::nd_range<1> mismatched_nd_rage(sycl::range(max_wg_size + 1), | ||
sycl::range(max_wg_size + 1)); | ||
|
||
SECTION("Task as separate lambda") { | ||
bool is_exception_thrown = false; | ||
try { | ||
queue | ||
.submit([&](sycl::handler& cgh) { | ||
cgh.parallel_for<kernel_separate_lambda>(mismatched_nd_rage, | ||
separate_lambda); | ||
}) | ||
.wait_and_throw(); | ||
} catch (const sycl::exception& e) { | ||
is_exception_thrown = true; | ||
INFO("Error code check"); | ||
CHECK(e.code() == errc_expected); | ||
andreyromanof marked this conversation as resolved.
Show resolved
Hide resolved
|
||
} | ||
CHECK(is_exception_expected == is_exception_thrown); | ||
} | ||
|
||
SECTION("Task as functor") { | ||
bool is_exception_thrown = false; | ||
try { | ||
queue | ||
.submit([&](sycl::handler& cgh) { | ||
cgh.parallel_for<kernel_functor>(mismatched_nd_rage, Functor{}); | ||
}) | ||
.wait_and_throw(); | ||
} catch (const sycl::exception& e) { | ||
is_exception_thrown = true; | ||
INFO("Error code check"); | ||
CHECK(e.code() == errc_expected); | ||
} | ||
CHECK(is_exception_expected == is_exception_thrown); | ||
} | ||
|
||
SECTION("Task as submission call") { | ||
bool is_exception_thrown = false; | ||
try { | ||
queue | ||
.submit([&](sycl::handler& cgh) { | ||
cgh.parallel_for<kernel_submission_call>( | ||
mismatched_nd_rage, | ||
[](sycl::nd_item<1>) | ||
[[sycl::reqd_work_group_size(testing_wg_size)]]{}); | ||
}) | ||
.wait_and_throw(); | ||
} catch (const sycl::exception& e) { | ||
is_exception_thrown = true; | ||
INFO("Error code check"); | ||
CHECK(e.code() == errc_expected); | ||
} | ||
CHECK(is_exception_expected == is_exception_thrown); | ||
} | ||
} | ||
} // namespace kernel_features_mismatched_nd_range_exception |
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.
From the spec:
You've marked two implementations in the CI as not passing (adding the test to the filter), but I believe they should be able to pass the test even if they don't implement the features, otherwise they wouldn't be able to pass conformance in the current CTS design.
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.
As I understand CI doesn't run tests, it's just built it. Here you can find output from the building of tests on CI. ComputeCpp and hipsycl failed building due to some errors.
From hipsycl logs, line 1970:
error: ‘errc’ is not a member of ‘sycl’
. Looks like enum with error codes is not implemented yet and attributes too (line 1966).Similar situation for computecpp.
But maybe I was wrong when I put these two implementations in the filter because both of them give the error message:
no known conversion from 'cl::sycl::nd_item<1>' to 'sycl::item<1>'
error, so it can be bug of tests itself.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.
It was my mistake. I changed to
nd_item
but now getting another error. Right now investigating it.