diff --git a/test_conformance/gl/common.h b/test_conformance/gl/common.h index d8587cf0..155deaeb 100644 --- a/test_conformance/gl/common.h +++ b/test_conformance/gl/common.h @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -18,17 +18,19 @@ #include "testBase.h" -typedef struct { - size_t width; - size_t height; - size_t depth; +typedef struct +{ + size_t width; + size_t height; + size_t depth; } sizevec_t; -struct format { - GLenum internal; - GLenum formattype; - GLenum datatype; - ExplicitType type; +struct format +{ + GLenum internal; + GLenum formattype; + GLenum datatype; + ExplicitType type; }; // These are the typically tested formats. @@ -78,6 +80,6 @@ int test_images_get_info_common(cl_device_id device, cl_context context, size_t ntargets, sizevec_t *sizes, size_t nsizes); -int is_rgb_101010_supported( cl_context context, GLenum gl_target ); +int is_rgb_101010_supported(cl_context context, GLenum gl_target); #endif // __COMMON_H__ diff --git a/test_conformance/gl/helpers.cpp b/test_conformance/gl/helpers.cpp index 16441a47..0ec75748 100644 --- a/test_conformance/gl/helpers.cpp +++ b/test_conformance/gl/helpers.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -14,113 +14,96 @@ // limitations under the License. // #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include +#include #endif -const char *get_kernel_suffix( cl_image_format *format ) +const char *get_kernel_suffix(cl_image_format *format) { - switch( format->image_channel_data_type ) + switch (format->image_channel_data_type) { case CL_UNORM_INT8: case CL_UNORM_INT16: - case CL_UNORM_INT24: + case CL_UNORM_INT24: case CL_SNORM_INT8: case CL_SNORM_INT16: case CL_HALF_FLOAT: case CL_FLOAT: - case CL_UNORM_INT_101010: - return "f"; + case CL_UNORM_INT_101010: return "f"; case CL_SIGNED_INT8: case CL_SIGNED_INT16: - case CL_SIGNED_INT32: - return "i"; + case CL_SIGNED_INT32: return "i"; case CL_UNSIGNED_INT8: case CL_UNSIGNED_INT16: - case CL_UNSIGNED_INT32: - return "ui"; + case CL_UNSIGNED_INT32: return "ui"; default: - log_error("Test error: unsupported kernel suffix for image_channel_data_type 0x%X\n",format->image_channel_data_type); + log_error("Test error: unsupported kernel suffix for " + "image_channel_data_type 0x%X\n", + format->image_channel_data_type); return ""; } } -ExplicitType get_read_kernel_type( cl_image_format *format ) +ExplicitType get_read_kernel_type(cl_image_format *format) { - switch( format->image_channel_data_type ) + switch (format->image_channel_data_type) { case CL_UNORM_INT8: case CL_UNORM_INT16: - case CL_UNORM_INT24: + case CL_UNORM_INT24: case CL_SNORM_INT8: case CL_SNORM_INT16: case CL_HALF_FLOAT: case CL_FLOAT: case CL_UNORM_INT_101010: #ifdef GL_VERSION_3_2 - case CL_DEPTH: + case CL_DEPTH: #endif return kFloat; case CL_SIGNED_INT8: case CL_SIGNED_INT16: - case CL_SIGNED_INT32: - return kInt; + case CL_SIGNED_INT32: return kInt; case CL_UNSIGNED_INT8: case CL_UNSIGNED_INT16: - case CL_UNSIGNED_INT32: - return kUInt; + case CL_UNSIGNED_INT32: return kUInt; default: - log_error("Test error: unsupported kernel suffix for image_channel_data_type 0x%X\n",format->image_channel_data_type); + log_error("Test error: unsupported kernel suffix for " + "image_channel_data_type 0x%X\n", + format->image_channel_data_type); return kNumExplicitTypes; } } -ExplicitType get_write_kernel_type( cl_image_format *format ) +ExplicitType get_write_kernel_type(cl_image_format *format) { - switch( format->image_channel_data_type ) + switch (format->image_channel_data_type) { - case CL_UNORM_INT8: - return kFloat; - case CL_UNORM_INT16: - return kFloat; - case CL_UNORM_INT24: - return kFloat; - case CL_SNORM_INT8: - return kFloat; - case CL_SNORM_INT16: - return kFloat; - case CL_HALF_FLOAT: - return kHalf; - case CL_FLOAT: - return kFloat; - case CL_SIGNED_INT8: - return kChar; - case CL_SIGNED_INT16: - return kShort; - case CL_SIGNED_INT32: - return kInt; - case CL_UNSIGNED_INT8: - return kUChar; - case CL_UNSIGNED_INT16: - return kUShort; - case CL_UNSIGNED_INT32: - return kUInt; - case CL_UNORM_INT_101010: - return kFloat; + case CL_UNORM_INT8: return kFloat; + case CL_UNORM_INT16: return kFloat; + case CL_UNORM_INT24: return kFloat; + case CL_SNORM_INT8: return kFloat; + case CL_SNORM_INT16: return kFloat; + case CL_HALF_FLOAT: return kHalf; + case CL_FLOAT: return kFloat; + case CL_SIGNED_INT8: return kChar; + case CL_SIGNED_INT16: return kShort; + case CL_SIGNED_INT32: return kInt; + case CL_UNSIGNED_INT8: return kUChar; + case CL_UNSIGNED_INT16: return kUShort; + case CL_UNSIGNED_INT32: return kUInt; + case CL_UNORM_INT_101010: return kFloat; #ifdef GL_VERSION_3_2 - case CL_DEPTH: - return kFloat; + case CL_DEPTH: return kFloat; #endif - default: - return kInt; + default: return kInt; } } -const char* get_write_conversion( cl_image_format *format, ExplicitType type ) +const char *get_write_conversion(cl_image_format *format, ExplicitType type) { - switch( format->image_channel_data_type ) + switch (format->image_channel_data_type) { case CL_UNORM_INT8: case CL_UNORM_INT16: @@ -130,250 +113,268 @@ const char* get_write_conversion( cl_image_format *format, ExplicitType type ) case CL_FLOAT: case CL_UNORM_INT_101010: case CL_UNORM_INT24: - if(type != kFloat) return "convert_float4"; + if (type != kFloat) return "convert_float4"; break; case CL_SIGNED_INT8: case CL_SIGNED_INT16: case CL_SIGNED_INT32: - if(type != kInt) return "convert_int4"; + if (type != kInt) return "convert_int4"; break; case CL_UNSIGNED_INT8: case CL_UNSIGNED_INT16: case CL_UNSIGNED_INT32: - if(type != kUInt) return "convert_uint4"; + if (type != kUInt) return "convert_uint4"; break; - default: - return ""; + default: return ""; } return ""; } -// The only three input types to this function are kInt, kUInt and kFloat, due to the way we set up our tests -// The output types, though, are pretty much anything valid for GL to receive +// The only three input types to this function are kInt, kUInt and kFloat, due +// to the way we set up our tests The output types, though, are pretty much +// anything valid for GL to receive -#define DOWNSCALE_INTEGER_CASE( enum, type, bitShift ) \ - case enum: \ - { \ - cl_##type *dst = new cl_##type[ numPixels * 4 ]; \ - for( size_t i = 0; i < numPixels * 4; i++ ) \ - dst[ i ] = src[ i ]; \ - return (char *)dst; \ +#define DOWNSCALE_INTEGER_CASE(enum, type, bitShift) \ + case enum: { \ + cl_##type *dst = new cl_##type[numPixels * 4]; \ + for (size_t i = 0; i < numPixels * 4; i++) dst[i] = src[i]; \ + return (char *)dst; \ } -#define UPSCALE_FLOAT_CASE( enum, type, typeMax ) \ - case enum: \ - { \ - cl_##type *dst = new cl_##type[ numPixels * 4 ]; \ - for( size_t i = 0; i < numPixels * 4; i++ ) \ - dst[ i ] = (cl_##type)( src[ i ] * typeMax ); \ - return (char *)dst; \ +#define UPSCALE_FLOAT_CASE(enum, type, typeMax) \ + case enum: { \ + cl_##type *dst = new cl_##type[numPixels * 4]; \ + for (size_t i = 0; i < numPixels * 4; i++) \ + dst[i] = (cl_##type)(src[i] * typeMax); \ + return (char *)dst; \ } -char * convert_to_expected( void * inputBuffer, size_t numPixels, ExplicitType inType, ExplicitType outType, size_t channelNum, GLenum glDataType ) +char *convert_to_expected(void *inputBuffer, size_t numPixels, + ExplicitType inType, ExplicitType outType, + size_t channelNum, GLenum glDataType) { #ifdef DEBUG - log_info( "- Converting from input type '%s' to output type '%s'\n", - get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_info("- Converting from input type '%s' to output type '%s'\n", + get_explicit_type_name(inType), get_explicit_type_name(outType)); #endif - if( inType == outType ) + if (inType == outType) { - char *outData = new char[ numPixels * channelNum * get_explicit_type_size(outType) ] ; // sizeof( cl_int ) ]; - if (glDataType == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) { - for (size_t i = 0; i < numPixels; ++i) { - ((cl_float*)outData)[i] = ((cl_float*)inputBuffer)[2 * i]; + char *outData = + new char[numPixels * channelNum + * get_explicit_type_size(outType)]; // sizeof( cl_int ) ]; + if (glDataType == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) + { + for (size_t i = 0; i < numPixels; ++i) + { + ((cl_float *)outData)[i] = ((cl_float *)inputBuffer)[2 * i]; } } - else { - memcpy( outData, inputBuffer, numPixels * channelNum * get_explicit_type_size(inType) ); + else + { + memcpy(outData, inputBuffer, + numPixels * channelNum * get_explicit_type_size(inType)); } return outData; } - else if( inType == kChar ) + else if (inType == kChar) { cl_char *src = (cl_char *)inputBuffer; - switch( outType ) + switch (outType) { - case kInt: - { - cl_int *outData = new cl_int[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kInt: { + cl_int *outData = new cl_int[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_int)((src[ i ])); + outData[i] = (cl_int)((src[i])); } return (char *)outData; } - case kFloat: - { - // If we're converting to float, then CL decided that we should be normalized - cl_float *outData = new cl_float[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kFloat: { + // If we're converting to float, then CL decided that we should + // be normalized + cl_float *outData = new cl_float[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_float)src[ i ] / 127.0f; + outData[i] = (cl_float)src[i] / 127.0f; } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } - else if( inType == kUChar ) + else if (inType == kUChar) { cl_uchar *src = (cl_uchar *)inputBuffer; - switch( outType ) + switch (outType) { - case kUInt: - { - cl_uint *outData = new cl_uint[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kUInt: { + cl_uint *outData = new cl_uint[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_uint)((src[ i ])); + outData[i] = (cl_uint)((src[i])); } return (char *)outData; } - case kFloat: - { - // If we're converting to float, then CL decided that we should be normalized - cl_float *outData = new cl_float[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kFloat: { + // If we're converting to float, then CL decided that we should + // be normalized + cl_float *outData = new cl_float[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_float)(src[ i ]) / 256.0f; + outData[i] = (cl_float)(src[i]) / 256.0f; } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } - else if( inType == kShort ) + else if (inType == kShort) { cl_short *src = (cl_short *)inputBuffer; - switch( outType ) + switch (outType) { - case kInt: - { - cl_int *outData = new cl_int[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kInt: { + cl_int *outData = new cl_int[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_int)((src[ i ])); + outData[i] = (cl_int)((src[i])); } return (char *)outData; } - case kFloat: - { - // If we're converting to float, then CL decided that we should be normalized - cl_float *outData = new cl_float[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kFloat: { + // If we're converting to float, then CL decided that we should + // be normalized + cl_float *outData = new cl_float[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_float)src[ i ] / 32768.0f; + outData[i] = (cl_float)src[i] / 32768.0f; } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } - else if( inType == kUShort ) + else if (inType == kUShort) { cl_ushort *src = (cl_ushort *)inputBuffer; - switch( outType ) + switch (outType) { - case kUInt: - { - cl_uint *outData = new cl_uint[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kUInt: { + cl_uint *outData = new cl_uint[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_uint)((src[ i ])); + outData[i] = (cl_uint)((src[i])); } return (char *)outData; } - case kFloat: - { - // If we're converting to float, then CL decided that we should be normalized - cl_float *outData = new cl_float[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kFloat: { + // If we're converting to float, then CL decided that we should + // be normalized + cl_float *outData = new cl_float[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_float)(src[ i ]) / 65535.0f; + outData[i] = (cl_float)(src[i]) / 65535.0f; } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } - else if( inType == kInt ) + else if (inType == kInt) { cl_int *src = (cl_int *)inputBuffer; - switch( outType ) + switch (outType) { - DOWNSCALE_INTEGER_CASE( kShort, short, 16 ) - DOWNSCALE_INTEGER_CASE( kChar, char, 24 ) - case kFloat: - { - // If we're converting to float, then CL decided that we should be normalized - cl_float *outData = new cl_float[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + DOWNSCALE_INTEGER_CASE(kShort, short, 16) + DOWNSCALE_INTEGER_CASE(kChar, char, 24) + case kFloat: { + // If we're converting to float, then CL decided that we should + // be normalized + cl_float *outData = new cl_float[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_float)fmaxf( (float)src[ i ] / 2147483647.f, -1.f ); + outData[i] = + (cl_float)fmaxf((float)src[i] / 2147483647.f, -1.f); } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } - else if( inType == kUInt ) + else if (inType == kUInt) { cl_uint *src = (cl_uint *)inputBuffer; - switch( outType ) + switch (outType) { - DOWNSCALE_INTEGER_CASE( kUShort, ushort, 16 ) - DOWNSCALE_INTEGER_CASE( kUChar, uchar, 24 ) - case kFloat: - { - // If we're converting to float, then CL decided that we should be normalized - cl_float *outData = new cl_float[ numPixels * channelNum ]; - const cl_float MaxValue = (glDataType == GL_UNSIGNED_INT_24_8) ? 16777215.f : 4294967295.f; - const cl_uint ShiftBits = (glDataType == GL_UNSIGNED_INT_24_8) ? 8 : 0; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + DOWNSCALE_INTEGER_CASE(kUShort, ushort, 16) + DOWNSCALE_INTEGER_CASE(kUChar, uchar, 24) + case kFloat: { + // If we're converting to float, then CL decided that we should + // be normalized + cl_float *outData = new cl_float[numPixels * channelNum]; + const cl_float MaxValue = (glDataType == GL_UNSIGNED_INT_24_8) + ? 16777215.f + : 4294967295.f; + const cl_uint ShiftBits = + (glDataType == GL_UNSIGNED_INT_24_8) ? 8 : 0; + for (size_t i = 0; i < numPixels * channelNum; i++) { - outData[ i ] = (cl_float)(src[ i ] >> ShiftBits) / MaxValue; + outData[i] = (cl_float)(src[i] >> ShiftBits) / MaxValue; } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } - else if( inType == kHalf ) + else if (inType == kHalf) { cl_half *src = (cl_half *)inputBuffer; - switch( outType ) + switch (outType) { - case kFloat: - { - cl_float *outData = new cl_float[ numPixels * channelNum ]; - for( size_t i = 0; i < numPixels * channelNum; i++ ) + case kFloat: { + cl_float *outData = new cl_float[numPixels * channelNum]; + for (size_t i = 0; i < numPixels * channelNum; i++) { outData[i] = cl_half_to_float(src[i]); } return (char *)outData; } default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } @@ -381,16 +382,18 @@ char * convert_to_expected( void * inputBuffer, size_t numPixels, ExplicitType i { cl_float *src = (cl_float *)inputBuffer; - switch( outType ) + switch (outType) { - UPSCALE_FLOAT_CASE( kChar, char, 127.f ) - UPSCALE_FLOAT_CASE( kUChar, uchar, 255.f ) - UPSCALE_FLOAT_CASE( kShort, short, 32767.f ) - UPSCALE_FLOAT_CASE( kUShort, ushort, 65535.f ) - UPSCALE_FLOAT_CASE( kInt, int, 2147483647.f ) - UPSCALE_FLOAT_CASE( kUInt, uint, 4294967295.f ) + UPSCALE_FLOAT_CASE(kChar, char, 127.f) + UPSCALE_FLOAT_CASE(kUChar, uchar, 255.f) + UPSCALE_FLOAT_CASE(kShort, short, 32767.f) + UPSCALE_FLOAT_CASE(kUShort, ushort, 65535.f) + UPSCALE_FLOAT_CASE(kInt, int, 2147483647.f) + UPSCALE_FLOAT_CASE(kUInt, uint, 4294967295.f) default: - log_error( "ERROR: Unsupported conversion from %s to %s!\n", get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); + log_error("ERROR: Unsupported conversion from %s to %s!\n", + get_explicit_type_name(inType), + get_explicit_type_name(outType)); return NULL; } } @@ -398,195 +401,256 @@ char * convert_to_expected( void * inputBuffer, size_t numPixels, ExplicitType i return NULL; } -int validate_integer_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum, size_t typeSize ) +int validate_integer_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t sampleNum, + size_t typeSize) { - return validate_integer_results( expectedResults, actualResults, width, height, sampleNum, 0, typeSize ); + return validate_integer_results(expectedResults, actualResults, width, + height, sampleNum, 0, typeSize); } -int validate_integer_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum, size_t typeSize ) +int validate_integer_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t depth, + size_t sampleNum, size_t typeSize) { char *expected = (char *)expectedResults; char *actual = (char *)actualResults; - for ( size_t s = 0; s < sampleNum; s++ ) - { - for( size_t z = 0; z < ( ( depth == 0 ) ? 1 : depth ); z++ ) + for (size_t s = 0; s < sampleNum; s++) { - for( size_t y = 0; y < height; y++ ) + for (size_t z = 0; z < ((depth == 0) ? 1 : depth); z++) { - for( size_t x = 0; x < width; x++ ) + for (size_t y = 0; y < height; y++) { - if( memcmp( expected, actual, typeSize * 4 ) != 0 ) + for (size_t x = 0; x < width; x++) { - char scratch[ 1024 ]; + if (memcmp(expected, actual, typeSize * 4) != 0) + { + char scratch[1024]; - if( depth == 0 ) - log_error( "ERROR: Data sample %d,%d,%d did not validate!\n", (int)x, (int)y, (int)s ); - else - log_error( "ERROR: Data sample %d,%d,%d,%d did not validate!\n", (int)x, (int)y, (int)z, (int)s ); - log_error( "\tExpected: %s\n", GetDataVectorString( expected, typeSize, 4, scratch ) ); - log_error( "\t Actual: %s\n", GetDataVectorString( actual, typeSize, 4, scratch ) ); - return -1; + if (depth == 0) + log_error("ERROR: Data sample %d,%d,%d did not " + "validate!\n", + (int)x, (int)y, (int)s); + else + log_error("ERROR: Data sample %d,%d,%d,%d did not " + "validate!\n", + (int)x, (int)y, (int)z, (int)s); + log_error("\tExpected: %s\n", + GetDataVectorString(expected, typeSize, 4, + scratch)); + log_error( + "\t Actual: %s\n", + GetDataVectorString(actual, typeSize, 4, scratch)); + return -1; + } + expected += typeSize * 4; + actual += typeSize * 4; } - expected += typeSize * 4; - actual += typeSize * 4; } } } - } return 0; } -int validate_float_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum, size_t channelNum ) +int validate_float_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t sampleNum, + size_t channelNum) { - return validate_float_results( expectedResults, actualResults, width, height, sampleNum, 0, channelNum ); + return validate_float_results(expectedResults, actualResults, width, height, + sampleNum, 0, channelNum); } -int validate_float_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum, size_t channelNum ) +int validate_float_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t depth, + size_t sampleNum, size_t channelNum) { cl_float *expected = (cl_float *)expectedResults; cl_float *actual = (cl_float *)actualResults; - for ( size_t s = 0; s < sampleNum; s++ ) - { - for( size_t z = 0; z < ( ( depth == 0 ) ? 1 : depth ); z++ ) + for (size_t s = 0; s < sampleNum; s++) { - for( size_t y = 0; y < height; y++ ) + for (size_t z = 0; z < ((depth == 0) ? 1 : depth); z++) { - for( size_t x = 0; x < width; x++ ) + for (size_t y = 0; y < height; y++) { - float err = 0.f; - for( size_t i = 0; i < channelNum; i++ ) + for (size_t x = 0; x < width; x++) { - float error = fabsf( expected[ i ] - actual[ i ] ); - if( error > err ) - err = error; - } + float err = 0.f; + for (size_t i = 0; i < channelNum; i++) + { + float error = fabsf(expected[i] - actual[i]); + if (error > err) err = error; + } - if( err > 1.f / 127.f ) // Max expected range of error if we converted from an 8-bit integer to a normalized float - { - if( depth == 0 ) - log_error( "ERROR: Data sample %d,%d,%d did not validate!\n", (int)x, (int)y, (int)s ); - else - log_error( "ERROR: Data sample %d,%d,%d,%d did not validate!\n", (int)x, (int)y, (int)z, (int)s ); + if (err > 1.f / 127.f) // Max expected range of error if we + // converted from an 8-bit integer to + // a normalized float + { + if (depth == 0) + log_error("ERROR: Data sample %d,%d,%d did not " + "validate!\n", + (int)x, (int)y, (int)s); + else + log_error("ERROR: Data sample %d,%d,%d,%d did not " + "validate!\n", + (int)x, (int)y, (int)z, (int)s); - if (channelNum == 4) - { - log_error( "\tExpected: %f %f %f %f\n", expected[ 0 ], expected[ 1 ], expected[ 2 ], expected[ 3 ] ); - log_error( "\t : %a %a %a %a\n", expected[ 0 ], expected[ 1 ], expected[ 2 ], expected[ 3 ] ); - log_error( "\t Actual: %f %f %f %f\n", actual[ 0 ], actual[ 1 ], actual[ 2 ], actual[ 3 ] ); - log_error( "\t : %a %a %a %a\n", actual[ 0 ], actual[ 1 ], actual[ 2 ], actual[ 3 ] ); - } - else if(channelNum == 1) - { - log_error( "\tExpected: %f\n", expected[ 0 ] ); - log_error( "\t : %a\n", expected[ 0 ] ); - log_error( "\t Actual: %f\n", actual[ 0 ] ); - log_error( "\t : %a\n", actual[ 0 ] ); - } - return -1; + if (channelNum == 4) + { + log_error("\tExpected: %f %f %f %f\n", expected[0], + expected[1], expected[2], expected[3]); + log_error("\t : %a %a %a %a\n", expected[0], + expected[1], expected[2], expected[3]); + log_error("\t Actual: %f %f %f %f\n", actual[0], + actual[1], actual[2], actual[3]); + log_error("\t : %a %a %a %a\n", actual[0], + actual[1], actual[2], actual[3]); + } + else if (channelNum == 1) + { + log_error("\tExpected: %f\n", expected[0]); + log_error("\t : %a\n", expected[0]); + log_error("\t Actual: %f\n", actual[0]); + log_error("\t : %a\n", actual[0]); + } + return -1; + } + expected += channelNum; + actual += channelNum; } - expected += channelNum; - actual += channelNum; } } } - } return 0; } -int validate_float_results_rgb_101010( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum ) +int validate_float_results_rgb_101010(void *expectedResults, + void *actualResults, size_t width, + size_t height, size_t sampleNum) { - return validate_float_results_rgb_101010( expectedResults, actualResults, width, height, sampleNum, 0 ); + return validate_float_results_rgb_101010(expectedResults, actualResults, + width, height, sampleNum, 0); } -int validate_float_results_rgb_101010( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum ) +int validate_float_results_rgb_101010(void *expectedResults, + void *actualResults, size_t width, + size_t height, size_t depth, + size_t sampleNum) { cl_float *expected = (cl_float *)expectedResults; cl_float *actual = (cl_float *)actualResults; - for ( size_t s = 0; s < sampleNum; s++ ) - { - for( size_t z = 0; z < ( ( depth == 0 ) ? 1 : depth ); z++ ) + for (size_t s = 0; s < sampleNum; s++) { - for( size_t y = 0; y < height; y++ ) + for (size_t z = 0; z < ((depth == 0) ? 1 : depth); z++) { - for( size_t x = 0; x < width; x++ ) + for (size_t y = 0; y < height; y++) { - float err = 0.f; - for( size_t i = 0; i < 3; i++ ) // skip the fourth channel + for (size_t x = 0; x < width; x++) { - float error = fabsf( expected[ i ] - actual[ i ] ); - if( error > err ) - err = error; - } + float err = 0.f; + for (size_t i = 0; i < 3; i++) // skip the fourth channel + { + float error = fabsf(expected[i] - actual[i]); + if (error > err) err = error; + } - if( err > 1.f / 127.f ) // Max expected range of error if we converted from an 8-bit integer to a normalized float - { - if( depth == 0 ) - log_error( "ERROR: Data sample %d,%d,%d did not validate!\n", (int)x, (int)y, (int)s ); - else - log_error( "ERROR: Data sample %d,%d,%d,%d did not validate!\n", (int)x, (int)y, (int)z, (int)s ); - log_error( "\tExpected: %f %f %f\n", expected[ 0 ], expected[ 1 ], expected[ 2 ] ); - log_error( "\t : %a %a %a\n", expected[ 0 ], expected[ 1 ], expected[ 2 ] ); - log_error( "\t Actual: %f %f %f\n", actual[ 0 ], actual[ 1 ], actual[ 2 ] ); - log_error( "\t : %a %a %a\n", actual[ 0 ], actual[ 1 ], actual[ 2 ] ); - return -1; + if (err > 1.f / 127.f) // Max expected range of error if we + // converted from an 8-bit integer to + // a normalized float + { + if (depth == 0) + log_error("ERROR: Data sample %d,%d,%d did not " + "validate!\n", + (int)x, (int)y, (int)s); + else + log_error("ERROR: Data sample %d,%d,%d,%d did not " + "validate!\n", + (int)x, (int)y, (int)z, (int)s); + log_error("\tExpected: %f %f %f\n", expected[0], + expected[1], expected[2]); + log_error("\t : %a %a %a\n", expected[0], + expected[1], expected[2]); + log_error("\t Actual: %f %f %f\n", actual[0], + actual[1], actual[2]); + log_error("\t : %a %a %a\n", actual[0], + actual[1], actual[2]); + return -1; + } + expected += 4; + actual += 4; } - expected += 4; - actual += 4; } } } - } return 0; } -int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type, GLuint expected_gl_name, - GLenum expected_cl_gl_texture_target, GLint expected_cl_gl_mipmap_level) +int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type, + GLuint expected_gl_name, + GLenum expected_cl_gl_texture_target, + GLint expected_cl_gl_mipmap_level) { - cl_gl_object_type object_type; - GLuint object_name; - GLenum texture_target; - GLint mipmap_level; + cl_gl_object_type object_type; + GLuint object_name; + GLenum texture_target; + GLint mipmap_level; int error; - error = (*clGetGLObjectInfo_ptr)(mem, &object_type, &object_name); - test_error( error, "clGetGLObjectInfo failed"); - if (object_type != expected_cl_gl_type) { - log_error("clGetGLObjectInfo did not return expected object type: expected %d, got %d.\n", expected_cl_gl_type, object_type); - return -1; - } - if (object_name != expected_gl_name) { - log_error("clGetGLObjectInfo did not return expected object name: expected %d, got %d.\n", expected_gl_name, object_name); - return -1; - } + error = (*clGetGLObjectInfo_ptr)(mem, &object_type, &object_name); + test_error(error, "clGetGLObjectInfo failed"); + if (object_type != expected_cl_gl_type) + { + log_error("clGetGLObjectInfo did not return expected object type: " + "expected %d, got %d.\n", + expected_cl_gl_type, object_type); + return -1; + } + if (object_name != expected_gl_name) + { + log_error("clGetGLObjectInfo did not return expected object name: " + "expected %d, got %d.\n", + expected_gl_name, object_name); + return -1; + } - // If we're dealing with a buffer or render buffer, we are done. + // If we're dealing with a buffer or render buffer, we are done. + + if (object_type == CL_GL_OBJECT_BUFFER + || object_type == CL_GL_OBJECT_RENDERBUFFER) + { + return 0; + } + + // Otherwise, it's a texture-based object and requires a bit more checking. + + error = (*clGetGLTextureInfo_ptr)(mem, CL_GL_TEXTURE_TARGET, + sizeof(texture_target), &texture_target, + NULL); + test_error(error, "clGetGLTextureInfo for CL_GL_TEXTURE_TARGET failed"); + + if (texture_target != expected_cl_gl_texture_target) + { + log_error("clGetGLTextureInfo did not return expected texture target: " + "expected %d, got %d.\n", + expected_cl_gl_texture_target, texture_target); + return -1; + } + + error = (*clGetGLTextureInfo_ptr)( + mem, CL_GL_MIPMAP_LEVEL, sizeof(mipmap_level), &mipmap_level, NULL); + test_error(error, "clGetGLTextureInfo for CL_GL_MIPMAP_LEVEL failed"); + + if (mipmap_level != expected_cl_gl_mipmap_level) + { + log_error("clGetGLTextureInfo did not return expected mipmap level: " + "expected %d, got %d.\n", + expected_cl_gl_mipmap_level, mipmap_level); + return -1; + } - if (object_type == CL_GL_OBJECT_BUFFER || object_type == CL_GL_OBJECT_RENDERBUFFER) { return 0; - } - - // Otherwise, it's a texture-based object and requires a bit more checking. - - error = (*clGetGLTextureInfo_ptr)(mem, CL_GL_TEXTURE_TARGET, sizeof(texture_target), &texture_target, NULL); - test_error( error, "clGetGLTextureInfo for CL_GL_TEXTURE_TARGET failed"); - - if (texture_target != expected_cl_gl_texture_target) { - log_error("clGetGLTextureInfo did not return expected texture target: expected %d, got %d.\n", expected_cl_gl_texture_target, texture_target); - return -1; - } - - error = (*clGetGLTextureInfo_ptr)(mem, CL_GL_MIPMAP_LEVEL, sizeof(mipmap_level), &mipmap_level, NULL); - test_error( error, "clGetGLTextureInfo for CL_GL_MIPMAP_LEVEL failed"); - - if (mipmap_level != expected_cl_gl_mipmap_level) { - log_error("clGetGLTextureInfo did not return expected mipmap level: expected %d, got %d.\n", expected_cl_gl_mipmap_level, mipmap_level); - return -1; - } - - return 0; } bool CheckGLIntegerExtensionSupport() @@ -595,22 +659,25 @@ bool CheckGLIntegerExtensionSupport() const GLubyte *glVersion = glGetString(GL_VERSION); const GLubyte *glExtensionList = glGetString(GL_EXTENSIONS); - // Check if the OpenGL vrsion is 3.0 or grater or GL_EXT_texture_integer is supported - return (((glVersion[0] - '0') >= 3) || (strstr((const char*)glExtensionList, "GL_EXT_texture_integer"))); + // Check if the OpenGL vrsion is 3.0 or grater or GL_EXT_texture_integer is + // supported + return ( + ((glVersion[0] - '0') >= 3) + || (strstr((const char *)glExtensionList, "GL_EXT_texture_integer"))); } -int is_rgb_101010_supported( cl_context context, GLenum gl_target ) +int is_rgb_101010_supported(cl_context context, GLenum gl_target) { - cl_image_format formatList[ 128 ]; + cl_image_format formatList[128]; cl_uint formatCount = 0; unsigned int i; int error; cl_mem_object_type image_type; - switch (get_base_gl_target(gl_target)) { - case GL_TEXTURE_1D: - image_type = CL_MEM_OBJECT_IMAGE1D; + switch (get_base_gl_target(gl_target)) + { + case GL_TEXTURE_1D: image_type = CL_MEM_OBJECT_IMAGE1D; case GL_TEXTURE_BUFFER: image_type = CL_MEM_OBJECT_IMAGE1D_BUFFER; break; @@ -627,27 +694,25 @@ int is_rgb_101010_supported( cl_context context, GLenum gl_target ) case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: image_type = CL_MEM_OBJECT_IMAGE2D; break; - case GL_TEXTURE_3D: - image_type = CL_MEM_OBJECT_IMAGE3D; - case GL_TEXTURE_1D_ARRAY: - image_type = CL_MEM_OBJECT_IMAGE1D_ARRAY; + case GL_TEXTURE_3D: image_type = CL_MEM_OBJECT_IMAGE3D; + case GL_TEXTURE_1D_ARRAY: image_type = CL_MEM_OBJECT_IMAGE1D_ARRAY; case GL_TEXTURE_2D_ARRAY: image_type = CL_MEM_OBJECT_IMAGE2D_ARRAY; break; - default: - image_type = CL_MEM_OBJECT_IMAGE2D; + default: image_type = CL_MEM_OBJECT_IMAGE2D; } - if ((error = clGetSupportedImageFormats(context, CL_MEM_READ_WRITE, - image_type, 128, formatList, - &formatCount ))) { + if ((error = + clGetSupportedImageFormats(context, CL_MEM_READ_WRITE, image_type, + 128, formatList, &formatCount))) + { return error; } // Check if the RGB 101010 format is supported - for( i = 0; i < formatCount; i++ ) + for (i = 0; i < formatCount; i++) { - if( formatList[ i ].image_channel_data_type == CL_UNORM_INT_101010 ) + if (formatList[i].image_channel_data_type == CL_UNORM_INT_101010) { return 1; } diff --git a/test_conformance/gl/main.cpp b/test_conformance/gl/main.cpp index 203e915e..32aafef5 100644 --- a/test_conformance/gl/main.cpp +++ b/test_conformance/gl/main.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -18,7 +18,7 @@ #include #include -#if !defined (__APPLE__) +#if !defined(__APPLE__) #include #endif @@ -31,348 +31,386 @@ #include #endif -static cl_context sCurrentContext = NULL; +static cl_context sCurrentContext = NULL; -#define TEST_FN_REDIRECT( fn ) ADD_TEST( redirect_##fn ) -#define TEST_FN_REDIRECTOR( fn ) \ -int test_redirect_##fn(cl_device_id device, cl_context context, cl_command_queue queue, int numElements ) \ -{ \ - int error; \ - clCommandQueueWrapper realQueue = clCreateCommandQueueWithProperties( sCurrentContext, device, 0, &error ); \ - test_error( error, "Unable to create command queue" ); \ - return test_##fn( device, sCurrentContext, realQueue, numElements ); \ -} +#define TEST_FN_REDIRECT(fn) ADD_TEST(redirect_##fn) +#define TEST_FN_REDIRECTOR(fn) \ + int test_redirect_##fn(cl_device_id device, cl_context context, \ + cl_command_queue queue, int numElements) \ + { \ + int error; \ + clCommandQueueWrapper realQueue = clCreateCommandQueueWithProperties( \ + sCurrentContext, device, 0, &error); \ + test_error(error, "Unable to create command queue"); \ + return test_##fn(device, sCurrentContext, realQueue, numElements); \ + } // buffers: -TEST_FN_REDIRECTOR( buffers ) -TEST_FN_REDIRECTOR( buffers_getinfo ) +TEST_FN_REDIRECTOR(buffers) +TEST_FN_REDIRECTOR(buffers_getinfo) // 1D images: -TEST_FN_REDIRECTOR( images_read_1D ) -TEST_FN_REDIRECTOR( images_write_1D ) -TEST_FN_REDIRECTOR( images_1D_getinfo ) +TEST_FN_REDIRECTOR(images_read_1D) +TEST_FN_REDIRECTOR(images_write_1D) +TEST_FN_REDIRECTOR(images_1D_getinfo) // 1D image arrays: -TEST_FN_REDIRECTOR( images_read_1Darray ) -TEST_FN_REDIRECTOR( images_write_1Darray ) -TEST_FN_REDIRECTOR( images_1Darray_getinfo ) +TEST_FN_REDIRECTOR(images_read_1Darray) +TEST_FN_REDIRECTOR(images_write_1Darray) +TEST_FN_REDIRECTOR(images_1Darray_getinfo) // 2D images: -TEST_FN_REDIRECTOR( images_read_2D ) -TEST_FN_REDIRECTOR( images_read_cube ) -TEST_FN_REDIRECTOR( images_write ) -TEST_FN_REDIRECTOR( images_write_cube ) -TEST_FN_REDIRECTOR( images_2D_getinfo ) -TEST_FN_REDIRECTOR( images_cube_getinfo ) +TEST_FN_REDIRECTOR(images_read_2D) +TEST_FN_REDIRECTOR(images_read_cube) +TEST_FN_REDIRECTOR(images_write) +TEST_FN_REDIRECTOR(images_write_cube) +TEST_FN_REDIRECTOR(images_2D_getinfo) +TEST_FN_REDIRECTOR(images_cube_getinfo) // 2D image arrays: -TEST_FN_REDIRECTOR( images_read_2Darray ) -TEST_FN_REDIRECTOR( images_write_2Darray ) -TEST_FN_REDIRECTOR( images_2Darray_getinfo ) +TEST_FN_REDIRECTOR(images_read_2Darray) +TEST_FN_REDIRECTOR(images_write_2Darray) +TEST_FN_REDIRECTOR(images_2Darray_getinfo) // 3D images: -TEST_FN_REDIRECTOR( images_read_3D ) -TEST_FN_REDIRECTOR( images_write_3D ) -TEST_FN_REDIRECTOR( images_3D_getinfo ) +TEST_FN_REDIRECTOR(images_read_3D) +TEST_FN_REDIRECTOR(images_write_3D) +TEST_FN_REDIRECTOR(images_3D_getinfo) #ifdef GL_VERSION_3_2 -TEST_FN_REDIRECTOR( images_read_texturebuffer ) -TEST_FN_REDIRECTOR( images_write_texturebuffer ) -TEST_FN_REDIRECTOR( images_texturebuffer_getinfo ) +TEST_FN_REDIRECTOR(images_read_texturebuffer) +TEST_FN_REDIRECTOR(images_write_texturebuffer) +TEST_FN_REDIRECTOR(images_texturebuffer_getinfo) // depth textures -TEST_FN_REDIRECTOR( images_read_2D_depth ) -TEST_FN_REDIRECTOR( images_write_2D_depth ) -TEST_FN_REDIRECTOR( images_read_2Darray_depth ) -TEST_FN_REDIRECTOR( images_write_2Darray_depth ) +TEST_FN_REDIRECTOR(images_read_2D_depth) +TEST_FN_REDIRECTOR(images_write_2D_depth) +TEST_FN_REDIRECTOR(images_read_2Darray_depth) +TEST_FN_REDIRECTOR(images_write_2Darray_depth) -TEST_FN_REDIRECTOR( images_read_2D_multisample ) -TEST_FN_REDIRECTOR( images_read_2Darray_multisample ) -TEST_FN_REDIRECTOR( image_methods_depth ) -TEST_FN_REDIRECTOR( image_methods_multisample ) +TEST_FN_REDIRECTOR(images_read_2D_multisample) +TEST_FN_REDIRECTOR(images_read_2Darray_multisample) +TEST_FN_REDIRECTOR(image_methods_depth) +TEST_FN_REDIRECTOR(image_methods_multisample) #endif // Renderbuffer-backed images: -TEST_FN_REDIRECTOR( renderbuffer_read ) -TEST_FN_REDIRECTOR( renderbuffer_write ) -TEST_FN_REDIRECTOR( renderbuffer_getinfo ) +TEST_FN_REDIRECTOR(renderbuffer_read) +TEST_FN_REDIRECTOR(renderbuffer_write) +TEST_FN_REDIRECTOR(renderbuffer_getinfo) -TEST_FN_REDIRECTOR( fence_sync ) +TEST_FN_REDIRECTOR(fence_sync) -test_definition test_list[] = { - TEST_FN_REDIRECT( buffers ), - TEST_FN_REDIRECT( buffers_getinfo ), +test_definition test_list[] = { TEST_FN_REDIRECT(buffers), + TEST_FN_REDIRECT(buffers_getinfo), - TEST_FN_REDIRECT( images_read_1D ), - TEST_FN_REDIRECT( images_write_1D ), - TEST_FN_REDIRECT( images_1D_getinfo ), + TEST_FN_REDIRECT(images_read_1D), + TEST_FN_REDIRECT(images_write_1D), + TEST_FN_REDIRECT(images_1D_getinfo), - TEST_FN_REDIRECT( images_read_1Darray ), - TEST_FN_REDIRECT( images_write_1Darray ), - TEST_FN_REDIRECT( images_1Darray_getinfo ), + TEST_FN_REDIRECT(images_read_1Darray), + TEST_FN_REDIRECT(images_write_1Darray), + TEST_FN_REDIRECT(images_1Darray_getinfo), - TEST_FN_REDIRECT( images_read_2D ), - TEST_FN_REDIRECT( images_write ), - TEST_FN_REDIRECT( images_2D_getinfo ), + TEST_FN_REDIRECT(images_read_2D), + TEST_FN_REDIRECT(images_write), + TEST_FN_REDIRECT(images_2D_getinfo), - TEST_FN_REDIRECT( images_read_cube ), - TEST_FN_REDIRECT( images_write_cube ), - TEST_FN_REDIRECT( images_cube_getinfo ), + TEST_FN_REDIRECT(images_read_cube), + TEST_FN_REDIRECT(images_write_cube), + TEST_FN_REDIRECT(images_cube_getinfo), - TEST_FN_REDIRECT( images_read_2Darray ), - TEST_FN_REDIRECT( images_write_2Darray), - TEST_FN_REDIRECT( images_2Darray_getinfo ), + TEST_FN_REDIRECT(images_read_2Darray), + TEST_FN_REDIRECT(images_write_2Darray), + TEST_FN_REDIRECT(images_2Darray_getinfo), - TEST_FN_REDIRECT( images_read_3D ), - TEST_FN_REDIRECT( images_write_3D ), - TEST_FN_REDIRECT( images_3D_getinfo ), + TEST_FN_REDIRECT(images_read_3D), + TEST_FN_REDIRECT(images_write_3D), + TEST_FN_REDIRECT(images_3D_getinfo), - TEST_FN_REDIRECT( renderbuffer_read ), - TEST_FN_REDIRECT( renderbuffer_write ), - TEST_FN_REDIRECT( renderbuffer_getinfo ) -}; + TEST_FN_REDIRECT(renderbuffer_read), + TEST_FN_REDIRECT(renderbuffer_write), + TEST_FN_REDIRECT(renderbuffer_getinfo) }; test_definition test_list32[] = { - TEST_FN_REDIRECT( images_read_texturebuffer ), - TEST_FN_REDIRECT( images_write_texturebuffer ), - TEST_FN_REDIRECT( images_texturebuffer_getinfo ), + TEST_FN_REDIRECT(images_read_texturebuffer), + TEST_FN_REDIRECT(images_write_texturebuffer), + TEST_FN_REDIRECT(images_texturebuffer_getinfo), - TEST_FN_REDIRECT( fence_sync ), - TEST_FN_REDIRECT( images_read_2D_depth ), - TEST_FN_REDIRECT( images_write_2D_depth ), - TEST_FN_REDIRECT( images_read_2Darray_depth ), - TEST_FN_REDIRECT( images_write_2Darray_depth ), - TEST_FN_REDIRECT( images_read_2D_multisample ), - TEST_FN_REDIRECT( images_read_2Darray_multisample ), - TEST_FN_REDIRECT( image_methods_depth ), - TEST_FN_REDIRECT( image_methods_multisample ) + TEST_FN_REDIRECT(fence_sync), + TEST_FN_REDIRECT(images_read_2D_depth), + TEST_FN_REDIRECT(images_write_2D_depth), + TEST_FN_REDIRECT(images_read_2Darray_depth), + TEST_FN_REDIRECT(images_write_2Darray_depth), + TEST_FN_REDIRECT(images_read_2D_multisample), + TEST_FN_REDIRECT(images_read_2Darray_multisample), + TEST_FN_REDIRECT(image_methods_depth), + TEST_FN_REDIRECT(image_methods_multisample) }; -const int test_num = ARRAY_SIZE( test_list ); -const int test_num32 = ARRAY_SIZE( test_list32 ); +const int test_num = ARRAY_SIZE(test_list); +const int test_num32 = ARRAY_SIZE(test_list32); int main(int argc, const char *argv[]) { - gTestRounding = true; - int error = 0; - int numErrors = 0; + gTestRounding = true; + int error = 0; + int numErrors = 0; - test_start(); - argc = parseCustomParam(argc, argv); - if (argc == -1) - { - return -1; - } - - cl_device_type requestedDeviceType = CL_DEVICE_TYPE_DEFAULT; - - /* Do we have a CPU/GPU specification? */ - if( argc > 1 ) - { - if( strcmp( argv[ argc - 1 ], "gpu" ) == 0 || strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_GPU" ) == 0 ) + test_start(); + argc = parseCustomParam(argc, argv); + if (argc == -1) { - requestedDeviceType = CL_DEVICE_TYPE_GPU; - argc--; - } - else if( strcmp( argv[ argc - 1 ], "cpu" ) == 0 || strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_CPU" ) == 0 ) - { - requestedDeviceType = CL_DEVICE_TYPE_CPU; - argc--; + return -1; } - else if( strcmp( argv[ argc - 1 ], "accelerator" ) == 0 || strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_ACCELERATOR" ) == 0 ) + + cl_device_type requestedDeviceType = CL_DEVICE_TYPE_DEFAULT; + + /* Do we have a CPU/GPU specification? */ + if (argc > 1) { - requestedDeviceType = CL_DEVICE_TYPE_ACCELERATOR; - argc--; + if (strcmp(argv[argc - 1], "gpu") == 0 + || strcmp(argv[argc - 1], "CL_DEVICE_TYPE_GPU") == 0) + { + requestedDeviceType = CL_DEVICE_TYPE_GPU; + argc--; + } + else if (strcmp(argv[argc - 1], "cpu") == 0 + || strcmp(argv[argc - 1], "CL_DEVICE_TYPE_CPU") == 0) + { + requestedDeviceType = CL_DEVICE_TYPE_CPU; + argc--; + } + else if (strcmp(argv[argc - 1], "accelerator") == 0 + || strcmp(argv[argc - 1], "CL_DEVICE_TYPE_ACCELERATOR") == 0) + { + requestedDeviceType = CL_DEVICE_TYPE_ACCELERATOR; + argc--; + } + else if (strcmp(argv[argc - 1], "CL_DEVICE_TYPE_DEFAULT") == 0) + { + requestedDeviceType = CL_DEVICE_TYPE_DEFAULT; + argc--; + } } - else if( strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_DEFAULT" ) == 0 ) + + if (argc > 1 && strcmp(argv[1], "-list") == 0) { - requestedDeviceType = CL_DEVICE_TYPE_DEFAULT; - argc--; - } - } + log_info("Available 2.x tests:\n"); + for (int i = 0; i < test_num; i++) + log_info("\t%s\n", test_list[i].name); - if( argc > 1 && strcmp( argv[ 1 ], "-list" ) == 0 ) - { - log_info( "Available 2.x tests:\n" ); - for( int i = 0; i < test_num; i++ ) - log_info( "\t%s\n", test_list[i].name ); + log_info("Available 3.2 tests:\n"); + for (int i = 0; i < test_num32; i++) + log_info("\t%s\n", test_list32[i].name); - log_info( "Available 3.2 tests:\n" ); - for( int i = 0; i < test_num32; i++ ) - log_info( "\t%s\n", test_list32[i].name ); - - log_info( "Note: Any 3.2 test names must follow 2.1 test names on the command line.\n" ); - log_info( "Use environment variables to specify desired device.\n" ); + log_info("Note: Any 3.2 test names must follow 2.1 test names on the " + "command line.\n"); + log_info("Use environment variables to specify desired device.\n"); return 0; } - // Check to see if any 2.x or 3.2 test names were specified on the command line. - unsigned first_32_testname = 0; + // Check to see if any 2.x or 3.2 test names were specified on the command + // line. + unsigned first_32_testname = 0; - for (int j=1; (jSupportsCLGLInterop( requestedDeviceType ); - if( supported == 0 ) { - log_info("Test not run because GL-CL interop is not supported for any devices of the requested type.\n"); - return 0; - } else if ( supported == -1 ) { - log_error("Unable to setup the test or failed to determine if CL-GL interop is supported.\n"); - return -1; - } - - // Initialize function pointers. - error = init_clgl_ext(); - if (error < 0) { - return error; - } - - // OpenGL tests for non-3.2 //////////////////////////////////////////////////////// - if ((argc == 1) || (first_32_testname != 1)) { - - // At least one device supports CL-GL interop, so init the test. - if( glEnv->Init( &argc, (char **)argv, CL_FALSE ) ) { - log_error("Failed to initialize the GL environment for this test.\n"); - return -1; + // Check if any devices of the requested type support CL/GL interop. + int supported = glEnv->SupportsCLGLInterop(requestedDeviceType); + if (supported == 0) + { + log_info("Test not run because GL-CL interop is not supported for any " + "devices of the requested type.\n"); + return 0; } - - // Create a context to use and then grab a device (or devices) from it - sCurrentContext = glEnv->CreateCLContext(); - if( sCurrentContext == NULL ) - { - log_error( "ERROR: Unable to obtain CL context from GL\n" ); - return -1; - } - - size_t numDevices = 0; - cl_device_id *deviceIDs; - - error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL, &numDevices); - if( error != CL_SUCCESS ) - { - print_error( error, "Unable to get device count from context" ); - return -1; - } - deviceIDs = (cl_device_id *)malloc(numDevices); - if (deviceIDs == NULL) { - print_error( error, "malloc failed" ); + else if (supported == -1) + { + log_error("Unable to setup the test or failed to determine if CL-GL " + "interop is supported.\n"); return -1; } - error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, numDevices, deviceIDs, NULL); - if( error != CL_SUCCESS ) { - print_error( error, "Unable to get device list from context" ); - return -1; + + // Initialize function pointers. + error = init_clgl_ext(); + if (error < 0) + { + return error; } - numDevices /= sizeof(cl_device_id); + // OpenGL tests for non-3.2 + // //////////////////////////////////////////////////////// + if ((argc == 1) || (first_32_testname != 1)) + { - if (numDevices < 1) { - log_error("No devices found.\n"); - return -1; - } - - // Execute tests. - int argc_ = (first_32_testname) ? first_32_testname : argc; - - for( size_t i = 0; i < numDevices; i++ ) { - log_info( "\nTesting OpenGL 2.x\n" ); - if( printDeviceHeader( deviceIDs[ i ] ) != CL_SUCCESS ) { - return -1; + // At least one device supports CL-GL interop, so init the test. + if (glEnv->Init(&argc, (char **)argv, CL_FALSE)) + { + log_error( + "Failed to initialize the GL environment for this test.\n"); + return -1; } - // Note: don't use the entire harness, because we have a different way of obtaining the device (via the context) - error = parseAndCallCommandLineTests( argc_, argv, deviceIDs[i], test_num, test_list, true, 0, 1024 ); - if( error != 0 ) - break; - } - - numErrors += error; - - // Clean-up. - free(deviceIDs); - clReleaseContext( sCurrentContext ); - //delete glEnv; - } - - // OpenGL 3.2 tests. //////////////////////////////////////////////////////// - if ((argc==1) || first_32_testname) { - - // At least one device supports CL-GL interop, so init the test. - if( glEnv->Init( &argc, (char **)argv, CL_TRUE ) ) { - log_error("Failed to initialize the GL environment for this test.\n"); - return -1; - } - - // Create a context to use and then grab a device (or devices) from it - sCurrentContext = glEnv->CreateCLContext(); - if( sCurrentContext == NULL ) { - log_error( "ERROR: Unable to obtain CL context from GL\n" ); - return -1; - } - - size_t numDevices = 0; - cl_device_id *deviceIDs; - - error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL, &numDevices); - if( error != CL_SUCCESS ) { - print_error( error, "Unable to get device count from context" ); - return -1; - } - deviceIDs = (cl_device_id *)malloc(numDevices); - if (deviceIDs == NULL) { - print_error( error, "malloc failed" ); - return -1; - } - error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, numDevices, deviceIDs, NULL); - if( error != CL_SUCCESS ) { - print_error( error, "Unable to get device list from context" ); - return -1; - } - - numDevices /= sizeof(cl_device_id); - - if (numDevices < 1) { - log_error("No devices found.\n"); - return -1; - } - - int argc_ = (first_32_testname) ? 1 + (argc - first_32_testname) : argc; - const char** argv_ = (first_32_testname) ? &argv[first_32_testname-1] : argv; - - // Execute the tests. - for( size_t i = 0; i < numDevices; i++ ) { - log_info( "\nTesting OpenGL 3.2\n" ); - if( printDeviceHeader( deviceIDs[ i ] ) != CL_SUCCESS ) { - return -1; + // Create a context to use and then grab a device (or devices) from it + sCurrentContext = glEnv->CreateCLContext(); + if (sCurrentContext == NULL) + { + log_error("ERROR: Unable to obtain CL context from GL\n"); + return -1; } - // Note: don't use the entire harness, because we have a different way of obtaining the device (via the context) - error = parseAndCallCommandLineTests( argc_, argv_, deviceIDs[i], test_num32, test_list32, true, 0, 1024 ); - if( error != 0 ) - break; + size_t numDevices = 0; + cl_device_id *deviceIDs; + + error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL, + &numDevices); + if (error != CL_SUCCESS) + { + print_error(error, "Unable to get device count from context"); + return -1; + } + deviceIDs = (cl_device_id *)malloc(numDevices); + if (deviceIDs == NULL) + { + print_error(error, "malloc failed"); + return -1; + } + error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES, + numDevices, deviceIDs, NULL); + if (error != CL_SUCCESS) + { + print_error(error, "Unable to get device list from context"); + return -1; + } + + numDevices /= sizeof(cl_device_id); + + if (numDevices < 1) + { + log_error("No devices found.\n"); + return -1; + } + + // Execute tests. + int argc_ = (first_32_testname) ? first_32_testname : argc; + + for (size_t i = 0; i < numDevices; i++) + { + log_info("\nTesting OpenGL 2.x\n"); + if (printDeviceHeader(deviceIDs[i]) != CL_SUCCESS) + { + return -1; + } + + // Note: don't use the entire harness, because we have a different + // way of obtaining the device (via the context) + error = parseAndCallCommandLineTests( + argc_, argv, deviceIDs[i], test_num, test_list, true, 0, 1024); + if (error != 0) break; + } + + numErrors += error; + + // Clean-up. + free(deviceIDs); + clReleaseContext(sCurrentContext); + // delete glEnv; } - numErrors += error; + // OpenGL 3.2 tests. + // //////////////////////////////////////////////////////// + if ((argc == 1) || first_32_testname) + { - // Clean-up. - free(deviceIDs); - clReleaseContext( sCurrentContext ); - delete glEnv; + // At least one device supports CL-GL interop, so init the test. + if (glEnv->Init(&argc, (char **)argv, CL_TRUE)) + { + log_error( + "Failed to initialize the GL environment for this test.\n"); + return -1; + } - } + // Create a context to use and then grab a device (or devices) from it + sCurrentContext = glEnv->CreateCLContext(); + if (sCurrentContext == NULL) + { + log_error("ERROR: Unable to obtain CL context from GL\n"); + return -1; + } - //All done. - return numErrors; + size_t numDevices = 0; + cl_device_id *deviceIDs; + + error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL, + &numDevices); + if (error != CL_SUCCESS) + { + print_error(error, "Unable to get device count from context"); + return -1; + } + deviceIDs = (cl_device_id *)malloc(numDevices); + if (deviceIDs == NULL) + { + print_error(error, "malloc failed"); + return -1; + } + error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES, + numDevices, deviceIDs, NULL); + if (error != CL_SUCCESS) + { + print_error(error, "Unable to get device list from context"); + return -1; + } + + numDevices /= sizeof(cl_device_id); + + if (numDevices < 1) + { + log_error("No devices found.\n"); + return -1; + } + + int argc_ = (first_32_testname) ? 1 + (argc - first_32_testname) : argc; + const char **argv_ = + (first_32_testname) ? &argv[first_32_testname - 1] : argv; + + // Execute the tests. + for (size_t i = 0; i < numDevices; i++) + { + log_info("\nTesting OpenGL 3.2\n"); + if (printDeviceHeader(deviceIDs[i]) != CL_SUCCESS) + { + return -1; + } + + // Note: don't use the entire harness, because we have a different + // way of obtaining the device (via the context) + error = parseAndCallCommandLineTests(argc_, argv_, deviceIDs[i], + test_num32, test_list32, true, + 0, 1024); + if (error != 0) break; + } + + numErrors += error; + + // Clean-up. + free(deviceIDs); + clReleaseContext(sCurrentContext); + delete glEnv; + } + + // All done. + return numErrors; } - diff --git a/test_conformance/gl/procs.h b/test_conformance/gl/procs.h index b14e22dc..111de7a6 100644 --- a/test_conformance/gl/procs.h +++ b/test_conformance/gl/procs.h @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -20,121 +20,134 @@ #pragma mark - #pragma Misc tests -extern int test_buffers( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_buffers(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_fence_sync( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_fence_sync(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); #pragma mark - #pragma mark Tead tests -extern int test_images_read_2D( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_2D(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_read_1D( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_1D(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_read_texturebuffer( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_texturebuffer(cl_device_id device, + cl_context context, + cl_command_queue queue, + int num_elements); -extern int test_images_read_1Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_1Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_read_2Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_2Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_read_cube( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_cube(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_read_3D( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_read_3D(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_renderbuffer_read( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_renderbuffer_read(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); #pragma mark - #pragma mark Write tests // 2D tests are the ones with no suffix: -extern int test_images_write( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_write_cube( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write_cube(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_renderbuffer_write( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_renderbuffer_write(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); // Here are the rest: -extern int test_images_write_1D( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write_1D(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_write_texturebuffer( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write_texturebuffer(cl_device_id device, + cl_context context, + cl_command_queue queue, + int num_elements); -extern int test_images_write_1Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write_1Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_write_2Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write_2Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); -extern int test_images_write_3D( cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements ); +extern int test_images_write_3D(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); #pragma mark - #pragma mark Get info test entry points -extern int test_buffers_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_buffers_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_1D_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_1D_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_texturebuffer_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_texturebuffer_getinfo(cl_device_id device, + cl_context context, + cl_command_queue queue, + int numElements); -extern int test_images_1Darray_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_1Darray_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_2D_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_2D_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_2Darray_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_2Darray_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_cube_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_cube_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_3D_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_3D_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_read_2D_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_read_2D_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_write_2D_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_write_2D_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); -extern int test_images_read_2Darray_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int ); +extern int test_images_read_2Darray_depth(cl_device_id device, + cl_context context, + cl_command_queue queue, int); -extern int test_images_write_2Darray_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_write_2Darray_depth(cl_device_id device, + cl_context context, + cl_command_queue queue, + int numElements); -extern int test_images_read_2D_multisample( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); +extern int test_images_read_2D_multisample(cl_device_id device, + cl_context context, + cl_command_queue queue, + int numElements); -extern int test_images_read_2Darray_multisample( cl_device_id device, cl_context context, - cl_command_queue queue, int ); +extern int test_images_read_2Darray_multisample(cl_device_id device, + cl_context context, + cl_command_queue queue, int); -extern int test_image_methods_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int ); +extern int test_image_methods_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int); -extern int test_image_methods_multisample( cl_device_id device, cl_context context, - cl_command_queue queue, int ); +extern int test_image_methods_multisample(cl_device_id device, + cl_context context, + cl_command_queue queue, int); -extern int test_renderbuffer_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ); \ No newline at end of file +extern int test_renderbuffer_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements); \ No newline at end of file diff --git a/test_conformance/gl/testBase.h b/test_conformance/gl/testBase.h index b6907621..7e187536 100644 --- a/test_conformance/gl/testBase.h +++ b/test_conformance/gl/testBase.h @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -23,7 +23,7 @@ #include #include -#if !defined (__APPLE__) +#if !defined(__APPLE__) #include #include "gl/gl_headers.h" #include @@ -40,25 +40,40 @@ #include "gl/helpers.h" -extern const char *get_kernel_suffix( cl_image_format *format ); -extern const char *get_write_conversion( cl_image_format *format, ExplicitType type); -extern ExplicitType get_read_kernel_type( cl_image_format *format ); -extern ExplicitType get_write_kernel_type( cl_image_format *format ); +extern const char *get_kernel_suffix(cl_image_format *format); +extern const char *get_write_conversion(cl_image_format *format, + ExplicitType type); +extern ExplicitType get_read_kernel_type(cl_image_format *format); +extern ExplicitType get_write_kernel_type(cl_image_format *format); -extern char * convert_to_expected( void * inputBuffer, size_t numPixels, ExplicitType inType, ExplicitType outType, size_t channelNum, GLenum glDataType = 0); -extern int validate_integer_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum, size_t typeSize ); -extern int validate_integer_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum, size_t typeSize ); -extern int validate_float_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum, size_t channelNum ); -extern int validate_float_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum, size_t channelNum ); -extern int validate_float_results_rgb_101010( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum ); -extern int validate_float_results_rgb_101010( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum ); +extern char *convert_to_expected(void *inputBuffer, size_t numPixels, + ExplicitType inType, ExplicitType outType, + size_t channelNum, GLenum glDataType = 0); +extern int validate_integer_results(void *expectedResults, void *actualResults, + size_t width, size_t height, + size_t sampleNum, size_t typeSize); +extern int validate_integer_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t depth, + size_t sampleNum, size_t typeSize); +extern int validate_float_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t sampleNum, + size_t channelNum); +extern int validate_float_results(void *expectedResults, void *actualResults, + size_t width, size_t height, size_t depth, + size_t sampleNum, size_t channelNum); +extern int validate_float_results_rgb_101010(void *expectedResults, + void *actualResults, size_t width, + size_t height, size_t sampleNum); +extern int validate_float_results_rgb_101010(void *expectedResults, + void *actualResults, size_t width, + size_t height, size_t depth, + size_t sampleNum); -extern int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type, GLuint expected_gl_name, - GLenum expected_cl_gl_texture_target, GLint expected_cl_gl_mipmap_level); +extern int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type, + GLuint expected_gl_name, + GLenum expected_cl_gl_texture_target, + GLint expected_cl_gl_mipmap_level); extern bool CheckGLIntegerExtensionSupport(); #endif // _testBase_h - - - diff --git a/test_conformance/gl/test_fence_sync.cpp b/test_conformance/gl/test_fence_sync.cpp index 35cc62de..088d7497 100644 --- a/test_conformance/gl/test_fence_sync.cpp +++ b/test_conformance/gl/test_fence_sync.cpp @@ -113,6 +113,7 @@ typedef cl_event(CL_API_CALL *clCreateEventFromGLsyncKHR_fn)( clCreateEventFromGLsyncKHR_fn clCreateEventFromGLsyncKHR_ptr; +// clang-format off static const char *updateBuffersKernel[] = { "__kernel void update( __global float4 * vertices, __global float4 " "*colors, int horizWrap, int rowIdx )\n" @@ -132,6 +133,7 @@ static const char *updateBuffersKernel[] = { " colors[ tid * 2 + 1 ] = colors[ tid * 2 + 0 ];\n" "}\n" }; +// clang-format on // Passthrough VertexShader static const char *vertexshader = "#version 150\n" diff --git a/test_conformance/gl/test_image_methods.cpp b/test_conformance/gl/test_image_methods.cpp index 7d055fb2..187f2e6e 100644 --- a/test_conformance/gl/test_image_methods.cpp +++ b/test_conformance/gl/test_image_methods.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -24,16 +24,17 @@ struct image_kernel_data cl_int width; cl_int height; cl_int depth; - cl_int arraySize; + cl_int arraySize; cl_int widthDim; cl_int heightDim; cl_int channelType; cl_int channelOrder; cl_int expectedChannelType; cl_int expectedChannelOrder; - cl_int numSamples; + cl_int numSamples; }; +// clang-format off static const char *methodTestKernelPattern = "%s" "typedef struct {\n" @@ -53,239 +54,260 @@ static const char *methodTestKernelPattern = "{\n" "%s%s%s%s%s%s%s%s%s%s%s" "}\n"; +// clang-format on static const char *arraySizeKernelLine = -" outData->arraySize = get_image_array_size( input );\n"; + " outData->arraySize = get_image_array_size( input );\n"; static const char *imageWidthKernelLine = -" outData->width = get_image_width( input );\n"; + " outData->width = get_image_width( input );\n"; static const char *imageHeightKernelLine = -" outData->height = get_image_height( input );\n"; + " outData->height = get_image_height( input );\n"; static const char *imageDimKernelLine = -" int2 dim = get_image_dim( input );\n"; -static const char *imageWidthDimKernelLine = -" outData->widthDim = dim.x;\n"; + " int2 dim = get_image_dim( input );\n"; +static const char *imageWidthDimKernelLine = " outData->widthDim = dim.x;\n"; static const char *imageHeightDimKernelLine = -" outData->heightDim = dim.y;\n"; + " outData->heightDim = dim.y;\n"; static const char *channelTypeKernelLine = -" outData->channelType = get_image_channel_data_type( input );\n"; + " outData->channelType = get_image_channel_data_type( input );\n"; static const char *channelTypeConstLine = -" outData->expectedChannelType = CLK_%s;\n"; + " outData->expectedChannelType = CLK_%s;\n"; static const char *channelOrderKernelLine = -" outData->channelOrder = get_image_channel_order( input );\n"; + " outData->channelOrder = get_image_channel_order( input );\n"; static const char *channelOrderConstLine = -" outData->expectedChannelOrder = CLK_%s;\n"; + " outData->expectedChannelOrder = CLK_%s;\n"; static const char *numSamplesKernelLine = -" outData->numSamples = get_image_num_samples( input );\n"; + " outData->numSamples = get_image_num_samples( input );\n"; static const char *enableMSAAKernelLine = -"#pragma OPENCL EXTENSION cl_khr_gl_msaa_sharing : enable\n"; + "#pragma OPENCL EXTENSION cl_khr_gl_msaa_sharing : enable\n"; -static int verify(cl_int input, cl_int kernelOutput, const char * description) +static int verify(cl_int input, cl_int kernelOutput, const char *description) { - if( kernelOutput != input ) - { - log_error( "ERROR: %s did not validate (expected %d, got %d)\n", description, input, kernelOutput); - return -1; - } - return 0; + if (kernelOutput != input) + { + log_error("ERROR: %s did not validate (expected %d, got %d)\n", + description, input, kernelOutput); + return -1; + } + return 0; } -extern int supportsMsaa(cl_context context, bool* supports_msaa); -extern int supportsDepth(cl_context context, bool* supports_depth); +extern int supportsMsaa(cl_context context, bool *supports_msaa); +extern int supportsDepth(cl_context context, bool *supports_depth); -int test_image_format_methods( cl_device_id device, cl_context context, cl_command_queue queue, - size_t width, size_t height, size_t arraySize, size_t samples, - GLenum target, format format, MTdata d ) +int test_image_format_methods(cl_device_id device, cl_context context, + cl_command_queue queue, size_t width, + size_t height, size_t arraySize, size_t samples, + GLenum target, format format, MTdata d) { - int error, result=0; + int error, result = 0; clProgramWrapper program; clKernelWrapper kernel; clMemWrapper image, outDataBuffer; - char programSrc[ 10240 ]; + char programSrc[10240]; - image_kernel_data outKernelData; + image_kernel_data outKernelData; #ifdef GL_VERSION_3_2 - if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE || - get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) + if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE + || get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) { bool supports_msaa; error = supportsMsaa(context, &supports_msaa); - if( error != 0 ) return error; + if (error != 0) return error; if (!supports_msaa) return 0; } - if (format.formattype == GL_DEPTH_COMPONENT || - format.formattype == GL_DEPTH_STENCIL) + if (format.formattype == GL_DEPTH_COMPONENT + || format.formattype == GL_DEPTH_STENCIL) { bool supports_depth; error = supportsDepth(context, &supports_depth); - if( error != 0 ) return error; + if (error != 0) return error; if (!supports_depth) return 0; } #endif - DetectFloatToHalfRoundingMode(queue); + DetectFloatToHalfRoundingMode(queue); - glTextureWrapper glTexture; - switch (get_base_gl_target(target)) { - case GL_TEXTURE_2D: - CreateGLTexture2D( width, height, target, - format.formattype, format.internal, format.datatype, - format.type, &glTexture, &error, false, d ); - break; - case GL_TEXTURE_2D_ARRAY: - CreateGLTexture2DArray( width, height, arraySize, target, - format.formattype, format.internal, format.datatype, - format.type, &glTexture, &error, false, d ); - break; - case GL_TEXTURE_2D_MULTISAMPLE: - CreateGLTexture2DMultisample( width, height, samples, target, - format.formattype, format.internal, format.datatype, - format.type, &glTexture, &error, false, d, false); - break; - case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: - CreateGLTexture2DArrayMultisample( width, height, arraySize, samples, target, - format.formattype, format.internal, format.datatype, - format.type, &glTexture, &error, false, d, false); - break; + glTextureWrapper glTexture; + switch (get_base_gl_target(target)) + { + case GL_TEXTURE_2D: + CreateGLTexture2D(width, height, target, format.formattype, + format.internal, format.datatype, format.type, + &glTexture, &error, false, d); + break; + case GL_TEXTURE_2D_ARRAY: + CreateGLTexture2DArray(width, height, arraySize, target, + format.formattype, format.internal, + format.datatype, format.type, &glTexture, + &error, false, d); + break; + case GL_TEXTURE_2D_MULTISAMPLE: + CreateGLTexture2DMultisample(width, height, samples, target, + format.formattype, format.internal, + format.datatype, format.type, + &glTexture, &error, false, d, false); + break; + case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: + CreateGLTexture2DArrayMultisample( + width, height, arraySize, samples, target, format.formattype, + format.internal, format.datatype, format.type, &glTexture, + &error, false, d, false); + break; - default: - log_error("Unsupported GL tex target (%s) passed to write test: " - "%s (%s):%d", GetGLTargetName(target), __FUNCTION__, - __FILE__, __LINE__); - } + default: + log_error("Unsupported GL tex target (%s) passed to write test: " + "%s (%s):%d", + GetGLTargetName(target), __FUNCTION__, __FILE__, + __LINE__); + } - // Check to see if the texture could not be created for some other reason like - // GL_FRAMEBUFFER_UNSUPPORTED - if (error == GL_FRAMEBUFFER_UNSUPPORTED) { - return 0; - } + // Check to see if the texture could not be created for some other reason + // like GL_FRAMEBUFFER_UNSUPPORTED + if (error == GL_FRAMEBUFFER_UNSUPPORTED) + { + return 0; + } // Construct testing source - log_info( " - Creating image %d by %d...\n", width, height ); - // Create a CL image from the supplied GL texture - image = (*clCreateFromGLTexture_ptr)( context, CL_MEM_READ_ONLY, - target, 0, glTexture, &error ); + log_info(" - Creating image %d by %d...\n", width, height); + // Create a CL image from the supplied GL texture + image = (*clCreateFromGLTexture_ptr)(context, CL_MEM_READ_ONLY, target, 0, + glTexture, &error); - if ( error != CL_SUCCESS ) { - print_error( error, "Unable to create CL image from GL texture" ); - GLint fmt; - glGetTexLevelParameteriv( target, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); - log_error( " Supplied GL texture was base format %s and internal " - "format %s\n", GetGLBaseFormatName( fmt ), GetGLFormatName( fmt ) ); - return error; - } - - cl_image_format imageFormat; - error = clGetImageInfo (image, CL_IMAGE_FORMAT, - sizeof(imageFormat), &imageFormat, NULL); - test_error(error, "Failed to get image format"); - - const char * imageType = 0; - bool doArraySize = false; - bool doImageWidth = false; - bool doImageHeight = false; - bool doImageChannelDataType = false; - bool doImageChannelOrder = false; - bool doImageDim = false; - bool doNumSamples = false; - bool doMSAA = false; - switch(target) { - case GL_TEXTURE_2D: - imageType = "image2d_depth_t"; - doImageWidth = true; - doImageHeight = true; - doImageChannelDataType = true; - doImageChannelOrder = true; - doImageDim = true; - break; - case GL_TEXTURE_2D_ARRAY: - imageType = "image2d_array_depth_t"; - doImageWidth = true; - doImageHeight = true; - doArraySize = true; - doImageChannelDataType = true; - doImageChannelOrder = true; - doImageDim = true; - doArraySize = true; - break; - case GL_TEXTURE_2D_MULTISAMPLE: - doNumSamples = true; - doMSAA = true; - if(format.formattype == GL_DEPTH_COMPONENT) { - doImageWidth = true; - imageType = "image2d_msaa_depth_t"; - } else { - imageType = "image2d_msaa_t"; - } - break; - case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: - doMSAA = true; - if(format.formattype == GL_DEPTH_COMPONENT) { - doImageWidth = true; - imageType = "image2d_msaa_array_depth_t"; - } else { - imageType = "image2d_array_msaa_t"; - } - break; - } - - - - char channelTypeConstKernelLine[512] = {0}; - char channelOrderConstKernelLine[512] = {0}; - const char* channelTypeName=0; - const char* channelOrderName=0; - if(doImageChannelDataType) { - channelTypeName = GetChannelTypeName( imageFormat.image_channel_data_type ); - if(channelTypeName && strlen(channelTypeName)) { - // replace CL_* with CLK_* - sprintf(channelTypeConstKernelLine, channelTypeConstLine, &channelTypeName[3]); + if (error != CL_SUCCESS) + { + print_error(error, "Unable to create CL image from GL texture"); + GLint fmt; + glGetTexLevelParameteriv(target, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt); + log_error(" Supplied GL texture was base format %s and internal " + "format %s\n", + GetGLBaseFormatName(fmt), GetGLFormatName(fmt)); + return error; } - } - if(doImageChannelOrder) { - channelOrderName = GetChannelOrderName( imageFormat.image_channel_order ); - if(channelOrderName && strlen(channelOrderName)) { - // replace CL_* with CLK_* - sprintf(channelOrderConstKernelLine, channelOrderConstLine, &channelOrderName[3]); + + cl_image_format imageFormat; + error = clGetImageInfo(image, CL_IMAGE_FORMAT, sizeof(imageFormat), + &imageFormat, NULL); + test_error(error, "Failed to get image format"); + + const char *imageType = 0; + bool doArraySize = false; + bool doImageWidth = false; + bool doImageHeight = false; + bool doImageChannelDataType = false; + bool doImageChannelOrder = false; + bool doImageDim = false; + bool doNumSamples = false; + bool doMSAA = false; + switch (target) + { + case GL_TEXTURE_2D: + imageType = "image2d_depth_t"; + doImageWidth = true; + doImageHeight = true; + doImageChannelDataType = true; + doImageChannelOrder = true; + doImageDim = true; + break; + case GL_TEXTURE_2D_ARRAY: + imageType = "image2d_array_depth_t"; + doImageWidth = true; + doImageHeight = true; + doArraySize = true; + doImageChannelDataType = true; + doImageChannelOrder = true; + doImageDim = true; + doArraySize = true; + break; + case GL_TEXTURE_2D_MULTISAMPLE: + doNumSamples = true; + doMSAA = true; + if (format.formattype == GL_DEPTH_COMPONENT) + { + doImageWidth = true; + imageType = "image2d_msaa_depth_t"; + } + else + { + imageType = "image2d_msaa_t"; + } + break; + case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: + doMSAA = true; + if (format.formattype == GL_DEPTH_COMPONENT) + { + doImageWidth = true; + imageType = "image2d_msaa_array_depth_t"; + } + else + { + imageType = "image2d_array_msaa_t"; + } + break; } - } - - // Create a program to run against - sprintf(programSrc, - methodTestKernelPattern, - ( doMSAA ) ? enableMSAAKernelLine : "", - imageType, - ( doArraySize ) ? arraySizeKernelLine : "", - ( doImageWidth ) ? imageWidthKernelLine : "", - ( doImageHeight ) ? imageHeightKernelLine : "", - ( doImageChannelDataType ) ? channelTypeKernelLine : "", - ( doImageChannelDataType ) ? channelTypeConstKernelLine : "", - ( doImageChannelOrder ) ? channelOrderKernelLine : "", - ( doImageChannelOrder ) ? channelOrderConstKernelLine : "", - ( doImageDim ) ? imageDimKernelLine : "", - ( doImageDim && doImageWidth ) ? imageWidthDimKernelLine : "", - ( doImageDim && doImageHeight ) ? imageHeightDimKernelLine : "", - ( doNumSamples ) ? numSamplesKernelLine : ""); - //log_info("-----------------------------------\n%s\n", programSrc); - error = clFinish(queue); - if (error) - print_error(error, "clFinish failed.\n"); + char channelTypeConstKernelLine[512] = { 0 }; + char channelOrderConstKernelLine[512] = { 0 }; + const char *channelTypeName = 0; + const char *channelOrderName = 0; + if (doImageChannelDataType) + { + channelTypeName = + GetChannelTypeName(imageFormat.image_channel_data_type); + if (channelTypeName && strlen(channelTypeName)) + { + // replace CL_* with CLK_* + sprintf(channelTypeConstKernelLine, channelTypeConstLine, + &channelTypeName[3]); + } + } + if (doImageChannelOrder) + { + channelOrderName = GetChannelOrderName(imageFormat.image_channel_order); + if (channelOrderName && strlen(channelOrderName)) + { + // replace CL_* with CLK_* + sprintf(channelOrderConstKernelLine, channelOrderConstLine, + &channelOrderName[3]); + } + } + + // Create a program to run against + sprintf(programSrc, methodTestKernelPattern, + (doMSAA) ? enableMSAAKernelLine : "", imageType, + (doArraySize) ? arraySizeKernelLine : "", + (doImageWidth) ? imageWidthKernelLine : "", + (doImageHeight) ? imageHeightKernelLine : "", + (doImageChannelDataType) ? channelTypeKernelLine : "", + (doImageChannelDataType) ? channelTypeConstKernelLine : "", + (doImageChannelOrder) ? channelOrderKernelLine : "", + (doImageChannelOrder) ? channelOrderConstKernelLine : "", + (doImageDim) ? imageDimKernelLine : "", + (doImageDim && doImageWidth) ? imageWidthDimKernelLine : "", + (doImageDim && doImageHeight) ? imageHeightDimKernelLine : "", + (doNumSamples) ? numSamplesKernelLine : ""); + + + // log_info("-----------------------------------\n%s\n", programSrc); + error = clFinish(queue); + if (error) print_error(error, "clFinish failed.\n"); const char *ptr = programSrc; - error = create_single_kernel_helper( context, &program, &kernel, 1, &ptr, "sample_kernel" ); - test_error( error, "Unable to create kernel to test against" ); + error = create_single_kernel_helper(context, &program, &kernel, 1, &ptr, + "sample_kernel"); + test_error(error, "Unable to create kernel to test against"); // Create an output buffer outDataBuffer = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(outKernelData), NULL, &error); - test_error( error, "Unable to create output buffer" ); + test_error(error, "Unable to create output buffer"); // Set up arguments and run - error = clSetKernelArg( kernel, 0, sizeof( image ), &image ); - test_error( error, "Unable to set kernel argument" ); - error = clSetKernelArg( kernel, 1, sizeof( outDataBuffer ), &outDataBuffer ); - test_error( error, "Unable to set kernel argument" ); + error = clSetKernelArg(kernel, 0, sizeof(image), &image); + test_error(error, "Unable to set kernel argument"); + error = clSetKernelArg(kernel, 1, sizeof(outDataBuffer), &outDataBuffer); + test_error(error, "Unable to set kernel argument"); // Finish and Acquire. glFinish(); @@ -294,119 +316,155 @@ int test_image_format_methods( cl_device_id device, cl_context context, cl_comma size_t threads[1] = { 1 }, localThreads[1] = { 1 }; - error = clEnqueueNDRangeKernel( queue, kernel, 1, NULL, threads, localThreads, 0, NULL, NULL ); - test_error( error, "Unable to run kernel" ); + error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, threads, + localThreads, 0, NULL, NULL); + test_error(error, "Unable to run kernel"); - error = clEnqueueReadBuffer( queue, outDataBuffer, CL_TRUE, 0, sizeof( outKernelData ), &outKernelData, 0, NULL, NULL ); - test_error( error, "Unable to read data buffer" ); + error = clEnqueueReadBuffer(queue, outDataBuffer, CL_TRUE, 0, + sizeof(outKernelData), &outKernelData, 0, NULL, + NULL); + test_error(error, "Unable to read data buffer"); // Verify the results now - if( doImageWidth ) - result |= verify(width, outKernelData.width, "width"); - if( doImageHeight) - result |= verify(height, outKernelData.height, "height"); - if( doImageDim && doImageWidth ) - result |= verify(width, outKernelData.widthDim, "width from get_image_dim"); - if( doImageDim && doImageHeight ) - result |= verify(height, outKernelData.heightDim, "height from get_image_dim"); - if( doImageChannelDataType ) - result |= verify(outKernelData.channelType, outKernelData.expectedChannelType, channelTypeName); - if( doImageChannelOrder ) - result |= verify(outKernelData.channelOrder, outKernelData.expectedChannelOrder, channelOrderName); - if( doArraySize ) - result |= verify(arraySize, outKernelData.arraySize, "array size"); - if( doNumSamples ) - result |= verify(samples, outKernelData.numSamples, "samples"); - if(result) { - log_error("Test image methods failed"); - } + if (doImageWidth) result |= verify(width, outKernelData.width, "width"); + if (doImageHeight) result |= verify(height, outKernelData.height, "height"); + if (doImageDim && doImageWidth) + result |= + verify(width, outKernelData.widthDim, "width from get_image_dim"); + if (doImageDim && doImageHeight) + result |= verify(height, outKernelData.heightDim, + "height from get_image_dim"); + if (doImageChannelDataType) + result |= verify(outKernelData.channelType, + outKernelData.expectedChannelType, channelTypeName); + if (doImageChannelOrder) + result |= verify(outKernelData.channelOrder, + outKernelData.expectedChannelOrder, channelOrderName); + if (doArraySize) + result |= verify(arraySize, outKernelData.arraySize, "array size"); + if (doNumSamples) + result |= verify(samples, outKernelData.numSamples, "samples"); + if (result) + { + log_error("Test image methods failed"); + } - clEventWrapper event; - error = (*clEnqueueReleaseGLObjects_ptr)( queue, 1, &image, 0, NULL, &event ); - test_error(error, "clEnqueueReleaseGLObjects failed"); + clEventWrapper event; + error = (*clEnqueueReleaseGLObjects_ptr)(queue, 1, &image, 0, NULL, &event); + test_error(error, "clEnqueueReleaseGLObjects failed"); - error = clWaitForEvents( 1, &event ); - test_error(error, "clWaitForEvents failed"); + error = clWaitForEvents(1, &event); + test_error(error, "clWaitForEvents failed"); return result; } -int test_image_methods_depth( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ){ - if (!is_extension_available(device, "cl_khr_gl_depth_images")) { - log_info("Test not run because 'cl_khr_gl_depth_images' extension is not supported by the tested device\n"); - return 0; - } +int test_image_methods_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) +{ + if (!is_extension_available(device, "cl_khr_gl_depth_images")) + { + log_info("Test not run because 'cl_khr_gl_depth_images' extension is " + "not supported by the tested device\n"); + return 0; + } int result = 0; - GLenum depth_targets[] = {GL_TEXTURE_2D, GL_TEXTURE_2D_ARRAY}; - size_t ntargets = sizeof(depth_targets) / sizeof(depth_targets[0]); - size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + GLenum depth_targets[] = { GL_TEXTURE_2D, GL_TEXTURE_2D_ARRAY }; + size_t ntargets = sizeof(depth_targets) / sizeof(depth_targets[0]); + size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - const size_t nsizes = 5; - sizevec_t sizes[nsizes]; - // Need to limit texture size according to GL device properties - GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, maxTextureLayers = 16, size; - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); - glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); + const size_t nsizes = 5; + sizevec_t sizes[nsizes]; + // Need to limit texture size according to GL device properties + GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, + maxTextureLayers = 16, size; + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); + glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); - size = min(maxTextureSize, maxTextureRectangleSize); + size = min(maxTextureSize, maxTextureRectangleSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); - } - - for (size_t i = 0; i < nsizes; i++) { - for(size_t itarget = 0; itarget < ntargets; ++itarget) { - for(size_t iformat = 0; iformat < nformats; ++iformat) - result |= test_image_format_methods(device, context, queue, sizes[i].width, sizes[i].height, (depth_targets[itarget] == GL_TEXTURE_2D_ARRAY) ? sizes[i].depth: 1, 0, - depth_targets[itarget], depth_formats[iformat], seed ); + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].height = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].depth = + random_in_range(2, min(maxTextureLayers, 1 << (i + 4)), seed); } - } - return result; -} -int test_image_methods_multisample( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ){ - if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) { - log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is not supported by the tested device\n"); - return 0; - } - - int result = 0; - GLenum targets[] = {GL_TEXTURE_2D_MULTISAMPLE, GL_TEXTURE_2D_MULTISAMPLE_ARRAY}; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - - const size_t nsizes = 5; - sizevec_t sizes[nsizes]; - GLint maxTextureLayers = 16, maxTextureSize = 4096; - glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - - RandomSeed seed( gRandomSeed ); - - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); + for (size_t i = 0; i < nsizes; i++) + { + for (size_t itarget = 0; itarget < ntargets; ++itarget) + { + for (size_t iformat = 0; iformat < nformats; ++iformat) + result |= test_image_format_methods( + device, context, queue, sizes[i].width, sizes[i].height, + (depth_targets[itarget] == GL_TEXTURE_2D_ARRAY) + ? sizes[i].depth + : 1, + 0, depth_targets[itarget], depth_formats[iformat], seed); + } + } + return result; +} + +int test_image_methods_multisample(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) +{ + if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) + { + log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is " + "not supported by the tested device\n"); + return 0; + } + + int result = 0; + GLenum targets[] = { GL_TEXTURE_2D_MULTISAMPLE, + GL_TEXTURE_2D_MULTISAMPLE_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + + const size_t nsizes = 5; + sizevec_t sizes[nsizes]; + GLint maxTextureLayers = 16, maxTextureSize = 4096; + glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + + RandomSeed seed(gRandomSeed); + + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].height = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].depth = + random_in_range(2, min(maxTextureLayers, 1 << (i + 4)), seed); + } + + glEnable(GL_MULTISAMPLE); + + for (size_t i = 0; i < nsizes; i++) + { + for (size_t itarget = 0; itarget < ntargets; ++itarget) + { + for (size_t iformat = 0; iformat < nformats; ++iformat) + { + GLint samples = get_gl_max_samples( + targets[itarget], common_formats[iformat].internal); + result |= test_image_format_methods( + device, context, queue, sizes[i].width, sizes[i].height, + (targets[ntargets] == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) + ? sizes[i].depth + : 1, + samples, targets[itarget], common_formats[iformat], seed); + } } - - glEnable(GL_MULTISAMPLE); - - for (size_t i = 0; i < nsizes; i++) { - for(size_t itarget = 0; itarget < ntargets; ++itarget) { - for(size_t iformat = 0; iformat < nformats; ++iformat) { - GLint samples = get_gl_max_samples(targets[itarget], common_formats[iformat].internal); - result |= test_image_format_methods(device, context, queue, sizes[i].width, sizes[i].height, (targets[ntargets] == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) ? sizes[i].depth: 1, - samples, targets[itarget], common_formats[iformat], seed ); - } } - } return result; } diff --git a/test_conformance/gl/test_images_1D.cpp b/test_conformance/gl/test_images_1D.cpp index 172dd4b5..4ccf86bc 100644 --- a/test_conformance/gl/test_images_1D.cpp +++ b/test_conformance/gl/test_images_1D.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,11 +16,11 @@ #include "common.h" #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif #include @@ -28,114 +28,116 @@ using namespace std; void calc_test_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - GLint maxTextureSize = 4096, maxTextureBufferSize = 4096, size; - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - glGetIntegerv(GL_MAX_TEXTURE_BUFFER_SIZE, &maxTextureBufferSize); + // Need to limit array size according to GL device properties + GLint maxTextureSize = 4096, maxTextureBufferSize = 4096, size; + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + glGetIntegerv(GL_MAX_TEXTURE_BUFFER_SIZE, &maxTextureBufferSize); - size = min(maxTextureSize, maxTextureBufferSize); + size = min(maxTextureSize, maxTextureBufferSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].height = 1; - sizes[i].depth = 1; - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].height = 1; + sizes[i].depth = 1; + } } -int test_images_read_1D( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_read_1D(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_1D }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_1D }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_test_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_write_1D( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_1D(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_1D }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GLenum targets[] = { GL_TEXTURE_1D }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_test_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_1D_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_1D_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_1D }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_1D }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_test_size_descriptors(sizes, nsizes); - return test_images_get_info_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_get_info_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, + nsizes); } -int test_images_read_texturebuffer( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_read_texturebuffer(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_BUFFER }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_BUFFER }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_test_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_write_texturebuffer( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_texturebuffer(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_BUFFER }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GLenum targets[] = { GL_TEXTURE_BUFFER }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_test_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_texturebuffer_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_texturebuffer_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_BUFFER }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_BUFFER }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_test_size_descriptors(sizes, nsizes); - return test_images_get_info_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_get_info_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, + nsizes); } - diff --git a/test_conformance/gl/test_images_1Darray.cpp b/test_conformance/gl/test_images_1Darray.cpp index 1914a457..daa2efa8 100644 --- a/test_conformance/gl/test_images_1Darray.cpp +++ b/test_conformance/gl/test_images_1Darray.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,75 +16,79 @@ #include "common.h" #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif #include using namespace std; void calc_1D_array_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - GLint maxTextureLayers = 16, maxTextureSize = 4096; - glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + // Need to limit array size according to GL device properties + GLint maxTextureLayers = 16, maxTextureSize = 4096; + glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); - sizes[i].depth = 1; - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].height = + random_in_range(2, min(maxTextureLayers, 1 << (i + 4)), seed); + sizes[i].depth = 1; + } } -int test_images_read_1Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int ) +int test_images_read_1Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_1D_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_1D_array_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_write_1Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_1Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_1D_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_1D_array_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_1Darray_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int ) +int test_images_1Darray_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_1D_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_1D_array_size_descriptors(sizes, nsizes); - return test_images_get_info_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_get_info_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, + nsizes); } \ No newline at end of file diff --git a/test_conformance/gl/test_images_2D.cpp b/test_conformance/gl/test_images_2D.cpp index fbff31c4..63ea31e2 100644 --- a/test_conformance/gl/test_images_2D.cpp +++ b/test_conformance/gl/test_images_2D.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,7 +16,7 @@ #include "testBase.h" #include "common.h" -#if defined( __APPLE__ ) +#if defined(__APPLE__) #include #else #include @@ -31,76 +31,77 @@ using namespace std; void calc_2D_test_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - // Need to limit texture size according to GL device properties - GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, size; - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); + // Need to limit array size according to GL device properties + // Need to limit texture size according to GL device properties + GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, size; + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); - size = min(maxTextureSize, maxTextureRectangleSize); + size = min(maxTextureSize, maxTextureRectangleSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].depth = 1; - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].height = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].depth = 1; + } } void calc_cube_test_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - // Need to limit texture size according to GL device properties - GLint maxQubeMapSize = 4096; - glGetIntegerv(GL_MAX_CUBE_MAP_TEXTURE_SIZE, &maxQubeMapSize); + // Need to limit array size according to GL device properties + // Need to limit texture size according to GL device properties + GLint maxQubeMapSize = 4096; + glGetIntegerv(GL_MAX_CUBE_MAP_TEXTURE_SIZE, &maxQubeMapSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = sizes[i].height = random_in_range( 2, min(maxQubeMapSize, 1<<(i+4)), seed ); - sizes[i].depth = 1; - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = sizes[i].height = + random_in_range(2, min(maxQubeMapSize, 1 << (i + 4)), seed); + sizes[i].depth = 1; + } } -int test_images_read_2D( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_read_2D(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_2D_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_2D_test_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_read_cube( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_read_cube(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { - GL_TEXTURE_CUBE_MAP_POSITIVE_X, - GL_TEXTURE_CUBE_MAP_POSITIVE_Y, - GL_TEXTURE_CUBE_MAP_POSITIVE_Z, - GL_TEXTURE_CUBE_MAP_NEGATIVE_X, - GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, - GL_TEXTURE_CUBE_MAP_NEGATIVE_Z }; + GLenum targets[] = { + GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y, + GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X, + GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_Z + }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_cube_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_cube_test_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } #pragma mark - @@ -108,81 +109,77 @@ int test_images_read_cube( cl_device_id device, cl_context context, #include "common.h" -int test_images_write( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_2D_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_2D_test_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_write_cube( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_cube(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { - GL_TEXTURE_CUBE_MAP_POSITIVE_X, - GL_TEXTURE_CUBE_MAP_POSITIVE_Y, - GL_TEXTURE_CUBE_MAP_POSITIVE_Z, - GL_TEXTURE_CUBE_MAP_NEGATIVE_X, - GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, - GL_TEXTURE_CUBE_MAP_NEGATIVE_Z - }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { + GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y, + GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X, + GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_Z + }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_cube_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_cube_test_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } #pragma mark - #pragma mark _2D get info tests -int test_images_2D_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_2D_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_2D_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_2D_test_size_descriptors(sizes, nsizes); - return test_images_get_info_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_get_info_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, + nsizes); } -int test_images_cube_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_cube_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { GLenum targets[] = { - GL_TEXTURE_CUBE_MAP_POSITIVE_X, - GL_TEXTURE_CUBE_MAP_POSITIVE_Y, - GL_TEXTURE_CUBE_MAP_POSITIVE_Z, - GL_TEXTURE_CUBE_MAP_NEGATIVE_X, - GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, - GL_TEXTURE_CUBE_MAP_NEGATIVE_Z - }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y, + GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X, + GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_Z + }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_cube_test_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_cube_test_size_descriptors(sizes, nsizes); - return test_images_get_info_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_get_info_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, + nsizes); } diff --git a/test_conformance/gl/test_images_2Darray.cpp b/test_conformance/gl/test_images_2Darray.cpp index bb7095d1..20ca7144 100644 --- a/test_conformance/gl/test_images_2Darray.cpp +++ b/test_conformance/gl/test_images_2Darray.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,11 +16,11 @@ #include "common.h" #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif #include @@ -28,66 +28,71 @@ using namespace std; void calc_2D_array_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - GLint maxTextureLayers = 16, maxTextureSize = 4096; - glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + // Need to limit array size according to GL device properties + GLint maxTextureLayers = 16, maxTextureSize = 4096; + glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].height = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].depth = + random_in_range(2, min(maxTextureLayers, 1 << (i + 4)), seed); + } } -int test_images_read_2Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int ) +int test_images_read_2Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_2D_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_2D_array_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_write_2Darray( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_2Darray(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - // FIXME: Query for 2D image array write support. + // FIXME: Query for 2D image array write support. - GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_2D_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_2D_array_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_2Darray_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int ) +int test_images_2Darray_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int) { - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_2D_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_2D_array_size_descriptors(sizes, nsizes); - return test_images_get_info_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_get_info_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, + nsizes); } \ No newline at end of file diff --git a/test_conformance/gl/test_images_3D.cpp b/test_conformance/gl/test_images_3D.cpp index 8abaa096..220cd0ca 100644 --- a/test_conformance/gl/test_images_3D.cpp +++ b/test_conformance/gl/test_images_3D.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,11 +16,11 @@ #include "testBase.h" #include "common.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif #include @@ -31,77 +31,85 @@ using namespace std; void calc_3D_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - GLint maxTextureSize = 2048; - glGetIntegerv(GL_MAX_3D_TEXTURE_SIZE, &maxTextureSize); + // Need to limit array size according to GL device properties + GLint maxTextureSize = 2048; + glGetIntegerv(GL_MAX_3D_TEXTURE_SIZE, &maxTextureSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].depth = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].height = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].depth = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + } } -int test_images_read_3D( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ) +int test_images_read_3D(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { GLenum targets[] = { GL_TEXTURE_3D }; - size_t ntargets = 1; + size_t ntargets = 1; - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_3D_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_3D_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } #pragma mark - #pragma marm _3D write test -int test_images_write_3D( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_3D(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - // TODO: Perhaps the expected behavior is to FAIL if 3D images are - // unsupported? + // TODO: Perhaps the expected behavior is to FAIL if 3D images are + // unsupported? - if (!is_extension_available(device, "cl_khr_3d_image_writes")) { - log_info("This device does not support 3D image writes. Skipping test.\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_3d_image_writes")) + { + log_info( + "This device does not support 3D image writes. Skipping test.\n"); + return 0; + } - GLenum targets[] = { GL_TEXTURE_3D }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + GLenum targets[] = { GL_TEXTURE_3D }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_3D_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_3D_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); } #pragma mark - #pragma mark _3D get info test -int test_images_3D_getinfo( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_3D_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - GLenum targets[] = { GL_TEXTURE_3D }; - size_t ntargets = 1; + GLenum targets[] = { GL_TEXTURE_3D }; + size_t ntargets = 1; - size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); + size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_3D_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_3D_size_descriptors(sizes, nsizes); return test_images_get_info_common(device, context, queue, common_formats, - nformats, targets, ntargets, sizes, nsizes); + nformats, targets, ntargets, sizes, + nsizes); } diff --git a/test_conformance/gl/test_images_depth.cpp b/test_conformance/gl/test_images_depth.cpp index f6cded47..05265cc6 100644 --- a/test_conformance/gl/test_images_depth.cpp +++ b/test_conformance/gl/test_images_depth.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,7 +16,7 @@ #include "testBase.h" #include "common.h" -#if defined( __APPLE__ ) +#if defined(__APPLE__) #include #else #include @@ -32,129 +32,140 @@ using namespace std; void calc_depth_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit texture size according to GL device properties - GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, size; - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); + // Need to limit texture size according to GL device properties + GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, size; + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); - size = min(maxTextureSize, maxTextureRectangleSize); + size = min(maxTextureSize, maxTextureRectangleSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].depth = 1; - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].height = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].depth = 1; + } } void calc_depth_array_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit texture size according to GL device properties - GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, maxTextureLayers = 16, size; - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); - glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); + // Need to limit texture size according to GL device properties + GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, + maxTextureLayers = 16, size; + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); + glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); - size = min(maxTextureSize, maxTextureRectangleSize); + size = min(maxTextureSize, maxTextureRectangleSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(size, 1<<(i+4)), seed ); - sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].height = random_in_range(2, min(size, 1 << (i + 4)), seed); + sizes[i].depth = + random_in_range(2, min(maxTextureLayers, 1 << (i + 4)), seed); + } } -int test_images_read_2D_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_read_2D_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - if (!is_extension_available(device, "cl_khr_gl_depth_images")) { - log_info("Test not run because 'cl_khr_gl_depth_images' extension is not supported by the tested device\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_gl_depth_images")) + { + log_info("Test not run because 'cl_khr_gl_depth_images' extension is " + "not supported by the tested device\n"); + return 0; + } - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_depth_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_depth_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, depth_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, depth_formats, + nformats, targets, ntargets, sizes, nsizes); } #pragma mark - #pragma mark _2D depth write tests -int test_images_write_2D_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_2D_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - if (!is_extension_available(device, "cl_khr_gl_depth_images")) { - log_info("Test not run because 'cl_khr_gl_depth_images' extension is not supported by the tested device\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_gl_depth_images")) + { + log_info("Test not run because 'cl_khr_gl_depth_images' extension is " + "not supported by the tested device\n"); + return 0; + } - GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_depth_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_depth_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, depth_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, depth_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_read_2Darray_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int ) +int test_images_read_2Darray_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int) { - if (!is_extension_available(device, "cl_khr_gl_depth_images")) { - log_info("Test not run because 'cl_khr_gl_depth_images' extension is not supported by the tested device\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_gl_depth_images")) + { + log_info("Test not run because 'cl_khr_gl_depth_images' extension is " + "not supported by the tested device\n"); + return 0; + } - size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_depth_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_depth_array_size_descriptors(sizes, nsizes); - return test_images_read_common(device, context, queue, depth_formats, - nformats, targets, ntargets, sizes, nsizes); + return test_images_read_common(device, context, queue, depth_formats, + nformats, targets, ntargets, sizes, nsizes); } -int test_images_write_2Darray_depth( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_write_2Darray_depth(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - if (!is_extension_available(device, "cl_khr_gl_depth_images")) { - log_info("Test not run because 'cl_khr_gl_depth_images' extension is not supported by the tested device\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_gl_depth_images")) + { + log_info("Test not run because 'cl_khr_gl_depth_images' extension is " + "not supported by the tested device\n"); + return 0; + } - // FIXME: Query for 2D image array write support. + // FIXME: Query for 2D image array write support. - GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); - size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + GLenum targets[] = { GL_TEXTURE_2D_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); + size_t nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - const size_t nsizes = 6; - sizevec_t sizes[nsizes]; - calc_depth_array_size_descriptors(sizes, nsizes); + const size_t nsizes = 6; + sizevec_t sizes[nsizes]; + calc_depth_array_size_descriptors(sizes, nsizes); - return test_images_write_common( device, context, queue, depth_formats, - nformats, targets, ntargets, sizes, nsizes ); + return test_images_write_common(device, context, queue, depth_formats, + nformats, targets, ntargets, sizes, nsizes); } - diff --git a/test_conformance/gl/test_images_getinfo_common.cpp b/test_conformance/gl/test_images_getinfo_common.cpp index 2322c269..836eb067 100644 --- a/test_conformance/gl/test_images_getinfo_common.cpp +++ b/test_conformance/gl/test_images_getinfo_common.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,74 +16,71 @@ #include "testBase.h" #include "common.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif -extern int supportsHalf(cl_context context, bool* supports_half); +extern int supportsHalf(cl_context context, bool *supports_half); -static int test_image_info( cl_context context, cl_command_queue queue, - GLenum glTarget, GLuint glTexture, size_t imageWidth, size_t imageHeight, - size_t imageDepth, cl_image_format *outFormat, ExplicitType *outType, - void **outResultBuffer ) +static int test_image_info(cl_context context, cl_command_queue queue, + GLenum glTarget, GLuint glTexture, size_t imageWidth, + size_t imageHeight, size_t imageDepth, + cl_image_format *outFormat, ExplicitType *outType, + void **outResultBuffer) { - clMemWrapper streams[ 2 ]; + clMemWrapper streams[2]; - int error; + int error; - // Create a CL image from the supplied GL texture - streams[ 0 ] = (*clCreateFromGLTexture_ptr)( context, CL_MEM_READ_ONLY, - glTarget, 0, glTexture, &error ); - if( error != CL_SUCCESS ) - { - print_error( error, "Unable to create CL image from GL texture" ); - GLint fmt; - glGetTexLevelParameteriv( glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); - log_error( " Supplied GL texture was format %s\n", GetGLFormatName( fmt ) ); - return error; - } + // Create a CL image from the supplied GL texture + streams[0] = (*clCreateFromGLTexture_ptr)(context, CL_MEM_READ_ONLY, + glTarget, 0, glTexture, &error); + if (error != CL_SUCCESS) + { + print_error(error, "Unable to create CL image from GL texture"); + GLint fmt; + glGetTexLevelParameteriv(glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt); + log_error(" Supplied GL texture was format %s\n", + GetGLFormatName(fmt)); + return error; + } - // Determine data type and format that CL came up with - error = clGetImageInfo( streams[ 0 ], CL_IMAGE_FORMAT, - sizeof( cl_image_format ), outFormat, NULL ); - test_error( error, "Unable to get CL image format" ); + // Determine data type and format that CL came up with + error = clGetImageInfo(streams[0], CL_IMAGE_FORMAT, sizeof(cl_image_format), + outFormat, NULL); + test_error(error, "Unable to get CL image format"); - cl_gl_object_type object_type; - switch (glTarget) { - case GL_TEXTURE_1D: - object_type = CL_GL_OBJECT_TEXTURE1D; - break; - case GL_TEXTURE_BUFFER: - object_type = CL_GL_OBJECT_TEXTURE_BUFFER; - break; - case GL_TEXTURE_1D_ARRAY: - object_type = CL_GL_OBJECT_TEXTURE1D_ARRAY; - break; - case GL_TEXTURE_2D: - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_CUBE_MAP_POSITIVE_X: - case GL_TEXTURE_CUBE_MAP_POSITIVE_Y: - case GL_TEXTURE_CUBE_MAP_POSITIVE_Z: - case GL_TEXTURE_CUBE_MAP_NEGATIVE_X: - case GL_TEXTURE_CUBE_MAP_NEGATIVE_Y: - case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: - object_type = CL_GL_OBJECT_TEXTURE2D; - break; - case GL_TEXTURE_2D_ARRAY: - object_type = CL_GL_OBJECT_TEXTURE2D_ARRAY; - break; - case GL_TEXTURE_3D: - object_type = CL_GL_OBJECT_TEXTURE3D; - break; - default: - log_error("Unsupported texture target."); - return 1; - } + cl_gl_object_type object_type; + switch (glTarget) + { + case GL_TEXTURE_1D: object_type = CL_GL_OBJECT_TEXTURE1D; break; + case GL_TEXTURE_BUFFER: + object_type = CL_GL_OBJECT_TEXTURE_BUFFER; + break; + case GL_TEXTURE_1D_ARRAY: + object_type = CL_GL_OBJECT_TEXTURE1D_ARRAY; + break; + case GL_TEXTURE_2D: + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_CUBE_MAP_POSITIVE_X: + case GL_TEXTURE_CUBE_MAP_POSITIVE_Y: + case GL_TEXTURE_CUBE_MAP_POSITIVE_Z: + case GL_TEXTURE_CUBE_MAP_NEGATIVE_X: + case GL_TEXTURE_CUBE_MAP_NEGATIVE_Y: + case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: + object_type = CL_GL_OBJECT_TEXTURE2D; + break; + case GL_TEXTURE_2D_ARRAY: + object_type = CL_GL_OBJECT_TEXTURE2D_ARRAY; + break; + case GL_TEXTURE_3D: object_type = CL_GL_OBJECT_TEXTURE3D; break; + default: log_error("Unsupported texture target."); return 1; + } - return CheckGLObjectInfo(streams[0], object_type, glTexture, glTarget, 0); + return CheckGLObjectInfo(streams[0], object_type, glTexture, glTarget, 0); } static int test_image_format_get_info(cl_context context, @@ -92,110 +89,119 @@ static int test_image_format_get_info(cl_context context, GLenum target, const format *fmt, MTdata data) { - int error = 0; + int error = 0; - // If we're testing a half float format, then we need to determine the - // rounding mode of this machine. Punt if we fail to do so. + // If we're testing a half float format, then we need to determine the + // rounding mode of this machine. Punt if we fail to do so. - if( fmt->type == kHalf ) - { - if( DetectFloatToHalfRoundingMode(queue) ) - return 0; - bool supports_half = false; - error = supportsHalf(context, &supports_half); - if( error != 0 ) - return error; - if (!supports_half) return 0; - } - - size_t w = width, h = height, d = depth; - - // Unpack the format and use it, along with the target, to create an - // appropriate GL texture. - - GLenum gl_fmt = fmt->formattype; - GLenum gl_internal_fmt = fmt->internal; - GLenum gl_type = fmt->datatype; - ExplicitType type = fmt->type; - - glTextureWrapper texture; - glBufferWrapper glbuf; - - // If we're testing a half float format, then we need to determine the - // rounding mode of this machine. Punt if we fail to do so. - - if( fmt->type == kHalf ) - if( DetectFloatToHalfRoundingMode(queue) ) - return 1; - - // Use the correct texture creation function depending on the target, and - // adjust width, height, depth as appropriate so subsequent size calculations - // succeed. - - switch (target) { - case GL_TEXTURE_1D: - h = 1; d = 1; - CreateGLTexture1D( width, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, false, data ); - break; - case GL_TEXTURE_BUFFER: - h = 1; d = 1; - CreateGLTextureBuffer( width, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &glbuf, &error, false, data ); - break; - case GL_TEXTURE_1D_ARRAY: - d = 1; - CreateGLTexture1DArray( width, height, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, false, data ); - break; - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_2D: - case GL_TEXTURE_CUBE_MAP_POSITIVE_X: - case GL_TEXTURE_CUBE_MAP_POSITIVE_Y: - case GL_TEXTURE_CUBE_MAP_POSITIVE_Z: - case GL_TEXTURE_CUBE_MAP_NEGATIVE_X: - case GL_TEXTURE_CUBE_MAP_NEGATIVE_Y: - case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: - d = 1; - CreateGLTexture2D( width, height, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, false, data ); - break; - case GL_TEXTURE_2D_ARRAY: - CreateGLTexture2DArray( width, height, depth, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, false, data ); - break; - case GL_TEXTURE_3D: - d = 1; - CreateGLTexture3D( width, height, depth, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, data, false ); - break; - default: - log_error("Unsupported texture target.\n"); - return 1; - } - - if ( error == -2 ) { - log_info("OpenGL texture couldn't be created, because a texture is too big. Skipping test.\n"); - return 0; - } - - if ( error != 0 ) { - if ((gl_fmt == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())) { - log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " - "Skipping test.\n"); - return 0; - } else { - return error; + if (fmt->type == kHalf) + { + if (DetectFloatToHalfRoundingMode(queue)) return 0; + bool supports_half = false; + error = supportsHalf(context, &supports_half); + if (error != 0) return error; + if (!supports_half) return 0; } - } - cl_image_format clFormat; - ExplicitType actualType; - char *outBuffer; + size_t w = width, h = height, d = depth; - // Perform the info check: - return test_image_info( context, queue, target, texture, w, h, d, &clFormat, - &actualType, (void **)&outBuffer ); + // Unpack the format and use it, along with the target, to create an + // appropriate GL texture. + + GLenum gl_fmt = fmt->formattype; + GLenum gl_internal_fmt = fmt->internal; + GLenum gl_type = fmt->datatype; + ExplicitType type = fmt->type; + + glTextureWrapper texture; + glBufferWrapper glbuf; + + // If we're testing a half float format, then we need to determine the + // rounding mode of this machine. Punt if we fail to do so. + + if (fmt->type == kHalf) + if (DetectFloatToHalfRoundingMode(queue)) return 1; + + // Use the correct texture creation function depending on the target, and + // adjust width, height, depth as appropriate so subsequent size + // calculations succeed. + + switch (target) + { + case GL_TEXTURE_1D: + h = 1; + d = 1; + CreateGLTexture1D(width, target, gl_fmt, gl_internal_fmt, gl_type, + type, &texture, &error, false, data); + break; + case GL_TEXTURE_BUFFER: + h = 1; + d = 1; + CreateGLTextureBuffer(width, target, gl_fmt, gl_internal_fmt, + gl_type, type, &texture, &glbuf, &error, + false, data); + break; + case GL_TEXTURE_1D_ARRAY: + d = 1; + CreateGLTexture1DArray(width, height, target, gl_fmt, + gl_internal_fmt, gl_type, type, &texture, + &error, false, data); + break; + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_2D: + case GL_TEXTURE_CUBE_MAP_POSITIVE_X: + case GL_TEXTURE_CUBE_MAP_POSITIVE_Y: + case GL_TEXTURE_CUBE_MAP_POSITIVE_Z: + case GL_TEXTURE_CUBE_MAP_NEGATIVE_X: + case GL_TEXTURE_CUBE_MAP_NEGATIVE_Y: + case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: + d = 1; + CreateGLTexture2D(width, height, target, gl_fmt, gl_internal_fmt, + gl_type, type, &texture, &error, false, data); + break; + case GL_TEXTURE_2D_ARRAY: + CreateGLTexture2DArray(width, height, depth, target, gl_fmt, + gl_internal_fmt, gl_type, type, &texture, + &error, false, data); + break; + case GL_TEXTURE_3D: + d = 1; + CreateGLTexture3D(width, height, depth, target, gl_fmt, + gl_internal_fmt, gl_type, type, &texture, &error, + data, false); + break; + default: log_error("Unsupported texture target.\n"); return 1; + } + + if (error == -2) + { + log_info("OpenGL texture couldn't be created, because a texture is too " + "big. Skipping test.\n"); + return 0; + } + + if (error != 0) + { + if ((gl_fmt == GL_RGBA_INTEGER_EXT) + && (!CheckGLIntegerExtensionSupport())) + { + log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " + "Skipping test.\n"); + return 0; + } + else + { + return error; + } + } + + cl_image_format clFormat; + ExplicitType actualType; + char *outBuffer; + + // Perform the info check: + return test_image_info(context, queue, target, texture, w, h, d, &clFormat, + &actualType, (void **)&outBuffer); } int test_images_get_info_common(cl_device_id device, cl_context context, @@ -204,60 +210,65 @@ int test_images_get_info_common(cl_device_id device, cl_context context, size_t ntargets, sizevec_t *sizes, size_t nsizes) { - int error = 0; - RandomSeed seed(gRandomSeed); + int error = 0; + RandomSeed seed(gRandomSeed); - // First, ensure this device supports images. + // First, ensure this device supports images. - if (checkForImageSupport(device)) { - log_info("Device does not support images. Skipping test.\n"); - return 0; - } - - size_t fidx, tidx, sidx; - - // Test each format on every target, every size. - - for ( fidx = 0; fidx < nformats; fidx++ ) { - for ( tidx = 0; tidx < ntargets; tidx++ ) { - - if ( formats[ fidx ].datatype == GL_UNSIGNED_INT_2_10_10_10_REV ) - { - // Check if the RGB 101010 format is supported - if ( is_rgb_101010_supported( context, targets[ tidx ] ) == 0 ) - break; // skip - } - - log_info( "Testing image info for GL format %s : %s : %s : %s\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - - for ( sidx = 0; sidx < nsizes; sidx++ ) { - - // Test this format + size: - - if ( test_image_format_get_info(context, queue, - sizes[sidx].width, sizes[sidx].height, sizes[sidx].depth, - targets[tidx], &formats[fidx], seed) ) - { - // We land here in the event of test failure. - - log_error( "ERROR: Image info test failed for %s : %s : %s : %s\n\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - error++; - - // Skip the other sizes for this format. - - break; - } - } + if (checkForImageSupport(device)) + { + log_info("Device does not support images. Skipping test.\n"); + return 0; } - } - return error; + size_t fidx, tidx, sidx; + + // Test each format on every target, every size. + + for (fidx = 0; fidx < nformats; fidx++) + { + for (tidx = 0; tidx < ntargets; tidx++) + { + + if (formats[fidx].datatype == GL_UNSIGNED_INT_2_10_10_10_REV) + { + // Check if the RGB 101010 format is supported + if (is_rgb_101010_supported(context, targets[tidx]) == 0) + break; // skip + } + + log_info("Testing image info for GL format %s : %s : %s : %s\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + + for (sidx = 0; sidx < nsizes; sidx++) + { + + // Test this format + size: + + if (test_image_format_get_info( + context, queue, sizes[sidx].width, sizes[sidx].height, + sizes[sidx].depth, targets[tidx], &formats[fidx], seed)) + { + // We land here in the event of test failure. + + log_error("ERROR: Image info test failed for %s : %s : %s " + ": %s\n\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + error++; + + // Skip the other sizes for this format. + + break; + } + } + } + } + + return error; } diff --git a/test_conformance/gl/test_images_multisample.cpp b/test_conformance/gl/test_images_multisample.cpp index 99f9ff2e..bfb04ab6 100644 --- a/test_conformance/gl/test_images_multisample.cpp +++ b/test_conformance/gl/test_images_multisample.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,11 +16,11 @@ #include "common.h" #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif #include @@ -29,90 +29,109 @@ using namespace std; void calc_2D_multisample_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit texture size according to GL device properties - GLint maxTextureSize = 4096; - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + // Need to limit texture size according to GL device properties + GLint maxTextureSize = 4096; + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].depth = 1; - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].height = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].depth = 1; + } } void calc_2D_array_multisample_size_descriptors(sizevec_t* sizes, size_t nsizes) { - // Need to limit array size according to GL device properties - GLint maxTextureLayers = 16, maxTextureSize = 4096; - glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); - glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); + // Need to limit array size according to GL device properties + GLint maxTextureLayers = 16, maxTextureSize = 4096; + glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); + glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Generate some random sizes (within reasonable ranges) - for (size_t i = 0; i < nsizes; i++) { - sizes[i].width = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].height = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); - sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); - } + // Generate some random sizes (within reasonable ranges) + for (size_t i = 0; i < nsizes; i++) + { + sizes[i].width = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].height = + random_in_range(2, min(maxTextureSize, 1 << (i + 4)), seed); + sizes[i].depth = + random_in_range(2, min(maxTextureLayers, 1 << (i + 4)), seed); + } } -int test_images_read_2D_multisample( cl_device_id device, cl_context context, - cl_command_queue queue, int numElements ) +int test_images_read_2D_multisample(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { - if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) { - log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is not supported by the tested device\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) + { + log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is " + "not supported by the tested device\n"); + return 0; + } - glEnable(GL_MULTISAMPLE); + glEnable(GL_MULTISAMPLE); - const size_t nsizes = 8; - sizevec_t sizes[nsizes]; - calc_2D_multisample_size_descriptors(sizes, nsizes); + const size_t nsizes = 8; + sizevec_t sizes[nsizes]; + calc_2D_multisample_size_descriptors(sizes, nsizes); - size_t nformats; + size_t nformats; - GLenum targets[] = { GL_TEXTURE_2D_MULTISAMPLE }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D_MULTISAMPLE }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - nformats = sizeof(common_formats) / sizeof(common_formats[0]); - int ret_common = test_images_read_common(device, context, queue, common_formats, nformats, targets, ntargets, sizes, nsizes); + nformats = sizeof(common_formats) / sizeof(common_formats[0]); + int ret_common = + test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); - nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - int ret_depth = test_images_read_common(device, context, queue, depth_formats, nformats, targets, ntargets, sizes, nsizes); + nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + int ret_depth = + test_images_read_common(device, context, queue, depth_formats, nformats, + targets, ntargets, sizes, nsizes); - return (ret_common) ? ret_common : ret_depth; + return (ret_common) ? ret_common : ret_depth; } -int test_images_read_2Darray_multisample( cl_device_id device, cl_context context, - cl_command_queue queue, int ) +int test_images_read_2Darray_multisample(cl_device_id device, + cl_context context, + cl_command_queue queue, int) { - if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) { - log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is not supported by the tested device\n"); - return 0; - } + if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) + { + log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is " + "not supported by the tested device\n"); + return 0; + } - glEnable(GL_MULTISAMPLE); + glEnable(GL_MULTISAMPLE); - const size_t nsizes = 4; - sizevec_t sizes[nsizes]; - calc_2D_array_multisample_size_descriptors(sizes, nsizes); + const size_t nsizes = 4; + sizevec_t sizes[nsizes]; + calc_2D_array_multisample_size_descriptors(sizes, nsizes); - size_t nformats; + size_t nformats; - GLenum targets[] = { GL_TEXTURE_2D_MULTISAMPLE_ARRAY }; - size_t ntargets = sizeof(targets) / sizeof(targets[0]); + GLenum targets[] = { GL_TEXTURE_2D_MULTISAMPLE_ARRAY }; + size_t ntargets = sizeof(targets) / sizeof(targets[0]); - nformats = sizeof(common_formats) / sizeof(common_formats[0]); - int ret_common = test_images_read_common(device, context, queue, common_formats, nformats, targets, ntargets, sizes, nsizes); + nformats = sizeof(common_formats) / sizeof(common_formats[0]); + int ret_common = + test_images_read_common(device, context, queue, common_formats, + nformats, targets, ntargets, sizes, nsizes); - nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); - int ret_depth = test_images_read_common(device, context, queue, depth_formats, nformats, targets, ntargets, sizes, nsizes); + nformats = sizeof(depth_formats) / sizeof(depth_formats[0]); + int ret_depth = + test_images_read_common(device, context, queue, depth_formats, nformats, + targets, ntargets, sizes, nsizes); - return (ret_common) ? ret_common : ret_depth; + return (ret_common) ? ret_common : ret_depth; } - diff --git a/test_conformance/gl/test_images_read_common.cpp b/test_conformance/gl/test_images_read_common.cpp index fe2a529b..dc6ad9fa 100644 --- a/test_conformance/gl/test_images_read_common.cpp +++ b/test_conformance/gl/test_images_read_common.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,17 +16,18 @@ #include "common.h" #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif -extern int supportsHalf(cl_context context, bool* supports_half); -extern int supportsMsaa(cl_context context, bool* supports_msaa); -extern int supportsDepth(cl_context context, bool* supports_depth); +extern int supportsHalf(cl_context context, bool *supports_half); +extern int supportsMsaa(cl_context context, bool *supports_msaa); +extern int supportsDepth(cl_context context, bool *supports_depth); +// clang-format off static const char *kernelpattern_image_read_1d = "__kernel void sample_test( read_only image1d_t source, sampler_t sampler, __global %s4 *results )\n" "{\n" @@ -167,481 +168,536 @@ static const char *kernelpattern_image_multisample_read_2darray_depth = " results[ offset ] = read_image%s( source, (int4)( tidX, tidY, tidZ, 1 ), sample );\n" " }\n" "}\n"; +// clang-format on -static const char* get_appropriate_kernel_for_target(GLenum target, cl_channel_order channel_order) { - - switch (get_base_gl_target(target)) { - case GL_TEXTURE_1D: - return kernelpattern_image_read_1d; - case GL_TEXTURE_BUFFER: - return kernelpattern_image_read_1d_buffer; - case GL_TEXTURE_1D_ARRAY: - return kernelpattern_image_read_1darray; - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_2D: - case GL_COLOR_ATTACHMENT0: - case GL_RENDERBUFFER: - case GL_TEXTURE_CUBE_MAP: -#ifdef GL_VERSION_3_2 - if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) - return kernelpattern_image_read_2d_depth; -#endif - return kernelpattern_image_read_2d; - case GL_TEXTURE_2D_ARRAY: -#ifdef GL_VERSION_3_2 - if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) - return kernelpattern_image_read_2darray_depth; -#endif - return kernelpattern_image_read_2darray; - case GL_TEXTURE_3D: - return kernelpattern_image_read_3d; - case GL_TEXTURE_2D_MULTISAMPLE: -#ifdef GL_VERSION_3_2 - if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) - return kernelpattern_image_multisample_read_2d_depth; -#endif - return kernelpattern_image_multisample_read_2d; - break; - case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: -#ifdef GL_VERSION_3_2 - if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) - return kernelpattern_image_multisample_read_2darray_depth; -#endif - return kernelpattern_image_multisample_read_2darray; - break; - default: - log_error("Unsupported texture target (%s); cannot determine " - "appropriate kernel.", GetGLTargetName(target)); - return NULL; - } -} - -int test_cl_image_read( cl_context context, cl_command_queue queue, - GLenum gl_target, cl_mem image, size_t width, size_t height, size_t depth, size_t sampleNum, - cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ) +static const char * +get_appropriate_kernel_for_target(GLenum target, cl_channel_order channel_order) { - clProgramWrapper program; - clKernelWrapper kernel; - clMemWrapper streams[ 2 ]; - int error; - char kernelSource[2048]; - char *programPtr; - - // Use the image created from the GL texture. - streams[ 0 ] = image; - - // Determine data type and format that CL came up with - error = clGetImageInfo( streams[ 0 ], CL_IMAGE_FORMAT, sizeof( cl_image_format ), outFormat, NULL ); - test_error( error, "Unable to get CL image format" ); - - // Determine the number of samples - cl_uint samples = 0; - error = clGetImageInfo( streams[ 0 ], CL_IMAGE_NUM_SAMPLES, sizeof( samples ), &samples, NULL ); - test_error( error, "Unable to get CL_IMAGE_NUM_SAMPLES" ); - - // Create the source - *outType = get_read_kernel_type( outFormat ); - size_t channelSize = get_explicit_type_size( *outType ); - - const char* source = get_appropriate_kernel_for_target(gl_target, outFormat->image_channel_order); - - sprintf( kernelSource, source, get_explicit_type_name( *outType ), - get_kernel_suffix( outFormat ) ); - - programPtr = kernelSource; - if( create_single_kernel_helper( context, &program, &kernel, 1, - (const char **)&programPtr, "sample_test", "" ) ) - { - return -1; - } - - // Create a vanilla output buffer - cl_device_id device; - error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device), &device, NULL); - test_error( error, "Unable to get queue device" ); - - cl_ulong maxAllocSize = 0; - error = clGetDeviceInfo( device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof( maxAllocSize ), &maxAllocSize, NULL ); - test_error( error, "Unable to get CL_DEVICE_MAX_MEM_ALLOC_SIZE" ); - - size_t buffer_bytes = channelSize * get_channel_order_channel_count(outFormat->image_channel_order) * width * height * depth * sampleNum; - if (buffer_bytes > maxAllocSize) { - log_info("Output buffer size %d is too large for device (max alloc size %d) Skipping...\n", - (int)buffer_bytes, (int)maxAllocSize); - return 1; - } - - streams[ 1 ] = clCreateBuffer( context, CL_MEM_READ_WRITE, buffer_bytes, NULL, &error ); - test_error( error, "Unable to create output buffer" ); - - /* Assign streams and execute */ - clSamplerWrapper sampler = clCreateSampler( context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error ); - test_error( error, "Unable to create sampler" ); - - error = clSetKernelArg( kernel, 0, sizeof( streams[ 0 ] ), &streams[ 0 ] ); - test_error( error, "Unable to set kernel arguments" ); - error = clSetKernelArg( kernel, 1, sizeof( sampler ), &sampler ); - test_error( error, "Unable to set kernel arguments" ); - error = clSetKernelArg( kernel, 2, sizeof( streams[ 1 ] ), &streams[ 1 ] ); - test_error( error, "Unable to set kernel arguments" ); - - glFinish(); - - error = (*clEnqueueAcquireGLObjects_ptr)( queue, 1, &streams[ 0 ], 0, NULL, NULL); - test_error( error, "Unable to acquire GL obejcts"); - - // The ND range we use is a function of the dimensionality of the image. - size_t global_range[3] = { width, height, depth }; - size_t *local_range = NULL; - int ndim = 1; - - switch (get_base_gl_target(gl_target)) { - case GL_TEXTURE_1D: - case GL_TEXTURE_BUFFER: - ndim = 1; - break; - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_2D: - case GL_TEXTURE_1D_ARRAY: - case GL_COLOR_ATTACHMENT0: - case GL_RENDERBUFFER: - case GL_TEXTURE_CUBE_MAP: - ndim = 2; - break; - case GL_TEXTURE_3D: - case GL_TEXTURE_2D_ARRAY: + switch (get_base_gl_target(target)) + { + case GL_TEXTURE_1D: return kernelpattern_image_read_1d; + case GL_TEXTURE_BUFFER: return kernelpattern_image_read_1d_buffer; + case GL_TEXTURE_1D_ARRAY: return kernelpattern_image_read_1darray; + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_2D: + case GL_COLOR_ATTACHMENT0: + case GL_RENDERBUFFER: + case GL_TEXTURE_CUBE_MAP: #ifdef GL_VERSION_3_2 - case GL_TEXTURE_2D_MULTISAMPLE: - case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: - ndim = 3; - break; + if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) + return kernelpattern_image_read_2d_depth; #endif - default: - log_error("Test error: Unsupported texture target.\n"); - return 1; - } - - // 2D and 3D images have a special way to set the local size (legacy). - // Otherwise, we let CL select by leaving local_range as NULL. - - if (gl_target == GL_TEXTURE_2D) { - local_range = (size_t*)malloc(sizeof(size_t) * ndim); - get_max_common_2D_work_group_size( context, kernel, global_range, local_range ); - - } else if (gl_target == GL_TEXTURE_3D) { - local_range = (size_t*)malloc(sizeof(size_t) * ndim); - get_max_common_3D_work_group_size( context, kernel, global_range, local_range ); - } - - error = clEnqueueNDRangeKernel( queue, kernel, ndim, NULL, global_range, - local_range, 0, NULL, NULL ); - test_error( error, "Unable to execute test kernel" ); - - error = (*clEnqueueReleaseGLObjects_ptr)( queue, 1, &streams[ 0 ], - 0, NULL, NULL ); - test_error(error, "clEnqueueReleaseGLObjects failed"); - - // Read results from the CL buffer - *outResultBuffer = (void *)( new char[ channelSize * get_channel_order_channel_count(outFormat->image_channel_order) * width * height * depth * sampleNum] ); - error = clEnqueueReadBuffer( queue, streams[ 1 ], CL_TRUE, 0, - channelSize * get_channel_order_channel_count(outFormat->image_channel_order) * width * height * depth * sampleNum, *outResultBuffer, 0, NULL, NULL ); - test_error( error, "Unable to read output CL buffer!" ); - - // free the ranges - if (local_range) free(local_range); - - return 0; -} - -static int test_image_read( cl_context context, cl_command_queue queue, - GLenum target, GLuint globj, size_t width, size_t height, size_t depth, size_t sampleNum, - cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ) -{ - int error; - - // Create a CL image from the supplied GL texture or renderbuffer. - cl_mem image; - if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { - image = (*clCreateFromGLRenderbuffer_ptr)( context, CL_MEM_READ_ONLY, globj, &error ); - } else { - image = (*clCreateFromGLTexture_ptr)( context, CL_MEM_READ_ONLY, - target, 0, globj, &error ); - } - - if( error != CL_SUCCESS ) { - if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { - print_error( error, "Unable to create CL image from GL renderbuffer" ); - } else { - print_error( error, "Unable to create CL image from GL texture" ); - GLint fmt; - glGetTexLevelParameteriv( target, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); - log_error( " Supplied GL texture was base format %s and internal " - "format %s\n", GetGLBaseFormatName( fmt ), GetGLFormatName( fmt ) ); + return kernelpattern_image_read_2d; + case GL_TEXTURE_2D_ARRAY: +#ifdef GL_VERSION_3_2 + if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) + return kernelpattern_image_read_2darray_depth; +#endif + return kernelpattern_image_read_2darray; + case GL_TEXTURE_3D: return kernelpattern_image_read_3d; + case GL_TEXTURE_2D_MULTISAMPLE: +#ifdef GL_VERSION_3_2 + if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) + return kernelpattern_image_multisample_read_2d_depth; +#endif + return kernelpattern_image_multisample_read_2d; + break; + case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: +#ifdef GL_VERSION_3_2 + if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) + return kernelpattern_image_multisample_read_2darray_depth; +#endif + return kernelpattern_image_multisample_read_2darray; + break; + default: + log_error("Unsupported texture target (%s); cannot determine " + "appropriate kernel.", + GetGLTargetName(target)); + return NULL; } - return error; - } +} - return test_cl_image_read( context, queue, target, image, - width, height, depth, sampleNum, outFormat, outType, outResultBuffer ); +int test_cl_image_read(cl_context context, cl_command_queue queue, + GLenum gl_target, cl_mem image, size_t width, + size_t height, size_t depth, size_t sampleNum, + cl_image_format *outFormat, ExplicitType *outType, + void **outResultBuffer) +{ + clProgramWrapper program; + clKernelWrapper kernel; + clMemWrapper streams[2]; + + int error; + char kernelSource[2048]; + char *programPtr; + + // Use the image created from the GL texture. + streams[0] = image; + + // Determine data type and format that CL came up with + error = clGetImageInfo(streams[0], CL_IMAGE_FORMAT, sizeof(cl_image_format), + outFormat, NULL); + test_error(error, "Unable to get CL image format"); + + // Determine the number of samples + cl_uint samples = 0; + error = clGetImageInfo(streams[0], CL_IMAGE_NUM_SAMPLES, sizeof(samples), + &samples, NULL); + test_error(error, "Unable to get CL_IMAGE_NUM_SAMPLES"); + + // Create the source + *outType = get_read_kernel_type(outFormat); + size_t channelSize = get_explicit_type_size(*outType); + + const char *source = get_appropriate_kernel_for_target( + gl_target, outFormat->image_channel_order); + + sprintf(kernelSource, source, get_explicit_type_name(*outType), + get_kernel_suffix(outFormat)); + + programPtr = kernelSource; + if (create_single_kernel_helper(context, &program, &kernel, 1, + (const char **)&programPtr, "sample_test", + "")) + { + return -1; + } + + // Create a vanilla output buffer + cl_device_id device; + error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device), + &device, NULL); + test_error(error, "Unable to get queue device"); + + cl_ulong maxAllocSize = 0; + error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, + sizeof(maxAllocSize), &maxAllocSize, NULL); + test_error(error, "Unable to get CL_DEVICE_MAX_MEM_ALLOC_SIZE"); + + size_t buffer_bytes = channelSize + * get_channel_order_channel_count(outFormat->image_channel_order) + * width * height * depth * sampleNum; + if (buffer_bytes > maxAllocSize) + { + log_info("Output buffer size %d is too large for device (max alloc " + "size %d) Skipping...\n", + (int)buffer_bytes, (int)maxAllocSize); + return 1; + } + + streams[1] = + clCreateBuffer(context, CL_MEM_READ_WRITE, buffer_bytes, NULL, &error); + test_error(error, "Unable to create output buffer"); + + /* Assign streams and execute */ + clSamplerWrapper sampler = clCreateSampler( + context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error); + test_error(error, "Unable to create sampler"); + + error = clSetKernelArg(kernel, 0, sizeof(streams[0]), &streams[0]); + test_error(error, "Unable to set kernel arguments"); + error = clSetKernelArg(kernel, 1, sizeof(sampler), &sampler); + test_error(error, "Unable to set kernel arguments"); + error = clSetKernelArg(kernel, 2, sizeof(streams[1]), &streams[1]); + test_error(error, "Unable to set kernel arguments"); + + glFinish(); + + error = + (*clEnqueueAcquireGLObjects_ptr)(queue, 1, &streams[0], 0, NULL, NULL); + test_error(error, "Unable to acquire GL obejcts"); + + // The ND range we use is a function of the dimensionality of the image. + size_t global_range[3] = { width, height, depth }; + size_t *local_range = NULL; + int ndim = 1; + + switch (get_base_gl_target(gl_target)) + { + case GL_TEXTURE_1D: + case GL_TEXTURE_BUFFER: ndim = 1; break; + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_2D: + case GL_TEXTURE_1D_ARRAY: + case GL_COLOR_ATTACHMENT0: + case GL_RENDERBUFFER: + case GL_TEXTURE_CUBE_MAP: ndim = 2; break; + case GL_TEXTURE_3D: + case GL_TEXTURE_2D_ARRAY: +#ifdef GL_VERSION_3_2 + case GL_TEXTURE_2D_MULTISAMPLE: + case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: ndim = 3; break; +#endif + default: + log_error("Test error: Unsupported texture target.\n"); + return 1; + } + + // 2D and 3D images have a special way to set the local size (legacy). + // Otherwise, we let CL select by leaving local_range as NULL. + + if (gl_target == GL_TEXTURE_2D) + { + local_range = (size_t *)malloc(sizeof(size_t) * ndim); + get_max_common_2D_work_group_size(context, kernel, global_range, + local_range); + } + else if (gl_target == GL_TEXTURE_3D) + { + local_range = (size_t *)malloc(sizeof(size_t) * ndim); + get_max_common_3D_work_group_size(context, kernel, global_range, + local_range); + } + + error = clEnqueueNDRangeKernel(queue, kernel, ndim, NULL, global_range, + local_range, 0, NULL, NULL); + test_error(error, "Unable to execute test kernel"); + + error = + (*clEnqueueReleaseGLObjects_ptr)(queue, 1, &streams[0], 0, NULL, NULL); + test_error(error, "clEnqueueReleaseGLObjects failed"); + + // Read results from the CL buffer + *outResultBuffer = (void *)(new char[channelSize + * get_channel_order_channel_count( + outFormat->image_channel_order) + * width * height * depth * sampleNum]); + error = clEnqueueReadBuffer( + queue, streams[1], CL_TRUE, 0, + channelSize + * get_channel_order_channel_count(outFormat->image_channel_order) + * width * height * depth * sampleNum, + *outResultBuffer, 0, NULL, NULL); + test_error(error, "Unable to read output CL buffer!"); + + // free the ranges + if (local_range) free(local_range); + + return 0; +} + +static int test_image_read(cl_context context, cl_command_queue queue, + GLenum target, GLuint globj, size_t width, + size_t height, size_t depth, size_t sampleNum, + cl_image_format *outFormat, ExplicitType *outType, + void **outResultBuffer) +{ + int error; + + // Create a CL image from the supplied GL texture or renderbuffer. + cl_mem image; + if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) + { + image = (*clCreateFromGLRenderbuffer_ptr)(context, CL_MEM_READ_ONLY, + globj, &error); + } + else + { + image = (*clCreateFromGLTexture_ptr)(context, CL_MEM_READ_ONLY, target, + 0, globj, &error); + } + + if (error != CL_SUCCESS) + { + if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) + { + print_error(error, + "Unable to create CL image from GL renderbuffer"); + } + else + { + print_error(error, "Unable to create CL image from GL texture"); + GLint fmt; + glGetTexLevelParameteriv(target, 0, GL_TEXTURE_INTERNAL_FORMAT, + &fmt); + log_error(" Supplied GL texture was base format %s and internal " + "format %s\n", + GetGLBaseFormatName(fmt), GetGLFormatName(fmt)); + } + return error; + } + + return test_cl_image_read(context, queue, target, image, width, height, + depth, sampleNum, outFormat, outType, + outResultBuffer); } static int test_image_format_read(cl_context context, cl_command_queue queue, size_t width, size_t height, size_t depth, GLenum target, const format *fmt, MTdata data) { - int error = 0; + int error = 0; - // Determine the maximum number of supported samples - GLint samples = 1; - if (target == GL_TEXTURE_2D_MULTISAMPLE || target == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) - samples = get_gl_max_samples(target, fmt->internal); + // Determine the maximum number of supported samples + GLint samples = 1; + if (target == GL_TEXTURE_2D_MULTISAMPLE + || target == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) + samples = get_gl_max_samples(target, fmt->internal); - // If we're testing a half float format, then we need to determine the - // rounding mode of this machine. Punt if we fail to do so. + // If we're testing a half float format, then we need to determine the + // rounding mode of this machine. Punt if we fail to do so. - if( fmt->type == kHalf ) - { - if( DetectFloatToHalfRoundingMode(queue) ) - return 1; - bool supports_half = false; - error = supportsHalf(context, &supports_half); - if( error != 0 ) - return error; - if (!supports_half) return 0; - } + if (fmt->type == kHalf) + { + if (DetectFloatToHalfRoundingMode(queue)) return 1; + bool supports_half = false; + error = supportsHalf(context, &supports_half); + if (error != 0) return error; + if (!supports_half) return 0; + } #ifdef GL_VERSION_3_2 - if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE || - get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) + if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE + || get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) { bool supports_msaa; error = supportsMsaa(context, &supports_msaa); - if( error != 0 ) return error; + if (error != 0) return error; if (!supports_msaa) return 0; } - if (fmt->formattype == GL_DEPTH_COMPONENT || - fmt->formattype == GL_DEPTH_STENCIL) + if (fmt->formattype == GL_DEPTH_COMPONENT + || fmt->formattype == GL_DEPTH_STENCIL) { bool supports_depth; error = supportsDepth(context, &supports_depth); - if( error != 0 ) return error; + if (error != 0) return error; if (!supports_depth) return 0; } #endif - size_t w = width, h = height, d = depth; + size_t w = width, h = height, d = depth; - // Unpack the format and use it, along with the target, to create an - // appropriate GL texture. + // Unpack the format and use it, along with the target, to create an + // appropriate GL texture. - GLenum gl_fmt = fmt->formattype; - GLenum gl_internal_fmt = fmt->internal; - GLenum gl_type = fmt->datatype; - ExplicitType type = fmt->type; + GLenum gl_fmt = fmt->formattype; + GLenum gl_internal_fmt = fmt->internal; + GLenum gl_type = fmt->datatype; + ExplicitType type = fmt->type; - // Required for most of the texture-backed cases: - glTextureWrapper texture; + // Required for most of the texture-backed cases: + glTextureWrapper texture; - // Required for the special case of TextureBuffer textures: - glBufferWrapper glbuf; + // Required for the special case of TextureBuffer textures: + glBufferWrapper glbuf; - // And these are required for the case of Renderbuffer images: - glFramebufferWrapper glFramebuffer; - glRenderbufferWrapper glRenderbuffer; + // And these are required for the case of Renderbuffer images: + glFramebufferWrapper glFramebuffer; + glRenderbufferWrapper glRenderbuffer; - void* buffer = NULL; + void *buffer = NULL; - // Use the correct texture creation function depending on the target, and - // adjust width, height, depth as appropriate so subsequent size calculations - // succeed. + // Use the correct texture creation function depending on the target, and + // adjust width, height, depth as appropriate so subsequent size + // calculations succeed. - switch (get_base_gl_target(target)) { - case GL_TEXTURE_1D: - h = 1; d = 1; - buffer = CreateGLTexture1D( width, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, true, data ); - break; - case GL_TEXTURE_BUFFER: - h = 1; d = 1; - buffer = CreateGLTextureBuffer(width, target, gl_fmt, gl_internal_fmt, - gl_type, type, &texture, &glbuf, &error, true, data); - break; - case GL_RENDERBUFFER: - case GL_COLOR_ATTACHMENT0: - d = 1; - buffer = CreateGLRenderbuffer(width, height, target, gl_fmt, - gl_internal_fmt, gl_type, type, &glFramebuffer, &glRenderbuffer, &error, - data, true); - break; - case GL_TEXTURE_2D: - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_CUBE_MAP: - d = 1; - buffer = CreateGLTexture2D(width, height, target, gl_fmt, gl_internal_fmt, - gl_type, type, &texture, &error, true, data); - break; - case GL_TEXTURE_1D_ARRAY: - d = 1; - buffer = CreateGLTexture1DArray( width, height, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, true, data ); - break; - case GL_TEXTURE_2D_ARRAY: - buffer = CreateGLTexture2DArray( width, height, depth, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, true, data ); - break; - case GL_TEXTURE_3D: - buffer = CreateGLTexture3D( width, height, depth, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, data, true ); - break; + switch (get_base_gl_target(target)) + { + case GL_TEXTURE_1D: + h = 1; + d = 1; + buffer = + CreateGLTexture1D(width, target, gl_fmt, gl_internal_fmt, + gl_type, type, &texture, &error, true, data); + break; + case GL_TEXTURE_BUFFER: + h = 1; + d = 1; + buffer = CreateGLTextureBuffer( + width, target, gl_fmt, gl_internal_fmt, gl_type, type, &texture, + &glbuf, &error, true, data); + break; + case GL_RENDERBUFFER: + case GL_COLOR_ATTACHMENT0: + d = 1; + buffer = CreateGLRenderbuffer( + width, height, target, gl_fmt, gl_internal_fmt, gl_type, type, + &glFramebuffer, &glRenderbuffer, &error, data, true); + break; + case GL_TEXTURE_2D: + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_CUBE_MAP: + d = 1; + buffer = CreateGLTexture2D(width, height, target, gl_fmt, + gl_internal_fmt, gl_type, type, &texture, + &error, true, data); + break; + case GL_TEXTURE_1D_ARRAY: + d = 1; + buffer = CreateGLTexture1DArray(width, height, target, gl_fmt, + gl_internal_fmt, gl_type, type, + &texture, &error, true, data); + break; + case GL_TEXTURE_2D_ARRAY: + buffer = CreateGLTexture2DArray(width, height, depth, target, + gl_fmt, gl_internal_fmt, gl_type, + type, &texture, &error, true, data); + break; + case GL_TEXTURE_3D: + buffer = CreateGLTexture3D(width, height, depth, target, gl_fmt, + gl_internal_fmt, gl_type, type, &texture, + &error, data, true); + break; #ifdef GL_VERSION_3_2 - case GL_TEXTURE_2D_MULTISAMPLE: - d = 1; - buffer = CreateGLTexture2DMultisample( width, height, samples, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, true, data, true ); - break; - case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: - buffer = CreateGLTexture2DArrayMultisample( width, height, depth, samples, target, gl_fmt, - gl_internal_fmt, gl_type, type, &texture, &error, true, data, true ); - break; + case GL_TEXTURE_2D_MULTISAMPLE: + d = 1; + buffer = CreateGLTexture2DMultisample( + width, height, samples, target, gl_fmt, gl_internal_fmt, + gl_type, type, &texture, &error, true, data, true); + break; + case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: + buffer = CreateGLTexture2DArrayMultisample( + width, height, depth, samples, target, gl_fmt, gl_internal_fmt, + gl_type, type, &texture, &error, true, data, true); + break; #endif - default: - log_error("Unsupported texture target."); - return 1; - } - - if ( error == -2 ) { - log_info("OpenGL texture couldn't be created, because a texture is too big. Skipping test.\n"); - return 0; - } - - // Check to see if the texture could not be created for some other reason like - // GL_FRAMEBUFFER_UNSUPPORTED - if (error == GL_FRAMEBUFFER_UNSUPPORTED) { - log_info("Skipping...\n"); - return 0; - } - - if ( error != 0 ) { - if ((gl_fmt == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())){ - log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " - "Skipping test.\n"); - return 0; - } else { - return error; - } - } - - BufferOwningPtr inputBuffer(buffer); - if( inputBuffer == NULL ) - return -1; - - cl_image_format clFormat; - ExplicitType actualType; - char *outBuffer; - - // Perform the read: - - GLuint globj = texture; - if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { - globj = glRenderbuffer; - } - - error = test_image_read( context, queue, target, globj, w, h, d, samples, &clFormat, - &actualType, (void **)&outBuffer ); - - if( error != 0 ) - return error; - - BufferOwningPtr actualResults(outBuffer); - if( actualResults == NULL ) - return -1; - - log_info( "- Read [%4d x %4d x %4d x %4d] : GL Texture : %s : %s : %s => CL Image : %s : %s \n", - (int)w, (int)h, (int)d, (int)samples, GetGLFormatName( gl_fmt ), GetGLFormatName( gl_internal_fmt ), - GetGLTypeName( gl_type ), GetChannelOrderName( clFormat.image_channel_order ), - GetChannelTypeName( clFormat.image_channel_data_type )); - - BufferOwningPtr convertedInputs; - - // We have to convert our input buffer to the returned type, so we can validate. - // This is necessary because OpenCL might not actually pick an internal format - // that actually matches our input format (for example, if it picks a normalized - // format, the results will come out as floats instead of going in as ints). - - if ( gl_type == GL_UNSIGNED_INT_2_10_10_10_REV ) - { - cl_uint *p = (cl_uint *)buffer; - float *inData = (float *)malloc( w * h * d * samples * sizeof(float) ); - - for( size_t i = 0; i < 4 * w * h * d * samples; i += 4 ) - { - inData[ i + 0 ] = (float)( ( p[ 0 ] >> 20 ) & 0x3ff ) / (float)1023; - inData[ i + 1 ] = (float)( ( p[ 0 ] >> 10 ) & 0x3ff ) / (float)1023; - inData[ i + 2 ] = (float)( p[ 0 ] & 0x3ff ) / (float)1023; - p++; + default: log_error("Unsupported texture target."); return 1; } - convertedInputs.reset( inData ); - if( convertedInputs == NULL ) - return -1; - } - else if ( gl_type == GL_DEPTH24_STENCIL8 ) - { - // GL_DEPTH24_STENCIL8 is treated as CL_UNORM_INT24 + CL_DEPTH_STENCIL where - // the stencil is ignored. - cl_uint *p = (cl_uint *)buffer; - float *inData = (float *)malloc( w * h * d * samples * sizeof(float) ); - - for( size_t i = 0; i < w * h * d * samples; i++ ) + if (error == -2) { - inData[ i ] = (float)((p[i] >> 8) & 0xffffff) / (float)0xfffffe; + log_info("OpenGL texture couldn't be created, because a texture is too " + "big. Skipping test.\n"); + return 0; } - convertedInputs.reset( inData ); - if( convertedInputs == NULL ) - return -1; - } - else if ( gl_type == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) - { - // GL_FLOAT_32_UNSIGNED_INT_24_8_REV is treated as a CL_FLOAT + - // unused 24 + CL_DEPTH_STENCIL; we check the float value and ignore the - // second word - - float *p = (float *)buffer; - float *inData = (float *)malloc( w * h * d * samples * sizeof(float) ); - - for( size_t i = 0; i < w * h * d * samples; i++ ) + // Check to see if the texture could not be created for some other reason + // like GL_FRAMEBUFFER_UNSUPPORTED + if (error == GL_FRAMEBUFFER_UNSUPPORTED) { - inData[ i ] = p[i*2]; + log_info("Skipping...\n"); + return 0; } - convertedInputs.reset( inData ); - if( convertedInputs == NULL ) - return -1; - } - else - { - convertedInputs.reset(convert_to_expected( inputBuffer, - w * h * d * samples, type, actualType, get_channel_order_channel_count(clFormat.image_channel_order) )); - if( convertedInputs == NULL ) - return -1; - } - - // Now we validate - if( actualType == kFloat ) - { - if ( clFormat.image_channel_data_type == CL_UNORM_INT_101010 ) + if (error != 0) { - return validate_float_results_rgb_101010( convertedInputs, actualResults, w, h, d, samples ); + if ((gl_fmt == GL_RGBA_INTEGER_EXT) + && (!CheckGLIntegerExtensionSupport())) + { + log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " + "Skipping test.\n"); + return 0; + } + else + { + return error; + } + } + + BufferOwningPtr inputBuffer(buffer); + if (inputBuffer == NULL) return -1; + + cl_image_format clFormat; + ExplicitType actualType; + char *outBuffer; + + // Perform the read: + + GLuint globj = texture; + if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) + { + globj = glRenderbuffer; + } + + error = test_image_read(context, queue, target, globj, w, h, d, samples, + &clFormat, &actualType, (void **)&outBuffer); + + if (error != 0) return error; + + BufferOwningPtr actualResults(outBuffer); + if (actualResults == NULL) return -1; + + log_info("- Read [%4d x %4d x %4d x %4d] : GL Texture : %s : %s : %s => CL " + "Image : %s : %s \n", + (int)w, (int)h, (int)d, (int)samples, GetGLFormatName(gl_fmt), + GetGLFormatName(gl_internal_fmt), GetGLTypeName(gl_type), + GetChannelOrderName(clFormat.image_channel_order), + GetChannelTypeName(clFormat.image_channel_data_type)); + + BufferOwningPtr convertedInputs; + + // We have to convert our input buffer to the returned type, so we can + // validate. This is necessary because OpenCL might not actually pick an + // internal format that actually matches our input format (for example, if + // it picks a normalized format, the results will come out as floats instead + // of going in as ints). + + if (gl_type == GL_UNSIGNED_INT_2_10_10_10_REV) + { + cl_uint *p = (cl_uint *)buffer; + float *inData = (float *)malloc(w * h * d * samples * sizeof(float)); + + for (size_t i = 0; i < 4 * w * h * d * samples; i += 4) + { + inData[i + 0] = (float)((p[0] >> 20) & 0x3ff) / (float)1023; + inData[i + 1] = (float)((p[0] >> 10) & 0x3ff) / (float)1023; + inData[i + 2] = (float)(p[0] & 0x3ff) / (float)1023; + p++; + } + + convertedInputs.reset(inData); + if (convertedInputs == NULL) return -1; + } + else if (gl_type == GL_DEPTH24_STENCIL8) + { + // GL_DEPTH24_STENCIL8 is treated as CL_UNORM_INT24 + CL_DEPTH_STENCIL + // where the stencil is ignored. + cl_uint *p = (cl_uint *)buffer; + float *inData = (float *)malloc(w * h * d * samples * sizeof(float)); + + for (size_t i = 0; i < w * h * d * samples; i++) + { + inData[i] = (float)((p[i] >> 8) & 0xffffff) / (float)0xfffffe; + } + + convertedInputs.reset(inData); + if (convertedInputs == NULL) return -1; + } + else if (gl_type == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) + { + // GL_FLOAT_32_UNSIGNED_INT_24_8_REV is treated as a CL_FLOAT + + // unused 24 + CL_DEPTH_STENCIL; we check the float value and ignore the + // second word + + float *p = (float *)buffer; + float *inData = (float *)malloc(w * h * d * samples * sizeof(float)); + + for (size_t i = 0; i < w * h * d * samples; i++) + { + inData[i] = p[i * 2]; + } + + convertedInputs.reset(inData); + if (convertedInputs == NULL) return -1; } else { - return validate_float_results( convertedInputs, actualResults, w, h, d, samples, get_channel_order_channel_count(clFormat.image_channel_order) ); + convertedInputs.reset(convert_to_expected( + inputBuffer, w * h * d * samples, type, actualType, + get_channel_order_channel_count(clFormat.image_channel_order))); + if (convertedInputs == NULL) return -1; + } + + // Now we validate + if (actualType == kFloat) + { + if (clFormat.image_channel_data_type == CL_UNORM_INT_101010) + { + return validate_float_results_rgb_101010( + convertedInputs, actualResults, w, h, d, samples); + } + else + { + return validate_float_results( + convertedInputs, actualResults, w, h, d, samples, + get_channel_order_channel_count(clFormat.image_channel_order)); + } + } + else + { + return validate_integer_results(convertedInputs, actualResults, w, h, d, + samples, + get_explicit_type_size(actualType)); } - } - else - { - return validate_integer_results( convertedInputs, actualResults, w, h, d, samples, get_explicit_type_size( actualType ) ); - } } int test_images_read_common(cl_device_id device, cl_context context, @@ -649,85 +705,98 @@ int test_images_read_common(cl_device_id device, cl_context context, size_t nformats, GLenum *targets, size_t ntargets, sizevec_t *sizes, size_t nsizes) { - int error = 0; - RandomSeed seed(gRandomSeed); + int error = 0; + RandomSeed seed(gRandomSeed); - // First, ensure this device supports images. + // First, ensure this device supports images. - if (checkForImageSupport(device)) { - log_info("Device does not support images. Skipping test.\n"); - return 0; - } - - size_t fidx, tidx, sidx; - - // Test each format on every target, every size. - - for ( fidx = 0; fidx < nformats; fidx++ ) { - for ( tidx = 0; tidx < ntargets; tidx++ ) { - - // Texture buffer only takes an internal format, so the level data passed - // by the test and used for verification must match the internal format - if ((targets[tidx] == GL_TEXTURE_BUFFER) && (GetGLFormat(formats[ fidx ].internal) != formats[fidx].formattype)) - continue; - - if ( formats[ fidx ].datatype == GL_UNSIGNED_INT_2_10_10_10_REV ) - { - // Check if the RGB 101010 format is supported - if ( is_rgb_101010_supported( context, targets[ tidx ] ) == 0 ) - break; // skip - } - - if (targets[tidx] != GL_TEXTURE_BUFFER) - log_info( "Testing image read for GL format %s : %s : %s : %s\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - else - log_info( "Testing image read for GL format %s : %s\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal )); - - for ( sidx = 0; sidx < nsizes; sidx++ ) { - - // Test this format + size: - int err; - if ((err = test_image_format_read(context, queue, - sizes[sidx].width, sizes[sidx].height, sizes[sidx].depth, - targets[tidx], &formats[fidx], seed) )) - { - // Negative return values are errors, positive mean the test was skipped - if (err < 0) { - - // We land here in the event of test failure. - - log_error( "ERROR: Image read test failed for %s : %s : %s : %s\n\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - error++; - } - - // Skip the other sizes for this format. - printf("Skipping remaining sizes for this format\n"); - - break; - } - } - - // Note a successful format test, if we passed every size. - - if( sidx == sizeof (sizes) / sizeof( sizes[0] ) ) { - log_info( "passed: Image read test for GL format %s : %s : %s : %s\n\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - } + if (checkForImageSupport(device)) + { + log_info("Device does not support images. Skipping test.\n"); + return 0; } - } - return error; + size_t fidx, tidx, sidx; + + // Test each format on every target, every size. + + for (fidx = 0; fidx < nformats; fidx++) + { + for (tidx = 0; tidx < ntargets; tidx++) + { + + // Texture buffer only takes an internal format, so the level data + // passed by the test and used for verification must match the + // internal format + if ((targets[tidx] == GL_TEXTURE_BUFFER) + && (GetGLFormat(formats[fidx].internal) + != formats[fidx].formattype)) + continue; + + if (formats[fidx].datatype == GL_UNSIGNED_INT_2_10_10_10_REV) + { + // Check if the RGB 101010 format is supported + if (is_rgb_101010_supported(context, targets[tidx]) == 0) + break; // skip + } + + if (targets[tidx] != GL_TEXTURE_BUFFER) + log_info("Testing image read for GL format %s : %s : %s : %s\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + else + log_info("Testing image read for GL format %s : %s\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal)); + + for (sidx = 0; sidx < nsizes; sidx++) + { + + // Test this format + size: + int err; + if ((err = test_image_format_read( + context, queue, sizes[sidx].width, sizes[sidx].height, + sizes[sidx].depth, targets[tidx], &formats[fidx], + seed))) + { + // Negative return values are errors, positive mean the test + // was skipped + if (err < 0) + { + + // We land here in the event of test failure. + + log_error("ERROR: Image read test failed for %s : %s : " + "%s : %s\n\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + error++; + } + + // Skip the other sizes for this format. + printf("Skipping remaining sizes for this format\n"); + + break; + } + } + + // Note a successful format test, if we passed every size. + + if (sidx == sizeof(sizes) / sizeof(sizes[0])) + { + log_info("passed: Image read test for GL format %s : %s : %s " + ": %s\n\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + } + } + } + + return error; } diff --git a/test_conformance/gl/test_images_write_common.cpp b/test_conformance/gl/test_images_write_common.cpp index 0dba83bb..4d721296 100644 --- a/test_conformance/gl/test_images_write_common.cpp +++ b/test_conformance/gl/test_images_write_common.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -17,16 +17,17 @@ #include "common.h" #include -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif #pragma mark - #pragma mark Write test kernels +// clang-format off static const char *kernelpattern_image_write_1D = "__kernel void sample_test( __global %s4 *source, write_only image1d_t dest )\n" "{\n" @@ -174,482 +175,534 @@ static const char * kernelpattern_image_write_2D_array_depth = #endif +// clang-format on #pragma mark - #pragma mark Utility functions -static const char* get_appropriate_write_kernel(GLenum target, - ExplicitType type, cl_channel_order channel_order) +static const char *get_appropriate_write_kernel(GLenum target, + ExplicitType type, + cl_channel_order channel_order) { - switch (get_base_gl_target(target)) { - case GL_TEXTURE_1D: + switch (get_base_gl_target(target)) + { + case GL_TEXTURE_1D: - if (type == kHalf) - return kernelpattern_image_write_1D_half; - else - return kernelpattern_image_write_1D; - break; - case GL_TEXTURE_BUFFER: - if (type == kHalf) - return kernelpattern_image_write_1D_buffer_half; - else - return kernelpattern_image_write_1D_buffer; - break; - case GL_TEXTURE_1D_ARRAY: - if (type == kHalf) - return kernelpattern_image_write_1Darray_half; - else - return kernelpattern_image_write_1Darray; - break; - case GL_COLOR_ATTACHMENT0: - case GL_RENDERBUFFER: - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_2D: - case GL_TEXTURE_CUBE_MAP: + if (type == kHalf) + return kernelpattern_image_write_1D_half; + else + return kernelpattern_image_write_1D; + break; + case GL_TEXTURE_BUFFER: + if (type == kHalf) + return kernelpattern_image_write_1D_buffer_half; + else + return kernelpattern_image_write_1D_buffer; + break; + case GL_TEXTURE_1D_ARRAY: + if (type == kHalf) + return kernelpattern_image_write_1Darray_half; + else + return kernelpattern_image_write_1Darray; + break; + case GL_COLOR_ATTACHMENT0: + case GL_RENDERBUFFER: + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_2D: + case GL_TEXTURE_CUBE_MAP: #ifdef GL_VERSION_3_2 - if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) - return kernelpattern_image_write_2D_depth; + if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) + return kernelpattern_image_write_2D_depth; #endif - if (type == kHalf) - return kernelpattern_image_write_2D_half; - else - return kernelpattern_image_write_2D; - break; + if (type == kHalf) + return kernelpattern_image_write_2D_half; + else + return kernelpattern_image_write_2D; + break; - case GL_TEXTURE_2D_ARRAY: + case GL_TEXTURE_2D_ARRAY: #ifdef GL_VERSION_3_2 - if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) - return kernelpattern_image_write_2D_array_depth; + if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) + return kernelpattern_image_write_2D_array_depth; #endif - if (type == kHalf) - return kernelpattern_image_write_2Darray_half; - else - return kernelpattern_image_write_2Darray; - break; + if (type == kHalf) + return kernelpattern_image_write_2Darray_half; + else + return kernelpattern_image_write_2Darray; + break; - case GL_TEXTURE_3D: - if (type == kHalf) - return kernelpattern_image_write_3D_half; - else - return kernelpattern_image_write_3D; - break; + case GL_TEXTURE_3D: + if (type == kHalf) + return kernelpattern_image_write_3D_half; + else + return kernelpattern_image_write_3D; + break; - default: - log_error("Unsupported GL tex target (%s) passed to write test: " - "%s (%s):%d", GetGLTargetName(target), __FUNCTION__, - __FILE__, __LINE__); - return NULL; - } + default: + log_error("Unsupported GL tex target (%s) passed to write test: " + "%s (%s):%d", + GetGLTargetName(target), __FUNCTION__, __FILE__, + __LINE__); + return NULL; + } } void set_dimensions_by_target(GLenum target, size_t *dims, size_t sizes[3], - size_t width, size_t height, size_t depth) + size_t width, size_t height, size_t depth) { - switch (get_base_gl_target(target)) { - case GL_TEXTURE_1D: - sizes[0] = width; - *dims = 1; - break; - - case GL_TEXTURE_BUFFER: - sizes[0] = width; - *dims = 1; - break; - - case GL_TEXTURE_1D_ARRAY: - sizes[0] = width; - sizes[1] = height; - *dims = 2; - break; - - case GL_COLOR_ATTACHMENT0: - case GL_RENDERBUFFER: - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_2D: - case GL_TEXTURE_CUBE_MAP: - - sizes[0] = width; - sizes[1] = height; - *dims = 2; - break; - - case GL_TEXTURE_2D_ARRAY: - sizes[0] = width; - sizes[1] = height; - sizes[2] = depth; - *dims = 3; - break; - - case GL_TEXTURE_3D: - sizes[0] = width; - sizes[1] = height; - sizes[2] = depth; - *dims = 3; - break; - - default: - log_error("Unsupported GL tex target (%s) passed to write test: " - "%s (%s):%d", GetGLTargetName(target), __FUNCTION__, - __FILE__, __LINE__); - } -} - -int test_cl_image_write( cl_context context, cl_command_queue queue, - GLenum target, cl_mem clImage, size_t width, size_t height, size_t depth, - cl_image_format *outFormat, ExplicitType *outType, void **outSourceBuffer, - MTdata d, bool supports_half ) -{ - size_t global_dims, global_sizes[3]; - clProgramWrapper program; - clKernelWrapper kernel; - clMemWrapper inStream; - char* programPtr; - int error; - char kernelSource[2048]; - - // What CL format did we get from the texture? - - error = clGetImageInfo(clImage, CL_IMAGE_FORMAT, sizeof(cl_image_format), - outFormat, NULL); - test_error(error, "Unable to get the CL image format"); - - // Create the kernel source. The target and the data type will influence - // which particular kernel we choose. - - *outType = get_write_kernel_type( outFormat ); - size_t channelSize = get_explicit_type_size(*outType); - - const char* appropriateKernel = get_appropriate_write_kernel(target, - *outType, outFormat->image_channel_order); - if (*outType == kHalf && !supports_half) { - log_info("cl_khr_fp16 isn't supported. Skip this test.\n"); - return 0; - } - - const char* suffix = get_kernel_suffix( outFormat ); - const char* convert = get_write_conversion( outFormat, *outType ); - - sprintf(kernelSource, appropriateKernel, get_explicit_type_name( *outType ), - get_explicit_type_name( *outType ), suffix, convert); - - programPtr = kernelSource; - if( create_single_kernel_helper_with_build_options( context, &program, &kernel, 1, - (const char **)&programPtr, "sample_test", "" ) ) - { - return -1; - } - - // Create an appropriately-sized output buffer. - - // Check to see if the output buffer will fit on the device - size_t bytes = channelSize * 4 * width * height * depth; - cl_ulong alloc_size = 0; - - cl_device_id device = NULL; - error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device), &device, NULL); - test_error( error, "Unable to query command queue for device" ); - - error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof(alloc_size), &alloc_size, NULL); - test_error( error, "Unable to device for max mem alloc size" ); - - if (bytes > alloc_size) { - log_info(" Skipping: Buffer size (%lu) is greater than CL_DEVICE_MAX_MEM_ALLOC_SIZE (%lu)\n", bytes, alloc_size); - *outSourceBuffer = NULL; - return 0; - } - - *outSourceBuffer = CreateRandomData(*outType, width * height * depth * 4, d); - - inStream = clCreateBuffer( context, CL_MEM_COPY_HOST_PTR, - channelSize * 4 * width * height * depth, *outSourceBuffer, &error ); - test_error( error, "Unable to create output buffer" ); - - clSamplerWrapper sampler = clCreateSampler(context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error); - test_error( error, "Unable to create sampler" ); - - error = clSetKernelArg( kernel, 0, sizeof( inStream ), &inStream ); - test_error( error, "Unable to set kernel arguments" ); - - error = clSetKernelArg( kernel, 1, sizeof( clImage ), &clImage ); - test_error( error, "Unable to set kernel arguments" ); - - // Flush and Acquire. - - glFinish(); - - error = (*clEnqueueAcquireGLObjects_ptr)( queue, 1, &clImage, 0, NULL, NULL); - test_error( error, "Unable to acquire GL obejcts"); - - // Execute ( letting OpenCL choose the local size ) - - // Setup the global dimensions and sizes based on the target type. - set_dimensions_by_target(target, &global_dims, global_sizes, - width, height, depth); - - error = clEnqueueNDRangeKernel( queue, kernel, global_dims, NULL, - global_sizes, NULL, 0, NULL, NULL ); - test_error( error, "Unable to execute test kernel" ); - - clEventWrapper event; - error = (*clEnqueueReleaseGLObjects_ptr)( queue, 1, &clImage, 0, NULL, &event ); - test_error(error, "clEnqueueReleaseGLObjects failed"); - - error = clWaitForEvents( 1, &event ); - test_error(error, "clWaitForEvents failed"); - - return 0; -} - -static int test_image_write( cl_context context, cl_command_queue queue, - GLenum glTarget, GLuint glTexture, size_t width, size_t height, size_t depth, - cl_image_format *outFormat, ExplicitType *outType, void **outSourceBuffer, - MTdata d, bool supports_half ) -{ - int error; - - // Create a CL image from the supplied GL texture - clMemWrapper image = (*clCreateFromGLTexture_ptr)( context, CL_MEM_WRITE_ONLY, - glTarget, 0, glTexture, &error ); - - if ( error != CL_SUCCESS ) { - print_error( error, "Unable to create CL image from GL texture" ); - GLint fmt; - glGetTexLevelParameteriv( glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); - log_error( " Supplied GL texture was base format %s and internal " - "format %s\n", GetGLBaseFormatName( fmt ), GetGLFormatName( fmt ) ); - return error; - } - - return test_cl_image_write( context, queue, glTarget, image, - width, height, depth, outFormat, outType, outSourceBuffer, d, supports_half ); -} - -int supportsHalf(cl_context context, bool* supports_half) -{ - int error; - cl_uint numDev; - - error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDev, NULL); - test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); - - cl_device_id* devices = new cl_device_id[numDev]; - error = clGetContextInfo(context, CL_CONTEXT_DEVICES, numDev * sizeof(cl_device_id), devices, NULL); - test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); - - *supports_half = is_extension_available(devices[0], "cl_khr_fp16"); - delete [] devices; - - return error; -} - -int supportsMsaa(cl_context context, bool* supports_msaa) -{ - int error; - cl_uint numDev; - - error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDev, NULL); - test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); - - cl_device_id* devices = new cl_device_id[numDev]; - error = clGetContextInfo(context, CL_CONTEXT_DEVICES, numDev * sizeof(cl_device_id), devices, NULL); - test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); - - *supports_msaa = is_extension_available(devices[0], "cl_khr_gl_msaa_sharing"); - delete [] devices; - - return error; -} - -int supportsDepth(cl_context context, bool* supports_depth) -{ - int error; - cl_uint numDev; - - error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDev, NULL); - test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); - - cl_device_id* devices = new cl_device_id[numDev]; - error = clGetContextInfo(context, CL_CONTEXT_DEVICES, numDev * sizeof(cl_device_id), devices, NULL); - test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); - - *supports_depth = is_extension_available(devices[0], "cl_khr_gl_depth_images"); - delete [] devices; - - return error; -} - -static int test_image_format_write( cl_context context, cl_command_queue queue, - size_t width, size_t height, size_t depth, GLenum target, GLenum format, - GLenum internalFormat, GLenum glType, ExplicitType type, MTdata d ) -{ - int error; - // If we're testing a half float format, then we need to determine the - // rounding mode of this machine. Punt if we fail to do so. - - if( type == kHalf ) - if( DetectFloatToHalfRoundingMode(queue) ) - return 1; - - // Create an appropriate GL texture or renderbuffer, given the target. - - glTextureWrapper glTexture; - glBufferWrapper glBuf; - glFramebufferWrapper glFramebuffer; - glRenderbufferWrapper glRenderbuffer; - switch (get_base_gl_target(target)) { - case GL_TEXTURE_1D: - CreateGLTexture1D( width, target, format, internalFormat, glType, - type, &glTexture, &error, false, d ); - break; - case GL_TEXTURE_BUFFER: - CreateGLTextureBuffer( width, target, format, internalFormat, glType, - type, &glTexture, &glBuf, &error, false, d ); - break; - case GL_TEXTURE_1D_ARRAY: - CreateGLTexture1DArray( width, height, target, format, internalFormat, - glType, type, &glTexture, &error, false, d ); - break; - case GL_TEXTURE_RECTANGLE_EXT: - case GL_TEXTURE_2D: - case GL_TEXTURE_CUBE_MAP: - CreateGLTexture2D( width, height, target, format, internalFormat, glType, - type, &glTexture, &error, false, d ); - break; - case GL_COLOR_ATTACHMENT0: - case GL_RENDERBUFFER: - CreateGLRenderbuffer(width, height, target, format, internalFormat, - glType, type, &glFramebuffer, &glRenderbuffer, &error, d, false); - case GL_TEXTURE_2D_ARRAY: - CreateGLTexture2DArray( width, height, depth, target, format, - internalFormat, glType, type, &glTexture, &error, false, d ); - break; - case GL_TEXTURE_3D: - CreateGLTexture3D( width, height, depth, target, format, - internalFormat, glType, type, &glTexture, &error, d, false ); - break; - - default: - log_error("Unsupported GL tex target (%s) passed to write test: " - "%s (%s):%d", GetGLTargetName(target), __FUNCTION__, - __FILE__, __LINE__); - } - - // If there was a problem during creation, make sure it isn't a known - // cause, and then complain. - if ( error == -2 ) { - log_info("OpenGL texture couldn't be created, because a texture is too big. Skipping test.\n"); - return 0; - } - - if ( error != 0 ) { - if ((format == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())){ - log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " - "Skipping test.\n"); - return 0; - } else { - return error; - } - } - - // Run and get the results - cl_image_format clFormat; - ExplicitType sourceType; - ExplicitType validationType; - void *outSourceBuffer = NULL; - - GLenum globj = glTexture; - if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { - globj = glRenderbuffer; - } - - bool supports_half = false; - error = supportsHalf(context, &supports_half); - if( error != 0 ) - return error; - - error = test_image_write( context, queue, target, globj, width, height, - depth, &clFormat, &sourceType, (void **)&outSourceBuffer, d, supports_half ); - - if( error != 0 || ((sourceType == kHalf ) && !supports_half)) { - if (outSourceBuffer) - free(outSourceBuffer); - return error; - } - - if (!outSourceBuffer) - return 0; - - // If actual source type was half, convert to float for validation. - - if ( sourceType == kHalf ) - validationType = kFloat; - else - validationType = sourceType; - - BufferOwningPtr validationSource; - - if ( clFormat.image_channel_data_type == CL_UNORM_INT_101010 ) - { - validationSource.reset( outSourceBuffer ); - } - else - { - validationSource.reset( convert_to_expected( outSourceBuffer, - width * height * depth, sourceType, validationType, get_channel_order_channel_count(clFormat.image_channel_order) ) ); - free(outSourceBuffer); - } - - log_info( "- Write for %s [%4ld x %4ld x %4ld] : GL Texture : %s : %s : %s =>" - " CL Image : %s : %s \n", - GetGLTargetName(target), - width, height, depth, - GetGLFormatName( format ), - GetGLFormatName( internalFormat ), - GetGLTypeName( glType), - GetChannelOrderName( clFormat.image_channel_order ), - GetChannelTypeName( clFormat.image_channel_data_type )); - - // Read the results from the GL texture. - - ExplicitType readType = type; - BufferOwningPtr glResults( ReadGLTexture( - target, glTexture, glBuf, width, format, - internalFormat, glType, readType, /* unused */ 1, 1 ) ); - if( glResults == NULL ) - return -1; - - // We have to convert our input buffer to the returned type, so we can validate. - BufferOwningPtr convertedGLResults; - if ( clFormat.image_channel_data_type != CL_UNORM_INT_101010 ) - { - convertedGLResults.reset( convert_to_expected( - glResults, width * height * depth, readType, validationType, get_channel_order_channel_count(clFormat.image_channel_order), glType )); - } - - // Validate. - - int valid = 0; - if (convertedGLResults) { - if( sourceType == kFloat || sourceType == kHalf ) + switch (get_base_gl_target(target)) { - if ( clFormat.image_channel_data_type == CL_UNORM_INT_101010 ) - { - valid = validate_float_results_rgb_101010( validationSource, glResults, width, height, depth, 1 ); - } - else - { - valid = validate_float_results( validationSource, convertedGLResults, - width, height, depth, 1, get_channel_order_channel_count(clFormat.image_channel_order) ); - } + case GL_TEXTURE_1D: + sizes[0] = width; + *dims = 1; + break; + + case GL_TEXTURE_BUFFER: + sizes[0] = width; + *dims = 1; + break; + + case GL_TEXTURE_1D_ARRAY: + sizes[0] = width; + sizes[1] = height; + *dims = 2; + break; + + case GL_COLOR_ATTACHMENT0: + case GL_RENDERBUFFER: + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_2D: + case GL_TEXTURE_CUBE_MAP: + + sizes[0] = width; + sizes[1] = height; + *dims = 2; + break; + + case GL_TEXTURE_2D_ARRAY: + sizes[0] = width; + sizes[1] = height; + sizes[2] = depth; + *dims = 3; + break; + + case GL_TEXTURE_3D: + sizes[0] = width; + sizes[1] = height; + sizes[2] = depth; + *dims = 3; + break; + + default: + log_error("Unsupported GL tex target (%s) passed to write test: " + "%s (%s):%d", + GetGLTargetName(target), __FUNCTION__, __FILE__, + __LINE__); + } +} + +int test_cl_image_write(cl_context context, cl_command_queue queue, + GLenum target, cl_mem clImage, size_t width, + size_t height, size_t depth, cl_image_format *outFormat, + ExplicitType *outType, void **outSourceBuffer, MTdata d, + bool supports_half) +{ + size_t global_dims, global_sizes[3]; + clProgramWrapper program; + clKernelWrapper kernel; + clMemWrapper inStream; + char *programPtr; + int error; + char kernelSource[2048]; + + // What CL format did we get from the texture? + + error = clGetImageInfo(clImage, CL_IMAGE_FORMAT, sizeof(cl_image_format), + outFormat, NULL); + test_error(error, "Unable to get the CL image format"); + + // Create the kernel source. The target and the data type will influence + // which particular kernel we choose. + + *outType = get_write_kernel_type(outFormat); + size_t channelSize = get_explicit_type_size(*outType); + + const char *appropriateKernel = get_appropriate_write_kernel( + target, *outType, outFormat->image_channel_order); + if (*outType == kHalf && !supports_half) + { + log_info("cl_khr_fp16 isn't supported. Skip this test.\n"); + return 0; + } + + const char *suffix = get_kernel_suffix(outFormat); + const char *convert = get_write_conversion(outFormat, *outType); + + sprintf(kernelSource, appropriateKernel, get_explicit_type_name(*outType), + get_explicit_type_name(*outType), suffix, convert); + + programPtr = kernelSource; + if (create_single_kernel_helper_with_build_options( + context, &program, &kernel, 1, (const char **)&programPtr, + "sample_test", "")) + { + return -1; + } + + // Create an appropriately-sized output buffer. + + // Check to see if the output buffer will fit on the device + size_t bytes = channelSize * 4 * width * height * depth; + cl_ulong alloc_size = 0; + + cl_device_id device = NULL; + error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device), + &device, NULL); + test_error(error, "Unable to query command queue for device"); + + error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, + sizeof(alloc_size), &alloc_size, NULL); + test_error(error, "Unable to device for max mem alloc size"); + + if (bytes > alloc_size) + { + log_info(" Skipping: Buffer size (%lu) is greater than " + "CL_DEVICE_MAX_MEM_ALLOC_SIZE (%lu)\n", + bytes, alloc_size); + *outSourceBuffer = NULL; + return 0; + } + + *outSourceBuffer = + CreateRandomData(*outType, width * height * depth * 4, d); + + inStream = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, + channelSize * 4 * width * height * depth, + *outSourceBuffer, &error); + test_error(error, "Unable to create output buffer"); + + clSamplerWrapper sampler = clCreateSampler( + context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error); + test_error(error, "Unable to create sampler"); + + error = clSetKernelArg(kernel, 0, sizeof(inStream), &inStream); + test_error(error, "Unable to set kernel arguments"); + + error = clSetKernelArg(kernel, 1, sizeof(clImage), &clImage); + test_error(error, "Unable to set kernel arguments"); + + // Flush and Acquire. + + glFinish(); + + error = (*clEnqueueAcquireGLObjects_ptr)(queue, 1, &clImage, 0, NULL, NULL); + test_error(error, "Unable to acquire GL obejcts"); + + // Execute ( letting OpenCL choose the local size ) + + // Setup the global dimensions and sizes based on the target type. + set_dimensions_by_target(target, &global_dims, global_sizes, width, height, + depth); + + error = clEnqueueNDRangeKernel(queue, kernel, global_dims, NULL, + global_sizes, NULL, 0, NULL, NULL); + test_error(error, "Unable to execute test kernel"); + + clEventWrapper event; + error = + (*clEnqueueReleaseGLObjects_ptr)(queue, 1, &clImage, 0, NULL, &event); + test_error(error, "clEnqueueReleaseGLObjects failed"); + + error = clWaitForEvents(1, &event); + test_error(error, "clWaitForEvents failed"); + + return 0; +} + +static int test_image_write(cl_context context, cl_command_queue queue, + GLenum glTarget, GLuint glTexture, size_t width, + size_t height, size_t depth, + cl_image_format *outFormat, ExplicitType *outType, + void **outSourceBuffer, MTdata d, + bool supports_half) +{ + int error; + + // Create a CL image from the supplied GL texture + clMemWrapper image = (*clCreateFromGLTexture_ptr)( + context, CL_MEM_WRITE_ONLY, glTarget, 0, glTexture, &error); + + if (error != CL_SUCCESS) + { + print_error(error, "Unable to create CL image from GL texture"); + GLint fmt; + glGetTexLevelParameteriv(glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt); + log_error(" Supplied GL texture was base format %s and internal " + "format %s\n", + GetGLBaseFormatName(fmt), GetGLFormatName(fmt)); + return error; + } + + return test_cl_image_write(context, queue, glTarget, image, width, height, + depth, outFormat, outType, outSourceBuffer, d, + supports_half); +} + +int supportsHalf(cl_context context, bool *supports_half) +{ + int error; + cl_uint numDev; + + error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), + &numDev, NULL); + test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); + + cl_device_id *devices = new cl_device_id[numDev]; + error = clGetContextInfo(context, CL_CONTEXT_DEVICES, + numDev * sizeof(cl_device_id), devices, NULL); + test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); + + *supports_half = is_extension_available(devices[0], "cl_khr_fp16"); + delete[] devices; + + return error; +} + +int supportsMsaa(cl_context context, bool *supports_msaa) +{ + int error; + cl_uint numDev; + + error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), + &numDev, NULL); + test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); + + cl_device_id *devices = new cl_device_id[numDev]; + error = clGetContextInfo(context, CL_CONTEXT_DEVICES, + numDev * sizeof(cl_device_id), devices, NULL); + test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); + + *supports_msaa = + is_extension_available(devices[0], "cl_khr_gl_msaa_sharing"); + delete[] devices; + + return error; +} + +int supportsDepth(cl_context context, bool *supports_depth) +{ + int error; + cl_uint numDev; + + error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), + &numDev, NULL); + test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); + + cl_device_id *devices = new cl_device_id[numDev]; + error = clGetContextInfo(context, CL_CONTEXT_DEVICES, + numDev * sizeof(cl_device_id), devices, NULL); + test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); + + *supports_depth = + is_extension_available(devices[0], "cl_khr_gl_depth_images"); + delete[] devices; + + return error; +} + +static int test_image_format_write(cl_context context, cl_command_queue queue, + size_t width, size_t height, size_t depth, + GLenum target, GLenum format, + GLenum internalFormat, GLenum glType, + ExplicitType type, MTdata d) +{ + int error; + // If we're testing a half float format, then we need to determine the + // rounding mode of this machine. Punt if we fail to do so. + + if (type == kHalf) + if (DetectFloatToHalfRoundingMode(queue)) return 1; + + // Create an appropriate GL texture or renderbuffer, given the target. + + glTextureWrapper glTexture; + glBufferWrapper glBuf; + glFramebufferWrapper glFramebuffer; + glRenderbufferWrapper glRenderbuffer; + switch (get_base_gl_target(target)) + { + case GL_TEXTURE_1D: + CreateGLTexture1D(width, target, format, internalFormat, glType, + type, &glTexture, &error, false, d); + break; + case GL_TEXTURE_BUFFER: + CreateGLTextureBuffer(width, target, format, internalFormat, glType, + type, &glTexture, &glBuf, &error, false, d); + break; + case GL_TEXTURE_1D_ARRAY: + CreateGLTexture1DArray(width, height, target, format, + internalFormat, glType, type, &glTexture, + &error, false, d); + break; + case GL_TEXTURE_RECTANGLE_EXT: + case GL_TEXTURE_2D: + case GL_TEXTURE_CUBE_MAP: + CreateGLTexture2D(width, height, target, format, internalFormat, + glType, type, &glTexture, &error, false, d); + break; + case GL_COLOR_ATTACHMENT0: + case GL_RENDERBUFFER: + CreateGLRenderbuffer(width, height, target, format, internalFormat, + glType, type, &glFramebuffer, &glRenderbuffer, + &error, d, false); + case GL_TEXTURE_2D_ARRAY: + CreateGLTexture2DArray(width, height, depth, target, format, + internalFormat, glType, type, &glTexture, + &error, false, d); + break; + case GL_TEXTURE_3D: + CreateGLTexture3D(width, height, depth, target, format, + internalFormat, glType, type, &glTexture, &error, + d, false); + break; + + default: + log_error("Unsupported GL tex target (%s) passed to write test: " + "%s (%s):%d", + GetGLTargetName(target), __FUNCTION__, __FILE__, + __LINE__); + } + + // If there was a problem during creation, make sure it isn't a known + // cause, and then complain. + if (error == -2) + { + log_info("OpenGL texture couldn't be created, because a texture is too " + "big. Skipping test.\n"); + return 0; + } + + if (error != 0) + { + if ((format == GL_RGBA_INTEGER_EXT) + && (!CheckGLIntegerExtensionSupport())) + { + log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " + "Skipping test.\n"); + return 0; + } + else + { + return error; + } + } + + // Run and get the results + cl_image_format clFormat; + ExplicitType sourceType; + ExplicitType validationType; + void *outSourceBuffer = NULL; + + GLenum globj = glTexture; + if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) + { + globj = glRenderbuffer; + } + + bool supports_half = false; + error = supportsHalf(context, &supports_half); + if (error != 0) return error; + + error = test_image_write(context, queue, target, globj, width, height, + depth, &clFormat, &sourceType, + (void **)&outSourceBuffer, d, supports_half); + + if (error != 0 || ((sourceType == kHalf) && !supports_half)) + { + if (outSourceBuffer) free(outSourceBuffer); + return error; + } + + if (!outSourceBuffer) return 0; + + // If actual source type was half, convert to float for validation. + + if (sourceType == kHalf) + validationType = kFloat; + else + validationType = sourceType; + + BufferOwningPtr validationSource; + + if (clFormat.image_channel_data_type == CL_UNORM_INT_101010) + { + validationSource.reset(outSourceBuffer); } else { - valid = validate_integer_results( validationSource, convertedGLResults, - width, height, depth, 1, get_explicit_type_size( readType ) ); + validationSource.reset(convert_to_expected( + outSourceBuffer, width * height * depth, sourceType, validationType, + get_channel_order_channel_count(clFormat.image_channel_order))); + free(outSourceBuffer); } - } - return valid; + log_info( + "- Write for %s [%4ld x %4ld x %4ld] : GL Texture : %s : %s : %s =>" + " CL Image : %s : %s \n", + GetGLTargetName(target), width, height, depth, GetGLFormatName(format), + GetGLFormatName(internalFormat), GetGLTypeName(glType), + GetChannelOrderName(clFormat.image_channel_order), + GetChannelTypeName(clFormat.image_channel_data_type)); + + // Read the results from the GL texture. + + ExplicitType readType = type; + BufferOwningPtr glResults( + ReadGLTexture(target, glTexture, glBuf, width, format, internalFormat, + glType, readType, /* unused */ 1, 1)); + if (glResults == NULL) return -1; + + // We have to convert our input buffer to the returned type, so we can + // validate. + BufferOwningPtr convertedGLResults; + if (clFormat.image_channel_data_type != CL_UNORM_INT_101010) + { + convertedGLResults.reset(convert_to_expected( + glResults, width * height * depth, readType, validationType, + get_channel_order_channel_count(clFormat.image_channel_order), + glType)); + } + + // Validate. + + int valid = 0; + if (convertedGLResults) + { + if (sourceType == kFloat || sourceType == kHalf) + { + if (clFormat.image_channel_data_type == CL_UNORM_INT_101010) + { + valid = validate_float_results_rgb_101010( + validationSource, glResults, width, height, depth, 1); + } + else + { + valid = + validate_float_results(validationSource, convertedGLResults, + width, height, depth, 1, + get_channel_order_channel_count( + clFormat.image_channel_order)); + } + } + else + { + valid = validate_integer_results( + validationSource, convertedGLResults, width, height, depth, 1, + get_explicit_type_size(readType)); + } + } + + return valid; } #pragma mark - @@ -664,152 +717,180 @@ int test_images_write_common(cl_device_id device, cl_context context, size_t nformats, GLenum *targets, size_t ntargets, sizevec_t *sizes, size_t nsizes) { - int err = 0; - int error = 0; - RandomSeed seed(gRandomSeed); + int err = 0; + int error = 0; + RandomSeed seed(gRandomSeed); - // First, ensure this device supports images. + // First, ensure this device supports images. - if (checkForImageSupport(device)) { - log_info("Device does not support images. Skipping test.\n"); - return 0; - } + if (checkForImageSupport(device)) + { + log_info("Device does not support images. Skipping test.\n"); + return 0; + } - // Get the value of CL_DEVICE_MAX_MEM_ALLOC_SIZE - cl_ulong max_individual_allocation_size = 0; - err = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, - sizeof(max_individual_allocation_size), - &max_individual_allocation_size, NULL); - if (err) { - log_error("ERROR: clGetDeviceInfo failed for CL_DEVICE_MAX_MEM_ALLOC_SIZE.\n"); - error++; - return error; - } + // Get the value of CL_DEVICE_MAX_MEM_ALLOC_SIZE + cl_ulong max_individual_allocation_size = 0; + err = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, + sizeof(max_individual_allocation_size), + &max_individual_allocation_size, NULL); + if (err) + { + log_error("ERROR: clGetDeviceInfo failed for " + "CL_DEVICE_MAX_MEM_ALLOC_SIZE.\n"); + error++; + return error; + } - size_t total_allocation_size; - size_t fidx, tidx, sidx; + size_t total_allocation_size; + size_t fidx, tidx, sidx; - for ( fidx = 0; fidx < nformats; fidx++ ) { - for ( tidx = 0; tidx < ntargets; tidx++ ) { + for (fidx = 0; fidx < nformats; fidx++) + { + for (tidx = 0; tidx < ntargets; tidx++) + { - // Texture buffer only takes an internal format, so the level data passed - // by the test and used for verification must match the internal format - if ((targets[tidx] == GL_TEXTURE_BUFFER) && (GetGLFormat(formats[ fidx ].internal) != formats[fidx].formattype)) - continue; + // Texture buffer only takes an internal format, so the level data + // passed by the test and used for verification must match the + // internal format + if ((targets[tidx] == GL_TEXTURE_BUFFER) + && (GetGLFormat(formats[fidx].internal) + != formats[fidx].formattype)) + continue; - if ( formats[ fidx ].datatype == GL_UNSIGNED_INT_2_10_10_10_REV ) - { - // Check if the RGB 101010 format is supported - if ( is_rgb_101010_supported( context, targets[ tidx ] ) == 0 ) - continue; // skip - } + if (formats[fidx].datatype == GL_UNSIGNED_INT_2_10_10_10_REV) + { + // Check if the RGB 101010 format is supported + if (is_rgb_101010_supported(context, targets[tidx]) == 0) + continue; // skip + } - if (formats[ fidx ].datatype == GL_UNSIGNED_INT_24_8) - { - //check if a implementation supports writing to the depth stencil formats - cl_image_format imageFormat = { CL_DEPTH_STENCIL, CL_UNORM_INT24 }; - if (!is_image_format_supported(context, CL_MEM_WRITE_ONLY, (targets[tidx] == GL_TEXTURE_2D || targets[tidx] == GL_TEXTURE_RECTANGLE) ? CL_MEM_OBJECT_IMAGE2D: CL_MEM_OBJECT_IMAGE2D_ARRAY, &imageFormat)) - continue; - } + if (formats[fidx].datatype == GL_UNSIGNED_INT_24_8) + { + // check if a implementation supports writing to the depth + // stencil formats + cl_image_format imageFormat = { CL_DEPTH_STENCIL, + CL_UNORM_INT24 }; + if (!is_image_format_supported( + context, CL_MEM_WRITE_ONLY, + (targets[tidx] == GL_TEXTURE_2D + || targets[tidx] == GL_TEXTURE_RECTANGLE) + ? CL_MEM_OBJECT_IMAGE2D + : CL_MEM_OBJECT_IMAGE2D_ARRAY, + &imageFormat)) + continue; + } - if (formats[ fidx ].datatype == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) - { - //check if a implementation supports writing to the depth stencil formats - cl_image_format imageFormat = { CL_DEPTH_STENCIL, CL_FLOAT}; - if (!is_image_format_supported(context, CL_MEM_WRITE_ONLY, (targets[tidx] == GL_TEXTURE_2D || targets[tidx] == GL_TEXTURE_RECTANGLE) ? CL_MEM_OBJECT_IMAGE2D: CL_MEM_OBJECT_IMAGE2D_ARRAY, &imageFormat)) - continue; - } + if (formats[fidx].datatype == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) + { + // check if a implementation supports writing to the depth + // stencil formats + cl_image_format imageFormat = { CL_DEPTH_STENCIL, CL_FLOAT }; + if (!is_image_format_supported( + context, CL_MEM_WRITE_ONLY, + (targets[tidx] == GL_TEXTURE_2D + || targets[tidx] == GL_TEXTURE_RECTANGLE) + ? CL_MEM_OBJECT_IMAGE2D + : CL_MEM_OBJECT_IMAGE2D_ARRAY, + &imageFormat)) + continue; + } - if (targets[tidx] != GL_TEXTURE_BUFFER) - log_info( "Testing image write for GL format %s : %s : %s : %s\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - else - log_info( "Testing image write for GL format %s : %s\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal )); + if (targets[tidx] != GL_TEXTURE_BUFFER) + log_info( + "Testing image write for GL format %s : %s : %s : %s\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + else + log_info("Testing image write for GL format %s : %s\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal)); - for (sidx = 0; sidx < nsizes; sidx++) { + for (sidx = 0; sidx < nsizes; sidx++) + { - // All tested formats are 4-channel formats - total_allocation_size = - sizes[sidx].width * sizes[sidx].height * sizes[sidx].depth * - 4 * get_explicit_type_size( formats[ fidx ].type ); + // All tested formats are 4-channel formats + total_allocation_size = sizes[sidx].width * sizes[sidx].height + * sizes[sidx].depth * 4 + * get_explicit_type_size(formats[fidx].type); - if (total_allocation_size > max_individual_allocation_size) { - log_info( "The requested allocation size (%gMB) is larger than the " - "maximum individual allocation size (%gMB)\n", - total_allocation_size/(1024.0*1024.0), - max_individual_allocation_size/(1024.0*1024.0)); - log_info( "Skipping write test for %s : %s : %s : %s " - " and size (%ld, %ld, %ld)\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ), - sizes[sidx].width, - sizes[sidx].height, - sizes[sidx].depth); - continue; - } + if (total_allocation_size > max_individual_allocation_size) + { + log_info("The requested allocation size (%gMB) is larger " + "than the " + "maximum individual allocation size (%gMB)\n", + total_allocation_size / (1024.0 * 1024.0), + max_individual_allocation_size + / (1024.0 * 1024.0)); + log_info("Skipping write test for %s : %s : %s : %s " + " and size (%ld, %ld, %ld)\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype), + sizes[sidx].width, sizes[sidx].height, + sizes[sidx].depth); + continue; + } #ifdef GL_VERSION_3_2 - if (get_base_gl_target(targets[ tidx ]) == GL_TEXTURE_2D_MULTISAMPLE || - get_base_gl_target(targets[ tidx ]) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) - { - bool supports_msaa; - int errorInGetInfo = supportsMsaa(context, &supports_msaa); - if (errorInGetInfo != 0) return errorInGetInfo; - if (!supports_msaa) return 0; - } - if (formats[ fidx ].formattype == GL_DEPTH_COMPONENT || - formats[ fidx ].formattype == GL_DEPTH_STENCIL) - { - bool supports_depth; - int errorInGetInfo = supportsDepth(context, &supports_depth); - if (errorInGetInfo != 0) return errorInGetInfo; - if (!supports_depth) return 0; - } + if (get_base_gl_target(targets[tidx]) + == GL_TEXTURE_2D_MULTISAMPLE + || get_base_gl_target(targets[tidx]) + == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) + { + bool supports_msaa; + int errorInGetInfo = supportsMsaa(context, &supports_msaa); + if (errorInGetInfo != 0) return errorInGetInfo; + if (!supports_msaa) return 0; + } + if (formats[fidx].formattype == GL_DEPTH_COMPONENT + || formats[fidx].formattype == GL_DEPTH_STENCIL) + { + bool supports_depth; + int errorInGetInfo = + supportsDepth(context, &supports_depth); + if (errorInGetInfo != 0) return errorInGetInfo; + if (!supports_depth) return 0; + } #endif - if( test_image_format_write( context, queue, - sizes[sidx].width, - sizes[sidx].height, - sizes[sidx].depth, - targets[ tidx ], - formats[ fidx ].formattype, - formats[ fidx ].internal, - formats[ fidx ].datatype, - formats[ fidx ].type, seed ) ) - { - log_error( "ERROR: Image write test failed for %s : %s : %s : %s " - " and size (%ld, %ld, %ld)\n\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ), - sizes[sidx].width, - sizes[sidx].height, - sizes[sidx].depth); + if (test_image_format_write( + context, queue, sizes[sidx].width, sizes[sidx].height, + sizes[sidx].depth, targets[tidx], + formats[fidx].formattype, formats[fidx].internal, + formats[fidx].datatype, formats[fidx].type, seed)) + { + log_error( + "ERROR: Image write test failed for %s : %s : %s : %s " + " and size (%ld, %ld, %ld)\n\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype), + sizes[sidx].width, sizes[sidx].height, + sizes[sidx].depth); - error++; - break; // Skip other sizes for this combination + error++; + break; // Skip other sizes for this combination + } + } + + // If we passed all sizes (check versus size loop count): + + if (sidx == nsizes) + { + log_info( + "passed: Image write for GL format %s : %s : %s : %s\n\n", + GetGLTargetName(targets[tidx]), + GetGLFormatName(formats[fidx].internal), + GetGLBaseFormatName(formats[fidx].formattype), + GetGLTypeName(formats[fidx].datatype)); + } } - } - - // If we passed all sizes (check versus size loop count): - - if (sidx == nsizes) { - log_info( "passed: Image write for GL format %s : %s : %s : %s\n\n", - GetGLTargetName( targets[ tidx ] ), - GetGLFormatName( formats[ fidx ].internal ), - GetGLBaseFormatName( formats[ fidx ].formattype ), - GetGLTypeName( formats[ fidx ].datatype ) ); - } } - } - return error; + return error; } diff --git a/test_conformance/gl/test_renderbuffer.cpp b/test_conformance/gl/test_renderbuffer.cpp index d75b4e84..422b5a3d 100644 --- a/test_conformance/gl/test_renderbuffer.cpp +++ b/test_conformance/gl/test_renderbuffer.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -15,93 +15,102 @@ // #include "testBase.h" -#if defined( __APPLE__ ) - #include +#if defined(__APPLE__) +#include #else - #include - #include +#include +#include #endif -#if defined (__linux__) -GLboolean -gluCheckExtension(const GLubyte *extension, const GLubyte *extensions) +#if defined(__linux__) +GLboolean gluCheckExtension(const GLubyte *extension, const GLubyte *extensions) { - const GLubyte *start; - GLubyte *where, *terminator; + const GLubyte *start; + GLubyte *where, *terminator; - /* Extension names should not have spaces. */ - where = (GLubyte *) strchr((const char*)extension, ' '); - if (where || *extension == '\0') + /* Extension names should not have spaces. */ + where = (GLubyte *)strchr((const char *)extension, ' '); + if (where || *extension == '\0') return 0; + /* It takes a bit of care to be fool-proof about parsing the + OpenGL extensions string. Don't be fooled by sub-strings, + etc. */ + start = extensions; + for (;;) + { + where = (GLubyte *)strstr((const char *)start, (const char *)extension); + if (!where) break; + terminator = where + strlen((const char *)extension); + if (where == start || *(where - 1) == ' ') + if (*terminator == ' ' || *terminator == '\0') return 1; + start = terminator; + } return 0; - /* It takes a bit of care to be fool-proof about parsing the - OpenGL extensions string. Don't be fooled by sub-strings, - etc. */ - start = extensions; - for (;;) { - where = (GLubyte *) strstr((const char *) start, (const char*) extension); - if (!where) - break; - terminator = where + strlen((const char*) extension); - if (where == start || *(where - 1) == ' ') - if (*terminator == ' ' || *terminator == '\0') - return 1; - start = terminator; - } - return 0; } #endif // This is defined in the write common code: -extern int test_cl_image_write( cl_context context, cl_command_queue queue, - GLenum target, cl_mem clImage, size_t width, size_t height, size_t depth, - cl_image_format *outFormat, ExplicitType *outType, void **outSourceBuffer, - MTdata d, bool supports_half ); +extern int test_cl_image_write(cl_context context, cl_command_queue queue, + GLenum target, cl_mem clImage, size_t width, + size_t height, size_t depth, + cl_image_format *outFormat, + ExplicitType *outType, void **outSourceBuffer, + MTdata d, bool supports_half); -extern int test_cl_image_read( cl_context context, cl_command_queue queue, - GLenum gl_target, cl_mem image, size_t width, size_t height, size_t depth, size_t sampleNum, - cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ); +extern int test_cl_image_read(cl_context context, cl_command_queue queue, + GLenum gl_target, cl_mem image, size_t width, + size_t height, size_t depth, size_t sampleNum, + cl_image_format *outFormat, ExplicitType *outType, + void **outResultBuffer); -extern int supportsHalf(cl_context context, bool* supports_half); +extern int supportsHalf(cl_context context, bool *supports_half); -static int test_attach_renderbuffer_read_image( cl_context context, cl_command_queue queue, GLenum glTarget, GLuint glRenderbuffer, - size_t imageWidth, size_t imageHeight, cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ) +static int test_attach_renderbuffer_read_image( + cl_context context, cl_command_queue queue, GLenum glTarget, + GLuint glRenderbuffer, size_t imageWidth, size_t imageHeight, + cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer) { int error; // Create a CL image from the supplied GL renderbuffer - cl_mem image = (*clCreateFromGLRenderbuffer_ptr)( context, CL_MEM_READ_ONLY, glRenderbuffer, &error ); - if( error != CL_SUCCESS ) + cl_mem image = (*clCreateFromGLRenderbuffer_ptr)(context, CL_MEM_READ_ONLY, + glRenderbuffer, &error); + if (error != CL_SUCCESS) { - print_error( error, "Unable to create CL image from GL renderbuffer" ); + print_error(error, "Unable to create CL image from GL renderbuffer"); return error; } - return test_cl_image_read( context, queue, glTarget, image, imageWidth, - imageHeight, 1, 1, outFormat, outType, outResultBuffer ); + return test_cl_image_read(context, queue, glTarget, image, imageWidth, + imageHeight, 1, 1, outFormat, outType, + outResultBuffer); } -int test_renderbuffer_read_image( cl_context context, cl_command_queue queue, - GLsizei width, GLsizei height, GLenum attachment, - GLenum format, GLenum internalFormat, - GLenum glType, ExplicitType type, MTdata d ) +int test_renderbuffer_read_image(cl_context context, cl_command_queue queue, + GLsizei width, GLsizei height, + GLenum attachment, GLenum format, + GLenum internalFormat, GLenum glType, + ExplicitType type, MTdata d) { int error; - if( type == kHalf ) - if( DetectFloatToHalfRoundingMode(queue) ) - return 1; + if (type == kHalf) + if (DetectFloatToHalfRoundingMode(queue)) return 1; // Create the GL renderbuffer glFramebufferWrapper glFramebuffer; glRenderbufferWrapper glRenderbuffer; - void *tmp = CreateGLRenderbuffer( width, height, attachment, format, internalFormat, glType, type, &glFramebuffer, &glRenderbuffer, &error, d, true ); + void *tmp = CreateGLRenderbuffer( + width, height, attachment, format, internalFormat, glType, type, + &glFramebuffer, &glRenderbuffer, &error, d, true); BufferOwningPtr inputBuffer(tmp); - if( error != 0 ) + if (error != 0) { - if ((format == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())) + if ((format == GL_RGBA_INTEGER_EXT) + && (!CheckGLIntegerExtensionSupport())) { - log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. Skipping test.\n"); + log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " + "Skipping test.\n"); return 0; } else @@ -114,14 +123,18 @@ int test_renderbuffer_read_image( cl_context context, cl_command_queue queue, cl_image_format clFormat; ExplicitType actualType; char *outBuffer; - error = test_attach_renderbuffer_read_image( context, queue, attachment, glRenderbuffer, width, height, &clFormat, &actualType, (void **)&outBuffer ); - if( error != 0 ) - return error; + error = test_attach_renderbuffer_read_image( + context, queue, attachment, glRenderbuffer, width, height, &clFormat, + &actualType, (void **)&outBuffer); + if (error != 0) return error; BufferOwningPtr actualResults(outBuffer); - log_info( "- Read [%4d x %4d] : GL renderbuffer : %s : %s : %s => CL Image : %s : %s \n", width, height, - GetGLFormatName( format ), GetGLFormatName( internalFormat ), GetGLTypeName( glType), - GetChannelOrderName( clFormat.image_channel_order ), GetChannelTypeName( clFormat.image_channel_data_type )); + log_info("- Read [%4d x %4d] : GL renderbuffer : %s : %s : %s => CL Image " + ": %s : %s \n", + width, height, GetGLFormatName(format), + GetGLFormatName(internalFormat), GetGLTypeName(glType), + GetChannelOrderName(clFormat.image_channel_order), + GetChannelTypeName(clFormat.image_channel_data_type)); #ifdef DEBUG log_info("- start read GL data -- \n"); @@ -129,63 +142,76 @@ int test_renderbuffer_read_image( cl_context context, cl_command_queue queue, log_info("- end read GL data -- \n"); #endif - // We have to convert our input buffer to the returned type, so we can validate. - BufferOwningPtr convertedInput(convert_to_expected( inputBuffer, width * height, type, actualType, get_channel_order_channel_count(clFormat.image_channel_order) )); + // We have to convert our input buffer to the returned type, so we can + // validate. + BufferOwningPtr convertedInput(convert_to_expected( + inputBuffer, width * height, type, actualType, + get_channel_order_channel_count(clFormat.image_channel_order))); #ifdef DEBUG log_info("- start input data -- \n"); - DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height, convertedInput); + DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height, + convertedInput); log_info("- end input data -- \n"); #endif #ifdef DEBUG log_info("- start converted data -- \n"); - DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height, actualResults); + DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height, + actualResults); log_info("- end converted data -- \n"); #endif // Now we validate int valid = 0; - if(convertedInput) { - if( actualType == kFloat ) - valid = validate_float_results( convertedInput, actualResults, width, height, 1, get_channel_order_channel_count(clFormat.image_channel_order) ); + if (convertedInput) + { + if (actualType == kFloat) + valid = validate_float_results( + convertedInput, actualResults, width, height, 1, + get_channel_order_channel_count(clFormat.image_channel_order)); else - valid = validate_integer_results( convertedInput, actualResults, width, height, 1, get_explicit_type_size( actualType ) ); + valid = validate_integer_results( + convertedInput, actualResults, width, height, 1, + get_explicit_type_size(actualType)); } return valid; } -int test_renderbuffer_read( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ) +int test_renderbuffer_read(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { GLenum attachments[] = { GL_COLOR_ATTACHMENT0_EXT }; - struct { + struct + { GLenum internal; GLenum format; GLenum datatype; ExplicitType type; } formats[] = { - { GL_RGBA, GL_BGRA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, - { GL_RGBA, GL_RGBA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, - { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, - { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, + { GL_RGBA, GL_BGRA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, + { GL_RGBA, GL_RGBA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, + { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, + { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, -// Renderbuffers with integer formats do not seem to work reliably across -// platforms/implementations. Disabling this in version 1.0 of CL conformance tests. + // Renderbuffers with integer formats do not seem to work reliably across + // platforms/implementations. Disabling this in version 1.0 of CL + // conformance tests. #ifdef TEST_INTEGER_FORMATS - { GL_RGBA8I_EXT, GL_RGBA_INTEGER_EXT, GL_BYTE, kChar }, - { GL_RGBA16I_EXT, GL_RGBA_INTEGER_EXT, GL_SHORT, kShort }, - { GL_RGBA32I_EXT, GL_RGBA_INTEGER_EXT, GL_INT, kInt }, - { GL_RGBA8UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_BYTE, kUChar }, - { GL_RGBA16UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_SHORT, kUShort }, - { GL_RGBA32UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_INT, kUInt }, + { GL_RGBA8I_EXT, GL_RGBA_INTEGER_EXT, GL_BYTE, kChar }, + { GL_RGBA16I_EXT, GL_RGBA_INTEGER_EXT, GL_SHORT, kShort }, + { GL_RGBA32I_EXT, GL_RGBA_INTEGER_EXT, GL_INT, kInt }, + { GL_RGBA8UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_BYTE, kUChar }, + { GL_RGBA16UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_SHORT, kUShort }, + { GL_RGBA32UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_INT, kUInt }, #endif - { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, - { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } + { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, + { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } }; size_t fmtIdx, attIdx; @@ -195,66 +221,70 @@ int test_renderbuffer_read( cl_device_id device, cl_context context, cl_command_ #else size_t iter = 6; #endif - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Check if images are supported - if (checkForImageSupport(device)) { - log_info("Device does not support images. Skipping test.\n"); - return 0; - } - - if( !gluCheckExtension( (const GLubyte *)"GL_EXT_framebuffer_object", glGetString( GL_EXTENSIONS ) ) ) + // Check if images are supported + if (checkForImageSupport(device)) { - log_info( "Renderbuffers are not supported by this OpenGL implementation; skipping test\n" ); + log_info("Device does not support images. Skipping test.\n"); + return 0; + } + + if (!gluCheckExtension((const GLubyte *)"GL_EXT_framebuffer_object", + glGetString(GL_EXTENSIONS))) + { + log_info("Renderbuffers are not supported by this OpenGL " + "implementation; skipping test\n"); return 0; } // Loop through a set of GL formats, testing a set of sizes against each one - for( fmtIdx = 0; fmtIdx < sizeof( formats ) / sizeof( formats[ 0 ] ); fmtIdx++ ) + for (fmtIdx = 0; fmtIdx < sizeof(formats) / sizeof(formats[0]); fmtIdx++) { - for( attIdx = 0; attIdx < sizeof( attachments ) / sizeof( attachments[ 0 ] ); attIdx++ ) + for (attIdx = 0; attIdx < sizeof(attachments) / sizeof(attachments[0]); + attIdx++) { size_t i; - log_info( "Testing renderbuffer read for %s : %s : %s : %s\n", - GetGLAttachmentName( attachments[ attIdx ] ), - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_info("Testing renderbuffer read for %s : %s : %s : %s\n", + GetGLAttachmentName(attachments[attIdx]), + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); - for( i = 0; i < iter; i++ ) + for (i = 0; i < iter; i++) { - GLsizei width = random_in_range( 16, 512, seed ); - GLsizei height = random_in_range( 16, 512, seed ); + GLsizei width = random_in_range(16, 512, seed); + GLsizei height = random_in_range(16, 512, seed); #ifdef DEBUG width = height = 4; #endif - if( test_renderbuffer_read_image( context, queue, width, height, - attachments[ attIdx ], - formats[ fmtIdx ].format, - formats[ fmtIdx ].internal, - formats[ fmtIdx ].datatype, - formats[ fmtIdx ].type, seed ) ) + if (test_renderbuffer_read_image( + context, queue, width, height, attachments[attIdx], + formats[fmtIdx].format, formats[fmtIdx].internal, + formats[fmtIdx].datatype, formats[fmtIdx].type, seed)) { - log_error( "ERROR: Renderbuffer read test failed for %s : %s : %s : %s\n\n", - GetGLAttachmentName( attachments[ attIdx ] ), - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_error("ERROR: Renderbuffer read test failed for %s : " + "%s : %s : %s\n\n", + GetGLAttachmentName(attachments[attIdx]), + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); error++; - break; // Skip other sizes for this combination + break; // Skip other sizes for this combination } } - if( i == iter ) + if (i == iter) { - log_info( "passed: Renderbuffer read test passed for %s : %s : %s : %s\n\n", - GetGLAttachmentName( attachments[ attIdx ] ), - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_info("passed: Renderbuffer read test passed for %s : %s : " + "%s : %s\n\n", + GetGLAttachmentName(attachments[attIdx]), + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); } } } @@ -265,43 +295,52 @@ int test_renderbuffer_read( cl_device_id device, cl_context context, cl_command_ #pragma mark -------------------- Write tests ------------------------- -int test_attach_renderbuffer_write_to_image( cl_context context, cl_command_queue queue, GLenum glTarget, GLuint glRenderbuffer, - size_t imageWidth, size_t imageHeight, cl_image_format *outFormat, ExplicitType *outType, MTdata d, void **outSourceBuffer, bool supports_half ) +int test_attach_renderbuffer_write_to_image( + cl_context context, cl_command_queue queue, GLenum glTarget, + GLuint glRenderbuffer, size_t imageWidth, size_t imageHeight, + cl_image_format *outFormat, ExplicitType *outType, MTdata d, + void **outSourceBuffer, bool supports_half) { int error; // Create a CL image from the supplied GL renderbuffer - clMemWrapper image = (*clCreateFromGLRenderbuffer_ptr)( context, CL_MEM_WRITE_ONLY, glRenderbuffer, &error ); - if( error != CL_SUCCESS ) + clMemWrapper image = (*clCreateFromGLRenderbuffer_ptr)( + context, CL_MEM_WRITE_ONLY, glRenderbuffer, &error); + if (error != CL_SUCCESS) { - print_error( error, "Unable to create CL image from GL renderbuffer" ); + print_error(error, "Unable to create CL image from GL renderbuffer"); return error; } - return test_cl_image_write( context, queue, glTarget, image, imageWidth, - imageHeight, 1, outFormat, outType, outSourceBuffer, d, supports_half ); + return test_cl_image_write(context, queue, glTarget, image, imageWidth, + imageHeight, 1, outFormat, outType, + outSourceBuffer, d, supports_half); } -int test_renderbuffer_image_write( cl_context context, cl_command_queue queue, - GLsizei width, GLsizei height, GLenum attachment, - GLenum format, GLenum internalFormat, - GLenum glType, ExplicitType type, MTdata d ) +int test_renderbuffer_image_write(cl_context context, cl_command_queue queue, + GLsizei width, GLsizei height, + GLenum attachment, GLenum format, + GLenum internalFormat, GLenum glType, + ExplicitType type, MTdata d) { int error; - if( type == kHalf ) - if( DetectFloatToHalfRoundingMode(queue) ) - return 1; + if (type == kHalf) + if (DetectFloatToHalfRoundingMode(queue)) return 1; // Create the GL renderbuffer glFramebufferWrapper glFramebuffer; glRenderbufferWrapper glRenderbuffer; - CreateGLRenderbuffer( width, height, attachment, format, internalFormat, glType, type, &glFramebuffer, &glRenderbuffer, &error, d, false ); - if( error != 0 ) + CreateGLRenderbuffer(width, height, attachment, format, internalFormat, + glType, type, &glFramebuffer, &glRenderbuffer, &error, + d, false); + if (error != 0) { - if ((format == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())) + if ((format == GL_RGBA_INTEGER_EXT) + && (!CheckGLIntegerExtensionSupport())) { - log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. Skipping test.\n"); + log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " + "Skipping test.\n"); return 0; } else @@ -318,27 +357,34 @@ int test_renderbuffer_image_write( cl_context context, cl_command_queue queue, bool supports_half = false; error = supportsHalf(context, &supports_half); - if( error != 0 ) - return error; + if (error != 0) return error; - error = test_attach_renderbuffer_write_to_image( context, queue, attachment, glRenderbuffer, width, height, &clFormat, &sourceType, d, (void **)&outSourceBuffer, supports_half ); - if( error != 0 || ((sourceType == kHalf ) && !supports_half)) - return error; + error = test_attach_renderbuffer_write_to_image( + context, queue, attachment, glRenderbuffer, width, height, &clFormat, + &sourceType, d, (void **)&outSourceBuffer, supports_half); + if (error != 0 || ((sourceType == kHalf) && !supports_half)) return error; // If actual source type was half, convert to float for validation. - if( sourceType == kHalf ) + if (sourceType == kHalf) validationType = kFloat; else validationType = sourceType; - BufferOwningPtr validationSource( convert_to_expected( outSourceBuffer, width * height, sourceType, validationType, get_channel_order_channel_count(clFormat.image_channel_order) ) ); + BufferOwningPtr validationSource(convert_to_expected( + outSourceBuffer, width * height, sourceType, validationType, + get_channel_order_channel_count(clFormat.image_channel_order))); - log_info( "- Write [%4d x %4d] : GL Renderbuffer : %s : %s : %s => CL Image : %s : %s \n", width, height, - GetGLFormatName( format ), GetGLFormatName( internalFormat ), GetGLTypeName( glType), - GetChannelOrderName( clFormat.image_channel_order ), GetChannelTypeName( clFormat.image_channel_data_type )); + log_info("- Write [%4d x %4d] : GL Renderbuffer : %s : %s : %s => CL Image " + ": %s : %s \n", + width, height, GetGLFormatName(format), + GetGLFormatName(internalFormat), GetGLTypeName(glType), + GetChannelOrderName(clFormat.image_channel_order), + GetChannelTypeName(clFormat.image_channel_data_type)); // Now read the results from the GL renderbuffer - BufferOwningPtr resultData( ReadGLRenderbuffer( glFramebuffer, glRenderbuffer, attachment, format, internalFormat, glType, type, width, height ) ); + BufferOwningPtr resultData( + ReadGLRenderbuffer(glFramebuffer, glRenderbuffer, attachment, format, + internalFormat, glType, type, width, height)); #ifdef DEBUG log_info("- start result data -- \n"); @@ -346,63 +392,76 @@ int test_renderbuffer_image_write( cl_context context, cl_command_queue queue, log_info("- end result data -- \n"); #endif - // We have to convert our input buffer to the returned type, so we can validate. - BufferOwningPtr convertedData( convert_to_expected( resultData, width * height, type, validationType, get_channel_order_channel_count(clFormat.image_channel_order) ) ); + // We have to convert our input buffer to the returned type, so we can + // validate. + BufferOwningPtr convertedData(convert_to_expected( + resultData, width * height, type, validationType, + get_channel_order_channel_count(clFormat.image_channel_order))); #ifdef DEBUG log_info("- start input data -- \n"); - DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height, validationSource); + DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height, + validationSource); log_info("- end input data -- \n"); #endif #ifdef DEBUG log_info("- start converted data -- \n"); - DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height, convertedData); + DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height, + convertedData); log_info("- end converted data -- \n"); #endif // Now we validate int valid = 0; - if(convertedData) { - if( sourceType == kFloat || sourceType == kHalf ) - valid = validate_float_results( validationSource, convertedData, width, height, 1, get_channel_order_channel_count(clFormat.image_channel_order) ); + if (convertedData) + { + if (sourceType == kFloat || sourceType == kHalf) + valid = validate_float_results( + validationSource, convertedData, width, height, 1, + get_channel_order_channel_count(clFormat.image_channel_order)); else - valid = validate_integer_results( validationSource, convertedData, width, height, 1, get_explicit_type_size( type ) ); + valid = validate_integer_results(validationSource, convertedData, + width, height, 1, + get_explicit_type_size(type)); } return valid; } -int test_renderbuffer_write( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ) +int test_renderbuffer_write(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { GLenum attachments[] = { GL_COLOR_ATTACHMENT0_EXT }; - struct { + struct + { GLenum internal; GLenum format; GLenum datatype; ExplicitType type; } formats[] = { - { GL_RGBA, GL_BGRA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, - { GL_RGBA, GL_RGBA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, - { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, - { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, + { GL_RGBA, GL_BGRA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, + { GL_RGBA, GL_RGBA, GL_UNSIGNED_INT_8_8_8_8_REV, kUChar }, + { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, + { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, -// Renderbuffers with integer formats do not seem to work reliably across -// platforms/implementations. Disabling this in version 1.0 of CL conformance tests. + // Renderbuffers with integer formats do not seem to work reliably across + // platforms/implementations. Disabling this in version 1.0 of CL + // conformance tests. #ifdef TEST_INTEGER_FORMATS - { GL_RGBA8I_EXT, GL_RGBA_INTEGER_EXT, GL_BYTE, kChar }, - { GL_RGBA16I_EXT, GL_RGBA_INTEGER_EXT, GL_SHORT, kShort }, - { GL_RGBA32I_EXT, GL_RGBA_INTEGER_EXT, GL_INT, kInt }, - { GL_RGBA8UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_BYTE, kUChar }, - { GL_RGBA16UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_SHORT, kUShort }, - { GL_RGBA32UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_INT, kUInt }, + { GL_RGBA8I_EXT, GL_RGBA_INTEGER_EXT, GL_BYTE, kChar }, + { GL_RGBA16I_EXT, GL_RGBA_INTEGER_EXT, GL_SHORT, kShort }, + { GL_RGBA32I_EXT, GL_RGBA_INTEGER_EXT, GL_INT, kInt }, + { GL_RGBA8UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_BYTE, kUChar }, + { GL_RGBA16UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_SHORT, kUShort }, + { GL_RGBA32UI_EXT, GL_RGBA_INTEGER_EXT, GL_UNSIGNED_INT, kUInt }, #endif - { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, - { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } + { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, + { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } }; size_t fmtIdx, attIdx; @@ -411,64 +470,68 @@ int test_renderbuffer_write( cl_device_id device, cl_context context, cl_command #ifdef DEBUG iter = 1; #endif - RandomSeed seed( gRandomSeed ); + RandomSeed seed(gRandomSeed); - // Check if images are supported - if (checkForImageSupport(device)) { - log_info("Device does not support images. Skipping test.\n"); - return 0; - } - - if( !gluCheckExtension( (const GLubyte *)"GL_EXT_framebuffer_object", glGetString( GL_EXTENSIONS ) ) ) + // Check if images are supported + if (checkForImageSupport(device)) { - log_info( "Renderbuffers are not supported by this OpenGL implementation; skipping test\n" ); + log_info("Device does not support images. Skipping test.\n"); + return 0; + } + + if (!gluCheckExtension((const GLubyte *)"GL_EXT_framebuffer_object", + glGetString(GL_EXTENSIONS))) + { + log_info("Renderbuffers are not supported by this OpenGL " + "implementation; skipping test\n"); return 0; } // Loop through a set of GL formats, testing a set of sizes against each one - for( fmtIdx = 0; fmtIdx < sizeof( formats ) / sizeof( formats[ 0 ] ); fmtIdx++ ) + for (fmtIdx = 0; fmtIdx < sizeof(formats) / sizeof(formats[0]); fmtIdx++) { - for( attIdx = 0; attIdx < sizeof( attachments ) / sizeof( attachments[ 0 ] ); attIdx++ ) + for (attIdx = 0; attIdx < sizeof(attachments) / sizeof(attachments[0]); + attIdx++) { - log_info( "Testing Renderbuffer write test for %s : %s : %s : %s\n", - GetGLAttachmentName( attachments[ attIdx ] ), - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_info("Testing Renderbuffer write test for %s : %s : %s : %s\n", + GetGLAttachmentName(attachments[attIdx]), + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); size_t i; - for( i = 0; i < iter; i++ ) + for (i = 0; i < iter; i++) { - GLsizei width = random_in_range( 16, 512, seed ); - GLsizei height = random_in_range( 16, 512, seed ); + GLsizei width = random_in_range(16, 512, seed); + GLsizei height = random_in_range(16, 512, seed); #ifdef DEBUG width = height = 4; #endif - if( test_renderbuffer_image_write( context, queue, width, height, - attachments[ attIdx ], - formats[ fmtIdx ].format, - formats[ fmtIdx ].internal, - formats[ fmtIdx ].datatype, - formats[ fmtIdx ].type, seed ) ) + if (test_renderbuffer_image_write( + context, queue, width, height, attachments[attIdx], + formats[fmtIdx].format, formats[fmtIdx].internal, + formats[fmtIdx].datatype, formats[fmtIdx].type, seed)) { - log_error( "ERROR: Renderbuffer write test failed for %s : %s : %s : %s\n\n", - GetGLAttachmentName( attachments[ attIdx ] ), - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_error("ERROR: Renderbuffer write test failed for %s : " + "%s : %s : %s\n\n", + GetGLAttachmentName(attachments[attIdx]), + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); error++; - break; // Skip other sizes for this combination + break; // Skip other sizes for this combination } } - if( i == iter ) + if (i == iter) { - log_info( "passed: Renderbuffer write test passed for %s : %s : %s : %s\n\n", - GetGLAttachmentName( attachments[ attIdx ] ), - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_info("passed: Renderbuffer write test passed for %s : %s : " + "%s : %s\n\n", + GetGLAttachmentName(attachments[attIdx]), + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); } } } diff --git a/test_conformance/gl/test_renderbuffer_info.cpp b/test_conformance/gl/test_renderbuffer_info.cpp index bb5ce848..d14b6032 100644 --- a/test_conformance/gl/test_renderbuffer_info.cpp +++ b/test_conformance/gl/test_renderbuffer_info.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -14,118 +14,125 @@ // limitations under the License. // #include "testBase.h" -#if defined( __APPLE__ ) +#if defined(__APPLE__) #include #else #include #include #endif -static int test_renderbuffer_object_info( cl_context context, cl_command_queue queue, - GLsizei width, GLsizei height, GLenum attachment, - GLenum format, GLenum internalFormat, - GLenum glType, ExplicitType type, MTdata d ) +static int test_renderbuffer_object_info(cl_context context, + cl_command_queue queue, GLsizei width, + GLsizei height, GLenum attachment, + GLenum format, GLenum internalFormat, + GLenum glType, ExplicitType type, + MTdata d) { int error; - if( type == kHalf ) - if( DetectFloatToHalfRoundingMode(queue) ) - return 1; + if (type == kHalf) + if (DetectFloatToHalfRoundingMode(queue)) return 1; // Create the GL render buffer glFramebufferWrapper glFramebuffer; glRenderbufferWrapper glRenderbuffer; - BufferOwningPtr inputBuffer(CreateGLRenderbuffer( width, height, attachment, format, internalFormat, glType, type, &glFramebuffer, &glRenderbuffer, &error, d, true )); - if( error != 0 ) - return error; + BufferOwningPtr inputBuffer(CreateGLRenderbuffer( + width, height, attachment, format, internalFormat, glType, type, + &glFramebuffer, &glRenderbuffer, &error, d, true)); + if (error != 0) return error; - clMemWrapper image = (*clCreateFromGLRenderbuffer_ptr)(context, CL_MEM_READ_ONLY, glRenderbuffer, &error); + clMemWrapper image = (*clCreateFromGLRenderbuffer_ptr)( + context, CL_MEM_READ_ONLY, glRenderbuffer, &error); test_error(error, "clCreateFromGLRenderbuffer failed"); - log_info( "- Given a GL format of %s, input type was %s, size was %d x %d\n", - GetGLFormatName( internalFormat ), - get_explicit_type_name( type ), (int)width, (int)height ); + log_info("- Given a GL format of %s, input type was %s, size was %d x %d\n", + GetGLFormatName(internalFormat), get_explicit_type_name(type), + (int)width, (int)height); // Verify the expected information here. - return CheckGLObjectInfo(image, CL_GL_OBJECT_RENDERBUFFER, (GLuint)glRenderbuffer, internalFormat, 0); + return CheckGLObjectInfo(image, CL_GL_OBJECT_RENDERBUFFER, + (GLuint)glRenderbuffer, internalFormat, 0); } -int test_renderbuffer_getinfo( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ) +int test_renderbuffer_getinfo(cl_device_id device, cl_context context, + cl_command_queue queue, int numElements) { GLenum attachments[] = { GL_COLOR_ATTACHMENT0_EXT }; - struct { + struct + { GLenum internal; GLenum format; GLenum datatype; ExplicitType type; - } formats[] = { - { GL_RGBA, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, - { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, - { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, - { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, - { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } - }; + } formats[] = { { GL_RGBA, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, + { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, + { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, + { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, + { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } }; size_t fmtIdx, tgtIdx; int error = 0; size_t iter = 6; RandomSeed seed(gRandomSeed); - // Check if images are supported - if (checkForImageSupport(device)) { - log_info("Device does not support images. Skipping test.\n"); - return 0; - } - - if( !gluCheckExtension( (const GLubyte *)"GL_EXT_framebuffer_object", glGetString( GL_EXTENSIONS ) ) ) + // Check if images are supported + if (checkForImageSupport(device)) { - log_info( "Renderbuffers are not supported by this OpenGL implementation; skipping test\n" ); + log_info("Device does not support images. Skipping test.\n"); + return 0; + } + + if (!gluCheckExtension((const GLubyte *)"GL_EXT_framebuffer_object", + glGetString(GL_EXTENSIONS))) + { + log_info("Renderbuffers are not supported by this OpenGL " + "implementation; skipping test\n"); return 0; } // Loop through a set of GL formats, testing a set of sizes against each one - for( fmtIdx = 0; fmtIdx < sizeof( formats ) / sizeof( formats[ 0 ] ); fmtIdx++ ) + for (fmtIdx = 0; fmtIdx < sizeof(formats) / sizeof(formats[0]); fmtIdx++) { - for( tgtIdx = 0; tgtIdx < sizeof( attachments ) / sizeof( attachments[ 0 ] ); tgtIdx++ ) + for (tgtIdx = 0; tgtIdx < sizeof(attachments) / sizeof(attachments[0]); + tgtIdx++) { - log_info( "Testing Renderbuffer object info for %s : %s : %s\n", - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLBaseFormatName( formats[ fmtIdx ].format ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_info("Testing Renderbuffer object info for %s : %s : %s\n", + GetGLFormatName(formats[fmtIdx].internal), + GetGLBaseFormatName(formats[fmtIdx].format), + GetGLTypeName(formats[fmtIdx].datatype)); size_t i; - for( i = 0; i < iter; i++ ) + for (i = 0; i < iter; i++) { - GLsizei width = random_in_range( 16, 512, seed ); - GLsizei height = random_in_range( 16, 512, seed ); + GLsizei width = random_in_range(16, 512, seed); + GLsizei height = random_in_range(16, 512, seed); - if( test_renderbuffer_object_info( context, queue, (int)width, (int)height, - attachments[ tgtIdx ], - formats[ fmtIdx ].format, - formats[ fmtIdx ].internal, - formats[ fmtIdx ].datatype, - formats[ fmtIdx ].type, seed ) ) + if (test_renderbuffer_object_info( + context, queue, (int)width, (int)height, + attachments[tgtIdx], formats[fmtIdx].format, + formats[fmtIdx].internal, formats[fmtIdx].datatype, + formats[fmtIdx].type, seed)) { - log_error( "ERROR: Renderbuffer write test failed for GL format %s : %s\n\n", - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); + log_error("ERROR: Renderbuffer write test failed for GL " + "format %s : %s\n\n", + GetGLFormatName(formats[fmtIdx].internal), + GetGLTypeName(formats[fmtIdx].datatype)); error++; - break; // Skip other sizes for this combination + break; // Skip other sizes for this combination } } - if( i == iter ) + if (i == iter) { - log_info( "passed: Renderbuffer write test passed for GL format %s : %s\n\n", - GetGLFormatName( formats[ fmtIdx ].internal ), - GetGLTypeName( formats[ fmtIdx ].datatype ) ); - + log_info("passed: Renderbuffer write test passed for GL format " + "%s : %s\n\n", + GetGLFormatName(formats[fmtIdx].internal), + GetGLTypeName(formats[fmtIdx].datatype)); } } } return error; } -