Skip to content
This repository was archived by the owner on Mar 28, 2023. It is now read-only.

[SYCL] Use float instead of double in some SpecConstants tests #1247

Merged
merged 2 commits into from
Sep 12, 2022
Merged

[SYCL] Use float instead of double in some SpecConstants tests #1247

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
921f55c
[SYCL] Use float instead of double in some SpecConstants tests
aelovikov-intel Sep 9, 2022
4751013
clang-format
aelovikov-intel Sep 9, 2022
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
46 changes: 23 additions & 23 deletions SYCL/SpecConstants/2020/handler-api.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -23,7 +23,7 @@

constexpr sycl::specialization_id<int> int_id;
constexpr sycl::specialization_id<int> int_id2(2);
constexpr sycl::specialization_id<double> double_id(3.14);
constexpr sycl::specialization_id<float> float_id(3.14);
constexpr sycl::specialization_id<custom_type> custom_type_id;

class TestDefaultValuesKernel;
Expand Down Expand Up @@ -71,19 +71,19 @@ int main() {
bool test_default_values(sycl::queue q) {
sycl::buffer<int> int_buffer(1);
sycl::buffer<int> int_buffer2(1);
sycl::buffer<double> double_buffer(1);
sycl::buffer<float> float_buffer(1);
sycl::buffer<custom_type> custom_type_buffer(1);

q.submit([&](sycl::handler &cgh) {
auto int_acc = int_buffer.get_access<sycl::access::mode::write>(cgh);
auto int_acc2 = int_buffer2.get_access<sycl::access::mode::write>(cgh);
auto double_acc = double_buffer.get_access<sycl::access::mode::write>(cgh);
auto float_acc = float_buffer.get_access<sycl::access::mode::write>(cgh);
auto custom_type_acc =
custom_type_buffer.get_access<sycl::access::mode::write>(cgh);
cgh.single_task<TestDefaultValuesKernel>([=](sycl::kernel_handler kh) {
int_acc[0] = kh.get_specialization_constant<int_id>();
int_acc2[0] = kh.get_specialization_constant<int_id2>();
double_acc[0] = kh.get_specialization_constant<double_id>();
float_acc[0] = kh.get_specialization_constant<float_id>();
custom_type_acc[0] = kh.get_specialization_constant<custom_type_id>();
});
});
Expand All @@ -98,8 +98,8 @@ bool test_default_values(sycl::queue q) {
if (!check_value(2, int_acc2[0], "integer specialization constant"))
return false;

auto double_acc = double_buffer.get_access<sycl::access::mode::read>();
if (!check_value(3.14, double_acc[0], "double specialization constant"))
auto float_acc = float_buffer.get_access<sycl::access::mode::read>();
if (!check_value(3.14f, float_acc[0], "float specialization constant"))
return false;

auto custom_type_acc =
Expand All @@ -120,8 +120,8 @@ bool test_set_and_get_on_host(sycl::queue q) {
"integer specializaiton constant before setting any value"))
++errors;

if (!check_value(3.14, cgh.get_specialization_constant<double_id>(),
"double specializaiton constant before setting any value"))
if (!check_value(3.14f, cgh.get_specialization_constant<float_id>(),
"float specializaiton constant before setting any value"))
++errors;

custom_type custom_type_ref;
Expand All @@ -131,20 +131,20 @@ bool test_set_and_get_on_host(sycl::queue q) {
++errors;

int new_int_value = 8;
double new_double_value = 3.0;
custom_type new_custom_type_value('b', 1.0, 12);
float new_float_value = 3.0f;
custom_type new_custom_type_value('b', 1.0f, 12);
cgh.set_specialization_constant<int_id>(new_int_value);
cgh.set_specialization_constant<double_id>(new_double_value);
cgh.set_specialization_constant<float_id>(new_float_value);
cgh.set_specialization_constant<custom_type_id>(new_custom_type_value);

if (!check_value(
new_int_value, cgh.get_specialization_constant<int_id>(),
"integer specializaiton constant after setting a new value"))
++errors;

if (!check_value(
new_double_value, cgh.get_specialization_constant<double_id>(),
"double specializaiton constant after setting a new value"))
if (!check_value(new_float_value,
cgh.get_specialization_constant<float_id>(),
"float specializaiton constant after setting a new value"))
++errors;

if (!check_value(
Expand All @@ -162,30 +162,30 @@ bool test_set_and_get_on_host(sycl::queue q) {
bool test_set_and_get_on_device(sycl::queue q) {
sycl::buffer<int> int_buffer(1);
sycl::buffer<int> int_buffer2(1);
sycl::buffer<double> double_buffer(1);
sycl::buffer<float> float_buffer(1);
sycl::buffer<custom_type> custom_type_buffer(1);

int new_int_value = 8;
int new_int_value2 = 0;
double new_double_value = 3.0;
custom_type new_custom_type_value('b', 1.0, 12);
float new_float_value = 3.0f;
custom_type new_custom_type_value('b', 1.0f, 12);

q.submit([&](sycl::handler &cgh) {
auto int_acc = int_buffer.get_access<sycl::access::mode::write>(cgh);
auto int_acc2 = int_buffer2.get_access<sycl::access::mode::write>(cgh);
auto double_acc = double_buffer.get_access<sycl::access::mode::write>(cgh);
auto float_acc = float_buffer.get_access<sycl::access::mode::write>(cgh);
auto custom_type_acc =
custom_type_buffer.get_access<sycl::access::mode::write>(cgh);

cgh.set_specialization_constant<int_id>(new_int_value);
cgh.set_specialization_constant<int_id2>(new_int_value2);
cgh.set_specialization_constant<double_id>(new_double_value);
cgh.set_specialization_constant<float_id>(new_float_value);
cgh.set_specialization_constant<custom_type_id>(new_custom_type_value);

cgh.single_task<TestSetAndGetOnDevice>([=](sycl::kernel_handler kh) {
int_acc[0] = kh.get_specialization_constant<int_id>();
int_acc2[0] = kh.get_specialization_constant<int_id2>();
double_acc[0] = kh.get_specialization_constant<double_id>();
float_acc[0] = kh.get_specialization_constant<float_id>();
custom_type_acc[0] = kh.get_specialization_constant<custom_type_id>();
});
});
Expand All @@ -200,9 +200,9 @@ bool test_set_and_get_on_device(sycl::queue q) {
"integer specialization constant"))
return false;

auto double_acc = double_buffer.get_access<sycl::access::mode::read>();
if (!check_value(new_double_value, double_acc[0],
"double specialization constant"))
auto float_acc = float_buffer.get_access<sycl::access::mode::read>();
if (!check_value(new_float_value, float_acc[0],
"float specialization constant"))
return false;

auto custom_type_acc =
Expand Down
54 changes: 27 additions & 27 deletions SYCL/SpecConstants/2020/kernel-bundle-api.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -21,7 +21,7 @@
#include "common.hpp"

constexpr sycl::specialization_id<int> int_id;
constexpr sycl::specialization_id<double> double_id(3.14);
constexpr sycl::specialization_id<float> float_id(3.14f);
constexpr sycl::specialization_id<custom_type> custom_type_id;

class TestDefaultValuesKernel;
Expand Down Expand Up @@ -76,7 +76,7 @@ bool test_default_values(sycl::queue q) {
}

sycl::buffer<int> int_buffer(1);
sycl::buffer<double> double_buffer(1);
sycl::buffer<float> float_buffer(1);
sycl::buffer<custom_type> custom_type_buffer(1);

auto input_bundle =
Expand All @@ -86,12 +86,12 @@ bool test_default_values(sycl::queue q) {
q.submit([&](sycl::handler &cgh) {
cgh.use_kernel_bundle(exec_bundle);
auto int_acc = int_buffer.get_access<sycl::access::mode::write>(cgh);
auto double_acc = double_buffer.get_access<sycl::access::mode::write>(cgh);
auto float_acc = float_buffer.get_access<sycl::access::mode::write>(cgh);
auto custom_type_acc =
custom_type_buffer.get_access<sycl::access::mode::write>(cgh);
cgh.single_task<TestDefaultValuesKernel>([=](sycl::kernel_handler kh) {
int_acc[0] = kh.get_specialization_constant<int_id>();
double_acc[0] = kh.get_specialization_constant<double_id>();
float_acc[0] = kh.get_specialization_constant<float_id>();
custom_type_acc[0] = kh.get_specialization_constant<custom_type_id>();
});
});
Expand All @@ -102,8 +102,8 @@ bool test_default_values(sycl::queue q) {
"integer specialization constant (defined without default value)"))
return false;

auto double_acc = double_buffer.get_access<sycl::access::mode::read>();
if (!check_value(3.14, double_acc[0], "double specialization constant"))
auto float_acc = float_buffer.get_access<sycl::access::mode::read>();
if (!check_value(3.14f, float_acc[0], "float specialization constant"))
return false;

auto custom_type_acc =
Expand Down Expand Up @@ -145,9 +145,9 @@ bool test_set_and_get_on_host(sycl::queue q) {
"integer specializaiton constant before setting any value"))
++errors;

if (!check_value(3.14,
input_bundle.get_specialization_constant<double_id>(),
"double specializaiton constant before setting any value"))
if (!check_value(3.14f,
input_bundle.get_specialization_constant<float_id>(),
"float specializaiton constant before setting any value"))
++errors;

custom_type custom_type_ref;
Expand All @@ -159,11 +159,11 @@ bool test_set_and_get_on_host(sycl::queue q) {

// Update values
int new_int_value = 42;
double new_double_value = 3.0;
custom_type new_custom_type_value('b', 1.0, 12);
float new_float_value = 3.0f;
custom_type new_custom_type_value('b', 1.0f, 12);

input_bundle.set_specialization_constant<int_id>(new_int_value);
input_bundle.set_specialization_constant<double_id>(new_double_value);
input_bundle.set_specialization_constant<float_id>(new_float_value);
input_bundle.set_specialization_constant<custom_type_id>(
new_custom_type_value);

Expand All @@ -173,9 +173,9 @@ bool test_set_and_get_on_host(sycl::queue q) {
"integer specializaiton constant after setting a new value"))
++errors;

if (!check_value(new_double_value,
input_bundle.get_specialization_constant<double_id>(),
"double specializaiton constant after setting a value"))
if (!check_value(new_float_value,
input_bundle.get_specialization_constant<float_id>(),
"float specializaiton constant after setting a value"))
++errors;

if (!check_value(
Expand All @@ -193,9 +193,9 @@ bool test_set_and_get_on_host(sycl::queue q) {
"integer specializaiton constant after build"))
++errors;

if (!check_value(new_double_value,
exec_bundle.get_specialization_constant<double_id>(),
"double specializaiton constant after build"))
if (!check_value(new_float_value,
exec_bundle.get_specialization_constant<float_id>(),
"float specializaiton constant after build"))
++errors;

if (!check_value(new_custom_type_value,
Expand All @@ -210,31 +210,31 @@ bool test_set_and_get_on_host(sycl::queue q) {

bool test_set_and_get_on_device(sycl::queue q) {
sycl::buffer<int> int_buffer(1);
sycl::buffer<double> double_buffer(1);
sycl::buffer<float> float_buffer(1);
sycl::buffer<custom_type> custom_type_buffer(1);

int new_int_value = 42;
double new_double_value = 3.0;
custom_type new_custom_type_value('b', 1.0, 12);
float new_float_value = 3.0f;
custom_type new_custom_type_value('b', 1.0f, 12);

auto input_bundle =
sycl::get_kernel_bundle<sycl::bundle_state::input>(q.get_context());
input_bundle.set_specialization_constant<int_id>(new_int_value);
input_bundle.set_specialization_constant<double_id>(new_double_value);
input_bundle.set_specialization_constant<float_id>(new_float_value);
input_bundle.set_specialization_constant<custom_type_id>(
new_custom_type_value);
auto exec_bundle = sycl::build(input_bundle);

q.submit([&](sycl::handler &cgh) {
cgh.use_kernel_bundle(exec_bundle);
auto int_acc = int_buffer.get_access<sycl::access::mode::write>(cgh);
auto double_acc = double_buffer.get_access<sycl::access::mode::write>(cgh);
auto float_acc = float_buffer.get_access<sycl::access::mode::write>(cgh);
auto custom_type_acc =
custom_type_buffer.get_access<sycl::access::mode::write>(cgh);

cgh.single_task<TestSetAndGetOnDevice>([=](sycl::kernel_handler kh) {
int_acc[0] = kh.get_specialization_constant<int_id>();
double_acc[0] = kh.get_specialization_constant<double_id>();
float_acc[0] = kh.get_specialization_constant<float_id>();
custom_type_acc[0] = kh.get_specialization_constant<custom_type_id>();
});
});
Expand All @@ -244,9 +244,9 @@ bool test_set_and_get_on_device(sycl::queue q) {
"integer specialization constant"))
return false;

auto double_acc = double_buffer.get_access<sycl::access::mode::read>();
if (!check_value(new_double_value, double_acc[0],
"double specialization constant"))
auto float_acc = float_buffer.get_access<sycl::access::mode::read>();
if (!check_value(new_float_value, float_acc[0],
"float specialization constant"))
return false;

auto custom_type_acc =
Expand Down