1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116
117
118
119
120
121
122
123
124
125
126
127
128
129
130
131
132
133
134
135
136
137
138
139
140
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155
156
157
158
159
160
161
162
163
164
165
166
167
168
169
170
171
172
173
174
175
176
177
178
179
180
181
182
183
184
185
186
187
188
189
190
191
192
193
194
195
196
197
198
199
200
201
202
203
204
205
206
207
208
209
210
211
212
213
214
215
216
217
218
219
220
221
222
223
224
225
226
227
228
229
230
231
232
233
234
235
236
237
238
239
240
241
242
243
244
245
246
247
248
|
/*
* Copyright (C) 2017-2020 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
*/
#include "opencl/test/unit_test/fixtures/hello_world_fixture.h"
#include "opencl/test/unit_test/fixtures/image_fixture.h"
#include "opencl/test/unit_test/mocks/mock_buffer.h"
class clEventWrapper {
public:
clEventWrapper() { mMem = NULL; }
clEventWrapper(cl_event mem) { mMem = mem; }
clEventWrapper(const clEventWrapper &rhs) : mMem(rhs.mMem) {
if (mMem != NULL)
clRetainEvent(mMem);
}
~clEventWrapper() {
if (mMem != NULL)
clReleaseEvent(mMem);
}
clEventWrapper &operator=(const cl_event &rhs) {
mMem = rhs;
return *this;
}
clEventWrapper &operator=(clEventWrapper rhs) {
std::swap(mMem, rhs.mMem);
return *this;
}
operator cl_event() const { return mMem; }
cl_event *operator&() { return &mMem; }
bool operator==(const cl_event &rhs) { return mMem == rhs; }
protected:
cl_event mMem;
};
using namespace NEO;
namespace ULT {
struct EnqueueWaitlistTest;
typedef HelloWorldTestWithParam<HelloWorldFixtureFactory> EnqueueWaitlistFixture;
typedef void (*ExecuteEnqueue)(EnqueueWaitlistTest *, uint32_t /*cl_uint*/, cl_event *, cl_event *, bool);
struct EnqueueWaitlistTest : public EnqueueWaitlistFixture,
public ::testing::TestWithParam<ExecuteEnqueue> {
public:
typedef CommandQueueHwFixture CommandQueueFixture;
using CommandQueueHwFixture::pCmdQ;
EnqueueWaitlistTest(void) {
buffer = nullptr;
}
void SetUp() override {
EnqueueWaitlistFixture::SetUp();
buffer = BufferHelper<>::create();
bufferNonZeroCopy = new UnalignedBuffer;
image = Image1dHelper<>::create(BufferDefaults::context);
imageNonZeroCopy = ImageHelper<ImageUseHostPtr<Image1dDefaults>>::create(BufferDefaults::context);
}
void TearDown() override {
buffer->decRefInternal();
bufferNonZeroCopy->decRefInternal();
image->decRefInternal();
imageNonZeroCopy->decRefInternal();
EnqueueWaitlistFixture::TearDown();
}
cl_int retVal = CL_SUCCESS;
cl_int error = CL_SUCCESS;
Buffer *buffer;
Buffer *bufferNonZeroCopy;
Image *image;
Image *imageNonZeroCopy;
void test_error(cl_int error, std::string str) {
EXPECT_EQ(CL_SUCCESS, error) << str << std::endl;
}
static void EnqueueNDRange(EnqueueWaitlistTest *test, cl_uint numWaits, cl_event *waits, cl_event *outEvent, bool blocking = false) {
size_t threadNum = 10;
size_t threads[1] = {threadNum};
cl_int error = clEnqueueNDRangeKernel(test->pCmdQ, test->pKernel, 1, NULL, threads, threads, numWaits, waits, outEvent);
test->test_error(error, "Unable to execute kernel");
return;
}
static void EnqueueMapBuffer(EnqueueWaitlistTest *test, cl_uint numWaits, cl_event *waits, cl_event *outEvent, bool blocking = false) {
cl_int error;
void *mappedPtr = clEnqueueMapBuffer(test->pCmdQ, test->buffer, blocking ? CL_TRUE : CL_FALSE, CL_MAP_READ, 0, test->buffer->getSize(), numWaits, waits, outEvent, &error);
EXPECT_NE(nullptr, mappedPtr);
test->test_error(error, "Unable to enqueue buffer map");
error = clEnqueueUnmapMemObject(test->pCmdQ, test->buffer, mappedPtr, 0, nullptr, nullptr);
return;
}
static void TwoEnqueueMapBuffer(EnqueueWaitlistTest *test, cl_uint numWaits, cl_event *waits, cl_event *outEvent, bool blocking = false) {
cl_int error;
void *mappedPtr = clEnqueueMapBuffer(test->pCmdQ, test->buffer, blocking ? CL_TRUE : CL_FALSE, CL_MAP_READ, 0, test->buffer->getSize(), numWaits, waits, outEvent, &error);
EXPECT_NE(nullptr, mappedPtr);
test->test_error(error, "Unable to enqueue buffer map");
void *mappedPtr2 = clEnqueueMapBuffer(test->pCmdQ, test->bufferNonZeroCopy, blocking ? CL_TRUE : CL_FALSE, CL_MAP_READ, 0, test->bufferNonZeroCopy->getSize(), 0, nullptr, nullptr, &error);
EXPECT_NE(nullptr, mappedPtr2);
test->test_error(error, "Unable to enqueue buffer map");
error = clEnqueueUnmapMemObject(test->pCmdQ, test->buffer, mappedPtr, 0, nullptr, nullptr);
error = clEnqueueUnmapMemObject(test->pCmdQ, test->bufferNonZeroCopy, mappedPtr2, 0, nullptr, nullptr);
return;
}
static void EnqueueUnMapBuffer(EnqueueWaitlistTest *test, cl_uint numWaits, cl_event *waits, cl_event *outEvent, bool blocking = false) {
cl_int error;
void *mappedPtr = clEnqueueMapBuffer(test->pCmdQ, test->buffer, CL_TRUE, CL_MAP_READ, 0, test->buffer->getSize(), 0, nullptr, nullptr, &error);
EXPECT_NE(nullptr, mappedPtr);
ASSERT_NE(test->buffer, nullptr);
error = clEnqueueUnmapMemObject(test->pCmdQ, test->buffer, mappedPtr, numWaits, waits, outEvent);
test->test_error(error, "Unable to unmap buffer");
return;
}
static void EnqueueMapImage(EnqueueWaitlistTest *test, cl_uint numWaits, cl_event *waits, cl_event *outEvent, bool blocking = false) {
cl_int error;
cl_image_desc desc = test->image->getImageDesc();
size_t origin[3] = {0, 0, 0}, region[3] = {desc.image_width, desc.image_height, 1};
size_t outPitch;
void *mappedPtr = clEnqueueMapImage(test->pCmdQ, test->image, blocking ? CL_TRUE : CL_FALSE, CL_MAP_READ, origin, region, &outPitch, NULL, numWaits, waits, outEvent, &error);
test->test_error(error, "Unable to enqueue image map");
EXPECT_NE(nullptr, mappedPtr);
test->test_error(error, "Unable to enqueue buffer map");
error = clEnqueueUnmapMemObject(test->pCmdQ, test->image, mappedPtr, 0, nullptr, nullptr);
return;
}
static void TwoEnqueueMapImage(EnqueueWaitlistTest *test, cl_uint numWaits, cl_event *waits, cl_event *outEvent, bool blocking = false) {
cl_int error;
cl_image_desc desc = test->image->getImageDesc();
size_t origin[3] = {0, 0, 0}, region[3] = {desc.image_width, desc.image_height, 1};
size_t outPitch;
size_t origin2[3] = {0, 0, 0}, region2[3] = {desc.image_width, desc.image_height, 1};
size_t outPitch2;
void *mappedPtr = clEnqueueMapImage(test->pCmdQ, test->image, blocking ? CL_TRUE : CL_FALSE, CL_MAP_READ, origin, region, &outPitch, NULL, numWaits, waits, outEvent, &error);
test->test_error(error, "Unable to enqueue image map");
EXPECT_NE(nullptr, mappedPtr);
test->test_error(error, "Unable to enqueue buffer map");
void *mappedPtr2 = clEnqueueMapImage(test->pCmdQ, test->imageNonZeroCopy, blocking ? CL_TRUE : CL_FALSE, CL_MAP_READ, origin2, region2, &outPitch2, NULL, 0, nullptr, nullptr, &error);
test->test_error(error, "Unable to enqueue image map");
EXPECT_NE(nullptr, mappedPtr2);
test->test_error(error, "Unable to enqueue buffer map");
error = clEnqueueUnmapMemObject(test->pCmdQ, test->image, mappedPtr, 0, nullptr, nullptr);
error = clEnqueueUnmapMemObject(test->pCmdQ, test->imageNonZeroCopy, mappedPtr2, 0, nullptr, nullptr);
return;
}
};
TEST_P(EnqueueWaitlistTest, BlockingWaitlist) {
// Set up a user event, which we use as a gate for the second event
clEventWrapper gateEvent = clCreateUserEvent(context, &error);
test_error(error, "Unable to set up user gate event");
// Set up the execution of the action with its actual event
clEventWrapper actualEvent;
// call the function to execute
GetParam()(this, 1, &gateEvent, &actualEvent, false);
// Now release the user event, which will allow our actual action to run
error = clSetUserEventStatus(gateEvent, CL_COMPLETE);
test_error(error, "Unable to trigger gate event");
// Now we wait for completion. Note that we can actually wait on the event itself, at least at first
error = clWaitForEvents(1, &actualEvent);
test_error(error, "Unable to wait for actual test event");
}
typedef EnqueueWaitlistTest EnqueueWaitlistTestTwoMapEnqueues;
TEST_P(EnqueueWaitlistTestTwoMapEnqueues, TestPreviousVirtualEvent) {
// Set up a user event, which we use as a gate for the second event
clEventWrapper gateEvent = clCreateUserEvent(context, &error);
test_error(error, "Unable to set up user gate event");
// Set up the execution of the action with its actual event
clEventWrapper actualEvent;
// call the function to execute
GetParam()(this, 1, &gateEvent, &actualEvent, false);
// Now release the user event, which will allow our actual action to run
error = clSetUserEventStatus(gateEvent, CL_COMPLETE);
// Now we wait for completion. Note that we can actually wait on the event itself, at least at first
error = clWaitForEvents(1, &actualEvent);
test_error(error, "Unable to wait for actual test event");
}
TEST_P(EnqueueWaitlistTest, BlockingWaitlistNoOutEvent) {
// Set up a user event, which we use as a gate for the second event
clEventWrapper gateEvent = clCreateUserEvent(context, &error);
test_error(error, "Unable to set up user gate event");
// call the function to execute
GetParam()(this, 1, &gateEvent, nullptr, false);
// Now release the user event, which will allow our actual action to run
error = clSetUserEventStatus(gateEvent, CL_COMPLETE);
test_error(error, "Unable to trigger gate event");
// Now we wait for completion. Note that we can actually wait on the event itself, at least at first
error = clFinish(pCmdQ);
test_error(error, "Finish FAILED");
}
ExecuteEnqueue Enqueues[] =
{
&EnqueueWaitlistTest::EnqueueNDRange,
&EnqueueWaitlistTest::EnqueueMapBuffer,
&EnqueueWaitlistTest::EnqueueUnMapBuffer,
&EnqueueWaitlistTest::EnqueueMapImage};
ExecuteEnqueue TwoEnqueueMap[] =
{
&EnqueueWaitlistTest::TwoEnqueueMapBuffer,
&EnqueueWaitlistTest::TwoEnqueueMapImage};
INSTANTIATE_TEST_CASE_P(
UnblockedEvent,
EnqueueWaitlistTest,
::testing::ValuesIn(Enqueues));
INSTANTIATE_TEST_CASE_P(
TwoEnqueueMap,
EnqueueWaitlistTestTwoMapEnqueues,
::testing::ValuesIn(TwoEnqueueMap));
} // namespace ULT
|
