Skip to content

Commit

Permalink
Skip testing mipmaps if extension is not available (#2085)
Browse files Browse the repository at this point in the history
'./test_image_streams test_mipmaps CL_FILTER_NEAREST' already skips many
tests, but doesn't do so for cl_ext_image_raw10_raw12, causing it to
fail.

Add a check similar to how it is done in the other tests already.

On a device which doesn't support mipmap'd access:
W/o fix
```
./test_image_streams test_mipmaps CL_FILTER_NEAREST
...
1Darray passed
2Darray...
-----------------------------------------------------
This device does not support cl_khr_mipmap_image.
Skipping mipmapped image test. 
-----------------------------------------------------

-----------------------------------------------------
This device does not support cl_khr_mipmap_image.
Skipping mipmapped image test. 
-----------------------------------------------------

2Darray passed
cl_image_requirements_size_ext_negative...
...
image_from_buffer_read_positive passed
cl_ext_image_raw10_raw12...
---- Supported 2D read formats for this device for cl_ext_image_raw10_raw12---- 
  CL_R    CL_UNSIGNED_INT_RAW10_EXT 0
  CL_R    CL_UNSIGNED_INT_RAW12_EXT 0
------------------------------------------- 
read_image (normalized float coords, uint results) *****************************
[CL_R    CL_UNSIGNED_INT_RAW10_EXT 1] - CL_FILTER_NEAREST - CL_ADDRESS_CLAMP_TO_EDGE - NORMALIZED
ERROR: clBuildProgram failed! (CL_BUILD_PROGRAM_FAILURE from <>/OpenCL-CTS/test_common/harness/kernelHelpers.cpp:844)
Build options:  -cl-std=CL3.0
Original source is: ------------
#pragma OPENCL EXTENSION cl_khr_mipmap_image: enable
__kernel void sample_kernel( read_only image2d_t input, sampler_t imageSampler, __global float *xOffsets, __global float *yOffsets, __global uint4 *results , float lod)
{
   int tidX = get_global_id(0), tidY = get_global_id(1);
   unsigned int lod_int = (unsigned int) lod;
   int width_lod = (get_image_width(input) >> lod_int) ?(get_image_width(input) >> lod_int):1 ;
   int offset = tidY*width_lod + tidX;
   float2 coords = (float2)( (float)( xOffsets[offset] ), (float)( yOffsets[offset] ) );
   results[offset] = read_imageui( input, imageSampler, coords , lod);
}Build not successful for device <>, status: CL_BUILD_ERROR
Build log for device <> is: ------------
<source>:1:26: warning: unsupported OpenCL extension 'cl_khr_mipmap_image' - ignoring
#pragma OPENCL EXTENSION cl_khr_mipmap_image: enable
                         ^

<source>:9:22: error: no matching function for call to 'read_imageui'
   results[offset] = read_imageui( input, imageSampler, coords , lod);
                     ^~~~~~~~~~~~

error: Compiler frontend failed (error code <unknown>)
```


With fix:
```
./test_image_streams test_mipmaps CL_FILTER_NEAREST
....
image_from_buffer_read_positive...
image_from_buffer_read_positive passed
cl_ext_image_raw10_raw12...
---- Supported 2D read formats for this device for cl_ext_image_raw10_raw12---- 
  CL_R    CL_UNSIGNED_INT_RAW10_EXT 0
  CL_R    CL_UNSIGNED_INT_RAW12_EXT 0
------------------------------------------- 
-----------------------------------------------------
This device does not support cl_khr_mipmap_image.
Skipping mipmapped image test. 
-----------------------------------------------------

cl_ext_image_raw10_raw12 passed
PASSED sub-test.
PASSED 18 of 18 tests.
```
  • Loading branch information
JorWag authored Sep 24, 2024
1 parent 661a7b0 commit 1cede6d
Showing 1 changed file with 12 additions and 0 deletions.
12 changes: 12 additions & 0 deletions test_conformance/images/kernel_read_write/test_loops.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -175,6 +175,18 @@ int test_read_image_formats(cl_device_id device, cl_context context,
bool flipFlop[2] = { false, true };
int normalizedIdx, floatCoordIdx;

if (gTestMipmaps)
{
if (0 == is_extension_available(device, "cl_khr_mipmap_image"))
{
log_info("-----------------------------------------------------\n");
log_info("This device does not support "
"cl_khr_mipmap_image.\nSkipping mipmapped image test. \n");
log_info(
"-----------------------------------------------------\n\n");
return 0;
}
}

// Use this run if we were told to only run a certain filter mode
if (gFilterModeToUse != (cl_filter_mode)-1
Expand Down

0 comments on commit 1cede6d

Please sign in to comment.