Skip to content

Commit 62fb6f6

Browse files
aarongreigaarongreig
committed
Add kernel execution test fixture.
Also add new kernel execution tests wherever relevant.
1 parent f85ee5b commit 62fb6f6

File tree

13 files changed

+292
-27
lines changed

13 files changed

+292
-27
lines changed

‎test/conformance/CMakeLists.txt

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -53,7 +53,6 @@ add_subdirectory(usm)
5353
add_subdirectory(event)
5454
add_subdirectory(queue)
5555
add_subdirectory(sampler)
56-
add_subdirectory(enqueue)
5756

5857
if(DEFINED UR_DPCXX)
5958
add_custom_target(generate_device_binaries)
@@ -71,4 +70,5 @@ if(DEFINED UR_DPCXX)
7170
add_subdirectory(device_code)
7271
add_subdirectory(kernel)
7372
add_subdirectory(program)
73+
add_subdirectory(enqueue)
7474
endif()

‎test/conformance/device_code/CMakeLists.txt

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -23,6 +23,8 @@ endfunction()
2323

2424
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/bar.cpp)
2525
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/fill.cpp)
26+
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/fill_2d.cpp)
27+
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/fill_3d.cpp)
2628
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/foo.cpp)
2729
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/image_copy.cpp)
2830
add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/mean.cpp)

‎test/conformance/device_code/fill.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -15,7 +15,7 @@ int main() {
1515
sycl_queue.submit([&](cl::sycl::handler &cgh) {
1616
auto A_acc = A_buff.get_access<cl::sycl::access::mode::write>(cgh);
1717
cgh.parallel_for<class fill>(cl::sycl::range<1>{array_size},
18-
[=](cl::sycl::item<1> itemId) {
18+
[A_acc, val](cl::sycl::item<1> itemId) {
1919
auto id = itemId.get_id(0);
2020
A_acc[id] = val;
2121
});

‎test/conformance/device_code/fill_2d.cpp

Lines changed: 27 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,27 @@
1+
// Copyright (C) 2023 Intel Corporation
2+
// SPDX-License-Identifier: MIT
3+
4+
#include <CL/sycl.hpp>
5+
6+
int main() {
7+
size_t nd_range_x = 8;
8+
size_t nd_range_y = 8;
9+
auto nd_range = cl::sycl::range<2>(nd_range_x, nd_range_y);
10+
11+
std::vector<uint32_t> A(nd_range_x * nd_range_y, 1);
12+
uint32_t val = 42;
13+
cl::sycl::queue sycl_queue;
14+
15+
auto work_range = cl::sycl::nd_range<2>(nd_range, cl::sycl::range<2>(1, 1));
16+
auto A_buff = cl::sycl::buffer<uint32_t>(
17+
A.data(), cl::sycl::range<1>(nd_range_x * nd_range_y));
18+
sycl_queue.submit([&](cl::sycl::handler &cgh) {
19+
auto A_acc = A_buff.get_access<cl::sycl::access::mode::write>(cgh);
20+
cgh.parallel_for<class fill_2d>(
21+
work_range, [A_acc, val](cl::sycl::nd_item<2> item_id) {
22+
auto id = item_id.get_global_linear_id();
23+
A_acc[id] = val;
24+
});
25+
});
26+
return 0;
27+
}

‎test/conformance/device_code/fill_3d.cpp

Lines changed: 28 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,28 @@
1+
// Copyright (C) 2023 Intel Corporation
2+
// SPDX-License-Identifier: MIT
3+
4+
#include <CL/sycl.hpp>
5+
6+
int main() {
7+
size_t nd_range_x = 4;
8+
size_t nd_range_y = 4;
9+
size_t nd_range_z = 4;
10+
auto nd_range = cl::sycl::range<3>(nd_range_x, nd_range_y, nd_range_z);
11+
12+
std::vector<uint32_t> A(nd_range_x * nd_range_y * nd_range_y, 1);
13+
uint32_t val = 42;
14+
cl::sycl::queue sycl_queue;
15+
16+
auto work_range = cl::sycl::nd_range<3>(nd_range, cl::sycl::range<3>(1, 1, 1));
17+
auto A_buff = cl::sycl::buffer<uint32_t>(
18+
A.data(), cl::sycl::range<1>(nd_range_x * nd_range_y));
19+
sycl_queue.submit([&](cl::sycl::handler &cgh) {
20+
auto A_acc = A_buff.get_access<cl::sycl::access::mode::write>(cgh);
21+
cgh.parallel_for<class fill_3d>(
22+
work_range, [A_acc, val](cl::sycl::nd_item<3> item_id) {
23+
auto id = item_id.get_global_linear_id();
24+
A_acc[id] = val;
25+
});
26+
});
27+
return 0;
28+
}

‎test/conformance/enqueue/CMakeLists.txt

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -3,7 +3,7 @@
33
# See LICENSE.TXT
44
# SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
55

6-
add_conformance_test_with_devices_environment(enqueue
6+
add_conformance_test_with_kernels_environment(enqueue
77
urEnqueueEventsWait.cpp
88
urEnqueueEventsWaitWithBarrier.cpp
99
urEnqueueKernelLaunch.cpp

‎test/conformance/enqueue/urEnqueueKernelLaunch.cpp

Lines changed: 118 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -2,3 +2,121 @@
22
// Part of the Unified-Runtime Project, under the Apache License v2.0 with LLVM Exceptions.
33
// See LICENSE.TXT
44
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
5+
6+
#include <uur/fixtures.h>
7+
8+
struct urEnqueueKernelLaunchTest : uur::urKernelExecutionTest {
9+
void SetUp() override {
10+
program_name = "fill";
11+
UUR_RETURN_ON_FATAL_FAILURE(urKernelExecutionTest::SetUp());
12+
}
13+
14+
uint32_t val = 42;
15+
size_t global_size = 32;
16+
size_t global_offset = 0;
17+
size_t n_dimensions = 1;
18+
};
19+
UUR_INSTANTIATE_DEVICE_TEST_SUITE_P(urEnqueueKernelLaunchTest);
20+
21+
TEST_P(urEnqueueKernelLaunchTest, Success) {
22+
ur_mem_handle_t buffer = nullptr;
23+
AddBuffer1DArg(sizeof(val) * global_size, &buffer);
24+
AddPodArg(val);
25+
ASSERT_SUCCESS(urEnqueueKernelLaunch(queue, kernel, n_dimensions,
26+
&global_offset, &global_size, nullptr,
27+
0, nullptr, nullptr));
28+
ASSERT_SUCCESS(urQueueFinish(queue));
29+
ValidateBuffer(buffer, sizeof(val) * global_size, val);
30+
}
31+
32+
TEST_P(urEnqueueKernelLaunchTest, InvalidNullHandleQueue) {
33+
ASSERT_EQ_RESULT(urEnqueueKernelLaunch(nullptr, kernel, n_dimensions,
34+
&global_offset, &global_size,
35+
nullptr, 0, nullptr, nullptr),
36+
UR_RESULT_ERROR_INVALID_NULL_HANDLE);
37+
}
38+
39+
TEST_P(urEnqueueKernelLaunchTest, InvalidNullHandleKernel) {
40+
ASSERT_EQ_RESULT(urEnqueueKernelLaunch(queue, nullptr, n_dimensions,
41+
&global_offset, &global_size,
42+
nullptr, 0, nullptr, nullptr),
43+
UR_RESULT_ERROR_INVALID_NULL_HANDLE);
44+
}
45+
46+
TEST_P(urEnqueueKernelLaunchTest, InvalidNullPtrEventWaitList) {
47+
ASSERT_EQ_RESULT(urEnqueueKernelLaunch(queue, kernel, n_dimensions,
48+
&global_offset, &global_size,
49+
nullptr, 1, nullptr, nullptr),
50+
UR_RESULT_ERROR_INVALID_EVENT_WAIT_LIST);
51+
52+
// does this make sense??
53+
ur_event_handle_t validEvent;
54+
ASSERT_SUCCESS(urEnqueueEventsWait(queue, 0, nullptr, &validEvent));
55+
56+
ASSERT_EQ_RESULT(urEnqueueKernelLaunch(queue, kernel, n_dimensions,
57+
&global_offset, &global_size,
58+
nullptr, 0, &validEvent, nullptr),
59+
UR_RESULT_ERROR_INVALID_EVENT_WAIT_LIST);
60+
}
61+
62+
TEST_P(urEnqueueKernelLaunchTest, InvalidWorkDimension) {
63+
uint32_t max_work_item_dimensions = 0;
64+
ASSERT_SUCCESS(urDeviceGetInfo(
65+
device, UR_DEVICE_INFO_MAX_WORK_ITEM_DIMENSIONS,
66+
sizeof(max_work_item_dimensions), &max_work_item_dimensions, nullptr));
67+
ASSERT_EQ_RESULT(urEnqueueKernelLaunch(queue, kernel,
68+
max_work_item_dimensions + 1,
69+
&global_offset, &global_size,
70+
nullptr, 0, nullptr, nullptr),
71+
UR_RESULT_ERROR_INVALID_WORK_DIMENSION);
72+
}
73+
74+
struct urEnqueueKernelLaunch2DTest : uur::urKernelExecutionTest {
75+
void SetUp() override {
76+
program_name = "fill_2d";
77+
UUR_RETURN_ON_FATAL_FAILURE(urKernelExecutionTest::SetUp());
78+
}
79+
80+
uint32_t val = 42;
81+
size_t global_size[2] = {8, 8};
82+
size_t global_offset[2] = {0, 0};
83+
size_t buffer_size = sizeof(val) * global_size[0] * global_size[1];
84+
size_t n_dimensions = 2;
85+
};
86+
UUR_INSTANTIATE_DEVICE_TEST_SUITE_P(urEnqueueKernelLaunch2DTest);
87+
88+
TEST_P(urEnqueueKernelLaunch2DTest, Success) {
89+
ur_mem_handle_t buffer = nullptr;
90+
AddBuffer1DArg(buffer_size, &buffer);
91+
AddPodArg(val);
92+
ASSERT_SUCCESS(urEnqueueKernelLaunch(queue, kernel, n_dimensions,
93+
global_offset, global_size, nullptr,
94+
0, nullptr, nullptr));
95+
ASSERT_SUCCESS(urQueueFinish(queue));
96+
ValidateBuffer(buffer, buffer_size, val);
97+
}
98+
99+
struct urEnqueueKernelLaunch3DTest : uur::urKernelExecutionTest {
100+
void SetUp() override {
101+
program_name = "fill_3d";
102+
UUR_RETURN_ON_FATAL_FAILURE(urKernelExecutionTest::SetUp());
103+
}
104+
105+
uint32_t val = 42;
106+
size_t global_size[3] = {4, 4, 4};
107+
size_t global_offset[3] = {0, 0, 0};
108+
size_t buffer_size = sizeof(val) * global_size[0] * global_size[1] * global_size[2];
109+
size_t n_dimensions = 3;
110+
};
111+
UUR_INSTANTIATE_DEVICE_TEST_SUITE_P(urEnqueueKernelLaunch3DTest);
112+
113+
TEST_P(urEnqueueKernelLaunch3DTest, Success) {
114+
ur_mem_handle_t buffer = nullptr;
115+
AddBuffer1DArg(buffer_size, &buffer);
116+
AddPodArg(val);
117+
ASSERT_SUCCESS(urEnqueueKernelLaunch(queue, kernel, n_dimensions,
118+
global_offset, global_size, nullptr,
119+
0, nullptr, nullptr));
120+
ASSERT_SUCCESS(urQueueFinish(queue));
121+
ValidateBuffer(buffer, buffer_size, val);
122+
}

‎test/conformance/kernel/urKernelSetArgPointer.cpp

Lines changed: 6 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -24,7 +24,7 @@ struct urKernelSetArgPointerTest : uur::urKernelTest {
2424
UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urKernelSetArgPointerTest);
2525

2626
TEST_P(urKernelSetArgPointerTest, SuccessHost) {
27-
bool host_supported = false;
27+
ur_device_usm_access_capability_flags_t host_supported = false;
2828
ASSERT_SUCCESS(uur::GetDeviceUSMHostSupport(device, host_supported));
2929
if (!host_supported) {
3030
GTEST_SKIP() << "Host USM is not supported.";
@@ -38,7 +38,7 @@ TEST_P(urKernelSetArgPointerTest, SuccessHost) {
3838
}
3939

4040
TEST_P(urKernelSetArgPointerTest, SuccessDevice) {
41-
bool device_supported = false;
41+
ur_device_usm_access_capability_flags_t device_supported = false;
4242
ASSERT_SUCCESS(uur::GetDeviceUSMDeviceSupport(device, device_supported));
4343
if (!device_supported) {
4444
GTEST_SKIP() << "Host USM is not supported.";
@@ -52,7 +52,7 @@ TEST_P(urKernelSetArgPointerTest, SuccessDevice) {
5252
}
5353

5454
TEST_P(urKernelSetArgPointerTest, SuccessShared) {
55-
bool shared_supported = false;
55+
ur_device_usm_access_capability_flags_t shared_supported = false;
5656
ASSERT_SUCCESS(
5757
uur::GetDeviceUSMSingleSharedSupport(device, shared_supported));
5858
if (!shared_supported) {
@@ -69,15 +69,15 @@ TEST_P(urKernelSetArgPointerTest, SuccessShared) {
6969
struct urKernelSetArgPointerNegativeTest : urKernelSetArgPointerTest {
7070
// Get any valid allocation we can to test validation of the other parameters.
7171
void SetUpAllocation() {
72-
bool host_supported = false;
72+
ur_device_usm_access_capability_flags_t host_supported = false;
7373
ASSERT_SUCCESS(uur::GetDeviceUSMHostSupport(device, host_supported));
7474
if (host_supported) {
7575
ASSERT_SUCCESS(urUSMHostAlloc(context, nullptr, nullptr,
7676
allocation_size, &allocation));
7777
return;
7878
}
7979

80-
bool device_supported = false;
80+
ur_device_usm_access_capability_flags_t device_supported = false;
8181
ASSERT_SUCCESS(
8282
uur::GetDeviceUSMDeviceSupport(device, device_supported));
8383
if (device_supported) {
@@ -86,7 +86,7 @@ struct urKernelSetArgPointerNegativeTest : urKernelSetArgPointerTest {
8686
return;
8787
}
8888

89-
bool shared_supported = false;
89+
ur_device_usm_access_capability_flags_t shared_supported = false;
9090
ASSERT_SUCCESS(
9191
uur::GetDeviceUSMSingleSharedSupport(device, shared_supported));
9292
if (shared_supported) {

‎test/conformance/kernel/urKernelSetExecInfo.cpp

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -41,7 +41,7 @@ TEST_P(urKernelSetExecInfoTest, InvalidNullPointerPropValue) {
4141

4242
struct urKernelSetExecInfoUSMPointersTest : uur::urKernelTest {
4343
void SetUp() {
44-
kernel_name = "fill";
44+
program_name = "fill";
4545
UUR_RETURN_ON_FATAL_FAILURE(urKernelTest::SetUp());
4646
}
4747

@@ -58,7 +58,7 @@ struct urKernelSetExecInfoUSMPointersTest : uur::urKernelTest {
5858
UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urKernelSetExecInfoUSMPointersTest);
5959

6060
TEST_P(urKernelSetExecInfoUSMPointersTest, SuccessHost) {
61-
bool host_supported = false;
61+
ur_device_usm_access_capability_flags_t host_supported = false;
6262
ASSERT_SUCCESS(uur::GetDeviceUSMHostSupport(device, host_supported));
6363
if (!host_supported) {
6464
GTEST_SKIP() << "Host USM is not supported.";
@@ -74,7 +74,7 @@ TEST_P(urKernelSetExecInfoUSMPointersTest, SuccessHost) {
7474
}
7575

7676
TEST_P(urKernelSetExecInfoUSMPointersTest, SuccessDevice) {
77-
bool device_supported = false;
77+
ur_device_usm_access_capability_flags_t device_supported = false;
7878
ASSERT_SUCCESS(uur::GetDeviceUSMDeviceSupport(device, device_supported));
7979
if (!device_supported) {
8080
GTEST_SKIP() << "Device USM is not supported.";
@@ -90,7 +90,7 @@ TEST_P(urKernelSetExecInfoUSMPointersTest, SuccessDevice) {
9090
}
9191

9292
TEST_P(urKernelSetExecInfoUSMPointersTest, SuccessShared) {
93-
bool shared_supported = false;
93+
ur_device_usm_access_capability_flags_t shared_supported = false;
9494
ASSERT_SUCCESS(
9595
uur::GetDeviceUSMSingleSharedSupport(device, shared_supported));
9696
if (!shared_supported) {

‎test/conformance/kernel/urKernelSetSpecializationConstants.cpp

Lines changed: 7 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -5,7 +5,7 @@
55

66
#include <uur/fixtures.h>
77

8-
struct urKernelSetSpecializationConstantsTest : uur::urKernelTest {
8+
struct urKernelSetSpecializationConstantsTest : uur::urKernelExecutionTest {
99
void SetUp() override {
1010
bool supports_kernel_spec_constant = false;
1111
ASSERT_SUCCESS(urDeviceGetInfo(
@@ -17,7 +17,7 @@ struct urKernelSetSpecializationConstantsTest : uur::urKernelTest {
1717
<< "Device does not support setting kernel spec constants.";
1818
}
1919
program_name = "spec_constant";
20-
UUR_RETURN_ON_FATAL_FAILURE(urKernelTest::SetUp());
20+
UUR_RETURN_ON_FATAL_FAILURE(urKernelExecutionTest::SetUp());
2121
}
2222

2323
uint32_t spec_value = 42;
@@ -28,7 +28,11 @@ UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urKernelSetSpecializationConstantsTest);
2828

2929
TEST_P(urKernelSetSpecializationConstantsTest, Success) {
3030
ASSERT_SUCCESS(urKernelSetSpecializationConstants(kernel, 1, &info));
31-
// TODO: Run the kernel to verify the spec constant was set.
31+
32+
ur_mem_handle_t buffer;
33+
AddBuffer1DArg(sizeof(spec_value), &buffer);
34+
Launch1DRange(1);
35+
ValidateBuffer<uint32_t>(buffer, sizeof(spec_value), spec_value);
3236
}
3337

3438
TEST_P(urKernelSetSpecializationConstantsTest, InvalidNullHandleKernel) {

0 commit comments

Comments
 (0)