1 /*
2  * Copyright (C) 2018-2021 Intel Corporation
3  *
4  * SPDX-License-Identifier: MIT
5  *
6  */
7 
8 #include "shared/source/command_stream/command_stream_receiver.h"
9 #include "shared/source/os_interface/os_context.h"
10 #include "shared/test/common/helpers/debug_manager_state_restore.h"
11 #include "shared/test/common/helpers/unit_test_helper.h"
12 #include "shared/test/common/libult/ult_command_stream_receiver.h"
13 #include "shared/test/common/mocks/mock_allocation_properties.h"
14 #include "shared/test/common/test_macros/test.h"
15 #include "shared/test/common/test_macros/test_checks_shared.h"
16 
17 #include "opencl/source/event/user_event.h"
18 #include "opencl/source/helpers/cl_memory_properties_helpers.h"
19 #include "opencl/test/unit_test/command_queue/command_enqueue_fixture.h"
20 #include "opencl/test/unit_test/command_queue/command_queue_fixture.h"
21 #include "opencl/test/unit_test/fixtures/cl_device_fixture.h"
22 #include "opencl/test/unit_test/fixtures/image_fixture.h"
23 #include "opencl/test/unit_test/mocks/mock_context.h"
24 #include "opencl/test/unit_test/mocks/mock_kernel.h"
25 
26 using namespace NEO;
27 
28 struct EnqueueMapImageTest : public ClDeviceFixture,
29                              public CommandQueueHwFixture,
30                              public ::testing::Test {
31     typedef CommandQueueHwFixture CommandQueueFixture;
32 
EnqueueMapImageTestEnqueueMapImageTest33     EnqueueMapImageTest() {
34     }
35 
SetUpEnqueueMapImageTest36     void SetUp() override {
37         REQUIRE_IMAGES_OR_SKIP(defaultHwInfo);
38 
39         ClDeviceFixture::SetUp();
40         CommandQueueFixture::SetUp(pClDevice, 0);
41         context = new MockContext(pClDevice);
42         image = ImageHelper<ImageUseHostPtr<Image2dDefaults>>::create(context);
43     }
44 
TearDownEnqueueMapImageTest45     void TearDown() override {
46         if (IsSkipped()) {
47             return;
48         }
49         delete image;
50         context->release();
51         CommandQueueFixture::TearDown();
52         ClDeviceFixture::TearDown();
53     }
54 
55     MockContext *context;
56     cl_int retVal = CL_INVALID_VALUE;
57     Image *image = nullptr;
58     char srcMemory[128];
59 };
60 
61 struct EnqueueMapImageParamsTest : public EnqueueMapImageTest,
62                                    public ::testing::WithParamInterface<uint32_t> {
63 };
64 
TEST_F(EnqueueMapImageTest,GivenTiledImageWhenMappingImageThenPointerIsReused)65 TEST_F(EnqueueMapImageTest, GivenTiledImageWhenMappingImageThenPointerIsReused) {
66     auto mapFlags = CL_MAP_READ;
67     const size_t origin[3] = {0, 0, 0};
68     const size_t region[3] = {1, 1, 1};
69 
70     auto mapAllocation = image->getMapAllocation(pClDevice->getRootDeviceIndex());
71     EXPECT_NE(nullptr, mapAllocation);
72 
73     auto ptr1 = pCmdQ->enqueueMapImage(
74         image, true, mapFlags, origin,
75         region, nullptr, nullptr, 0,
76         nullptr, nullptr, retVal);
77     EXPECT_EQ(CL_SUCCESS, retVal);
78     EXPECT_NE(nullptr, image->getHostPtr());
79     mapAllocation = image->getMapAllocation(pClDevice->getRootDeviceIndex());
80     EXPECT_NE(nullptr, mapAllocation);
81 
82     auto ptr2 = pCmdQ->enqueueMapImage(
83         image, true, mapFlags, origin,
84         region, nullptr, nullptr, 0,
85         nullptr, nullptr, retVal);
86     EXPECT_EQ(CL_SUCCESS, retVal);
87 
88     EXPECT_EQ(ptr1, ptr2);
89 
90     retVal = pCmdQ->enqueueUnmapMemObject(image, ptr1, 0, nullptr, nullptr);
91     EXPECT_EQ(CL_SUCCESS, retVal);
92 }
93 
HWTEST_F(EnqueueMapImageTest,givenAllocatedMapPtrAndMapWithDifferentOriginIsCalledThenReturnDifferentPointers)94 HWTEST_F(EnqueueMapImageTest, givenAllocatedMapPtrAndMapWithDifferentOriginIsCalledThenReturnDifferentPointers) {
95     if (!UnitTestHelper<FamilyType>::tiledImagesSupported) {
96         GTEST_SKIP();
97     }
98     std::unique_ptr<Image> img(Image2dHelper<Image2dDefaults>::create(context));
99     auto mapFlags = CL_MAP_READ;
100     const size_t origin1[3] = {0, 0, 0};
101     const size_t origin2[3] = {2, 2, 0};
102     const size_t region[3] = {1, 1, 1};
103 
104     auto ptr1 = pCmdQ->enqueueMapImage(img.get(), true, mapFlags, origin1,
105                                        region, nullptr, nullptr, 0,
106                                        nullptr, nullptr, retVal);
107     EXPECT_EQ(CL_SUCCESS, retVal);
108 
109     auto ptr2 = pCmdQ->enqueueMapImage(img.get(), true, mapFlags, origin2,
110                                        region, nullptr, nullptr, 0,
111                                        nullptr, nullptr, retVal);
112     EXPECT_EQ(CL_SUCCESS, retVal);
113 
114     EXPECT_NE(ptr1, ptr2);
115     EXPECT_NE(nullptr, img->getAllocatedMapPtr());
116 
117     size_t mapOffset = img->getSurfaceFormatInfo().surfaceFormat.ImageElementSizeInBytes * origin2[0] +
118                        img->getHostPtrRowPitch() * origin2[1];
119     EXPECT_EQ(ptr2, ptrOffset(ptr1, mapOffset));
120 }
121 
122 typedef EnqueueMapImageParamsTest MipMapMapImageParamsTest;
123 
TEST_P(MipMapMapImageParamsTest,givenAllocatedMapPtrWhenMapsWithDifferentMipMapsAreCalledThenReturnDifferentPointers)124 TEST_P(MipMapMapImageParamsTest, givenAllocatedMapPtrWhenMapsWithDifferentMipMapsAreCalledThenReturnDifferentPointers) {
125     auto image_type = (cl_mem_object_type)GetParam();
126     cl_int retVal = CL_SUCCESS;
127     cl_image_desc imageDesc = {};
128     imageDesc.image_type = image_type;
129     imageDesc.num_mip_levels = 10;
130     imageDesc.image_width = 4;
131     imageDesc.image_height = 1;
132     imageDesc.image_depth = 1;
133     const size_t origin1[4] = {0, 0, 0, 0};
134     size_t origin2[4] = {0, 0, 0, 0};
135     std::unique_ptr<Image> image;
136     size_t mapOffset = 16u;
137     switch (image_type) {
138     case CL_MEM_OBJECT_IMAGE1D:
139         origin2[1] = 1;
140         image = std::unique_ptr<Image>(ImageHelper<Image1dDefaults>::create(context, &imageDesc));
141         break;
142     case CL_MEM_OBJECT_IMAGE1D_ARRAY:
143         origin2[2] = 1;
144         imageDesc.image_array_size = 2;
145         image = std::unique_ptr<Image>(ImageHelper<Image1dArrayDefaults>::create(context, &imageDesc));
146         break;
147     case CL_MEM_OBJECT_IMAGE2D:
148         origin2[2] = 1;
149         image = std::unique_ptr<Image>(ImageHelper<Image2dDefaults>::create(context, &imageDesc));
150         break;
151     case CL_MEM_OBJECT_IMAGE2D_ARRAY:
152         origin2[3] = 1;
153         imageDesc.image_array_size = 2;
154         image = std::unique_ptr<Image>(ImageHelper<Image2dArrayDefaults>::create(context, &imageDesc));
155         break;
156     case CL_MEM_OBJECT_IMAGE3D:
157         origin2[3] = 1;
158         image = std::unique_ptr<Image>(ImageHelper<Image3dDefaults>::create(context, &imageDesc));
159         break;
160     }
161     EXPECT_NE(nullptr, image.get());
162 
163     auto mapFlags = CL_MAP_READ;
164     const size_t region[3] = {1, 1, 1};
165 
166     auto ptr1 = pCmdQ->enqueueMapImage(image.get(), true, mapFlags, origin1,
167                                        region, nullptr, nullptr, 0,
168                                        nullptr, nullptr, retVal);
169     EXPECT_EQ(CL_SUCCESS, retVal);
170 
171     auto ptr2 = pCmdQ->enqueueMapImage(image.get(), true, mapFlags, origin2,
172                                        region, nullptr, nullptr, 0,
173                                        nullptr, nullptr, retVal);
174     EXPECT_EQ(CL_SUCCESS, retVal);
175 
176     EXPECT_NE(ptr1, ptr2);
177     if (image->mappingOnCpuAllowed() == false) {
178         EXPECT_NE(nullptr, image->getAllocatedMapPtr());
179     }
180 
181     EXPECT_EQ(ptr2, ptrOffset(ptr1, mapOffset));
182 }
183 
184 INSTANTIATE_TEST_CASE_P(MipMapMapImageParamsTest_givenAllocatedMapPtrAndMapWithDifferentMipMapsIsCalledThenReturnDifferentPointers,
185                         MipMapMapImageParamsTest, ::testing::Values(CL_MEM_OBJECT_IMAGE1D, CL_MEM_OBJECT_IMAGE1D_ARRAY, CL_MEM_OBJECT_IMAGE2D, CL_MEM_OBJECT_IMAGE2D_ARRAY, CL_MEM_OBJECT_IMAGE3D));
186 
187 template <typename GfxFamily>
188 struct mockedImage : public ImageHw<GfxFamily> {
189     using ImageHw<GfxFamily>::ImageHw;
setAllocatedMapPtrmockedImage190     void setAllocatedMapPtr(void *allocatedMapPtr) override {
191         ownershipTaken = this->hasOwnership();
192         MemObj::setAllocatedMapPtr(allocatedMapPtr);
193     }
194     bool ownershipTaken = false;
195 };
196 
HWTEST_F(EnqueueMapImageTest,givenTiledImageWhenMapImageIsCalledThenStorageIsSetWithImageMutexTaken)197 HWTEST_F(EnqueueMapImageTest, givenTiledImageWhenMapImageIsCalledThenStorageIsSetWithImageMutexTaken) {
198     if (!UnitTestHelper<FamilyType>::tiledImagesSupported) {
199         GTEST_SKIP();
200     }
201     auto imageFormat = image->getImageFormat();
202     auto imageDesc = image->getImageDesc();
203     auto graphicsAllocation = image->getGraphicsAllocation(pClDevice->getRootDeviceIndex());
204     auto surfaceFormatInfo = image->getSurfaceFormatInfo();
205 
206     mockedImage<FamilyType> mockImage(context,
207                                       {},
208                                       0,
209                                       0,
210                                       4096u,
211                                       nullptr,
212                                       nullptr,
213                                       imageFormat,
214                                       imageDesc,
215                                       false,
216                                       GraphicsAllocationHelper::toMultiGraphicsAllocation(graphicsAllocation),
217                                       true,
218                                       0,
219                                       0,
220                                       surfaceFormatInfo,
221                                       nullptr);
222 
223     mockImage.createFunction = image->createFunction;
224 
225     auto mapAllocation = mockImage.getMapAllocation(pClDevice->getRootDeviceIndex());
226     EXPECT_EQ(nullptr, mapAllocation);
227     EXPECT_EQ(nullptr, mockImage.getHostPtr());
228 
229     auto mapFlags = CL_MAP_READ;
230     const size_t origin[3] = {0, 0, 0};
231     const size_t region[3] = {1, 1, 1};
232 
233     auto apiMapPtr = pCmdQ->enqueueMapImage(
234         &mockImage, true, mapFlags, origin,
235         region, nullptr, nullptr, 0,
236         nullptr, nullptr, retVal);
237     EXPECT_TRUE(mockImage.ownershipTaken);
238 
239     auto mapPtr = mockImage.getAllocatedMapPtr();
240     EXPECT_EQ(apiMapPtr, mapPtr);
241     mapAllocation = mockImage.getMapAllocation(pClDevice->getRootDeviceIndex());
242     EXPECT_NE(nullptr, mapAllocation);
243     EXPECT_EQ(apiMapPtr, mapAllocation->getUnderlyingBuffer());
244 
245     auto osContextId = pCmdQ->getGpgpuCommandStreamReceiver().getOsContext().getContextId();
246     auto expectedTaskCount = pCmdQ->getGpgpuCommandStreamReceiver().peekTaskCount();
247     auto actualMapAllocationTaskCount = mapAllocation->getTaskCount(osContextId);
248     EXPECT_EQ(expectedTaskCount, actualMapAllocationTaskCount);
249 
250     pDevice->getMemoryManager()->freeGraphicsMemory(mockImage.getMapAllocation(pClDevice->getRootDeviceIndex()));
251     mockImage.releaseAllocatedMapPtr();
252 }
253 
TEST_F(EnqueueMapImageTest,WhenMappingImageThenCpuAndGpuAddressAreEqualWhenZeroCopyIsUsed)254 TEST_F(EnqueueMapImageTest, WhenMappingImageThenCpuAndGpuAddressAreEqualWhenZeroCopyIsUsed) {
255     auto mapFlags = CL_MAP_READ;
256     const size_t origin[3] = {0, 0, 0};
257     const size_t region[3] = {1, 1, 1};
258     size_t imageRowPitch = 0;
259     size_t imageSlicePitch = 0;
260     auto ptr = pCmdQ->enqueueMapImage(
261         image,
262         true,
263         mapFlags,
264         origin,
265         region,
266         &imageRowPitch,
267         &imageSlicePitch,
268         0,
269         nullptr,
270         nullptr,
271         retVal);
272     if (image->isMemObjZeroCopy()) {
273         EXPECT_EQ(image->getCpuAddress(), ptr);
274     } else {
275         EXPECT_NE(image->getCpuAddress(), ptr);
276     }
277     size_t imageRowPitchRef = 0;
278     image->getImageInfo(CL_IMAGE_ROW_PITCH, sizeof(imageRowPitchRef), &imageRowPitchRef, nullptr);
279     EXPECT_EQ(imageRowPitch, imageRowPitchRef);
280 
281     size_t imageSlicePitchRef = 0;
282     image->getImageInfo(CL_IMAGE_SLICE_PITCH, sizeof(imageSlicePitchRef), &imageSlicePitchRef, nullptr);
283     EXPECT_EQ(imageSlicePitch, imageSlicePitchRef);
284 }
285 
TEST_F(EnqueueMapImageTest,GivenCmdqAndValidArgsWhenMappingImageThenSuccessIsReturned)286 TEST_F(EnqueueMapImageTest, GivenCmdqAndValidArgsWhenMappingImageThenSuccessIsReturned) {
287     auto mapFlags = CL_MAP_READ;
288     const size_t origin[3] = {0, 0, 0};
289     const size_t region[3] = {1, 1, 1};
290     size_t imageRowPitch = 0;
291     size_t imageSlicePitch = 0;
292     auto ptr = pCmdQ->enqueueMapImage(
293         image,
294         true,
295         mapFlags,
296         origin,
297         region,
298         &imageRowPitch,
299         &imageSlicePitch,
300         0,
301         nullptr,
302         nullptr,
303         retVal);
304     EXPECT_NE(nullptr, ptr);
305     EXPECT_EQ(CL_SUCCESS, retVal);
306 
307     size_t imageRowPitchRef = 0;
308     image->getImageInfo(CL_IMAGE_ROW_PITCH, sizeof(imageRowPitchRef), &imageRowPitchRef, nullptr);
309     EXPECT_EQ(imageRowPitch, imageRowPitchRef);
310 
311     size_t imageSlicePitchRef = 0;
312     image->getImageInfo(CL_IMAGE_SLICE_PITCH, sizeof(imageSlicePitchRef), &imageSlicePitchRef, nullptr);
313     EXPECT_EQ(imageSlicePitch, imageSlicePitchRef);
314 }
315 
HWTEST_F(EnqueueMapImageTest,givenNonReadOnlyMapWithOutEventWhenMappedThenSetEventAndIncraseTaskCountFromWriteImage)316 HWTEST_F(EnqueueMapImageTest, givenNonReadOnlyMapWithOutEventWhenMappedThenSetEventAndIncraseTaskCountFromWriteImage) {
317     if (!UnitTestHelper<FamilyType>::tiledImagesSupported) {
318         GTEST_SKIP();
319     }
320     DebugManagerStateRestore dbgRestore;
321     DebugManager.flags.EnableAsyncEventsHandler.set(false);
322     cl_event mapEventReturned = nullptr;
323     cl_event unmapEventReturned = nullptr;
324     uint32_t tagHW = 0;
325     auto mapFlags = CL_MAP_WRITE;
326     const size_t origin[3] = {0, 0, 0};
327     const size_t region[3] = {1, 1, 1};
328     size_t imageRowPitch = 0;
329     size_t imageSlicePitch = 0;
330     size_t GWS = 1;
331 
332     MockKernelWithInternals kernel(*pClDevice);
333     *pTagMemory = tagHW;
334     auto &commandStreamReceiver = pCmdQ->getGpgpuCommandStreamReceiver();
335     auto tag_address = commandStreamReceiver.getTagAddress();
336     EXPECT_TRUE(pTagMemory == tag_address);
337 
338     struct E2Clb {
339         static void CL_CALLBACK SignalEv2(cl_event e, cl_int status, void *data) {
340             uint32_t *pTagMem = static_cast<uint32_t *>(data);
341             *pTagMem = 4;
342         }
343     };
344 
345     uint32_t taskCount = commandStreamReceiver.peekTaskCount();
346     EXPECT_EQ(1u, taskCount);
347 
348     // enqueue something that can be finished...
349     retVal = clEnqueueNDRangeKernel(pCmdQ, kernel.mockMultiDeviceKernel, 1, 0, &GWS, nullptr, 0, nullptr, nullptr);
350     EXPECT_EQ(retVal, CL_SUCCESS);
351 
352     *pTagMemory = tagHW += 3;
353     auto ptr = pCmdQ->enqueueMapImage(
354         image,
355         false,
356         mapFlags,
357         origin,
358         region,
359         &imageRowPitch,
360         &imageSlicePitch,
361         0,
362         nullptr,
363         &mapEventReturned,
364         retVal);
365 
366     EXPECT_NE(nullptr, ptr);
367     EXPECT_EQ(CL_SUCCESS, retVal);
368     auto mapEvent = castToObject<Event>(mapEventReturned);
369     EXPECT_TRUE(CL_COMMAND_MAP_IMAGE == mapEvent->getCommandType());
370 
371     taskCount = commandStreamReceiver.peekTaskCount();
372     EXPECT_EQ(3u, taskCount);
373 
374     clSetEventCallback(mapEventReturned, CL_COMPLETE, E2Clb::SignalEv2, (void *)pTagMemory);
375 
376     retVal = clWaitForEvents(1, &mapEventReturned);
377     EXPECT_EQ(CL_SUCCESS, retVal);
378     EXPECT_EQ(4u, *pTagMemory);
379     taskCount = commandStreamReceiver.peekTaskCount();
380     EXPECT_EQ(3u, taskCount);
381 
382     (*pTagMemory)++;
383     retVal = clEnqueueUnmapMemObject(
384         pCmdQ,
385         image,
386         ptr,
387         0,
388         nullptr,
389         &unmapEventReturned);
390     EXPECT_EQ(CL_SUCCESS, retVal);
391     auto unmapEvent = castToObject<Event>(unmapEventReturned);
392     EXPECT_TRUE(CL_COMMAND_UNMAP_MEM_OBJECT == unmapEvent->getCommandType());
393 
394     retVal = clWaitForEvents(1, &unmapEventReturned);
395 
396     taskCount = commandStreamReceiver.peekTaskCount();
397     EXPECT_EQ(4u, taskCount);
398 
399     clReleaseEvent(mapEventReturned);
400     clReleaseEvent(unmapEventReturned);
401 }
402 
HWTEST_F(EnqueueMapImageTest,givenReadOnlyMapWithOutEventWhenMappedThenSetEventAndDontIncraseTaskCountFromWriteImage)403 HWTEST_F(EnqueueMapImageTest, givenReadOnlyMapWithOutEventWhenMappedThenSetEventAndDontIncraseTaskCountFromWriteImage) {
404     if (!UnitTestHelper<FamilyType>::tiledImagesSupported) {
405         GTEST_SKIP();
406     }
407     DebugManagerStateRestore dbgRestore;
408     DebugManager.flags.EnableAsyncEventsHandler.set(false);
409     cl_event mapEventReturned = nullptr;
410     cl_event unmapEventReturned = nullptr;
411     auto mapFlags = CL_MAP_READ;
412     const size_t origin[3] = {0, 0, 0};
413     const size_t region[3] = {1, 1, 1};
414     *pTagMemory = 5;
415 
416     auto &commandStreamReceiver = pCmdQ->getGpgpuCommandStreamReceiver();
417 
418     EXPECT_EQ(1u, commandStreamReceiver.peekTaskCount());
419 
420     auto ptr = pCmdQ->enqueueMapImage(image, false, mapFlags, origin, region, nullptr, nullptr, 0,
421                                       nullptr, &mapEventReturned, retVal);
422 
423     EXPECT_NE(nullptr, ptr);
424     EXPECT_EQ(CL_SUCCESS, retVal);
425     EXPECT_EQ(2u, commandStreamReceiver.peekTaskCount());
426 
427     auto mapEvent = castToObject<Event>(mapEventReturned);
428     EXPECT_TRUE(CL_COMMAND_MAP_IMAGE == mapEvent->getCommandType());
429 
430     retVal = clWaitForEvents(1, &mapEventReturned);
431     EXPECT_EQ(CL_SUCCESS, retVal);
432 
433     retVal = clEnqueueUnmapMemObject(pCmdQ, image, ptr, 0, nullptr, &unmapEventReturned);
434     EXPECT_EQ(CL_SUCCESS, retVal);
435 
436     EXPECT_EQ(2u, commandStreamReceiver.peekTaskCount());
437 
438     auto unmapEvent = castToObject<Event>(unmapEventReturned);
439     EXPECT_TRUE(CL_COMMAND_UNMAP_MEM_OBJECT == unmapEvent->getCommandType());
440 
441     retVal = clWaitForEvents(1, &unmapEventReturned);
442     EXPECT_EQ(CL_SUCCESS, retVal);
443 
444     clReleaseEvent(mapEventReturned);
445     clReleaseEvent(unmapEventReturned);
446 }
447 
HWTEST_F(EnqueueMapImageTest,GivenPtrToReturnEventWhenMappingImageThenEventIsNotNull)448 HWTEST_F(EnqueueMapImageTest, GivenPtrToReturnEventWhenMappingImageThenEventIsNotNull) {
449     if (!UnitTestHelper<FamilyType>::tiledImagesSupported) {
450         GTEST_SKIP();
451     }
452     cl_event eventReturned = nullptr;
453     auto mapFlags = CL_MAP_READ;
454     const size_t origin[3] = {0, 0, 0};
455     const size_t region[3] = {1, 1, 1};
456     size_t imageRowPitch = 0;
457     size_t imageSlicePitch = 0;
458     uint32_t forceTaskCount = 100;
459 
460     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
461     commandStreamReceiver.taskCount = forceTaskCount;
462 
463     auto ptr = clEnqueueMapImage(
464         pCmdQ,
465         image,
466         CL_FALSE,
467         mapFlags,
468         origin,
469         region,
470         &imageRowPitch,
471         &imageSlicePitch,
472         0,
473         nullptr,
474         &eventReturned,
475         &retVal);
476 
477     EXPECT_NE(nullptr, ptr);
478     EXPECT_EQ(CL_SUCCESS, retVal);
479     EXPECT_NE(nullptr, eventReturned);
480 
481     auto eventObject = castToObject<Event>(eventReturned);
482     EXPECT_EQ(forceTaskCount + 1, eventObject->peekTaskCount());
483     EXPECT_TRUE(eventObject->updateStatusAndCheckCompletion());
484 
485     retVal = clEnqueueUnmapMemObject(
486         pCmdQ,
487         image,
488         ptr,
489         0,
490         nullptr,
491         nullptr);
492     EXPECT_EQ(CL_SUCCESS, retVal);
493 
494     clReleaseEvent(eventReturned);
495 }
496 
HWTEST_F(EnqueueMapImageTest,givenZeroCopyImageWhenItIsMappedAndReturnsEventThenEventHasCorrectProperties)497 HWTEST_F(EnqueueMapImageTest, givenZeroCopyImageWhenItIsMappedAndReturnsEventThenEventHasCorrectProperties) {
498     cl_event eventReturned = nullptr;
499     auto mapFlags = CL_MAP_READ;
500     const size_t origin[3] = {0, 0, 0};
501     const size_t region[3] = {1, 1, 1};
502     size_t imageRowPitch = 0;
503     size_t imageSlicePitch = 0;
504     uint32_t forceTaskCount = 100;
505 
506     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
507     commandStreamReceiver.taskCount = forceTaskCount;
508 
509     std::unique_ptr<Image> zero_copy_image(ImageHelper<ImageWriteOnly<Image1dDefaults>>::create(context));
510 
511     ASSERT_TRUE(zero_copy_image->isMemObjZeroCopy());
512     pCmdQ->taskCount = 40u;
513 
514     auto ptr = clEnqueueMapImage(
515         pCmdQ,
516         zero_copy_image.get(),
517         CL_FALSE,
518         mapFlags,
519         origin,
520         region,
521         &imageRowPitch,
522         &imageSlicePitch,
523         0,
524         nullptr,
525         &eventReturned,
526         &retVal);
527 
528     EXPECT_NE(nullptr, ptr);
529     EXPECT_EQ(CL_SUCCESS, retVal);
530     EXPECT_NE(nullptr, eventReturned);
531 
532     EXPECT_EQ(ptr, zero_copy_image->getCpuAddressForMemoryTransfer());
533 
534     auto eventObject = castToObject<Event>(eventReturned);
535     EXPECT_EQ(pCmdQ->taskCount, eventObject->peekTaskCount());
536     EXPECT_TRUE(eventObject->updateStatusAndCheckCompletion());
537 
538     retVal = clEnqueueUnmapMemObject(
539         pCmdQ,
540         zero_copy_image.get(),
541         ptr,
542         0,
543         nullptr,
544         nullptr);
545     EXPECT_EQ(CL_SUCCESS, retVal);
546 
547     clReleaseEvent(eventReturned);
548 }
549 
TEST_F(EnqueueMapImageTest,GivenNonZeroCopyImageWhenMappedWithOffsetThenCorrectPointerIsReturned)550 TEST_F(EnqueueMapImageTest, GivenNonZeroCopyImageWhenMappedWithOffsetThenCorrectPointerIsReturned) {
551     auto mapFlags = CL_MAP_WRITE;
552     const size_t origin[3] = {1, 0, 0};
553     const size_t region[3] = {1, 1, 1};
554     size_t imageRowPitch = 0;
555     size_t imageSlicePitch = 0;
556 
557     Image *nonZeroCopyImage = ImageHelper<ImageUseHostPtr<Image1dDefaults>>::create(context);
558 
559     EXPECT_FALSE(nonZeroCopyImage->isMemObjZeroCopy());
560 
561     auto ptr = clEnqueueMapImage(
562         pCmdQ,
563         nonZeroCopyImage,
564         CL_TRUE,
565         mapFlags,
566         origin,
567         region,
568         &imageRowPitch,
569         &imageSlicePitch,
570         0,
571         nullptr,
572         nullptr,
573         &retVal);
574 
575     float *HostPtrOffseted = (float *)Image1dDefaults::hostPtr + 1; //
576 
577     EXPECT_NE(nullptr, ptr);
578 
579     if (!image->isTiledAllocation()) {
580         EXPECT_EQ(HostPtrOffseted, ptr); // Returned pointer should be offseted
581     }
582 
583     EXPECT_EQ(CL_SUCCESS, retVal);
584 
585     retVal = clEnqueueUnmapMemObject(
586         pCmdQ,
587         nonZeroCopyImage,
588         ptr,
589         0,
590         nullptr,
591         nullptr);
592 
593     EXPECT_EQ(CL_SUCCESS, retVal);
594 
595     delete nonZeroCopyImage;
596 }
597 
HWTEST_F(EnqueueMapImageTest,givenSharingHandlerWhenNonReadOnlyMapAndUnmapOnNonTiledImageIsCalledThenMakeGpuCopy)598 HWTEST_F(EnqueueMapImageTest, givenSharingHandlerWhenNonReadOnlyMapAndUnmapOnNonTiledImageIsCalledThenMakeGpuCopy) {
599     std::unique_ptr<Image> image(ImageHelper<ImageUseHostPtr<Image1dDefaults>>::create(context));
600     ASSERT_NE(nullptr, image);
601     image->setSharingHandler(new SharingHandler());
602     EXPECT_FALSE(image->isTiledAllocation());
603 
604     auto &csr = pDevice->getUltCommandStreamReceiver<FamilyType>();
605     csr.taskCount = 1;
606     csr.taskLevel = 1;
607     pCmdQ->taskCount = 1;
608     pCmdQ->taskLevel = 1;
609 
610     size_t origin[] = {0, 0, 0};
611     size_t region[] = {1, 1, 1};
612     void *data = clEnqueueMapImage(pCmdQ, image.get(), CL_TRUE, CL_MAP_WRITE, origin, region, nullptr, nullptr, 0, NULL, NULL, &retVal);
613     EXPECT_NE(nullptr, data);
614     EXPECT_EQ(CL_SUCCESS, retVal);
615     EXPECT_EQ(2u, pCmdQ->taskCount);
616     EXPECT_EQ(2u, pCmdQ->taskLevel);
617 
618     retVal = clEnqueueUnmapMemObject(pCmdQ, image.get(), data, 0, NULL, NULL);
619     EXPECT_EQ(3u, pCmdQ->taskCount);
620     EXPECT_EQ(3u, pCmdQ->taskLevel);
621 }
622 
HWTEST_F(EnqueueMapImageTest,givenSharingHandlerWhenReadOnlyMapAndUnmapOnNonTiledImageIsCalledThenMakeGpuCopy)623 HWTEST_F(EnqueueMapImageTest, givenSharingHandlerWhenReadOnlyMapAndUnmapOnNonTiledImageIsCalledThenMakeGpuCopy) {
624     std::unique_ptr<Image> image(ImageHelper<ImageUseHostPtr<Image1dDefaults>>::create(context));
625     ASSERT_NE(nullptr, image);
626     image->setSharingHandler(new SharingHandler());
627     EXPECT_FALSE(image->isTiledAllocation());
628 
629     auto &csr = pDevice->getUltCommandStreamReceiver<FamilyType>();
630     csr.taskCount = 1;
631     csr.taskLevel = 1;
632     pCmdQ->taskCount = 1;
633     pCmdQ->taskLevel = 1;
634 
635     size_t origin[] = {0, 0, 0};
636     size_t region[] = {1, 1, 1};
637     void *data = clEnqueueMapImage(pCmdQ, image.get(), CL_TRUE, CL_MAP_READ, origin, region, nullptr, nullptr, 0, NULL, NULL, &retVal);
638     EXPECT_NE(nullptr, data);
639     EXPECT_EQ(CL_SUCCESS, retVal);
640     EXPECT_EQ(2u, pCmdQ->taskCount);
641     EXPECT_EQ(2u, pCmdQ->taskLevel);
642 
643     retVal = clEnqueueUnmapMemObject(pCmdQ, image.get(), data, 0, NULL, NULL);
644     EXPECT_EQ(2u, pCmdQ->taskCount);
645     EXPECT_EQ(2u, pCmdQ->taskLevel);
646 }
647 
HWTEST_F(EnqueueMapImageTest,givenImageWithouUsetHostPtrFlagWhenMappedOnCpuThenSetAllMapProperties)648 HWTEST_F(EnqueueMapImageTest, givenImageWithouUsetHostPtrFlagWhenMappedOnCpuThenSetAllMapProperties) {
649     std::unique_ptr<Image> image(ImageHelper<Image1dDefaults>::create(context));
650     ASSERT_NE(nullptr, image);
651     EXPECT_TRUE(image->mappingOnCpuAllowed());
652 
653     size_t origin[] = {2, 0, 0};
654     size_t region[] = {2, 1, 1};
655     void *mappedPtr = clEnqueueMapImage(pCmdQ, image.get(), CL_TRUE, CL_MAP_READ, origin, region, nullptr, nullptr, 0, NULL, NULL, &retVal);
656     EXPECT_NE(nullptr, mappedPtr);
657 
658     MapInfo mappedInfo;
659     auto success = image->findMappedPtr(mappedPtr, mappedInfo);
660     EXPECT_TRUE(success);
661     EXPECT_NE(nullptr, mappedInfo.ptr);
662 
663     EXPECT_EQ(origin[0], mappedInfo.offset[0]);
664     EXPECT_EQ(origin[1], mappedInfo.offset[1]);
665     EXPECT_EQ(origin[2], mappedInfo.offset[2]);
666 
667     EXPECT_EQ(region[0], mappedInfo.size[0]);
668     EXPECT_EQ(region[1], mappedInfo.size[1]);
669     EXPECT_EQ(region[2], mappedInfo.size[2]);
670 
671     auto expectedPtr = ptrOffset(image->getCpuAddressForMapping(), image->calculateOffsetForMapping(mappedInfo.offset));
672 
673     EXPECT_EQ(mappedPtr, expectedPtr);
674 }
675 
HWTEST_F(EnqueueMapImageTest,givenImageWithUseHostPtrFlagWhenMappedOnCpuThenSetAllMapProperties)676 HWTEST_F(EnqueueMapImageTest, givenImageWithUseHostPtrFlagWhenMappedOnCpuThenSetAllMapProperties) {
677     std::unique_ptr<Image> image(ImageHelper<ImageUseHostPtr<Image1dDefaults>>::create(context));
678     ASSERT_NE(nullptr, image);
679     EXPECT_TRUE(image->mappingOnCpuAllowed());
680 
681     size_t origin[] = {2, 0, 0};
682     size_t region[] = {2, 1, 1};
683     void *mappedPtr = clEnqueueMapImage(pCmdQ, image.get(), CL_TRUE, CL_MAP_READ, origin, region, nullptr, nullptr, 0, NULL, NULL, &retVal);
684     EXPECT_NE(nullptr, mappedPtr);
685 
686     MapInfo mappedInfo;
687     auto success = image->findMappedPtr(mappedPtr, mappedInfo);
688     EXPECT_TRUE(success);
689     EXPECT_NE(nullptr, mappedInfo.ptr);
690 
691     EXPECT_EQ(origin[0], mappedInfo.offset[0]);
692     EXPECT_EQ(origin[1], mappedInfo.offset[1]);
693     EXPECT_EQ(origin[2], mappedInfo.offset[2]);
694 
695     EXPECT_EQ(region[0], mappedInfo.size[0]);
696     EXPECT_EQ(region[1], mappedInfo.size[1]);
697     EXPECT_EQ(region[2], mappedInfo.size[2]);
698 
699     auto expectedPtr = ptrOffset(image->getCpuAddressForMapping(), image->calculateOffsetForMapping(mappedInfo.offset));
700 
701     EXPECT_EQ(mappedPtr, expectedPtr);
702 }
703 
TEST_F(EnqueueMapImageTest,givenBlockedCommandQueueWhenBlockingMapWith2DImageIsEnqueuedAndEventAsynchrounouslyCompletedThenEnqueueFinishesWithoutStall)704 TEST_F(EnqueueMapImageTest, givenBlockedCommandQueueWhenBlockingMapWith2DImageIsEnqueuedAndEventAsynchrounouslyCompletedThenEnqueueFinishesWithoutStall) {
705     auto mapFlags = CL_MAP_READ;
706     const size_t origin[3] = {0, 0, 0};
707     const size_t region[3] = {1, 1, 1};
708     size_t imageRowPitch = 0;
709     size_t imageSlicePitch = 0;
710 
711     class MockEventWithSetCompleteOnUpdate : public Event {
712       public:
713         MockEventWithSetCompleteOnUpdate(CommandQueue *cmdQueue, cl_command_type cmdType,
714                                          uint32_t taskLevel, uint32_t taskCount) : Event(cmdQueue, cmdType, taskLevel, taskCount) {
715         }
716         void updateExecutionStatus() override {
717             setStatus(CL_COMPLETE);
718         }
719     };
720     MockEventWithSetCompleteOnUpdate blockingEvent(pCmdQ, CL_COMMAND_NDRANGE_KERNEL, 0, 1);
721 
722     cl_event blockingClEvent = &blockingEvent;
723 
724     int32_t initialRefCountCmdQ = pCmdQ->getRefInternalCount();
725 
726     auto ptr = pCmdQ->enqueueMapImage(
727         image,
728         true,
729         mapFlags,
730         origin,
731         region,
732         &imageRowPitch,
733         &imageSlicePitch,
734         1,
735         &blockingClEvent,
736         nullptr,
737         retVal);
738     EXPECT_NE(nullptr, ptr);
739     EXPECT_EQ(CL_SUCCESS, retVal);
740     EXPECT_EQ(initialRefCountCmdQ, pCmdQ->getRefInternalCount());
741 }
742 
TEST_F(EnqueueMapImageTest,givenBlockedCommandQueueWhenBlockingMapWith1DImageIsEnqueuedAndEventAsynchrounouslyCompletedThenEnqueueFinishesWithoutStall)743 TEST_F(EnqueueMapImageTest, givenBlockedCommandQueueWhenBlockingMapWith1DImageIsEnqueuedAndEventAsynchrounouslyCompletedThenEnqueueFinishesWithoutStall) {
744     auto mapFlags = CL_MAP_READ;
745     const size_t origin[3] = {0, 0, 0};
746     const size_t region[3] = {1, 1, 1};
747     size_t imageRowPitch = 0;
748     size_t imageSlicePitch = 0;
749 
750     Image *image1D = ImageHelper<ImageUseHostPtr<Image1dDefaults>>::create(context);
751 
752     ASSERT_NE(nullptr, image1D);
753     class MockEventWithSetCompleteOnUpdate : public Event {
754       public:
755         MockEventWithSetCompleteOnUpdate(CommandQueue *cmdQueue, cl_command_type cmdType,
756                                          uint32_t taskLevel, uint32_t taskCount) : Event(cmdQueue, cmdType, taskLevel, taskCount) {
757         }
758         void updateExecutionStatus() override {
759             setStatus(CL_COMPLETE);
760         }
761     };
762     MockEventWithSetCompleteOnUpdate blockingEvent(pCmdQ, CL_COMMAND_NDRANGE_KERNEL, 0, 1);
763 
764     cl_event blockingClEvent = &blockingEvent;
765 
766     int32_t initialRefCountCmdQ = pCmdQ->getRefInternalCount();
767 
768     auto ptr = pCmdQ->enqueueMapImage(
769         image1D,
770         true,
771         mapFlags,
772         origin,
773         region,
774         &imageRowPitch,
775         &imageSlicePitch,
776         1,
777         &blockingClEvent,
778         nullptr,
779         retVal);
780     EXPECT_NE(nullptr, ptr);
781     EXPECT_EQ(CL_SUCCESS, retVal);
782 
783     EXPECT_EQ(initialRefCountCmdQ, pCmdQ->getRefInternalCount());
784     delete image1D;
785 }
786 
TEST_F(EnqueueMapImageTest,givenBlockedCommandQueueWhenBlockingCpuMapIsCalledThenReturnRowPitchAndSlicePitch)787 TEST_F(EnqueueMapImageTest, givenBlockedCommandQueueWhenBlockingCpuMapIsCalledThenReturnRowPitchAndSlicePitch) {
788     const size_t origin[3] = {0, 0, 0};
789     const size_t region[3] = {1, 1, 1};
790     size_t retImageRowPitch = 0;
791     size_t retImageSlicePitch = 0;
792 
793     struct MyMockUserEvent : public UserEvent {
794         MyMockUserEvent() : UserEvent(nullptr) {}
795         void updateExecutionStatus() override {
796             setStatus(CL_COMPLETE);
797         }
798     };
799 
800     std::unique_ptr<Image> image(ImageHelper<Image1dArrayDefaults>::create(context));
801     EXPECT_TRUE(image->mappingOnCpuAllowed());
802 
803     MyMockUserEvent blockingEvent;
804     cl_event blockingClEvent = &blockingEvent;
805 
806     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
807                            &retImageRowPitch, &retImageSlicePitch,
808                            1, &blockingClEvent, nullptr, retVal);
809     EXPECT_EQ(CL_SUCCESS, retVal);
810 
811     EXPECT_NE(0u, retImageRowPitch);
812     EXPECT_NE(0u, retImageSlicePitch);
813 
814     image.reset(ImageHelper<Image1dDefaults>::create(context));
815     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
816                            &retImageRowPitch, &retImageSlicePitch,
817                            1, &blockingClEvent, nullptr, retVal);
818     EXPECT_EQ(CL_SUCCESS, retVal);
819 
820     EXPECT_NE(0u, retImageRowPitch);
821     EXPECT_EQ(0u, retImageSlicePitch);
822 }
823 
TEST_F(EnqueueMapImageTest,givenZeroCopyImageWhenMappedOnCpuThenReturnImageRowAndSlicePitch)824 TEST_F(EnqueueMapImageTest, givenZeroCopyImageWhenMappedOnCpuThenReturnImageRowAndSlicePitch) {
825     const size_t origin[3] = {0, 0, 0};
826     const size_t region[3] = {1, 1, 1};
827     size_t retImageRowPitch = 0;
828     size_t retImageSlicePitch = 0;
829 
830     std::unique_ptr<Image> image(ImageHelper<Image1dArrayDefaults>::create(context));
831     EXPECT_TRUE(image->mappingOnCpuAllowed());
832     EXPECT_TRUE(image->isMemObjZeroCopy());
833 
834     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
835                            &retImageRowPitch, &retImageSlicePitch,
836                            0, nullptr, nullptr, retVal);
837     EXPECT_EQ(CL_SUCCESS, retVal);
838 
839     EXPECT_EQ(image->getImageDesc().image_row_pitch, retImageRowPitch);
840     EXPECT_EQ(image->getImageDesc().image_slice_pitch, retImageSlicePitch);
841 }
842 
TEST_F(EnqueueMapImageTest,givenNonZeroCopyImageWhenMappedOnCpuThenReturnHostRowAndSlicePitch)843 TEST_F(EnqueueMapImageTest, givenNonZeroCopyImageWhenMappedOnCpuThenReturnHostRowAndSlicePitch) {
844     const size_t origin[3] = {0, 0, 0};
845     const size_t region[3] = {1, 1, 1};
846     size_t retImageRowPitch = 0;
847     size_t retImageSlicePitch = 0;
848 
849     std::unique_ptr<Image> image(ImageHelper<ImageUseHostPtr<Image1dArrayDefaults>>::create(context));
850     EXPECT_TRUE(image->mappingOnCpuAllowed());
851     EXPECT_FALSE(image->isMemObjZeroCopy());
852 
853     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
854                            &retImageRowPitch, &retImageSlicePitch,
855                            0, nullptr, nullptr, retVal);
856     EXPECT_EQ(CL_SUCCESS, retVal);
857 
858     EXPECT_EQ(image->getHostPtrRowPitch(), retImageRowPitch);
859     EXPECT_EQ(image->getHostPtrSlicePitch(), retImageSlicePitch);
860 }
861 
TEST_F(EnqueueMapImageTest,givenZeroCopyImageWhenMappedOnGpuThenReturnHostRowAndSlicePitch)862 TEST_F(EnqueueMapImageTest, givenZeroCopyImageWhenMappedOnGpuThenReturnHostRowAndSlicePitch) {
863     const size_t origin[3] = {0, 0, 0};
864     const size_t region[3] = {1, 1, 1};
865     size_t retImageRowPitch = 0;
866     size_t retImageSlicePitch = 0;
867 
868     std::unique_ptr<Image> image(ImageHelper<Image1dArrayDefaults>::create(context));
869     image->setSharingHandler(new SharingHandler());
870     EXPECT_FALSE(image->mappingOnCpuAllowed());
871     EXPECT_TRUE(image->isMemObjZeroCopy());
872 
873     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
874                            &retImageRowPitch, &retImageSlicePitch,
875                            0, nullptr, nullptr, retVal);
876     EXPECT_EQ(CL_SUCCESS, retVal);
877 
878     EXPECT_EQ(image->getHostPtrRowPitch(), retImageRowPitch);
879     EXPECT_EQ(image->getHostPtrSlicePitch(), retImageSlicePitch);
880 }
881 
TEST_F(EnqueueMapImageTest,givenNonZeroCopyImageWhenMappedOnGpuThenReturnHostRowAndSlicePitch)882 TEST_F(EnqueueMapImageTest, givenNonZeroCopyImageWhenMappedOnGpuThenReturnHostRowAndSlicePitch) {
883     const size_t origin[3] = {0, 0, 0};
884     const size_t region[3] = {1, 1, 1};
885     size_t retImageRowPitch = 0;
886     size_t retImageSlicePitch = 0;
887 
888     std::unique_ptr<Image> image(ImageHelper<ImageUseHostPtr<Image1dArrayDefaults>>::create(context));
889     image->setSharingHandler(new SharingHandler());
890     EXPECT_FALSE(image->mappingOnCpuAllowed());
891     EXPECT_FALSE(image->isMemObjZeroCopy());
892 
893     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
894                            &retImageRowPitch, &retImageSlicePitch,
895                            0, nullptr, nullptr, retVal);
896     EXPECT_EQ(CL_SUCCESS, retVal);
897 
898     EXPECT_EQ(image->getHostPtrRowPitch(), retImageRowPitch);
899     EXPECT_EQ(image->getHostPtrSlicePitch(), retImageSlicePitch);
900 }
901 
TEST_F(EnqueueMapImageTest,givenMipMapImageWhenMappedThenReturnHostRowAndSlicePitch)902 TEST_F(EnqueueMapImageTest, givenMipMapImageWhenMappedThenReturnHostRowAndSlicePitch) {
903     const size_t origin[4] = {0, 0, 0, 1};
904     const size_t region[3] = {1, 1, 1};
905     size_t retImageRowPitch = 0;
906     size_t retImageSlicePitch = 0;
907 
908     cl_image_desc imageDesc = {};
909     imageDesc.image_type = CL_MEM_OBJECT_IMAGE3D;
910     imageDesc.num_mip_levels = 10;
911     imageDesc.image_width = 4;
912     imageDesc.image_height = 4;
913     imageDesc.image_depth = 4;
914 
915     std::unique_ptr<Image> image(ImageHelper<Image3dDefaults>::create(context, &imageDesc));
916     image->setSharingHandler(new SharingHandler());
917     EXPECT_FALSE(image->mappingOnCpuAllowed());
918 
919     pCmdQ->enqueueMapImage(image.get(), true, CL_MAP_READ, origin, region,
920                            &retImageRowPitch, &retImageSlicePitch,
921                            0, nullptr, nullptr, retVal);
922     EXPECT_EQ(CL_SUCCESS, retVal);
923 
924     EXPECT_EQ(image->getHostPtrRowPitch(), retImageRowPitch);
925     EXPECT_EQ(image->getHostPtrSlicePitch(), retImageSlicePitch);
926 }
927 
TEST_F(EnqueueMapImageTest,givenImage1DArrayWhenEnqueueMapImageIsCalledThenReturnRowAndSlicePitchAreEqual)928 TEST_F(EnqueueMapImageTest, givenImage1DArrayWhenEnqueueMapImageIsCalledThenReturnRowAndSlicePitchAreEqual) {
929     class MockImage : public Image {
930       public:
931         MockImage(Context *context, cl_mem_flags flags, GraphicsAllocation *allocation, const ClSurfaceFormatInfo &surfaceFormat,
932                   const cl_image_format &imageFormat, const cl_image_desc &imageDesc)
933             : Image(context, ClMemoryPropertiesHelper::createMemoryProperties(flags, 0, 0, &context->getDevice(0)->getDevice()), flags, 0,
934                     0, nullptr, nullptr,
935                     imageFormat, imageDesc,
936                     true,
937                     GraphicsAllocationHelper::toMultiGraphicsAllocation(allocation),
938                     false, 0, 0,
939                     surfaceFormat, nullptr) {
940         }
941 
942         void setImageArg(void *memory, bool isMediaBlockImage, uint32_t mipLevel, uint32_t rootDeviceIndex, bool useGlobalAtomics) override {}
943         void setMediaImageArg(void *memory, uint32_t rootDeviceIndex) override {}
944         void setMediaSurfaceRotation(void *memory) override {}
945         void setSurfaceMemoryObjectControlState(void *memory, uint32_t value) override {}
946         void transformImage2dArrayTo3d(void *memory) override {}
947         void transformImage3dTo2dArray(void *memory) override {}
948     };
949 
950     const size_t origin[3] = {0, 0, 0};
951     const size_t region[3] = {1, 1, 1};
952     size_t retImageRowPitch = 0;
953     size_t retImageSlicePitch = 0;
954 
955     cl_mem_flags flags = CL_MEM_READ_ONLY;
956 
957     cl_image_desc imageDesc = {};
958     imageDesc.image_type = CL_MEM_OBJECT_IMAGE1D_ARRAY;
959     imageDesc.image_width = 329;
960     imageDesc.image_array_size = 48;
961     imageDesc.image_row_pitch = 2688;
962     imageDesc.image_slice_pitch = 10752;
963     imageDesc.num_mip_levels = 0;
964 
965     size_t imgSize = imageDesc.image_slice_pitch * imageDesc.image_array_size;
966 
967     cl_image_format imageFormat = {};
968     imageFormat.image_channel_order = CL_RGBA;
969     imageFormat.image_channel_data_type = CL_UNSIGNED_INT16;
970 
971     const ClSurfaceFormatInfo *surfaceFormat = Image::getSurfaceFormatFromTable(flags, &imageFormat, context->getDevice(0)->getHardwareInfo().capabilityTable.supportsOcl21Features);
972     auto allocation = context->getMemoryManager()->allocateGraphicsMemoryWithProperties(MockAllocationProperties{context->getDevice(0)->getRootDeviceIndex(), imgSize});
973     ASSERT_NE(allocation, nullptr);
974 
975     MockImage image(context, flags, allocation, *surfaceFormat, imageFormat, imageDesc);
976 
977     EXPECT_TRUE(image.mappingOnCpuAllowed());
978     EXPECT_TRUE(image.isMemObjZeroCopy());
979 
980     pCmdQ->enqueueMapImage(&image, true, CL_MAP_READ, origin, region,
981                            &retImageRowPitch, &retImageSlicePitch,
982                            0, nullptr, nullptr, retVal);
983     EXPECT_EQ(CL_SUCCESS, retVal);
984 
985     EXPECT_EQ(retImageRowPitch, retImageSlicePitch);
986 }
987 
988 struct EnqueueMapImageTypeTest : public CommandEnqueueFixture,
989                                  public ::testing::Test {
990 
991     typedef CommandQueueHwFixture CommandQueueFixture;
992     using CommandQueueHwFixture::pCmdQ;
993 
EnqueueMapImageTypeTestEnqueueMapImageTypeTest994     EnqueueMapImageTypeTest(void) {
995     }
996 
SetUpEnqueueMapImageTypeTest997     void SetUp() override {
998         CommandEnqueueFixture::SetUp();
999         image = ImageHelper<ImageUseHostPtr<Image2dDefaults>>::create(&context);
1000     }
1001 
TearDownEnqueueMapImageTypeTest1002     void TearDown() override {
1003         delete image;
1004         CommandEnqueueFixture::TearDown();
1005     }
1006 
1007   protected:
1008     template <typename FamilyType>
enqueueMapImageEnqueueMapImageTypeTest1009     void enqueueMapImage(cl_bool blocking = CL_TRUE) {
1010         typedef ImageUseHostPtr<Image2dDefaults> Traits;
1011 
1012         size_t imageRowPitch;
1013         size_t imageSlicePitch;
1014         size_t origin[3] = {0, 0, 0};
1015         size_t region[3] = {Traits::imageDesc.image_width, Traits::imageDesc.image_height, Traits::imageDesc.image_depth};
1016 
1017         cl_int retVal = 0;
1018         auto mappedPtr = pCmdQ->enqueueMapImage(
1019             image,
1020             blocking,
1021             Traits::flags,
1022             origin,
1023             region,
1024             &imageRowPitch,
1025             &imageSlicePitch,
1026             0,
1027             nullptr,
1028             nullptr,
1029             retVal);
1030         EXPECT_NE(nullptr, mappedPtr);
1031         EXPECT_EQ(CL_SUCCESS, retVal);
1032 
1033         parseCommands<FamilyType>(*pCmdQ);
1034     }
1035     MockContext context;
1036     Image *image = nullptr;
1037 };
1038 
HWCMDTEST_F(IGFX_GEN8_CORE,EnqueueMapImageTypeTest,GiveRequirementForPipeControlWorkaroundWhenMappingImageThenAdditionalPipeControlIsProgrammed)1039 HWCMDTEST_F(IGFX_GEN8_CORE, EnqueueMapImageTypeTest, GiveRequirementForPipeControlWorkaroundWhenMappingImageThenAdditionalPipeControlIsProgrammed) {
1040     typedef typename FamilyType::PIPE_CONTROL PIPE_CONTROL;
1041 
1042     // Set taskCount to 1 to call finish on map operation
1043     pCmdQ->taskCount = 1;
1044     bool blocking = true;
1045     enqueueMapImage<FamilyType>(blocking);
1046 
1047     auto itorWalker = find<typename FamilyType::GPGPU_WALKER *>(cmdList.begin(), cmdList.end());
1048 
1049     auto itorCmd = find<PIPE_CONTROL *>(itorWalker, cmdList.end());
1050     auto *cmd = (PIPE_CONTROL *)*itorCmd;
1051     EXPECT_NE(cmdList.end(), itorCmd);
1052 
1053     if (UnitTestHelper<FamilyType>::isPipeControlWArequired(pDevice->getHardwareInfo())) {
1054         // SKL: two PIPE_CONTROLs following GPGPU_WALKER: first has DcFlush and second has Write HwTag
1055         EXPECT_FALSE(cmd->getDcFlushEnable());
1056         // Move to next PPC
1057         auto itorCmdP = ++((GenCmdList::iterator)itorCmd);
1058         EXPECT_NE(cmdList.end(), itorCmdP);
1059         auto itorCmd2 = find<PIPE_CONTROL *>(itorCmdP, cmdList.end());
1060         cmd = (PIPE_CONTROL *)*itorCmd2;
1061         EXPECT_TRUE(cmd->getDcFlushEnable());
1062     } else {
1063         // single PIPE_CONTROL following GPGPU_WALKER has DcFlush and Write HwTag
1064         EXPECT_TRUE(cmd->getDcFlushEnable());
1065     }
1066 }
1067