ahmed/OpenCL-CTS
1
0
Fork 0
mirror of https://github.com/KhronosGroup/OpenCL-CTS.git synced 2026-03-26 00:39:03 +00:00
Code Packages Projects Releases Wiki Activity

[NFC] clang-format gl (#1612)

Browse Source 
Add some clang-format off/on comments to keep kernel code readable.

Signed-off-by: Sven van Haastregt <sven.vanhaastregt@arm.com>
This commit is contained in:
Sven van Haastregt
2023-02-06 15:09:04 +00:00
committed by GitHub
parent 2318cedb21
commit f46cca0f8f
19 changed files with 3497 additions and 3027 deletions
Download Patch File Download Diff File Expand all files Collapse all files

8
test_conformance/gl/common.h
View File

@@ -18,13 +18,15 @@
#include "testBase.h" #include "testBase.h"
typedef struct { typedef struct
{
size_t width; size_t width;
size_t height; size_t height;
size_t depth; size_t depth;
} sizevec_t; } sizevec_t;
struct format { struct format
{
GLenum internal; GLenum internal;
GLenum formattype; GLenum formattype;
GLenum datatype; GLenum datatype;
@@ -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 ntargets, sizevec_t *sizes,
size_t nsizes); 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__ #endif // __COMMON_H__

605
test_conformance/gl/helpers.cpp
View File

@@ -14,15 +14,15 @@
// limitations under the License. // limitations under the License.
// //
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#endif #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_INT8:
case CL_UNORM_INT16: case CL_UNORM_INT16:
@@ -31,25 +31,24 @@ const char *get_kernel_suffix( cl_image_format *format )
case CL_SNORM_INT16: case CL_SNORM_INT16:
case CL_HALF_FLOAT: case CL_HALF_FLOAT:
case CL_FLOAT: case CL_FLOAT:
case CL_UNORM_INT_101010: case CL_UNORM_INT_101010: return "f";
return "f";
case CL_SIGNED_INT8: case CL_SIGNED_INT8:
case CL_SIGNED_INT16: case CL_SIGNED_INT16:
case CL_SIGNED_INT32: case CL_SIGNED_INT32: return "i";
return "i";
case CL_UNSIGNED_INT8: case CL_UNSIGNED_INT8:
case CL_UNSIGNED_INT16: case CL_UNSIGNED_INT16:
case CL_UNSIGNED_INT32: case CL_UNSIGNED_INT32: return "ui";
return "ui";
default: 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 ""; 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_INT8:
case CL_UNORM_INT16: case CL_UNORM_INT16:
@@ -65,62 +64,46 @@ ExplicitType get_read_kernel_type( cl_image_format *format )
return kFloat; return kFloat;
case CL_SIGNED_INT8: case CL_SIGNED_INT8:
case CL_SIGNED_INT16: case CL_SIGNED_INT16:
case CL_SIGNED_INT32: case CL_SIGNED_INT32: return kInt;
return kInt;
case CL_UNSIGNED_INT8: case CL_UNSIGNED_INT8:
case CL_UNSIGNED_INT16: case CL_UNSIGNED_INT16:
case CL_UNSIGNED_INT32: case CL_UNSIGNED_INT32: return kUInt;
return kUInt;
default: 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; 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: case CL_UNORM_INT8: return kFloat;
return kFloat; case CL_UNORM_INT16: return kFloat;
case CL_UNORM_INT16: case CL_UNORM_INT24: return kFloat;
return kFloat; case CL_SNORM_INT8: return kFloat;
case CL_UNORM_INT24: case CL_SNORM_INT16: return kFloat;
return kFloat; case CL_HALF_FLOAT: return kHalf;
case CL_SNORM_INT8: case CL_FLOAT: return kFloat;
return kFloat; case CL_SIGNED_INT8: return kChar;
case CL_SNORM_INT16: case CL_SIGNED_INT16: return kShort;
return kFloat; case CL_SIGNED_INT32: return kInt;
case CL_HALF_FLOAT: case CL_UNSIGNED_INT8: return kUChar;
return kHalf; case CL_UNSIGNED_INT16: return kUShort;
case CL_FLOAT: case CL_UNSIGNED_INT32: return kUInt;
return kFloat; case CL_UNORM_INT_101010: 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 #ifdef GL_VERSION_3_2
case CL_DEPTH: case CL_DEPTH: return kFloat;
return kFloat;
#endif #endif
default: default: return kInt;
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_INT8:
case CL_UNORM_INT16: case CL_UNORM_INT16:
@@ -130,250 +113,268 @@ const char* get_write_conversion( cl_image_format *format, ExplicitType type )
case CL_FLOAT: case CL_FLOAT:
case CL_UNORM_INT_101010: case CL_UNORM_INT_101010:
case CL_UNORM_INT24: case CL_UNORM_INT24:
if(type != kFloat) return "convert_float4"; if (type != kFloat) return "convert_float4";
break; break;
case CL_SIGNED_INT8: case CL_SIGNED_INT8:
case CL_SIGNED_INT16: case CL_SIGNED_INT16:
case CL_SIGNED_INT32: case CL_SIGNED_INT32:
if(type != kInt) return "convert_int4"; if (type != kInt) return "convert_int4";
break; break;
case CL_UNSIGNED_INT8: case CL_UNSIGNED_INT8:
case CL_UNSIGNED_INT16: case CL_UNSIGNED_INT16:
case CL_UNSIGNED_INT32: case CL_UNSIGNED_INT32:
if(type != kUInt) return "convert_uint4"; if (type != kUInt) return "convert_uint4";
break; break;
default: default: return "";
return "";
} }
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 only three input types to this function are kInt, kUInt and kFloat, due
// The output types, though, are pretty much anything valid for GL to receive // 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 ) \ #define DOWNSCALE_INTEGER_CASE(enum, type, bitShift) \
case enum: \ case enum: { \
{ \ cl_##type *dst = new cl_##type[numPixels * 4]; \
cl_##type *dst = new cl_##type[ numPixels * 4 ]; \ for (size_t i = 0; i < numPixels * 4; i++) dst[i] = src[i]; \
for( size_t i = 0; i < numPixels * 4; i++ ) \
dst[ i ] = src[ i ]; \
return (char *)dst; \ return (char *)dst; \
} }
#define UPSCALE_FLOAT_CASE( enum, type, typeMax ) \ #define UPSCALE_FLOAT_CASE(enum, type, typeMax) \
case enum: \ case enum: { \
{ \ cl_##type *dst = new cl_##type[numPixels * 4]; \
cl_##type *dst = new cl_##type[ numPixels * 4 ]; \ for (size_t i = 0; i < numPixels * 4; i++) \
for( size_t i = 0; i < numPixels * 4; i++ ) \ dst[i] = (cl_##type)(src[i] * typeMax); \
dst[ i ] = (cl_##type)( src[ i ] * typeMax ); \
return (char *)dst; \ 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 #ifdef DEBUG
log_info( "- Converting from input type '%s' to output type '%s'\n", log_info("- Converting from input type '%s' to output type '%s'\n",
get_explicit_type_name( inType ), get_explicit_type_name( outType ) ); get_explicit_type_name(inType), get_explicit_type_name(outType));
#endif #endif
if( inType == outType ) if (inType == outType)
{ {
char *outData = new char[ numPixels * channelNum * get_explicit_type_size(outType) ] ; // sizeof( cl_int ) ]; char *outData =
if (glDataType == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) { new char[numPixels * channelNum
for (size_t i = 0; i < numPixels; ++i) { * get_explicit_type_size(outType)]; // sizeof( cl_int ) ];
((cl_float*)outData)[i] = ((cl_float*)inputBuffer)[2 * i]; 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 { else
memcpy( outData, inputBuffer, numPixels * channelNum * get_explicit_type_size(inType) ); {
memcpy(outData, inputBuffer,
numPixels * channelNum * get_explicit_type_size(inType));
} }
return outData; return outData;
} }
else if( inType == kChar ) else if (inType == kChar)
{ {
cl_char *src = (cl_char *)inputBuffer; cl_char *src = (cl_char *)inputBuffer;
switch( outType ) switch (outType)
{ {
case kInt: case kInt: {
cl_int *outData = new cl_int[numPixels * channelNum];
for (size_t i = 0; i < numPixels * channelNum; i++)
{ {
cl_int *outData = new cl_int[ numPixels * channelNum ]; outData[i] = (cl_int)((src[i]));
for( size_t i = 0; i < numPixels * channelNum; i++ )
{
outData[ i ] = (cl_int)((src[ i ]));
} }
return (char *)outData; return (char *)outData;
} }
case kFloat: 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++)
{ {
// If we're converting to float, then CL decided that we should be normalized outData[i] = (cl_float)src[i] / 127.0f;
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;
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
else if( inType == kUChar ) else if (inType == kUChar)
{ {
cl_uchar *src = (cl_uchar *)inputBuffer; cl_uchar *src = (cl_uchar *)inputBuffer;
switch( outType ) switch (outType)
{ {
case kUInt: case kUInt: {
cl_uint *outData = new cl_uint[numPixels * channelNum];
for (size_t i = 0; i < numPixels * channelNum; i++)
{ {
cl_uint *outData = new cl_uint[ numPixels * channelNum ]; outData[i] = (cl_uint)((src[i]));
for( size_t i = 0; i < numPixels * channelNum; i++ )
{
outData[ i ] = (cl_uint)((src[ i ]));
} }
return (char *)outData; return (char *)outData;
} }
case kFloat: 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++)
{ {
// If we're converting to float, then CL decided that we should be normalized outData[i] = (cl_float)(src[i]) / 256.0f;
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;
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
else if( inType == kShort ) else if (inType == kShort)
{ {
cl_short *src = (cl_short *)inputBuffer; cl_short *src = (cl_short *)inputBuffer;
switch( outType ) switch (outType)
{ {
case kInt: case kInt: {
cl_int *outData = new cl_int[numPixels * channelNum];
for (size_t i = 0; i < numPixels * channelNum; i++)
{ {
cl_int *outData = new cl_int[ numPixels * channelNum ]; outData[i] = (cl_int)((src[i]));
for( size_t i = 0; i < numPixels * channelNum; i++ )
{
outData[ i ] = (cl_int)((src[ i ]));
} }
return (char *)outData; return (char *)outData;
} }
case kFloat: 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++)
{ {
// If we're converting to float, then CL decided that we should be normalized outData[i] = (cl_float)src[i] / 32768.0f;
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;
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
else if( inType == kUShort ) else if (inType == kUShort)
{ {
cl_ushort *src = (cl_ushort *)inputBuffer; cl_ushort *src = (cl_ushort *)inputBuffer;
switch( outType ) switch (outType)
{ {
case kUInt: case kUInt: {
cl_uint *outData = new cl_uint[numPixels * channelNum];
for (size_t i = 0; i < numPixels * channelNum; i++)
{ {
cl_uint *outData = new cl_uint[ numPixels * channelNum ]; outData[i] = (cl_uint)((src[i]));
for( size_t i = 0; i < numPixels * channelNum; i++ )
{
outData[ i ] = (cl_uint)((src[ i ]));
} }
return (char *)outData; return (char *)outData;
} }
case kFloat: 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++)
{ {
// If we're converting to float, then CL decided that we should be normalized outData[i] = (cl_float)(src[i]) / 65535.0f;
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;
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
else if( inType == kInt ) else if (inType == kInt)
{ {
cl_int *src = (cl_int *)inputBuffer; cl_int *src = (cl_int *)inputBuffer;
switch( outType ) switch (outType)
{ {
DOWNSCALE_INTEGER_CASE( kShort, short, 16 ) DOWNSCALE_INTEGER_CASE(kShort, short, 16)
DOWNSCALE_INTEGER_CASE( kChar, char, 24 ) DOWNSCALE_INTEGER_CASE(kChar, char, 24)
case kFloat: 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++)
{ {
// If we're converting to float, then CL decided that we should be normalized outData[i] =
cl_float *outData = new cl_float[ numPixels * channelNum ]; (cl_float)fmaxf((float)src[i] / 2147483647.f, -1.f);
for( size_t i = 0; i < numPixels * channelNum; i++ )
{
outData[ i ] = (cl_float)fmaxf( (float)src[ i ] / 2147483647.f, -1.f );
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
else if( inType == kUInt ) else if (inType == kUInt)
{ {
cl_uint *src = (cl_uint *)inputBuffer; cl_uint *src = (cl_uint *)inputBuffer;
switch( outType ) switch (outType)
{ {
DOWNSCALE_INTEGER_CASE( kUShort, ushort, 16 ) DOWNSCALE_INTEGER_CASE(kUShort, ushort, 16)
DOWNSCALE_INTEGER_CASE( kUChar, uchar, 24 ) DOWNSCALE_INTEGER_CASE(kUChar, uchar, 24)
case kFloat: 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++)
{ {
// If we're converting to float, then CL decided that we should be normalized outData[i] = (cl_float)(src[i] >> ShiftBits) / MaxValue;
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;
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
else if( inType == kHalf ) else if (inType == kHalf)
{ {
cl_half *src = (cl_half *)inputBuffer; cl_half *src = (cl_half *)inputBuffer;
switch( outType ) switch (outType)
{ {
case kFloat: case kFloat: {
{ cl_float *outData = new cl_float[numPixels * channelNum];
cl_float *outData = new cl_float[ numPixels * channelNum ]; for (size_t i = 0; i < numPixels * channelNum; i++)
for( size_t i = 0; i < numPixels * channelNum; i++ )
{ {
outData[i] = cl_half_to_float(src[i]); outData[i] = cl_half_to_float(src[i]);
} }
return (char *)outData; return (char *)outData;
} }
default: 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; return NULL;
} }
} }
@@ -381,16 +382,18 @@ char * convert_to_expected( void * inputBuffer, size_t numPixels, ExplicitType i
{ {
cl_float *src = (cl_float *)inputBuffer; cl_float *src = (cl_float *)inputBuffer;
switch( outType ) switch (outType)
{ {
UPSCALE_FLOAT_CASE( kChar, char, 127.f ) UPSCALE_FLOAT_CASE(kChar, char, 127.f)
UPSCALE_FLOAT_CASE( kUChar, uchar, 255.f ) UPSCALE_FLOAT_CASE(kUChar, uchar, 255.f)
UPSCALE_FLOAT_CASE( kShort, short, 32767.f ) UPSCALE_FLOAT_CASE(kShort, short, 32767.f)
UPSCALE_FLOAT_CASE( kUShort, ushort, 65535.f ) UPSCALE_FLOAT_CASE(kUShort, ushort, 65535.f)
UPSCALE_FLOAT_CASE( kInt, int, 2147483647.f ) UPSCALE_FLOAT_CASE(kInt, int, 2147483647.f)
UPSCALE_FLOAT_CASE( kUInt, uint, 4294967295.f ) UPSCALE_FLOAT_CASE(kUInt, uint, 4294967295.f)
default: 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; return NULL;
} }
} }
@@ -398,33 +401,46 @@ char * convert_to_expected( void * inputBuffer, size_t numPixels, ExplicitType i
return NULL; 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 *expected = (char *)expectedResults;
char *actual = (char *)actualResults; char *actual = (char *)actualResults;
for ( size_t s = 0; s < sampleNum; s++ ) for (size_t s = 0; s < sampleNum; s++)
{ {
for( size_t z = 0; z < ( ( depth == 0 ) ? 1 : depth ); z++ ) for (size_t z = 0; z < ((depth == 0) ? 1 : depth); z++)
{ {
for( size_t y = 0; y < height; y++ ) for (size_t y = 0; y < height; y++)
{ {
for( size_t x = 0; x < width; x++ ) for (size_t x = 0; x < width; x++)
{ {
if( memcmp( expected, actual, typeSize * 4 ) != 0 ) if (memcmp(expected, actual, typeSize * 4) != 0)
{ {
char scratch[ 1024 ]; char scratch[1024];
if( depth == 0 ) if (depth == 0)
log_error( "ERROR: Data sample %d,%d,%d did not validate!\n", (int)x, (int)y, (int)s ); log_error("ERROR: Data sample %d,%d,%d did not "
"validate!\n",
(int)x, (int)y, (int)s);
else else
log_error( "ERROR: Data sample %d,%d,%d,%d did not validate!\n", (int)x, (int)y, (int)z, (int)s ); log_error("ERROR: Data sample %d,%d,%d,%d did not "
log_error( "\tExpected: %s\n", GetDataVectorString( expected, typeSize, 4, scratch ) ); "validate!\n",
log_error( "\t Actual: %s\n", GetDataVectorString( actual, typeSize, 4, scratch ) ); (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; return -1;
} }
expected += typeSize * 4; expected += typeSize * 4;
@@ -437,51 +453,65 @@ int validate_integer_results( void *expectedResults, void *actualResults, size_t
return 0; 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 *expected = (cl_float *)expectedResults;
cl_float *actual = (cl_float *)actualResults; cl_float *actual = (cl_float *)actualResults;
for ( size_t s = 0; s < sampleNum; s++ ) for (size_t s = 0; s < sampleNum; s++)
{ {
for( size_t z = 0; z < ( ( depth == 0 ) ? 1 : depth ); z++ ) for (size_t z = 0; z < ((depth == 0) ? 1 : depth); z++)
{ {
for( size_t y = 0; y < height; y++ ) for (size_t y = 0; y < height; y++)
{ {
for( size_t x = 0; x < width; x++ ) for (size_t x = 0; x < width; x++)
{ {
float err = 0.f; float err = 0.f;
for( size_t i = 0; i < channelNum; i++ ) for (size_t i = 0; i < channelNum; i++)
{ {
float error = fabsf( expected[ i ] - actual[ i ] ); float error = fabsf(expected[i] - actual[i]);
if( error > err ) if (error > err) err = error;
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 (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 ) if (depth == 0)
log_error( "ERROR: Data sample %d,%d,%d did not validate!\n", (int)x, (int)y, (int)s ); log_error("ERROR: Data sample %d,%d,%d did not "
"validate!\n",
(int)x, (int)y, (int)s);
else else
log_error( "ERROR: Data sample %d,%d,%d,%d did not validate!\n", (int)x, (int)y, (int)z, (int)s ); log_error("ERROR: Data sample %d,%d,%d,%d did not "
"validate!\n",
(int)x, (int)y, (int)z, (int)s);
if (channelNum == 4) if (channelNum == 4)
{ {
log_error( "\tExpected: %f %f %f %f\n", expected[ 0 ], expected[ 1 ], expected[ 2 ], expected[ 3 ] ); log_error("\tExpected: %f %f %f %f\n", expected[0],
log_error( "\t : %a %a %a %a\n", expected[ 0 ], expected[ 1 ], expected[ 2 ], expected[ 3 ] ); 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", expected[0],
log_error( "\t : %a %a %a %a\n", actual[ 0 ], actual[ 1 ], actual[ 2 ], actual[ 3 ] ); 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) else if (channelNum == 1)
{ {
log_error( "\tExpected: %f\n", expected[ 0 ] ); log_error("\tExpected: %f\n", expected[0]);
log_error( "\t : %a\n", expected[ 0 ] ); log_error("\t : %a\n", expected[0]);
log_error( "\t Actual: %f\n", actual[ 0 ] ); log_error("\t Actual: %f\n", actual[0]);
log_error( "\t : %a\n", actual[ 0 ] ); log_error("\t : %a\n", actual[0]);
} }
return -1; return -1;
} }
@@ -495,41 +525,56 @@ int validate_float_results( void *expectedResults, void *actualResults, size_t w
return 0; 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 *expected = (cl_float *)expectedResults;
cl_float *actual = (cl_float *)actualResults; cl_float *actual = (cl_float *)actualResults;
for ( size_t s = 0; s < sampleNum; s++ ) for (size_t s = 0; s < sampleNum; s++)
{ {
for( size_t z = 0; z < ( ( depth == 0 ) ? 1 : depth ); z++ ) for (size_t z = 0; z < ((depth == 0) ? 1 : depth); z++)
{ {
for( size_t y = 0; y < height; y++ ) for (size_t y = 0; y < height; y++)
{ {
for( size_t x = 0; x < width; x++ ) for (size_t x = 0; x < width; x++)
{ {
float err = 0.f; float err = 0.f;
for( size_t i = 0; i < 3; i++ ) // skip the fourth channel for (size_t i = 0; i < 3; i++) // skip the fourth channel
{ {
float error = fabsf( expected[ i ] - actual[ i ] ); float error = fabsf(expected[i] - actual[i]);
if( error > err ) if (error > err) err = error;
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 (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 ) if (depth == 0)
log_error( "ERROR: Data sample %d,%d,%d did not validate!\n", (int)x, (int)y, (int)s ); log_error("ERROR: Data sample %d,%d,%d did not "
"validate!\n",
(int)x, (int)y, (int)s);
else else
log_error( "ERROR: Data sample %d,%d,%d,%d did not validate!\n", (int)x, (int)y, (int)z, (int)s ); log_error("ERROR: Data sample %d,%d,%d,%d did not "
log_error( "\tExpected: %f %f %f\n", expected[ 0 ], expected[ 1 ], expected[ 2 ] ); "validate!\n",
log_error( "\t : %a %a %a\n", expected[ 0 ], expected[ 1 ], expected[ 2 ] ); (int)x, (int)y, (int)z, (int)s);
log_error( "\t Actual: %f %f %f\n", actual[ 0 ], actual[ 1 ], actual[ 2 ] ); log_error("\tExpected: %f %f %f\n", expected[0],
log_error( "\t : %a %a %a\n", actual[ 0 ], actual[ 1 ], actual[ 2 ] ); 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; return -1;
} }
expected += 4; expected += 4;
@@ -542,8 +587,10 @@ int validate_float_results_rgb_101010( void *expectedResults, void *actualResult
return 0; return 0;
} }
int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type, GLuint expected_gl_name, int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type,
GLenum expected_cl_gl_texture_target, GLint expected_cl_gl_mipmap_level) GLuint expected_gl_name,
GLenum expected_cl_gl_texture_target,
GLint expected_cl_gl_mipmap_level)
{ {
cl_gl_object_type object_type; cl_gl_object_type object_type;
GLuint object_name; GLuint object_name;
@@ -552,37 +599,54 @@ int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type, GLuint
int error; int error;
error = (*clGetGLObjectInfo_ptr)(mem, &object_type, &object_name); error = (*clGetGLObjectInfo_ptr)(mem, &object_type, &object_name);
test_error( error, "clGetGLObjectInfo failed"); test_error(error, "clGetGLObjectInfo failed");
if (object_type != expected_cl_gl_type) { 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); {
log_error("clGetGLObjectInfo did not return expected object type: "
"expected %d, got %d.\n",
expected_cl_gl_type, object_type);
return -1; return -1;
} }
if (object_name != expected_gl_name) { 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); {
log_error("clGetGLObjectInfo did not return expected object name: "
"expected %d, got %d.\n",
expected_gl_name, object_name);
return -1; 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) { if (object_type == CL_GL_OBJECT_BUFFER
|| object_type == CL_GL_OBJECT_RENDERBUFFER)
{
return 0; return 0;
} }
// Otherwise, it's a texture-based object and requires a bit more checking. // 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); error = (*clGetGLTextureInfo_ptr)(mem, CL_GL_TEXTURE_TARGET,
test_error( error, "clGetGLTextureInfo for CL_GL_TEXTURE_TARGET failed"); sizeof(texture_target), &texture_target,
NULL);
test_error(error, "clGetGLTextureInfo for CL_GL_TEXTURE_TARGET failed");
if (texture_target != expected_cl_gl_texture_target) { 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); {
log_error("clGetGLTextureInfo did not return expected texture target: "
"expected %d, got %d.\n",
expected_cl_gl_texture_target, texture_target);
return -1; return -1;
} }
error = (*clGetGLTextureInfo_ptr)(mem, CL_GL_MIPMAP_LEVEL, sizeof(mipmap_level), &mipmap_level, NULL); error = (*clGetGLTextureInfo_ptr)(
test_error( error, "clGetGLTextureInfo for CL_GL_MIPMAP_LEVEL failed"); 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) { 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); {
log_error("clGetGLTextureInfo did not return expected mipmap level: "
"expected %d, got %d.\n",
expected_cl_gl_mipmap_level, mipmap_level);
return -1; return -1;
} }
@@ -595,22 +659,25 @@ bool CheckGLIntegerExtensionSupport()
const GLubyte *glVersion = glGetString(GL_VERSION); const GLubyte *glVersion = glGetString(GL_VERSION);
const GLubyte *glExtensionList = glGetString(GL_EXTENSIONS); const GLubyte *glExtensionList = glGetString(GL_EXTENSIONS);
// Check if the OpenGL vrsion is 3.0 or grater or GL_EXT_texture_integer is supported // Check if the OpenGL vrsion is 3.0 or grater or GL_EXT_texture_integer is
return (((glVersion[0] - '0') >= 3) || (strstr((const char*)glExtensionList, "GL_EXT_texture_integer"))); // 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; cl_uint formatCount = 0;
unsigned int i; unsigned int i;
int error; int error;
cl_mem_object_type image_type; cl_mem_object_type image_type;
switch (get_base_gl_target(gl_target)) { switch (get_base_gl_target(gl_target))
case GL_TEXTURE_1D: {
image_type = CL_MEM_OBJECT_IMAGE1D; case GL_TEXTURE_1D: image_type = CL_MEM_OBJECT_IMAGE1D;
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER:
image_type = CL_MEM_OBJECT_IMAGE1D_BUFFER; image_type = CL_MEM_OBJECT_IMAGE1D_BUFFER;
break; break;
@@ -627,27 +694,25 @@ int is_rgb_101010_supported( cl_context context, GLenum gl_target )
case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z:
image_type = CL_MEM_OBJECT_IMAGE2D; image_type = CL_MEM_OBJECT_IMAGE2D;
break; break;
case GL_TEXTURE_3D: case GL_TEXTURE_3D: image_type = CL_MEM_OBJECT_IMAGE3D;
image_type = CL_MEM_OBJECT_IMAGE3D; case GL_TEXTURE_1D_ARRAY: image_type = CL_MEM_OBJECT_IMAGE1D_ARRAY;
case GL_TEXTURE_1D_ARRAY:
image_type = CL_MEM_OBJECT_IMAGE1D_ARRAY;
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
image_type = CL_MEM_OBJECT_IMAGE2D_ARRAY; image_type = CL_MEM_OBJECT_IMAGE2D_ARRAY;
break; break;
default: default: image_type = CL_MEM_OBJECT_IMAGE2D;
image_type = CL_MEM_OBJECT_IMAGE2D;
} }
if ((error = clGetSupportedImageFormats(context, CL_MEM_READ_WRITE, if ((error =
image_type, 128, formatList, clGetSupportedImageFormats(context, CL_MEM_READ_WRITE, image_type,
&formatCount ))) { 128, formatList, &formatCount)))
{
return error; return error;
} }
// Check if the RGB 101010 format is supported // 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; return 1;
} }

352
test_conformance/gl/main.cpp
View File

@@ -18,7 +18,7 @@
#include <stdio.h> #include <stdio.h>
#include <string.h> #include <string.h>
#if !defined (__APPLE__) #if !defined(__APPLE__)
#include <CL/cl.h> #include <CL/cl.h>
#endif #endif
@@ -34,124 +34,124 @@
static cl_context sCurrentContext = NULL; static cl_context sCurrentContext = NULL;
#define TEST_FN_REDIRECT( fn ) ADD_TEST( redirect_##fn ) #define TEST_FN_REDIRECT(fn) ADD_TEST(redirect_##fn)
#define TEST_FN_REDIRECTOR( fn ) \ #define TEST_FN_REDIRECTOR(fn) \
int test_redirect_##fn(cl_device_id device, cl_context context, cl_command_queue queue, int numElements ) \ int test_redirect_##fn(cl_device_id device, cl_context context, \
{ \ cl_command_queue queue, int numElements) \
{ \
int error; \ int error; \
clCommandQueueWrapper realQueue = clCreateCommandQueueWithProperties( sCurrentContext, device, 0, &error ); \ clCommandQueueWrapper realQueue = clCreateCommandQueueWithProperties( \
test_error( error, "Unable to create command queue" ); \ sCurrentContext, device, 0, &error); \
return test_##fn( device, sCurrentContext, realQueue, numElements ); \ test_error(error, "Unable to create command queue"); \
} return test_##fn(device, sCurrentContext, realQueue, numElements); \
}
// buffers: // buffers:
TEST_FN_REDIRECTOR( buffers ) TEST_FN_REDIRECTOR(buffers)
TEST_FN_REDIRECTOR( buffers_getinfo ) TEST_FN_REDIRECTOR(buffers_getinfo)
// 1D images: // 1D images:
TEST_FN_REDIRECTOR( images_read_1D ) TEST_FN_REDIRECTOR(images_read_1D)
TEST_FN_REDIRECTOR( images_write_1D ) TEST_FN_REDIRECTOR(images_write_1D)
TEST_FN_REDIRECTOR( images_1D_getinfo ) TEST_FN_REDIRECTOR(images_1D_getinfo)
// 1D image arrays: // 1D image arrays:
TEST_FN_REDIRECTOR( images_read_1Darray ) TEST_FN_REDIRECTOR(images_read_1Darray)
TEST_FN_REDIRECTOR( images_write_1Darray ) TEST_FN_REDIRECTOR(images_write_1Darray)
TEST_FN_REDIRECTOR( images_1Darray_getinfo ) TEST_FN_REDIRECTOR(images_1Darray_getinfo)
// 2D images: // 2D images:
TEST_FN_REDIRECTOR( images_read_2D ) TEST_FN_REDIRECTOR(images_read_2D)
TEST_FN_REDIRECTOR( images_read_cube ) TEST_FN_REDIRECTOR(images_read_cube)
TEST_FN_REDIRECTOR( images_write ) TEST_FN_REDIRECTOR(images_write)
TEST_FN_REDIRECTOR( images_write_cube ) TEST_FN_REDIRECTOR(images_write_cube)
TEST_FN_REDIRECTOR( images_2D_getinfo ) TEST_FN_REDIRECTOR(images_2D_getinfo)
TEST_FN_REDIRECTOR( images_cube_getinfo ) TEST_FN_REDIRECTOR(images_cube_getinfo)
// 2D image arrays: // 2D image arrays:
TEST_FN_REDIRECTOR( images_read_2Darray ) TEST_FN_REDIRECTOR(images_read_2Darray)
TEST_FN_REDIRECTOR( images_write_2Darray ) TEST_FN_REDIRECTOR(images_write_2Darray)
TEST_FN_REDIRECTOR( images_2Darray_getinfo ) TEST_FN_REDIRECTOR(images_2Darray_getinfo)
// 3D images: // 3D images:
TEST_FN_REDIRECTOR( images_read_3D ) TEST_FN_REDIRECTOR(images_read_3D)
TEST_FN_REDIRECTOR( images_write_3D ) TEST_FN_REDIRECTOR(images_write_3D)
TEST_FN_REDIRECTOR( images_3D_getinfo ) TEST_FN_REDIRECTOR(images_3D_getinfo)
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
TEST_FN_REDIRECTOR( images_read_texturebuffer ) TEST_FN_REDIRECTOR(images_read_texturebuffer)
TEST_FN_REDIRECTOR( images_write_texturebuffer ) TEST_FN_REDIRECTOR(images_write_texturebuffer)
TEST_FN_REDIRECTOR( images_texturebuffer_getinfo ) TEST_FN_REDIRECTOR(images_texturebuffer_getinfo)
// depth textures // depth textures
TEST_FN_REDIRECTOR( images_read_2D_depth ) TEST_FN_REDIRECTOR(images_read_2D_depth)
TEST_FN_REDIRECTOR( images_write_2D_depth ) TEST_FN_REDIRECTOR(images_write_2D_depth)
TEST_FN_REDIRECTOR( images_read_2Darray_depth ) TEST_FN_REDIRECTOR(images_read_2Darray_depth)
TEST_FN_REDIRECTOR( images_write_2Darray_depth ) TEST_FN_REDIRECTOR(images_write_2Darray_depth)
TEST_FN_REDIRECTOR( images_read_2D_multisample ) TEST_FN_REDIRECTOR(images_read_2D_multisample)
TEST_FN_REDIRECTOR( images_read_2Darray_multisample ) TEST_FN_REDIRECTOR(images_read_2Darray_multisample)
TEST_FN_REDIRECTOR( image_methods_depth ) TEST_FN_REDIRECTOR(image_methods_depth)
TEST_FN_REDIRECTOR( image_methods_multisample ) TEST_FN_REDIRECTOR(image_methods_multisample)
#endif #endif
// Renderbuffer-backed images: // Renderbuffer-backed images:
TEST_FN_REDIRECTOR( renderbuffer_read ) TEST_FN_REDIRECTOR(renderbuffer_read)
TEST_FN_REDIRECTOR( renderbuffer_write ) TEST_FN_REDIRECTOR(renderbuffer_write)
TEST_FN_REDIRECTOR( renderbuffer_getinfo ) TEST_FN_REDIRECTOR(renderbuffer_getinfo)
TEST_FN_REDIRECTOR( fence_sync ) TEST_FN_REDIRECTOR(fence_sync)
test_definition test_list[] = { test_definition test_list[] = { TEST_FN_REDIRECT(buffers),
TEST_FN_REDIRECT( buffers ), TEST_FN_REDIRECT(buffers_getinfo),
TEST_FN_REDIRECT( buffers_getinfo ),
TEST_FN_REDIRECT( images_read_1D ), TEST_FN_REDIRECT(images_read_1D),
TEST_FN_REDIRECT( images_write_1D ), TEST_FN_REDIRECT(images_write_1D),
TEST_FN_REDIRECT( images_1D_getinfo ), TEST_FN_REDIRECT(images_1D_getinfo),
TEST_FN_REDIRECT( images_read_1Darray ), TEST_FN_REDIRECT(images_read_1Darray),
TEST_FN_REDIRECT( images_write_1Darray ), TEST_FN_REDIRECT(images_write_1Darray),
TEST_FN_REDIRECT( images_1Darray_getinfo ), TEST_FN_REDIRECT(images_1Darray_getinfo),
TEST_FN_REDIRECT( images_read_2D ), TEST_FN_REDIRECT(images_read_2D),
TEST_FN_REDIRECT( images_write ), TEST_FN_REDIRECT(images_write),
TEST_FN_REDIRECT( images_2D_getinfo ), TEST_FN_REDIRECT(images_2D_getinfo),
TEST_FN_REDIRECT( images_read_cube ), TEST_FN_REDIRECT(images_read_cube),
TEST_FN_REDIRECT( images_write_cube ), TEST_FN_REDIRECT(images_write_cube),
TEST_FN_REDIRECT( images_cube_getinfo ), TEST_FN_REDIRECT(images_cube_getinfo),
TEST_FN_REDIRECT( images_read_2Darray ), TEST_FN_REDIRECT(images_read_2Darray),
TEST_FN_REDIRECT( images_write_2Darray), TEST_FN_REDIRECT(images_write_2Darray),
TEST_FN_REDIRECT( images_2Darray_getinfo ), TEST_FN_REDIRECT(images_2Darray_getinfo),
TEST_FN_REDIRECT( images_read_3D ), TEST_FN_REDIRECT(images_read_3D),
TEST_FN_REDIRECT( images_write_3D ), TEST_FN_REDIRECT(images_write_3D),
TEST_FN_REDIRECT( images_3D_getinfo ), TEST_FN_REDIRECT(images_3D_getinfo),
TEST_FN_REDIRECT( renderbuffer_read ), TEST_FN_REDIRECT(renderbuffer_read),
TEST_FN_REDIRECT( renderbuffer_write ), TEST_FN_REDIRECT(renderbuffer_write),
TEST_FN_REDIRECT( renderbuffer_getinfo ) TEST_FN_REDIRECT(renderbuffer_getinfo) };
};
test_definition test_list32[] = { test_definition test_list32[] = {
TEST_FN_REDIRECT( images_read_texturebuffer ), TEST_FN_REDIRECT(images_read_texturebuffer),
TEST_FN_REDIRECT( images_write_texturebuffer ), TEST_FN_REDIRECT(images_write_texturebuffer),
TEST_FN_REDIRECT( images_texturebuffer_getinfo ), TEST_FN_REDIRECT(images_texturebuffer_getinfo),
TEST_FN_REDIRECT( fence_sync ), TEST_FN_REDIRECT(fence_sync),
TEST_FN_REDIRECT( images_read_2D_depth ), TEST_FN_REDIRECT(images_read_2D_depth),
TEST_FN_REDIRECT( images_write_2D_depth ), TEST_FN_REDIRECT(images_write_2D_depth),
TEST_FN_REDIRECT( images_read_2Darray_depth ), TEST_FN_REDIRECT(images_read_2Darray_depth),
TEST_FN_REDIRECT( images_write_2Darray_depth ), TEST_FN_REDIRECT(images_write_2Darray_depth),
TEST_FN_REDIRECT( images_read_2D_multisample ), TEST_FN_REDIRECT(images_read_2D_multisample),
TEST_FN_REDIRECT( images_read_2Darray_multisample ), TEST_FN_REDIRECT(images_read_2Darray_multisample),
TEST_FN_REDIRECT( image_methods_depth ), TEST_FN_REDIRECT(image_methods_depth),
TEST_FN_REDIRECT( image_methods_multisample ) TEST_FN_REDIRECT(image_methods_multisample)
}; };
const int test_num = ARRAY_SIZE( test_list ); const int test_num = ARRAY_SIZE(test_list);
const int test_num32 = ARRAY_SIZE( test_list32 ); const int test_num32 = ARRAY_SIZE(test_list32);
int main(int argc, const char *argv[]) int main(int argc, const char *argv[])
{ {
@@ -169,52 +169,58 @@ int main(int argc, const char *argv[])
cl_device_type requestedDeviceType = CL_DEVICE_TYPE_DEFAULT; cl_device_type requestedDeviceType = CL_DEVICE_TYPE_DEFAULT;
/* Do we have a CPU/GPU specification? */ /* Do we have a CPU/GPU specification? */
if( argc > 1 ) if (argc > 1)
{ {
if( strcmp( argv[ argc - 1 ], "gpu" ) == 0 || strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_GPU" ) == 0 ) if (strcmp(argv[argc - 1], "gpu") == 0
|| strcmp(argv[argc - 1], "CL_DEVICE_TYPE_GPU") == 0)
{ {
requestedDeviceType = CL_DEVICE_TYPE_GPU; requestedDeviceType = CL_DEVICE_TYPE_GPU;
argc--; argc--;
} }
else if( strcmp( argv[ argc - 1 ], "cpu" ) == 0 || strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_CPU" ) == 0 ) else if (strcmp(argv[argc - 1], "cpu") == 0
|| strcmp(argv[argc - 1], "CL_DEVICE_TYPE_CPU") == 0)
{ {
requestedDeviceType = CL_DEVICE_TYPE_CPU; requestedDeviceType = CL_DEVICE_TYPE_CPU;
argc--; argc--;
} }
else if( strcmp( argv[ argc - 1 ], "accelerator" ) == 0 || strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_ACCELERATOR" ) == 0 ) else if (strcmp(argv[argc - 1], "accelerator") == 0
|| strcmp(argv[argc - 1], "CL_DEVICE_TYPE_ACCELERATOR") == 0)
{ {
requestedDeviceType = CL_DEVICE_TYPE_ACCELERATOR; requestedDeviceType = CL_DEVICE_TYPE_ACCELERATOR;
argc--; argc--;
} }
else if( strcmp( argv[ argc - 1 ], "CL_DEVICE_TYPE_DEFAULT" ) == 0 ) else if (strcmp(argv[argc - 1], "CL_DEVICE_TYPE_DEFAULT") == 0)
{ {
requestedDeviceType = CL_DEVICE_TYPE_DEFAULT; requestedDeviceType = CL_DEVICE_TYPE_DEFAULT;
argc--; argc--;
} }
} }
if( argc > 1 && strcmp( argv[ 1 ], "-list" ) == 0 ) if (argc > 1 && strcmp(argv[1], "-list") == 0)
{ {
log_info( "Available 2.x tests:\n" ); log_info("Available 2.x tests:\n");
for( int i = 0; i < test_num; i++ ) for (int i = 0; i < test_num; i++)
log_info( "\t%s\n", test_list[i].name ); log_info("\t%s\n", test_list[i].name);
log_info( "Available 3.2 tests:\n" ); log_info("Available 3.2 tests:\n");
for( int i = 0; i < test_num32; i++ ) for (int i = 0; i < test_num32; i++)
log_info( "\t%s\n", test_list32[i].name ); 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("Note: Any 3.2 test names must follow 2.1 test names on the "
log_info( "Use environment variables to specify desired device.\n" ); "command line.\n");
log_info("Use environment variables to specify desired device.\n");
return 0; return 0;
} }
// Check to see if any 2.x or 3.2 test names were specified on the command line. // Check to see if any 2.x or 3.2 test names were specified on the command
// line.
unsigned first_32_testname = 0; unsigned first_32_testname = 0;
for (int j=1; (j<argc) && (!first_32_testname); ++j) for (int j = 1; (j < argc) && (!first_32_testname); ++j)
for (int i = 0; i < test_num32; ++i) for (int i = 0; i < test_num32; ++i)
if (strcmp(test_list32[i].name, argv[j]) == 0) { if (strcmp(test_list32[i].name, argv[j]) == 0)
{
first_32_testname = j; first_32_testname = j;
break; break;
} }
@@ -223,61 +229,76 @@ int main(int argc, const char *argv[])
GLEnvironment *glEnv = GLEnvironment::Instance(); GLEnvironment *glEnv = GLEnvironment::Instance();
// Check if any devices of the requested type support CL/GL interop. // Check if any devices of the requested type support CL/GL interop.
int supported = glEnv->SupportsCLGLInterop( requestedDeviceType ); int supported = glEnv->SupportsCLGLInterop(requestedDeviceType);
if( supported == 0 ) { if (supported == 0)
log_info("Test not run because GL-CL interop is not supported for any devices of the requested type.\n"); {
log_info("Test not run because GL-CL interop is not supported for any "
"devices of the requested type.\n");
return 0; return 0;
} else if ( supported == -1 ) { }
log_error("Unable to setup the test or failed to determine if CL-GL interop is supported.\n"); else if (supported == -1)
{
log_error("Unable to setup the test or failed to determine if CL-GL "
"interop is supported.\n");
return -1; return -1;
} }
// Initialize function pointers. // Initialize function pointers.
error = init_clgl_ext(); error = init_clgl_ext();
if (error < 0) { if (error < 0)
{
return error; return error;
} }
// OpenGL tests for non-3.2 //////////////////////////////////////////////////////// // OpenGL tests for non-3.2
if ((argc == 1) || (first_32_testname != 1)) { // ////////////////////////////////////////////////////////
if ((argc == 1) || (first_32_testname != 1))
{
// At least one device supports CL-GL interop, so init the test. // At least one device supports CL-GL interop, so init the test.
if( glEnv->Init( &argc, (char **)argv, CL_FALSE ) ) { if (glEnv->Init(&argc, (char **)argv, CL_FALSE))
log_error("Failed to initialize the GL environment for this test.\n"); {
log_error(
"Failed to initialize the GL environment for this test.\n");
return -1; return -1;
} }
// Create a context to use and then grab a device (or devices) from it // Create a context to use and then grab a device (or devices) from it
sCurrentContext = glEnv->CreateCLContext(); sCurrentContext = glEnv->CreateCLContext();
if( sCurrentContext == NULL ) if (sCurrentContext == NULL)
{ {
log_error( "ERROR: Unable to obtain CL context from GL\n" ); log_error("ERROR: Unable to obtain CL context from GL\n");
return -1; return -1;
} }
size_t numDevices = 0; size_t numDevices = 0;
cl_device_id *deviceIDs; cl_device_id *deviceIDs;
error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL, &numDevices); error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL,
if( error != CL_SUCCESS ) &numDevices);
if (error != CL_SUCCESS)
{ {
print_error( error, "Unable to get device count from context" ); print_error(error, "Unable to get device count from context");
return -1; return -1;
} }
deviceIDs = (cl_device_id *)malloc(numDevices); deviceIDs = (cl_device_id *)malloc(numDevices);
if (deviceIDs == NULL) { if (deviceIDs == NULL)
print_error( error, "malloc failed" ); {
print_error(error, "malloc failed");
return -1; return -1;
} }
error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, numDevices, deviceIDs, NULL); error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES,
if( error != CL_SUCCESS ) { numDevices, deviceIDs, NULL);
print_error( error, "Unable to get device list from context" ); if (error != CL_SUCCESS)
{
print_error(error, "Unable to get device list from context");
return -1; return -1;
} }
numDevices /= sizeof(cl_device_id); numDevices /= sizeof(cl_device_id);
if (numDevices < 1) { if (numDevices < 1)
{
log_error("No devices found.\n"); log_error("No devices found.\n");
return -1; return -1;
} }
@@ -285,94 +306,111 @@ int main(int argc, const char *argv[])
// Execute tests. // Execute tests.
int argc_ = (first_32_testname) ? first_32_testname : argc; int argc_ = (first_32_testname) ? first_32_testname : argc;
for( size_t i = 0; i < numDevices; i++ ) { for (size_t i = 0; i < numDevices; i++)
log_info( "\nTesting OpenGL 2.x\n" ); {
if( printDeviceHeader( deviceIDs[ i ] ) != CL_SUCCESS ) { log_info("\nTesting OpenGL 2.x\n");
if (printDeviceHeader(deviceIDs[i]) != CL_SUCCESS)
{
return -1; return -1;
} }
// Note: don't use the entire harness, because we have a different way of obtaining the device (via the context) // Note: don't use the entire harness, because we have a different
error = parseAndCallCommandLineTests( argc_, argv, deviceIDs[i], test_num, test_list, true, 0, 1024 ); // way of obtaining the device (via the context)
if( error != 0 ) error = parseAndCallCommandLineTests(
break; argc_, argv, deviceIDs[i], test_num, test_list, true, 0, 1024);
if (error != 0) break;
} }
numErrors += error; numErrors += error;
// Clean-up. // Clean-up.
free(deviceIDs); free(deviceIDs);
clReleaseContext( sCurrentContext ); clReleaseContext(sCurrentContext);
//delete glEnv; // delete glEnv;
} }
// OpenGL 3.2 tests. //////////////////////////////////////////////////////// // OpenGL 3.2 tests.
if ((argc==1) || first_32_testname) { // ////////////////////////////////////////////////////////
if ((argc == 1) || first_32_testname)
{
// At least one device supports CL-GL interop, so init the test. // At least one device supports CL-GL interop, so init the test.
if( glEnv->Init( &argc, (char **)argv, CL_TRUE ) ) { if (glEnv->Init(&argc, (char **)argv, CL_TRUE))
log_error("Failed to initialize the GL environment for this test.\n"); {
log_error(
"Failed to initialize the GL environment for this test.\n");
return -1; return -1;
} }
// Create a context to use and then grab a device (or devices) from it // Create a context to use and then grab a device (or devices) from it
sCurrentContext = glEnv->CreateCLContext(); sCurrentContext = glEnv->CreateCLContext();
if( sCurrentContext == NULL ) { if (sCurrentContext == NULL)
log_error( "ERROR: Unable to obtain CL context from GL\n" ); {
log_error("ERROR: Unable to obtain CL context from GL\n");
return -1; return -1;
} }
size_t numDevices = 0; size_t numDevices = 0;
cl_device_id *deviceIDs; cl_device_id *deviceIDs;
error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL, &numDevices); error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES, 0, NULL,
if( error != CL_SUCCESS ) { &numDevices);
print_error( error, "Unable to get device count from context" ); if (error != CL_SUCCESS)
{
print_error(error, "Unable to get device count from context");
return -1; return -1;
} }
deviceIDs = (cl_device_id *)malloc(numDevices); deviceIDs = (cl_device_id *)malloc(numDevices);
if (deviceIDs == NULL) { if (deviceIDs == NULL)
print_error( error, "malloc failed" ); {
print_error(error, "malloc failed");
return -1; return -1;
} }
error = clGetContextInfo( sCurrentContext, CL_CONTEXT_DEVICES, numDevices, deviceIDs, NULL); error = clGetContextInfo(sCurrentContext, CL_CONTEXT_DEVICES,
if( error != CL_SUCCESS ) { numDevices, deviceIDs, NULL);
print_error( error, "Unable to get device list from context" ); if (error != CL_SUCCESS)
{
print_error(error, "Unable to get device list from context");
return -1; return -1;
} }
numDevices /= sizeof(cl_device_id); numDevices /= sizeof(cl_device_id);
if (numDevices < 1) { if (numDevices < 1)
{
log_error("No devices found.\n"); log_error("No devices found.\n");
return -1; return -1;
} }
int argc_ = (first_32_testname) ? 1 + (argc - first_32_testname) : argc; int argc_ = (first_32_testname) ? 1 + (argc - first_32_testname) : argc;
const char** argv_ = (first_32_testname) ? &argv[first_32_testname-1] : argv; const char **argv_ =
(first_32_testname) ? &argv[first_32_testname - 1] : argv;
// Execute the tests. // Execute the tests.
for( size_t i = 0; i < numDevices; i++ ) { for (size_t i = 0; i < numDevices; i++)
log_info( "\nTesting OpenGL 3.2\n" ); {
if( printDeviceHeader( deviceIDs[ i ] ) != CL_SUCCESS ) { log_info("\nTesting OpenGL 3.2\n");
if (printDeviceHeader(deviceIDs[i]) != CL_SUCCESS)
{
return -1; return -1;
} }
// Note: don't use the entire harness, because we have a different way of obtaining the device (via the context) // Note: don't use the entire harness, because we have a different
error = parseAndCallCommandLineTests( argc_, argv_, deviceIDs[i], test_num32, test_list32, true, 0, 1024 ); // way of obtaining the device (via the context)
if( error != 0 ) error = parseAndCallCommandLineTests(argc_, argv_, deviceIDs[i],
break; test_num32, test_list32, true,
0, 1024);
if (error != 0) break;
} }
numErrors += error; numErrors += error;
// Clean-up. // Clean-up.
free(deviceIDs); free(deviceIDs);
clReleaseContext( sCurrentContext ); clReleaseContext(sCurrentContext);
delete glEnv; delete glEnv;
} }
//All done. // All done.
return numErrors; return numErrors;
} }

153
test_conformance/gl/procs.h
View File

@@ -20,121 +20,134 @@
#pragma mark - #pragma mark -
#pragma Misc tests #pragma Misc tests
extern int test_buffers( cl_device_id device, cl_context context, extern int test_buffers(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_fence_sync( cl_device_id device, cl_context context, extern int test_fence_sync(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
#pragma mark - #pragma mark -
#pragma mark Tead tests #pragma mark Tead tests
extern int test_images_read_2D( cl_device_id device, cl_context context, extern int test_images_read_2D(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_read_1D( cl_device_id device, cl_context context, extern int test_images_read_1D(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_read_texturebuffer( cl_device_id device, cl_context context, extern int test_images_read_texturebuffer(cl_device_id device,
cl_command_queue queue, int num_elements ); cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_images_read_1Darray( cl_device_id device, cl_context context, extern int test_images_read_1Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_read_2Darray( cl_device_id device, cl_context context, extern int test_images_read_2Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_read_cube( cl_device_id device, cl_context context, extern int test_images_read_cube(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_read_3D( cl_device_id device, cl_context context, extern int test_images_read_3D(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_renderbuffer_read( cl_device_id device, cl_context context, extern int test_renderbuffer_read(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
#pragma mark - #pragma mark -
#pragma mark Write tests #pragma mark Write tests
// 2D tests are the ones with no suffix: // 2D tests are the ones with no suffix:
extern int test_images_write( cl_device_id device, cl_context context, extern int test_images_write(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_write_cube( cl_device_id device, cl_context context, extern int test_images_write_cube(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_renderbuffer_write( cl_device_id device, cl_context context, extern int test_renderbuffer_write(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
// Here are the rest: // Here are the rest:
extern int test_images_write_1D( cl_device_id device, cl_context context, extern int test_images_write_1D(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_write_texturebuffer( cl_device_id device, cl_context context, extern int test_images_write_texturebuffer(cl_device_id device,
cl_command_queue queue, int num_elements ); cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_images_write_1Darray( cl_device_id device, cl_context context, extern int test_images_write_1Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_write_2Darray( cl_device_id device, cl_context context, extern int test_images_write_2Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
extern int test_images_write_3D( cl_device_id device, cl_context context, extern int test_images_write_3D(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements ); cl_command_queue queue, int num_elements);
#pragma mark - #pragma mark -
#pragma mark Get info test entry points #pragma mark Get info test entry points
extern int test_buffers_getinfo( cl_device_id device, cl_context context, extern int test_buffers_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_1D_getinfo( cl_device_id device, cl_context context, extern int test_images_1D_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_texturebuffer_getinfo( cl_device_id device, cl_context context, extern int test_images_texturebuffer_getinfo(cl_device_id device,
cl_command_queue queue, int numElements ); cl_context context,
cl_command_queue queue,
int numElements);
extern int test_images_1Darray_getinfo( cl_device_id device, cl_context context, extern int test_images_1Darray_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_2D_getinfo( cl_device_id device, cl_context context, extern int test_images_2D_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_2Darray_getinfo( cl_device_id device, cl_context context, extern int test_images_2Darray_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_cube_getinfo( cl_device_id device, cl_context context, extern int test_images_cube_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_3D_getinfo( cl_device_id device, cl_context context, extern int test_images_3D_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_read_2D_depth( cl_device_id device, cl_context context, extern int test_images_read_2D_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_write_2D_depth( cl_device_id device, cl_context context, extern int test_images_write_2D_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);
extern int test_images_read_2Darray_depth( cl_device_id device, cl_context context, extern int test_images_read_2Darray_depth(cl_device_id device,
cl_command_queue queue, int ); cl_context context,
cl_command_queue queue, int);
extern int test_images_write_2Darray_depth( cl_device_id device, cl_context context, extern int test_images_write_2Darray_depth(cl_device_id device,
cl_command_queue queue, int numElements ); cl_context context,
cl_command_queue queue,
int numElements);
extern int test_images_read_2D_multisample( cl_device_id device, cl_context context, extern int test_images_read_2D_multisample(cl_device_id device,
cl_command_queue queue, int numElements ); cl_context context,
cl_command_queue queue,
int numElements);
extern int test_images_read_2Darray_multisample( cl_device_id device, cl_context context, extern int test_images_read_2Darray_multisample(cl_device_id device,
cl_command_queue queue, int ); cl_context context,
cl_command_queue queue, int);
extern int test_image_methods_depth( cl_device_id device, cl_context context, extern int test_image_methods_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int ); cl_command_queue queue, int);
extern int test_image_methods_multisample( cl_device_id device, cl_context context, extern int test_image_methods_multisample(cl_device_id device,
cl_command_queue queue, int ); cl_context context,
cl_command_queue queue, int);
extern int test_renderbuffer_getinfo( cl_device_id device, cl_context context, extern int test_renderbuffer_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ); cl_command_queue queue, int numElements);

49
test_conformance/gl/testBase.h
View File

@@ -23,7 +23,7 @@
#include <sys/types.h> #include <sys/types.h>
#include <sys/stat.h> #include <sys/stat.h>
#if !defined (__APPLE__) #if !defined(__APPLE__)
#include <CL/cl.h> #include <CL/cl.h>
#include "gl/gl_headers.h" #include "gl/gl_headers.h"
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
@@ -40,25 +40,40 @@
#include "gl/helpers.h" #include "gl/helpers.h"
extern const char *get_kernel_suffix( 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 const char *get_write_conversion(cl_image_format *format,
extern ExplicitType get_read_kernel_type( cl_image_format *format ); ExplicitType type);
extern ExplicitType get_write_kernel_type( cl_image_format *format ); 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 char *convert_to_expected(void *inputBuffer, size_t numPixels,
extern int validate_integer_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum, size_t typeSize ); ExplicitType inType, ExplicitType outType,
extern int validate_integer_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum, size_t typeSize ); size_t channelNum, GLenum glDataType = 0);
extern int validate_float_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum, size_t channelNum ); extern int validate_integer_results(void *expectedResults, void *actualResults,
extern int validate_float_results( void *expectedResults, void *actualResults, size_t width, size_t height, size_t depth, size_t sampleNum, size_t channelNum ); size_t width, size_t height,
extern int validate_float_results_rgb_101010( void *expectedResults, void *actualResults, size_t width, size_t height, size_t sampleNum ); size_t sampleNum, size_t typeSize);
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 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, extern int CheckGLObjectInfo(cl_mem mem, cl_gl_object_type expected_cl_gl_type,
GLenum expected_cl_gl_texture_target, GLint expected_cl_gl_mipmap_level); GLuint expected_gl_name,
GLenum expected_cl_gl_texture_target,
GLint expected_cl_gl_mipmap_level);
extern bool CheckGLIntegerExtensionSupport(); extern bool CheckGLIntegerExtensionSupport();
#endif // _testBase_h #endif // _testBase_h

2
test_conformance/gl/test_fence_sync.cpp
View File

@@ -113,6 +113,7 @@ typedef cl_event(CL_API_CALL *clCreateEventFromGLsyncKHR_fn)(
clCreateEventFromGLsyncKHR_fn clCreateEventFromGLsyncKHR_ptr; clCreateEventFromGLsyncKHR_fn clCreateEventFromGLsyncKHR_ptr;
// clang-format off
static const char *updateBuffersKernel[] = { static const char *updateBuffersKernel[] = {
"__kernel void update( __global float4 * vertices, __global float4 " "__kernel void update( __global float4 * vertices, __global float4 "
"*colors, int horizWrap, int rowIdx )\n" "*colors, int horizWrap, int rowIdx )\n"
@@ -132,6 +133,7 @@ static const char *updateBuffersKernel[] = {
" colors[ tid * 2 + 1 ] = colors[ tid * 2 + 0 ];\n" " colors[ tid * 2 + 1 ] = colors[ tid * 2 + 0 ];\n"
"}\n" "}\n"
}; };
// clang-format on
// Passthrough VertexShader // Passthrough VertexShader
static const char *vertexshader = "#version 150\n" static const char *vertexshader = "#version 150\n"

362
test_conformance/gl/test_image_methods.cpp
View File

@@ -34,6 +34,7 @@ struct image_kernel_data
cl_int numSamples; cl_int numSamples;
}; };
// clang-format off
static const char *methodTestKernelPattern = static const char *methodTestKernelPattern =
"%s" "%s"
"typedef struct {\n" "typedef struct {\n"
@@ -53,134 +54,144 @@ static const char *methodTestKernelPattern =
"{\n" "{\n"
"%s%s%s%s%s%s%s%s%s%s%s" "%s%s%s%s%s%s%s%s%s%s%s"
"}\n"; "}\n";
// clang-format on
static const char *arraySizeKernelLine = static const char *arraySizeKernelLine =
" outData->arraySize = get_image_array_size( input );\n"; " outData->arraySize = get_image_array_size( input );\n";
static const char *imageWidthKernelLine = static const char *imageWidthKernelLine =
" outData->width = get_image_width( input );\n"; " outData->width = get_image_width( input );\n";
static const char *imageHeightKernelLine = static const char *imageHeightKernelLine =
" outData->height = get_image_height( input );\n"; " outData->height = get_image_height( input );\n";
static const char *imageDimKernelLine = static const char *imageDimKernelLine =
" int2 dim = get_image_dim( input );\n"; " int2 dim = get_image_dim( input );\n";
static const char *imageWidthDimKernelLine = static const char *imageWidthDimKernelLine = " outData->widthDim = dim.x;\n";
" outData->widthDim = dim.x;\n";
static const char *imageHeightDimKernelLine = static const char *imageHeightDimKernelLine =
" outData->heightDim = dim.y;\n"; " outData->heightDim = dim.y;\n";
static const char *channelTypeKernelLine = 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 = static const char *channelTypeConstLine =
" outData->expectedChannelType = CLK_%s;\n"; " outData->expectedChannelType = CLK_%s;\n";
static const char *channelOrderKernelLine = static const char *channelOrderKernelLine =
" outData->channelOrder = get_image_channel_order( input );\n"; " outData->channelOrder = get_image_channel_order( input );\n";
static const char *channelOrderConstLine = static const char *channelOrderConstLine =
" outData->expectedChannelOrder = CLK_%s;\n"; " outData->expectedChannelOrder = CLK_%s;\n";
static const char *numSamplesKernelLine = static const char *numSamplesKernelLine =
" outData->numSamples = get_image_num_samples( input );\n"; " outData->numSamples = get_image_num_samples( input );\n";
static const char *enableMSAAKernelLine = 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 ) if (kernelOutput != input)
{ {
log_error( "ERROR: %s did not validate (expected %d, got %d)\n", description, input, kernelOutput); log_error("ERROR: %s did not validate (expected %d, got %d)\n",
description, input, kernelOutput);
return -1; return -1;
} }
return 0; return 0;
} }
extern int supportsMsaa(cl_context context, bool* supports_msaa); extern int supportsMsaa(cl_context context, bool *supports_msaa);
extern int supportsDepth(cl_context context, bool* supports_depth); 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, int test_image_format_methods(cl_device_id device, cl_context context,
size_t width, size_t height, size_t arraySize, size_t samples, cl_command_queue queue, size_t width,
GLenum target, format format, MTdata d ) 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; clProgramWrapper program;
clKernelWrapper kernel; clKernelWrapper kernel;
clMemWrapper image, outDataBuffer; clMemWrapper image, outDataBuffer;
char programSrc[ 10240 ]; char programSrc[10240];
image_kernel_data outKernelData; image_kernel_data outKernelData;
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE || if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE
get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) || get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY)
{ {
bool supports_msaa; bool supports_msaa;
error = supportsMsaa(context, &supports_msaa); error = supportsMsaa(context, &supports_msaa);
if( error != 0 ) return error; if (error != 0) return error;
if (!supports_msaa) return 0; if (!supports_msaa) return 0;
} }
if (format.formattype == GL_DEPTH_COMPONENT || if (format.formattype == GL_DEPTH_COMPONENT
format.formattype == GL_DEPTH_STENCIL) || format.formattype == GL_DEPTH_STENCIL)
{ {
bool supports_depth; bool supports_depth;
error = supportsDepth(context, &supports_depth); error = supportsDepth(context, &supports_depth);
if( error != 0 ) return error; if (error != 0) return error;
if (!supports_depth) return 0; if (!supports_depth) return 0;
} }
#endif #endif
DetectFloatToHalfRoundingMode(queue); DetectFloatToHalfRoundingMode(queue);
glTextureWrapper glTexture; glTextureWrapper glTexture;
switch (get_base_gl_target(target)) { switch (get_base_gl_target(target))
{
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
CreateGLTexture2D( width, height, target, CreateGLTexture2D(width, height, target, format.formattype,
format.formattype, format.internal, format.datatype, format.internal, format.datatype, format.type,
format.type, &glTexture, &error, false, d ); &glTexture, &error, false, d);
break; break;
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
CreateGLTexture2DArray( width, height, arraySize, target, CreateGLTexture2DArray(width, height, arraySize, target,
format.formattype, format.internal, format.datatype, format.formattype, format.internal,
format.type, &glTexture, &error, false, d ); format.datatype, format.type, &glTexture,
&error, false, d);
break; break;
case GL_TEXTURE_2D_MULTISAMPLE: case GL_TEXTURE_2D_MULTISAMPLE:
CreateGLTexture2DMultisample( width, height, samples, target, CreateGLTexture2DMultisample(width, height, samples, target,
format.formattype, format.internal, format.datatype, format.formattype, format.internal,
format.type, &glTexture, &error, false, d, false); format.datatype, format.type,
&glTexture, &error, false, d, false);
break; break;
case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: case GL_TEXTURE_2D_MULTISAMPLE_ARRAY:
CreateGLTexture2DArrayMultisample( width, height, arraySize, samples, target, CreateGLTexture2DArrayMultisample(
format.formattype, format.internal, format.datatype, width, height, arraySize, samples, target, format.formattype,
format.type, &glTexture, &error, false, d, false); format.internal, format.datatype, format.type, &glTexture,
&error, false, d, false);
break; break;
default: default:
log_error("Unsupported GL tex target (%s) passed to write test: " log_error("Unsupported GL tex target (%s) passed to write test: "
"%s (%s):%d", GetGLTargetName(target), __FUNCTION__, "%s (%s):%d",
__FILE__, __LINE__); GetGLTargetName(target), __FUNCTION__, __FILE__,
__LINE__);
} }
// Check to see if the texture could not be created for some other reason like // Check to see if the texture could not be created for some other reason
// GL_FRAMEBUFFER_UNSUPPORTED // like GL_FRAMEBUFFER_UNSUPPORTED
if (error == GL_FRAMEBUFFER_UNSUPPORTED) { if (error == GL_FRAMEBUFFER_UNSUPPORTED)
{
return 0; return 0;
} }
// Construct testing source // Construct testing source
log_info( " - Creating image %d by %d...\n", width, height ); log_info(" - Creating image %d by %d...\n", width, height);
// Create a CL image from the supplied GL texture // Create a CL image from the supplied GL texture
image = (*clCreateFromGLTexture_ptr)( context, CL_MEM_READ_ONLY, image = (*clCreateFromGLTexture_ptr)(context, CL_MEM_READ_ONLY, target, 0,
target, 0, glTexture, &error ); glTexture, &error);
if ( error != CL_SUCCESS ) { if (error != CL_SUCCESS)
print_error( error, "Unable to create CL image from GL texture" ); {
print_error(error, "Unable to create CL image from GL texture");
GLint fmt; GLint fmt;
glGetTexLevelParameteriv( target, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); glGetTexLevelParameteriv(target, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt);
log_error( " Supplied GL texture was base format %s and internal " log_error(" Supplied GL texture was base format %s and internal "
"format %s\n", GetGLBaseFormatName( fmt ), GetGLFormatName( fmt ) ); "format %s\n",
GetGLBaseFormatName(fmt), GetGLFormatName(fmt));
return error; return error;
} }
cl_image_format imageFormat; cl_image_format imageFormat;
error = clGetImageInfo (image, CL_IMAGE_FORMAT, error = clGetImageInfo(image, CL_IMAGE_FORMAT, sizeof(imageFormat),
sizeof(imageFormat), &imageFormat, NULL); &imageFormat, NULL);
test_error(error, "Failed to get image format"); test_error(error, "Failed to get image format");
const char * imageType = 0; const char *imageType = 0;
bool doArraySize = false; bool doArraySize = false;
bool doImageWidth = false; bool doImageWidth = false;
bool doImageHeight = false; bool doImageHeight = false;
@@ -189,7 +200,8 @@ int test_image_format_methods( cl_device_id device, cl_context context, cl_comma
bool doImageDim = false; bool doImageDim = false;
bool doNumSamples = false; bool doNumSamples = false;
bool doMSAA = false; bool doMSAA = false;
switch(target) { switch (target)
{
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
imageType = "image2d_depth_t"; imageType = "image2d_depth_t";
doImageWidth = true; doImageWidth = true;
@@ -211,81 +223,91 @@ int test_image_format_methods( cl_device_id device, cl_context context, cl_comma
case GL_TEXTURE_2D_MULTISAMPLE: case GL_TEXTURE_2D_MULTISAMPLE:
doNumSamples = true; doNumSamples = true;
doMSAA = true; doMSAA = true;
if(format.formattype == GL_DEPTH_COMPONENT) { if (format.formattype == GL_DEPTH_COMPONENT)
{
doImageWidth = true; doImageWidth = true;
imageType = "image2d_msaa_depth_t"; imageType = "image2d_msaa_depth_t";
} else { }
else
{
imageType = "image2d_msaa_t"; imageType = "image2d_msaa_t";
} }
break; break;
case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: case GL_TEXTURE_2D_MULTISAMPLE_ARRAY:
doMSAA = true; doMSAA = true;
if(format.formattype == GL_DEPTH_COMPONENT) { if (format.formattype == GL_DEPTH_COMPONENT)
{
doImageWidth = true; doImageWidth = true;
imageType = "image2d_msaa_array_depth_t"; imageType = "image2d_msaa_array_depth_t";
} else { }
else
{
imageType = "image2d_array_msaa_t"; imageType = "image2d_array_msaa_t";
} }
break; break;
} }
char channelTypeConstKernelLine[512] = { 0 };
char channelTypeConstKernelLine[512] = {0}; char channelOrderConstKernelLine[512] = { 0 };
char channelOrderConstKernelLine[512] = {0}; const char *channelTypeName = 0;
const char* channelTypeName=0; const char *channelOrderName = 0;
const char* channelOrderName=0; if (doImageChannelDataType)
if(doImageChannelDataType) { {
channelTypeName = GetChannelTypeName( imageFormat.image_channel_data_type ); channelTypeName =
if(channelTypeName && strlen(channelTypeName)) { GetChannelTypeName(imageFormat.image_channel_data_type);
if (channelTypeName && strlen(channelTypeName))
{
// replace CL_* with CLK_* // replace CL_* with CLK_*
sprintf(channelTypeConstKernelLine, channelTypeConstLine, &channelTypeName[3]); sprintf(channelTypeConstKernelLine, channelTypeConstLine,
&channelTypeName[3]);
} }
} }
if(doImageChannelOrder) { if (doImageChannelOrder)
channelOrderName = GetChannelOrderName( imageFormat.image_channel_order ); {
if(channelOrderName && strlen(channelOrderName)) { channelOrderName = GetChannelOrderName(imageFormat.image_channel_order);
if (channelOrderName && strlen(channelOrderName))
{
// replace CL_* with CLK_* // replace CL_* with CLK_*
sprintf(channelOrderConstKernelLine, channelOrderConstLine, &channelOrderName[3]); sprintf(channelOrderConstKernelLine, channelOrderConstLine,
&channelOrderName[3]);
} }
} }
// Create a program to run against // Create a program to run against
sprintf(programSrc, sprintf(programSrc, methodTestKernelPattern,
methodTestKernelPattern, (doMSAA) ? enableMSAAKernelLine : "", imageType,
( doMSAA ) ? enableMSAAKernelLine : "", (doArraySize) ? arraySizeKernelLine : "",
imageType, (doImageWidth) ? imageWidthKernelLine : "",
( doArraySize ) ? arraySizeKernelLine : "", (doImageHeight) ? imageHeightKernelLine : "",
( doImageWidth ) ? imageWidthKernelLine : "", (doImageChannelDataType) ? channelTypeKernelLine : "",
( doImageHeight ) ? imageHeightKernelLine : "", (doImageChannelDataType) ? channelTypeConstKernelLine : "",
( doImageChannelDataType ) ? channelTypeKernelLine : "", (doImageChannelOrder) ? channelOrderKernelLine : "",
( doImageChannelDataType ) ? channelTypeConstKernelLine : "", (doImageChannelOrder) ? channelOrderConstKernelLine : "",
( doImageChannelOrder ) ? channelOrderKernelLine : "", (doImageDim) ? imageDimKernelLine : "",
( doImageChannelOrder ) ? channelOrderConstKernelLine : "", (doImageDim && doImageWidth) ? imageWidthDimKernelLine : "",
( doImageDim ) ? imageDimKernelLine : "", (doImageDim && doImageHeight) ? imageHeightDimKernelLine : "",
( doImageDim && doImageWidth ) ? imageWidthDimKernelLine : "", (doNumSamples) ? numSamplesKernelLine : "");
( doImageDim && doImageHeight ) ? imageHeightDimKernelLine : "",
( doNumSamples ) ? numSamplesKernelLine : "");
//log_info("-----------------------------------\n%s\n", programSrc); // log_info("-----------------------------------\n%s\n", programSrc);
error = clFinish(queue); error = clFinish(queue);
if (error) if (error) print_error(error, "clFinish failed.\n");
print_error(error, "clFinish failed.\n");
const char *ptr = programSrc; const char *ptr = programSrc;
error = create_single_kernel_helper( context, &program, &kernel, 1, &ptr, "sample_kernel" ); error = create_single_kernel_helper(context, &program, &kernel, 1, &ptr,
test_error( error, "Unable to create kernel to test against" ); "sample_kernel");
test_error(error, "Unable to create kernel to test against");
// Create an output buffer // Create an output buffer
outDataBuffer = clCreateBuffer(context, CL_MEM_READ_WRITE, outDataBuffer = clCreateBuffer(context, CL_MEM_READ_WRITE,
sizeof(outKernelData), NULL, &error); 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 // Set up arguments and run
error = clSetKernelArg( kernel, 0, sizeof( image ), &image ); error = clSetKernelArg(kernel, 0, sizeof(image), &image);
test_error( error, "Unable to set kernel argument" ); test_error(error, "Unable to set kernel argument");
error = clSetKernelArg( kernel, 1, sizeof( outDataBuffer ), &outDataBuffer ); error = clSetKernelArg(kernel, 1, sizeof(outDataBuffer), &outDataBuffer);
test_error( error, "Unable to set kernel argument" ); test_error(error, "Unable to set kernel argument");
// Finish and Acquire. // Finish and Acquire.
glFinish(); glFinish();
@@ -294,91 +316,115 @@ int test_image_format_methods( cl_device_id device, cl_context context, cl_comma
size_t threads[1] = { 1 }, localThreads[1] = { 1 }; size_t threads[1] = { 1 }, localThreads[1] = { 1 };
error = clEnqueueNDRangeKernel( queue, kernel, 1, NULL, threads, localThreads, 0, NULL, NULL ); error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, threads,
test_error( error, "Unable to run kernel" ); localThreads, 0, NULL, NULL);
test_error(error, "Unable to run kernel");
error = clEnqueueReadBuffer( queue, outDataBuffer, CL_TRUE, 0, sizeof( outKernelData ), &outKernelData, 0, NULL, NULL ); error = clEnqueueReadBuffer(queue, outDataBuffer, CL_TRUE, 0,
test_error( error, "Unable to read data buffer" ); sizeof(outKernelData), &outKernelData, 0, NULL,
NULL);
test_error(error, "Unable to read data buffer");
// Verify the results now // Verify the results now
if( doImageWidth ) if (doImageWidth) result |= verify(width, outKernelData.width, "width");
result |= verify(width, outKernelData.width, "width"); if (doImageHeight) result |= verify(height, outKernelData.height, "height");
if( doImageHeight) if (doImageDim && doImageWidth)
result |= verify(height, outKernelData.height, "height"); result |=
if( doImageDim && doImageWidth ) verify(width, outKernelData.widthDim, "width from get_image_dim");
result |= verify(width, outKernelData.widthDim, "width from get_image_dim"); if (doImageDim && doImageHeight)
if( doImageDim && doImageHeight ) result |= verify(height, outKernelData.heightDim,
result |= verify(height, outKernelData.heightDim, "height from get_image_dim"); "height from get_image_dim");
if( doImageChannelDataType ) if (doImageChannelDataType)
result |= verify(outKernelData.channelType, outKernelData.expectedChannelType, channelTypeName); result |= verify(outKernelData.channelType,
if( doImageChannelOrder ) outKernelData.expectedChannelType, channelTypeName);
result |= verify(outKernelData.channelOrder, outKernelData.expectedChannelOrder, channelOrderName); if (doImageChannelOrder)
if( doArraySize ) result |= verify(outKernelData.channelOrder,
outKernelData.expectedChannelOrder, channelOrderName);
if (doArraySize)
result |= verify(arraySize, outKernelData.arraySize, "array size"); result |= verify(arraySize, outKernelData.arraySize, "array size");
if( doNumSamples ) if (doNumSamples)
result |= verify(samples, outKernelData.numSamples, "samples"); result |= verify(samples, outKernelData.numSamples, "samples");
if(result) { if (result)
{
log_error("Test image methods failed"); log_error("Test image methods failed");
} }
clEventWrapper event; clEventWrapper event;
error = (*clEnqueueReleaseGLObjects_ptr)( queue, 1, &image, 0, NULL, &event ); error = (*clEnqueueReleaseGLObjects_ptr)(queue, 1, &image, 0, NULL, &event);
test_error(error, "clEnqueueReleaseGLObjects failed"); test_error(error, "clEnqueueReleaseGLObjects failed");
error = clWaitForEvents( 1, &event ); error = clWaitForEvents(1, &event);
test_error(error, "clWaitForEvents failed"); test_error(error, "clWaitForEvents failed");
return result; return result;
} }
int test_image_methods_depth( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ){ int test_image_methods_depth(cl_device_id device, cl_context context,
if (!is_extension_available(device, "cl_khr_gl_depth_images")) { cl_command_queue queue, int numElements)
log_info("Test not run because 'cl_khr_gl_depth_images' extension is not supported by the tested device\n"); {
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; return 0;
} }
int result = 0; int result = 0;
GLenum depth_targets[] = {GL_TEXTURE_2D, GL_TEXTURE_2D_ARRAY}; GLenum depth_targets[] = { GL_TEXTURE_2D, GL_TEXTURE_2D_ARRAY };
size_t ntargets = sizeof(depth_targets) / sizeof(depth_targets[0]); size_t ntargets = sizeof(depth_targets) / sizeof(depth_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 = 5; const size_t nsizes = 5;
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
// Need to limit texture size according to GL device properties // Need to limit texture size according to GL device properties
GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, maxTextureLayers = 16, size; GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096,
maxTextureLayers = 16, size;
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize);
glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); 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) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width = random_in_range(2, min(size, 1 << (i + 4)), seed);
sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 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 i = 0; i < nsizes; i++)
for(size_t itarget = 0; itarget < ntargets; ++itarget) { {
for(size_t iformat = 0; iformat < nformats; ++iformat) for (size_t itarget = 0; itarget < ntargets; ++itarget)
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 ); 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; return result;
} }
int test_image_methods_multisample( cl_device_id device, cl_context context, cl_command_queue queue, int numElements ){ int test_image_methods_multisample(cl_device_id device, cl_context context,
if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) { cl_command_queue queue, int numElements)
log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is not supported by the tested device\n"); {
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; return 0;
} }
int result = 0; int result = 0;
GLenum targets[] = {GL_TEXTURE_2D_MULTISAMPLE, GL_TEXTURE_2D_MULTISAMPLE_ARRAY}; GLenum targets[] = { GL_TEXTURE_2D_MULTISAMPLE,
GL_TEXTURE_2D_MULTISAMPLE_ARRAY };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); 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]);
@@ -388,23 +434,35 @@ int test_image_methods_multisample( cl_device_id device, cl_context context, cl_
glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers);
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width =
sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); 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); glEnable(GL_MULTISAMPLE);
for (size_t i = 0; i < nsizes; i++) { for (size_t i = 0; i < nsizes; i++)
for(size_t itarget = 0; itarget < ntargets; ++itarget) { {
for(size_t iformat = 0; iformat < nformats; ++iformat) { for (size_t itarget = 0; itarget < ntargets; ++itarget)
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, for (size_t iformat = 0; iformat < nformats; ++iformat)
samples, targets[itarget], common_formats[iformat], seed ); {
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);
} }
} }
} }

100
test_conformance/gl/test_images_1D.cpp
View File

@@ -16,11 +16,11 @@
#include "common.h" #include "common.h"
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#include <algorithm> #include <algorithm>
@@ -35,18 +35,19 @@ void calc_test_size_descriptors(sizevec_t* sizes, size_t nsizes)
size = min(maxTextureSize, maxTextureBufferSize); size = min(maxTextureSize, maxTextureBufferSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { for (size_t i = 0; i < nsizes; i++)
sizes[i].width = random_in_range( 2, min(size, 1<<(i+4)), seed ); {
sizes[i].width = random_in_range(2, min(size, 1 << (i + 4)), seed);
sizes[i].height = 1; sizes[i].height = 1;
sizes[i].depth = 1; sizes[i].depth = 1;
} }
} }
int test_images_read_1D( cl_device_id device, cl_context context, int test_images_read_1D(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) 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]);
@@ -61,8 +62,8 @@ int test_images_read_1D( cl_device_id device, cl_context context,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_write_1D( cl_device_id device, cl_context context, int test_images_write_1D(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_1D }; GLenum targets[] = { GL_TEXTURE_1D };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -72,28 +73,29 @@ int test_images_write_1D( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_test_size_descriptors(sizes, nsizes); calc_test_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, 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 )
{
size_t nformats = sizeof(common_formats) / sizeof(common_formats[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);
return test_images_get_info_common( device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_read_texturebuffer( cl_device_id device, cl_context context, int test_images_1D_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{
size_t nformats = sizeof(common_formats) / sizeof(common_formats[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);
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)
{ {
size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]); size_t nformats = sizeof(common_formats) / sizeof(common_formats[0]);
@@ -108,8 +110,8 @@ int test_images_read_texturebuffer( cl_device_id device, cl_context context,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_write_texturebuffer( cl_device_id device, cl_context context, int test_images_write_texturebuffer(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_BUFFER }; GLenum targets[] = { GL_TEXTURE_BUFFER };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -119,23 +121,23 @@ int test_images_write_texturebuffer( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_test_size_descriptors(sizes, nsizes); calc_test_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, 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 )
{
size_t nformats = sizeof(common_formats) / sizeof(common_formats[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);
return test_images_get_info_common( device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
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]);
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);
return test_images_get_info_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes,
nsizes);
}

38
test_conformance/gl/test_images_1Darray.cpp
View File

@@ -16,11 +16,11 @@
#include "common.h" #include "common.h"
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#include <algorithm> #include <algorithm>
@@ -32,18 +32,21 @@ void calc_1D_array_size_descriptors(sizevec_t* sizes, size_t nsizes)
glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers);
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].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; sizes[i].depth = 1;
} }
} }
int test_images_read_1Darray( cl_device_id device, cl_context context, int test_images_read_1Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int ) 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]);
@@ -58,8 +61,8 @@ int test_images_read_1Darray( cl_device_id device, cl_context context,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_write_1Darray( cl_device_id device, cl_context context, int test_images_write_1Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_1D_ARRAY }; GLenum targets[] = { GL_TEXTURE_1D_ARRAY };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -69,12 +72,12 @@ int test_images_write_1Darray( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_1D_array_size_descriptors(sizes, nsizes); calc_1D_array_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, return test_images_write_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_1Darray_getinfo( cl_device_id device, cl_context context, int test_images_1Darray_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int ) 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]);
@@ -86,5 +89,6 @@ int test_images_1Darray_getinfo( cl_device_id device, cl_context context,
calc_1D_array_size_descriptors(sizes, nsizes); calc_1D_array_size_descriptors(sizes, nsizes);
return test_images_get_info_common(device, context, queue, common_formats, return test_images_get_info_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes,
nsizes);
} }

85
test_conformance/gl/test_images_2D.cpp
View File

@@ -16,7 +16,7 @@
#include "testBase.h" #include "testBase.h"
#include "common.h" #include "common.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
@@ -39,12 +39,13 @@ void calc_2D_test_size_descriptors(sizevec_t* sizes, size_t nsizes)
size = min(maxTextureSize, maxTextureRectangleSize); size = min(maxTextureSize, maxTextureRectangleSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].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; sizes[i].depth = 1;
} }
} }
@@ -56,17 +57,19 @@ void calc_cube_test_size_descriptors(sizevec_t* sizes, size_t nsizes)
GLint maxQubeMapSize = 4096; GLint maxQubeMapSize = 4096;
glGetIntegerv(GL_MAX_CUBE_MAP_TEXTURE_SIZE, &maxQubeMapSize); glGetIntegerv(GL_MAX_CUBE_MAP_TEXTURE_SIZE, &maxQubeMapSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width = sizes[i].height =
random_in_range(2, min(maxQubeMapSize, 1 << (i + 4)), seed);
sizes[i].depth = 1; sizes[i].depth = 1;
} }
} }
int test_images_read_2D( cl_device_id device, cl_context context, int test_images_read_2D(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -81,16 +84,14 @@ int test_images_read_2D( cl_device_id device, cl_context context,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_read_cube( cl_device_id device, cl_context context, int test_images_read_cube(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GLenum targets[] = {
GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y,
GL_TEXTURE_CUBE_MAP_POSITIVE_Y, GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X,
GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_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 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]);
@@ -108,8 +109,8 @@ int test_images_read_cube( cl_device_id device, cl_context context,
#include "common.h" #include "common.h"
int test_images_write( cl_device_id device, cl_context context, int test_images_write(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -119,22 +120,19 @@ int test_images_write( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_2D_test_size_descriptors(sizes, nsizes); calc_2D_test_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, return test_images_write_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_write_cube( cl_device_id device, cl_context context, int test_images_write_cube(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) 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[] = { GLenum targets[] = {
GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y,
GL_TEXTURE_CUBE_MAP_POSITIVE_Y, GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X,
GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_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 ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -142,15 +140,15 @@ int test_images_write_cube( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_cube_test_size_descriptors(sizes, nsizes); calc_cube_test_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, return test_images_write_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }
#pragma mark - #pragma mark -
#pragma mark _2D get info tests #pragma mark _2D get info tests
int test_images_2D_getinfo( cl_device_id device, cl_context context, int test_images_2D_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -162,19 +160,17 @@ int test_images_2D_getinfo( cl_device_id device, cl_context context,
calc_2D_test_size_descriptors(sizes, nsizes); calc_2D_test_size_descriptors(sizes, nsizes);
return test_images_get_info_common(device, context, queue, common_formats, return test_images_get_info_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes,
nsizes);
} }
int test_images_cube_getinfo( cl_device_id device, cl_context context, int test_images_cube_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GLenum targets[] = {
GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y,
GL_TEXTURE_CUBE_MAP_POSITIVE_Y, GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X,
GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_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 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]);
@@ -184,5 +180,6 @@ int test_images_cube_getinfo( cl_device_id device, cl_context context,
calc_cube_test_size_descriptors(sizes, nsizes); calc_cube_test_size_descriptors(sizes, nsizes);
return test_images_get_info_common(device, context, queue, common_formats, return test_images_get_info_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes,
nsizes);
} }

41
test_conformance/gl/test_images_2Darray.cpp
View File

@@ -16,11 +16,11 @@
#include "common.h" #include "common.h"
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#include <algorithm> #include <algorithm>
@@ -33,18 +33,22 @@ void calc_2D_array_size_descriptors(sizevec_t* sizes, size_t nsizes)
glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers);
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width =
sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); 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, int test_images_read_2Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int ) 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]);
@@ -59,8 +63,8 @@ int test_images_read_2Darray( cl_device_id device, cl_context context,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_write_2Darray( cl_device_id device, cl_context context, int test_images_write_2Darray(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
// FIXME: Query for 2D image array write support. // FIXME: Query for 2D image array write support.
@@ -72,12 +76,12 @@ int test_images_write_2Darray( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_2D_array_size_descriptors(sizes, nsizes); calc_2D_array_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, return test_images_write_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_2Darray_getinfo( cl_device_id device, cl_context context, int test_images_2Darray_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int ) 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]);
@@ -89,5 +93,6 @@ int test_images_2Darray_getinfo( cl_device_id device, cl_context context,
calc_2D_array_size_descriptors(sizes, nsizes); calc_2D_array_size_descriptors(sizes, nsizes);
return test_images_get_info_common(device, context, queue, common_formats, return test_images_get_info_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes,
nsizes);
} }

46
test_conformance/gl/test_images_3D.cpp
View File

@@ -16,11 +16,11 @@
#include "testBase.h" #include "testBase.h"
#include "common.h" #include "common.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#include <algorithm> #include <algorithm>
@@ -35,17 +35,22 @@ void calc_3D_size_descriptors(sizevec_t* sizes, size_t nsizes)
GLint maxTextureSize = 2048; GLint maxTextureSize = 2048;
glGetIntegerv(GL_MAX_3D_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_3D_TEXTURE_SIZE, &maxTextureSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width =
sizes[i].depth = random_in_range( 2, min(maxTextureSize, 1<<(i+4)), seed ); 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 }; GLenum targets[] = { GL_TEXTURE_3D };
size_t ntargets = 1; size_t ntargets = 1;
@@ -63,14 +68,16 @@ int test_images_read_3D( cl_device_id device, cl_context context, cl_command_que
#pragma mark - #pragma mark -
#pragma marm _3D write test #pragma marm _3D write test
int test_images_write_3D( cl_device_id device, cl_context context, int test_images_write_3D(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
// TODO: Perhaps the expected behavior is to FAIL if 3D images are // TODO: Perhaps the expected behavior is to FAIL if 3D images are
// unsupported? // unsupported?
if (!is_extension_available(device, "cl_khr_3d_image_writes")) { if (!is_extension_available(device, "cl_khr_3d_image_writes"))
log_info("This device does not support 3D image writes. Skipping test.\n"); {
log_info(
"This device does not support 3D image writes. Skipping test.\n");
return 0; return 0;
} }
@@ -83,15 +90,15 @@ int test_images_write_3D( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_3D_size_descriptors(sizes, nsizes); calc_3D_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, common_formats, return test_images_write_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }
#pragma mark - #pragma mark -
#pragma mark _3D get info test #pragma mark _3D get info test
int test_images_3D_getinfo( cl_device_id device, cl_context context, int test_images_3D_getinfo(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
GLenum targets[] = { GL_TEXTURE_3D }; GLenum targets[] = { GL_TEXTURE_3D };
size_t ntargets = 1; size_t ntargets = 1;
@@ -103,5 +110,6 @@ int test_images_3D_getinfo( cl_device_id device, cl_context context,
calc_3D_size_descriptors(sizes, nsizes); calc_3D_size_descriptors(sizes, nsizes);
return test_images_get_info_common(device, context, queue, common_formats, return test_images_get_info_common(device, context, queue, common_formats,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes,
nsizes);
} }

77
test_conformance/gl/test_images_depth.cpp
View File

@@ -16,7 +16,7 @@
#include "testBase.h" #include "testBase.h"
#include "common.h" #include "common.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
@@ -39,12 +39,13 @@ void calc_depth_size_descriptors(sizevec_t* sizes, size_t nsizes)
size = min(maxTextureSize, maxTextureRectangleSize); size = min(maxTextureSize, maxTextureRectangleSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].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; sizes[i].depth = 1;
} }
} }
@@ -52,32 +53,37 @@ void calc_depth_size_descriptors(sizevec_t* sizes, size_t nsizes)
void calc_depth_array_size_descriptors(sizevec_t* sizes, size_t nsizes) void calc_depth_array_size_descriptors(sizevec_t* sizes, size_t nsizes)
{ {
// Need to limit texture size according to GL device properties // Need to limit texture size according to GL device properties
GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096, maxTextureLayers = 16, size; GLint maxTextureSize = 4096, maxTextureRectangleSize = 4096,
maxTextureLayers = 16, size;
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize); glGetIntegerv(GL_MAX_RECTANGLE_TEXTURE_SIZE_EXT, &maxTextureRectangleSize);
glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); 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) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width = random_in_range(2, min(size, 1 << (i + 4)), seed);
sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 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, int test_images_read_2D_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
if (!is_extension_available(device, "cl_khr_gl_depth_images")) { 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"); {
log_info("Test not run because 'cl_khr_gl_depth_images' extension is "
"not supported by the tested device\n");
return 0; return 0;
} }
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT }; GLenum targets[] = { GL_TEXTURE_2D, GL_TEXTURE_RECTANGLE_EXT };
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
@@ -96,11 +102,13 @@ int test_images_read_2D_depth( cl_device_id device, cl_context context,
#pragma mark _2D depth write tests #pragma mark _2D depth write tests
int test_images_write_2D_depth( cl_device_id device, cl_context context, int test_images_write_2D_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
if (!is_extension_available(device, "cl_khr_gl_depth_images")) { 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"); {
log_info("Test not run because 'cl_khr_gl_depth_images' extension is "
"not supported by the tested device\n");
return 0; return 0;
} }
@@ -112,15 +120,17 @@ int test_images_write_2D_depth( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_depth_size_descriptors(sizes, nsizes); calc_depth_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, depth_formats, return test_images_write_common(device, context, queue, depth_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_read_2Darray_depth( cl_device_id device, cl_context context, int test_images_read_2Darray_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int ) cl_command_queue queue, int)
{ {
if (!is_extension_available(device, "cl_khr_gl_depth_images")) { 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"); {
log_info("Test not run because 'cl_khr_gl_depth_images' extension is "
"not supported by the tested device\n");
return 0; return 0;
} }
@@ -136,11 +146,13 @@ int test_images_read_2Darray_depth( cl_device_id device, cl_context context,
nformats, targets, ntargets, sizes, nsizes); nformats, targets, ntargets, sizes, nsizes);
} }
int test_images_write_2Darray_depth( cl_device_id device, cl_context context, int test_images_write_2Darray_depth(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
if (!is_extension_available(device, "cl_khr_gl_depth_images")) { 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"); {
log_info("Test not run because 'cl_khr_gl_depth_images' extension is "
"not supported by the tested device\n");
return 0; return 0;
} }
@@ -154,7 +166,6 @@ int test_images_write_2Darray_depth( cl_device_id device, cl_context context,
sizevec_t sizes[nsizes]; sizevec_t sizes[nsizes];
calc_depth_array_size_descriptors(sizes, nsizes); calc_depth_array_size_descriptors(sizes, nsizes);
return test_images_write_common( device, context, queue, depth_formats, return test_images_write_common(device, context, queue, depth_formats,
nformats, targets, ntargets, sizes, nsizes ); nformats, targets, ntargets, sizes, nsizes);
} }

177
test_conformance/gl/test_images_getinfo_common.cpp
View File

@@ -16,46 +16,47 @@
#include "testBase.h" #include "testBase.h"
#include "common.h" #include "common.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #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, static int test_image_info(cl_context context, cl_command_queue queue,
GLenum glTarget, GLuint glTexture, size_t imageWidth, size_t imageHeight, GLenum glTarget, GLuint glTexture, size_t imageWidth,
size_t imageDepth, cl_image_format *outFormat, ExplicitType *outType, size_t imageHeight, size_t imageDepth,
void **outResultBuffer ) 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 // Create a CL image from the supplied GL texture
streams[ 0 ] = (*clCreateFromGLTexture_ptr)( context, CL_MEM_READ_ONLY, streams[0] = (*clCreateFromGLTexture_ptr)(context, CL_MEM_READ_ONLY,
glTarget, 0, glTexture, &error ); glTarget, 0, glTexture, &error);
if( error != CL_SUCCESS ) if (error != CL_SUCCESS)
{ {
print_error( error, "Unable to create CL image from GL texture" ); print_error(error, "Unable to create CL image from GL texture");
GLint fmt; GLint fmt;
glGetTexLevelParameteriv( glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); glGetTexLevelParameteriv(glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt);
log_error( " Supplied GL texture was format %s\n", GetGLFormatName( fmt ) ); log_error(" Supplied GL texture was format %s\n",
GetGLFormatName(fmt));
return error; return error;
} }
// Determine data type and format that CL came up with // Determine data type and format that CL came up with
error = clGetImageInfo( streams[ 0 ], CL_IMAGE_FORMAT, error = clGetImageInfo(streams[0], CL_IMAGE_FORMAT, sizeof(cl_image_format),
sizeof( cl_image_format ), outFormat, NULL ); outFormat, NULL);
test_error( error, "Unable to get CL image format" ); test_error(error, "Unable to get CL image format");
cl_gl_object_type object_type; cl_gl_object_type object_type;
switch (glTarget) { switch (glTarget)
case GL_TEXTURE_1D: {
object_type = CL_GL_OBJECT_TEXTURE1D; case GL_TEXTURE_1D: object_type = CL_GL_OBJECT_TEXTURE1D; break;
break;
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER:
object_type = CL_GL_OBJECT_TEXTURE_BUFFER; object_type = CL_GL_OBJECT_TEXTURE_BUFFER;
break; break;
@@ -75,12 +76,8 @@ static int test_image_info( cl_context context, cl_command_queue queue,
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
object_type = CL_GL_OBJECT_TEXTURE2D_ARRAY; object_type = CL_GL_OBJECT_TEXTURE2D_ARRAY;
break; break;
case GL_TEXTURE_3D: case GL_TEXTURE_3D: object_type = CL_GL_OBJECT_TEXTURE3D; break;
object_type = CL_GL_OBJECT_TEXTURE3D; default: log_error("Unsupported texture target."); return 1;
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);
@@ -97,14 +94,12 @@ static int test_image_format_get_info(cl_context context,
// If we're testing a half float format, then we need to determine the // 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. // rounding mode of this machine. Punt if we fail to do so.
if( fmt->type == kHalf ) if (fmt->type == kHalf)
{ {
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 0;
return 0;
bool supports_half = false; bool supports_half = false;
error = supportsHalf(context, &supports_half); error = supportsHalf(context, &supports_half);
if( error != 0 ) if (error != 0) return error;
return error;
if (!supports_half) return 0; if (!supports_half) return 0;
} }
@@ -124,29 +119,33 @@ static int test_image_format_get_info(cl_context context,
// If we're testing a half float format, then we need to determine the // 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. // rounding mode of this machine. Punt if we fail to do so.
if( fmt->type == kHalf ) if (fmt->type == kHalf)
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 1;
return 1;
// Use the correct texture creation function depending on the target, and // Use the correct texture creation function depending on the target, and
// adjust width, height, depth as appropriate so subsequent size calculations // adjust width, height, depth as appropriate so subsequent size
// succeed. // calculations succeed.
switch (target) { switch (target)
{
case GL_TEXTURE_1D: case GL_TEXTURE_1D:
h = 1; d = 1; h = 1;
CreateGLTexture1D( width, target, gl_fmt, d = 1;
gl_internal_fmt, gl_type, type, &texture, &error, false, data ); CreateGLTexture1D(width, target, gl_fmt, gl_internal_fmt, gl_type,
type, &texture, &error, false, data);
break; break;
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER:
h = 1; d = 1; h = 1;
CreateGLTextureBuffer( width, target, gl_fmt, d = 1;
gl_internal_fmt, gl_type, type, &texture, &glbuf, &error, false, data ); CreateGLTextureBuffer(width, target, gl_fmt, gl_internal_fmt,
gl_type, type, &texture, &glbuf, &error,
false, data);
break; break;
case GL_TEXTURE_1D_ARRAY: case GL_TEXTURE_1D_ARRAY:
d = 1; d = 1;
CreateGLTexture1DArray( width, height, target, gl_fmt, CreateGLTexture1DArray(width, height, target, gl_fmt,
gl_internal_fmt, gl_type, type, &texture, &error, false, data ); gl_internal_fmt, gl_type, type, &texture,
&error, false, data);
break; break;
case GL_TEXTURE_RECTANGLE_EXT: case GL_TEXTURE_RECTANGLE_EXT:
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
@@ -157,34 +156,41 @@ static int test_image_format_get_info(cl_context context,
case GL_TEXTURE_CUBE_MAP_NEGATIVE_Y: case GL_TEXTURE_CUBE_MAP_NEGATIVE_Y:
case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z: case GL_TEXTURE_CUBE_MAP_NEGATIVE_Z:
d = 1; d = 1;
CreateGLTexture2D( width, height, target, gl_fmt, CreateGLTexture2D(width, height, target, gl_fmt, gl_internal_fmt,
gl_internal_fmt, gl_type, type, &texture, &error, false, data ); gl_type, type, &texture, &error, false, data);
break; break;
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
CreateGLTexture2DArray( width, height, depth, target, gl_fmt, CreateGLTexture2DArray(width, height, depth, target, gl_fmt,
gl_internal_fmt, gl_type, type, &texture, &error, false, data ); gl_internal_fmt, gl_type, type, &texture,
&error, false, data);
break; break;
case GL_TEXTURE_3D: case GL_TEXTURE_3D:
d = 1; d = 1;
CreateGLTexture3D( width, height, depth, target, gl_fmt, CreateGLTexture3D(width, height, depth, target, gl_fmt,
gl_internal_fmt, gl_type, type, &texture, &error, data, false ); gl_internal_fmt, gl_type, type, &texture, &error,
data, false);
break; break;
default: default: log_error("Unsupported texture target.\n"); return 1;
log_error("Unsupported texture target.\n");
return 1;
} }
if ( error == -2 ) { if (error == -2)
log_info("OpenGL texture couldn't be created, because a texture is too big. Skipping test.\n"); {
log_info("OpenGL texture couldn't be created, because a texture is too "
"big. Skipping test.\n");
return 0; return 0;
} }
if ( error != 0 ) { if (error != 0)
if ((gl_fmt == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())) { {
if ((gl_fmt == GL_RGBA_INTEGER_EXT)
&& (!CheckGLIntegerExtensionSupport()))
{
log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. "
"Skipping test.\n"); "Skipping test.\n");
return 0; return 0;
} else { }
else
{
return error; return error;
} }
} }
@@ -194,8 +200,8 @@ static int test_image_format_get_info(cl_context context,
char *outBuffer; char *outBuffer;
// Perform the info check: // Perform the info check:
return test_image_info( context, queue, target, texture, w, h, d, &clFormat, return test_image_info(context, queue, target, texture, w, h, d, &clFormat,
&actualType, (void **)&outBuffer ); &actualType, (void **)&outBuffer);
} }
int test_images_get_info_common(cl_device_id device, cl_context context, int test_images_get_info_common(cl_device_id device, cl_context context,
@@ -209,7 +215,8 @@ int test_images_get_info_common(cl_device_id device, cl_context context,
// First, ensure this device supports images. // First, ensure this device supports images.
if (checkForImageSupport(device)) { if (checkForImageSupport(device))
{
log_info("Device does not support images. Skipping test.\n"); log_info("Device does not support images. Skipping test.\n");
return 0; return 0;
} }
@@ -218,37 +225,41 @@ int test_images_get_info_common(cl_device_id device, cl_context context,
// Test each format on every target, every size. // Test each format on every target, every size.
for ( fidx = 0; fidx < nformats; fidx++ ) { for (fidx = 0; fidx < nformats; fidx++)
for ( tidx = 0; tidx < ntargets; tidx++ ) { {
for (tidx = 0; tidx < ntargets; tidx++)
{
if ( formats[ fidx ].datatype == GL_UNSIGNED_INT_2_10_10_10_REV ) if (formats[fidx].datatype == GL_UNSIGNED_INT_2_10_10_10_REV)
{ {
// Check if the RGB 101010 format is supported // Check if the RGB 101010 format is supported
if ( is_rgb_101010_supported( context, targets[ tidx ] ) == 0 ) if (is_rgb_101010_supported(context, targets[tidx]) == 0)
break; // skip break; // skip
} }
log_info( "Testing image info for GL format %s : %s : %s : %s\n", log_info("Testing image info for GL format %s : %s : %s : %s\n",
GetGLTargetName( targets[ tidx ] ), GetGLTargetName(targets[tidx]),
GetGLFormatName( formats[ fidx ].internal ), GetGLFormatName(formats[fidx].internal),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLTypeName(formats[fidx].datatype));
for ( sidx = 0; sidx < nsizes; sidx++ ) { for (sidx = 0; sidx < nsizes; sidx++)
{
// Test this format + size: // Test this format + size:
if ( test_image_format_get_info(context, queue, if (test_image_format_get_info(
sizes[sidx].width, sizes[sidx].height, sizes[sidx].depth, context, queue, sizes[sidx].width, sizes[sidx].height,
targets[tidx], &formats[fidx], seed) ) sizes[sidx].depth, targets[tidx], &formats[fidx], seed))
{ {
// We land here in the event of test failure. // We land here in the event of test failure.
log_error( "ERROR: Image info test failed for %s : %s : %s : %s\n\n", log_error("ERROR: Image info test failed for %s : %s : %s "
GetGLTargetName( targets[ tidx ] ), ": %s\n\n",
GetGLFormatName( formats[ fidx ].internal ), GetGLTargetName(targets[tidx]),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLFormatName(formats[fidx].internal),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName(formats[fidx].datatype));
error++; error++;
// Skip the other sizes for this format. // Skip the other sizes for this format.

71
test_conformance/gl/test_images_multisample.cpp
View File

@@ -16,11 +16,11 @@
#include "common.h" #include "common.h"
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#include <algorithm> #include <algorithm>
@@ -33,12 +33,15 @@ void calc_2D_multisample_size_descriptors(sizevec_t* sizes, size_t nsizes)
GLint maxTextureSize = 4096; GLint maxTextureSize = 4096;
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].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; sizes[i].depth = 1;
} }
} }
@@ -50,21 +53,27 @@ void calc_2D_array_multisample_size_descriptors(sizevec_t* sizes, size_t nsizes)
glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers); glGetIntegerv(GL_MAX_ARRAY_TEXTURE_LAYERS, &maxTextureLayers);
glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize); glGetIntegerv(GL_MAX_TEXTURE_SIZE, &maxTextureSize);
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Generate some random sizes (within reasonable ranges) // Generate some random sizes (within reasonable ranges)
for (size_t i = 0; i < nsizes; i++) { 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].width =
sizes[i].depth = random_in_range( 2, min(maxTextureLayers, 1<<(i+4)), seed ); 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, int test_images_read_2D_multisample(cl_device_id device, cl_context context,
cl_command_queue queue, int numElements ) cl_command_queue queue, int numElements)
{ {
if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) { 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"); {
log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is "
"not supported by the tested device\n");
return 0; return 0;
} }
@@ -80,19 +89,26 @@ int test_images_read_2D_multisample( cl_device_id device, cl_context context,
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
nformats = sizeof(common_formats) / sizeof(common_formats[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); 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]); 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); 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, int test_images_read_2Darray_multisample(cl_device_id device,
cl_command_queue queue, int ) cl_context context,
cl_command_queue queue, int)
{ {
if (!is_extension_available(device, "cl_khr_gl_msaa_sharing")) { 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"); {
log_info("Test not run because 'cl_khr_gl_msaa_sharing' extension is "
"not supported by the tested device\n");
return 0; return 0;
} }
@@ -108,11 +124,14 @@ int test_images_read_2Darray_multisample( cl_device_id device, cl_context contex
size_t ntargets = sizeof(targets) / sizeof(targets[0]); size_t ntargets = sizeof(targets) / sizeof(targets[0]);
nformats = sizeof(common_formats) / sizeof(common_formats[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); 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]); 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); 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;
} }

535
test_conformance/gl/test_images_read_common.cpp
View File

@@ -16,17 +16,18 @@
#include "common.h" #include "common.h"
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
extern int supportsHalf(cl_context context, bool* supports_half); extern int supportsHalf(cl_context context, bool *supports_half);
extern int supportsMsaa(cl_context context, bool* supports_msaa); extern int supportsMsaa(cl_context context, bool *supports_msaa);
extern int supportsDepth(cl_context context, bool* supports_depth); extern int supportsDepth(cl_context context, bool *supports_depth);
// clang-format off
static const char *kernelpattern_image_read_1d = static const char *kernelpattern_image_read_1d =
"__kernel void sample_test( read_only image1d_t source, sampler_t sampler, __global %s4 *results )\n" "__kernel void sample_test( read_only image1d_t source, sampler_t sampler, __global %s4 *results )\n"
"{\n" "{\n"
@@ -167,155 +168,166 @@ static const char *kernelpattern_image_multisample_read_2darray_depth =
" results[ offset ] = read_image%s( source, (int4)( tidX, tidY, tidZ, 1 ), sample );\n" " results[ offset ] = read_image%s( source, (int4)( tidX, tidY, tidZ, 1 ), sample );\n"
" }\n" " }\n"
"}\n"; "}\n";
// clang-format on
static const char* get_appropriate_kernel_for_target(GLenum target, cl_channel_order channel_order) { static const char *
get_appropriate_kernel_for_target(GLenum target, cl_channel_order channel_order)
{
switch (get_base_gl_target(target)) { switch (get_base_gl_target(target))
case GL_TEXTURE_1D: {
return kernelpattern_image_read_1d; case GL_TEXTURE_1D: return kernelpattern_image_read_1d;
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER: return kernelpattern_image_read_1d_buffer;
return kernelpattern_image_read_1d_buffer; case GL_TEXTURE_1D_ARRAY: return kernelpattern_image_read_1darray;
case GL_TEXTURE_1D_ARRAY:
return kernelpattern_image_read_1darray;
case GL_TEXTURE_RECTANGLE_EXT: case GL_TEXTURE_RECTANGLE_EXT:
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
case GL_COLOR_ATTACHMENT0: case GL_COLOR_ATTACHMENT0:
case GL_RENDERBUFFER: case GL_RENDERBUFFER:
case GL_TEXTURE_CUBE_MAP: case GL_TEXTURE_CUBE_MAP:
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL)
return kernelpattern_image_read_2d_depth; return kernelpattern_image_read_2d_depth;
#endif #endif
return kernelpattern_image_read_2d; return kernelpattern_image_read_2d;
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL)
return kernelpattern_image_read_2darray_depth; return kernelpattern_image_read_2darray_depth;
#endif #endif
return kernelpattern_image_read_2darray; return kernelpattern_image_read_2darray;
case GL_TEXTURE_3D: case GL_TEXTURE_3D: return kernelpattern_image_read_3d;
return kernelpattern_image_read_3d;
case GL_TEXTURE_2D_MULTISAMPLE: case GL_TEXTURE_2D_MULTISAMPLE:
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL)
return kernelpattern_image_multisample_read_2d_depth; return kernelpattern_image_multisample_read_2d_depth;
#endif #endif
return kernelpattern_image_multisample_read_2d; return kernelpattern_image_multisample_read_2d;
break; break;
case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: case GL_TEXTURE_2D_MULTISAMPLE_ARRAY:
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if(channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL) if (channel_order == CL_DEPTH || channel_order == CL_DEPTH_STENCIL)
return kernelpattern_image_multisample_read_2darray_depth; return kernelpattern_image_multisample_read_2darray_depth;
#endif #endif
return kernelpattern_image_multisample_read_2darray; return kernelpattern_image_multisample_read_2darray;
break; break;
default: default:
log_error("Unsupported texture target (%s); cannot determine " log_error("Unsupported texture target (%s); cannot determine "
"appropriate kernel.", GetGLTargetName(target)); "appropriate kernel.",
GetGLTargetName(target));
return NULL; return NULL;
} }
} }
int test_cl_image_read( cl_context context, cl_command_queue queue, 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, GLenum gl_target, cl_mem image, size_t width,
cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ) size_t height, size_t depth, size_t sampleNum,
cl_image_format *outFormat, ExplicitType *outType,
void **outResultBuffer)
{ {
clProgramWrapper program; clProgramWrapper program;
clKernelWrapper kernel; clKernelWrapper kernel;
clMemWrapper streams[ 2 ]; clMemWrapper streams[2];
int error; int error;
char kernelSource[2048]; char kernelSource[2048];
char *programPtr; char *programPtr;
// Use the image created from the GL texture. // Use the image created from the GL texture.
streams[ 0 ] = image; streams[0] = image;
// Determine data type and format that CL came up with // Determine data type and format that CL came up with
error = clGetImageInfo( streams[ 0 ], CL_IMAGE_FORMAT, sizeof( cl_image_format ), outFormat, NULL ); error = clGetImageInfo(streams[0], CL_IMAGE_FORMAT, sizeof(cl_image_format),
test_error( error, "Unable to get CL image format" ); outFormat, NULL);
test_error(error, "Unable to get CL image format");
// Determine the number of samples // Determine the number of samples
cl_uint samples = 0; cl_uint samples = 0;
error = clGetImageInfo( streams[ 0 ], CL_IMAGE_NUM_SAMPLES, sizeof( samples ), &samples, NULL ); error = clGetImageInfo(streams[0], CL_IMAGE_NUM_SAMPLES, sizeof(samples),
test_error( error, "Unable to get CL_IMAGE_NUM_SAMPLES" ); &samples, NULL);
test_error(error, "Unable to get CL_IMAGE_NUM_SAMPLES");
// Create the source // Create the source
*outType = get_read_kernel_type( outFormat ); *outType = get_read_kernel_type(outFormat);
size_t channelSize = get_explicit_type_size( *outType ); size_t channelSize = get_explicit_type_size(*outType);
const char* source = get_appropriate_kernel_for_target(gl_target, outFormat->image_channel_order); const char *source = get_appropriate_kernel_for_target(
gl_target, outFormat->image_channel_order);
sprintf( kernelSource, source, get_explicit_type_name( *outType ), sprintf(kernelSource, source, get_explicit_type_name(*outType),
get_kernel_suffix( outFormat ) ); get_kernel_suffix(outFormat));
programPtr = kernelSource; programPtr = kernelSource;
if( create_single_kernel_helper( context, &program, &kernel, 1, if (create_single_kernel_helper(context, &program, &kernel, 1,
(const char **)&programPtr, "sample_test", "" ) ) (const char **)&programPtr, "sample_test",
""))
{ {
return -1; return -1;
} }
// Create a vanilla output buffer // Create a vanilla output buffer
cl_device_id device; cl_device_id device;
error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device), &device, NULL); error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device),
test_error( error, "Unable to get queue device" ); &device, NULL);
test_error(error, "Unable to get queue device");
cl_ulong maxAllocSize = 0; cl_ulong maxAllocSize = 0;
error = clGetDeviceInfo( device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof( maxAllocSize ), &maxAllocSize, NULL ); error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
test_error( error, "Unable to get 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; size_t buffer_bytes = channelSize
if (buffer_bytes > maxAllocSize) { * get_channel_order_channel_count(outFormat->image_channel_order)
log_info("Output buffer size %d is too large for device (max alloc size %d) Skipping...\n", * 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); (int)buffer_bytes, (int)maxAllocSize);
return 1; return 1;
} }
streams[ 1 ] = clCreateBuffer( context, CL_MEM_READ_WRITE, buffer_bytes, NULL, &error ); streams[1] =
test_error( error, "Unable to create output buffer" ); clCreateBuffer(context, CL_MEM_READ_WRITE, buffer_bytes, NULL, &error);
test_error(error, "Unable to create output buffer");
/* Assign streams and execute */ /* Assign streams and execute */
clSamplerWrapper sampler = clCreateSampler( context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error ); clSamplerWrapper sampler = clCreateSampler(
test_error( error, "Unable to create sampler" ); 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 ] ); error = clSetKernelArg(kernel, 0, sizeof(streams[0]), &streams[0]);
test_error( error, "Unable to set kernel arguments" ); test_error(error, "Unable to set kernel arguments");
error = clSetKernelArg( kernel, 1, sizeof( sampler ), &sampler ); error = clSetKernelArg(kernel, 1, sizeof(sampler), &sampler);
test_error( error, "Unable to set kernel arguments" ); test_error(error, "Unable to set kernel arguments");
error = clSetKernelArg( kernel, 2, sizeof( streams[ 1 ] ), &streams[ 1 ] ); error = clSetKernelArg(kernel, 2, sizeof(streams[1]), &streams[1]);
test_error( error, "Unable to set kernel arguments" ); test_error(error, "Unable to set kernel arguments");
glFinish(); glFinish();
error = (*clEnqueueAcquireGLObjects_ptr)( queue, 1, &streams[ 0 ], 0, NULL, NULL); error =
test_error( error, "Unable to acquire GL obejcts"); (*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. // The ND range we use is a function of the dimensionality of the image.
size_t global_range[3] = { width, height, depth }; size_t global_range[3] = { width, height, depth };
size_t *local_range = NULL; size_t *local_range = NULL;
int ndim = 1; int ndim = 1;
switch (get_base_gl_target(gl_target)) { switch (get_base_gl_target(gl_target))
{
case GL_TEXTURE_1D: case GL_TEXTURE_1D:
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER: ndim = 1; break;
ndim = 1;
break;
case GL_TEXTURE_RECTANGLE_EXT: case GL_TEXTURE_RECTANGLE_EXT:
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
case GL_TEXTURE_1D_ARRAY: case GL_TEXTURE_1D_ARRAY:
case GL_COLOR_ATTACHMENT0: case GL_COLOR_ATTACHMENT0:
case GL_RENDERBUFFER: case GL_RENDERBUFFER:
case GL_TEXTURE_CUBE_MAP: case GL_TEXTURE_CUBE_MAP: ndim = 2; break;
ndim = 2;
break;
case GL_TEXTURE_3D: case GL_TEXTURE_3D:
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
case GL_TEXTURE_2D_MULTISAMPLE: case GL_TEXTURE_2D_MULTISAMPLE:
case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: ndim = 3; break;
ndim = 3;
break;
#endif #endif
default: default:
log_error("Test error: Unsupported texture target.\n"); log_error("Test error: Unsupported texture target.\n");
@@ -325,28 +337,39 @@ int test_cl_image_read( cl_context context, cl_command_queue queue,
// 2D and 3D images have a special way to set the local size (legacy). // 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. // Otherwise, we let CL select by leaving local_range as NULL.
if (gl_target == GL_TEXTURE_2D) { 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 ); local_range = (size_t *)malloc(sizeof(size_t) * ndim);
get_max_common_2D_work_group_size(context, kernel, global_range,
} else if (gl_target == GL_TEXTURE_3D) { local_range);
local_range = (size_t*)malloc(sizeof(size_t) * ndim); }
get_max_common_3D_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, error = clEnqueueNDRangeKernel(queue, kernel, ndim, NULL, global_range,
local_range, 0, NULL, NULL ); local_range, 0, NULL, NULL);
test_error( error, "Unable to execute test kernel" ); test_error(error, "Unable to execute test kernel");
error = (*clEnqueueReleaseGLObjects_ptr)( queue, 1, &streams[ 0 ], error =
0, NULL, NULL ); (*clEnqueueReleaseGLObjects_ptr)(queue, 1, &streams[0], 0, NULL, NULL);
test_error(error, "clEnqueueReleaseGLObjects failed"); test_error(error, "clEnqueueReleaseGLObjects failed");
// Read results from the CL buffer // Read results from the CL buffer
*outResultBuffer = (void *)( new char[ channelSize * get_channel_order_channel_count(outFormat->image_channel_order) * width * height * depth * sampleNum] ); *outResultBuffer = (void *)(new char[channelSize
error = clEnqueueReadBuffer( queue, streams[ 1 ], CL_TRUE, 0, * get_channel_order_channel_count(
channelSize * get_channel_order_channel_count(outFormat->image_channel_order) * width * height * depth * sampleNum, *outResultBuffer, 0, NULL, NULL ); outFormat->image_channel_order)
test_error( error, "Unable to read output CL buffer!" ); * 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 // free the ranges
if (local_range) free(local_range); if (local_range) free(local_range);
@@ -354,36 +377,50 @@ int test_cl_image_read( cl_context context, cl_command_queue queue,
return 0; return 0;
} }
static int test_image_read( cl_context context, cl_command_queue queue, 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, GLenum target, GLuint globj, size_t width,
cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ) size_t height, size_t depth, size_t sampleNum,
cl_image_format *outFormat, ExplicitType *outType,
void **outResultBuffer)
{ {
int error; int error;
// Create a CL image from the supplied GL texture or renderbuffer. // Create a CL image from the supplied GL texture or renderbuffer.
cl_mem image; cl_mem image;
if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0)
image = (*clCreateFromGLRenderbuffer_ptr)( context, CL_MEM_READ_ONLY, globj, &error ); {
} else { image = (*clCreateFromGLRenderbuffer_ptr)(context, CL_MEM_READ_ONLY,
image = (*clCreateFromGLTexture_ptr)( context, CL_MEM_READ_ONLY, globj, &error);
target, 0, globj, &error ); }
else
{
image = (*clCreateFromGLTexture_ptr)(context, CL_MEM_READ_ONLY, target,
0, globj, &error);
} }
if( error != CL_SUCCESS ) { if (error != CL_SUCCESS)
if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { {
print_error( error, "Unable to create CL image from GL renderbuffer" ); if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0)
} else { {
print_error( error, "Unable to create CL image from GL texture" ); 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; GLint fmt;
glGetTexLevelParameteriv( target, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); glGetTexLevelParameteriv(target, 0, GL_TEXTURE_INTERNAL_FORMAT,
log_error( " Supplied GL texture was base format %s and internal " &fmt);
"format %s\n", GetGLBaseFormatName( fmt ), GetGLFormatName( fmt ) ); log_error(" Supplied GL texture was base format %s and internal "
"format %s\n",
GetGLBaseFormatName(fmt), GetGLFormatName(fmt));
} }
return error; return error;
} }
return test_cl_image_read( context, queue, target, image, return test_cl_image_read(context, queue, target, image, width, height,
width, height, depth, sampleNum, outFormat, outType, outResultBuffer ); depth, sampleNum, outFormat, outType,
outResultBuffer);
} }
static int test_image_format_read(cl_context context, cl_command_queue queue, static int test_image_format_read(cl_context context, cl_command_queue queue,
@@ -394,37 +431,36 @@ static int test_image_format_read(cl_context context, cl_command_queue queue,
// Determine the maximum number of supported samples // Determine the maximum number of supported samples
GLint samples = 1; GLint samples = 1;
if (target == GL_TEXTURE_2D_MULTISAMPLE || target == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) if (target == GL_TEXTURE_2D_MULTISAMPLE
|| target == GL_TEXTURE_2D_MULTISAMPLE_ARRAY)
samples = get_gl_max_samples(target, fmt->internal); samples = get_gl_max_samples(target, fmt->internal);
// If we're testing a half float format, then we need to determine the // 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. // rounding mode of this machine. Punt if we fail to do so.
if( fmt->type == kHalf ) if (fmt->type == kHalf)
{ {
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 1;
return 1;
bool supports_half = false; bool supports_half = false;
error = supportsHalf(context, &supports_half); error = supportsHalf(context, &supports_half);
if( error != 0 ) if (error != 0) return error;
return error;
if (!supports_half) return 0; if (!supports_half) return 0;
} }
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE || if (get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE
get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) || get_base_gl_target(target) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY)
{ {
bool supports_msaa; bool supports_msaa;
error = supportsMsaa(context, &supports_msaa); error = supportsMsaa(context, &supports_msaa);
if( error != 0 ) return error; if (error != 0) return error;
if (!supports_msaa) return 0; if (!supports_msaa) return 0;
} }
if (fmt->formattype == GL_DEPTH_COMPONENT || if (fmt->formattype == GL_DEPTH_COMPONENT
fmt->formattype == GL_DEPTH_STENCIL) || fmt->formattype == GL_DEPTH_STENCIL)
{ {
bool supports_depth; bool supports_depth;
error = supportsDepth(context, &supports_depth); error = supportsDepth(context, &supports_depth);
if( error != 0 ) return error; if (error != 0) return error;
if (!supports_depth) return 0; if (!supports_depth) return 0;
} }
#endif #endif
@@ -448,91 +484,107 @@ static int test_image_format_read(cl_context context, cl_command_queue queue,
glFramebufferWrapper glFramebuffer; glFramebufferWrapper glFramebuffer;
glRenderbufferWrapper glRenderbuffer; glRenderbufferWrapper glRenderbuffer;
void* buffer = NULL; void *buffer = NULL;
// Use the correct texture creation function depending on the target, and // Use the correct texture creation function depending on the target, and
// adjust width, height, depth as appropriate so subsequent size calculations // adjust width, height, depth as appropriate so subsequent size
// succeed. // calculations succeed.
switch (get_base_gl_target(target)) { switch (get_base_gl_target(target))
{
case GL_TEXTURE_1D: case GL_TEXTURE_1D:
h = 1; d = 1; h = 1;
buffer = CreateGLTexture1D( width, target, gl_fmt, d = 1;
gl_internal_fmt, gl_type, type, &texture, &error, true, data ); buffer =
CreateGLTexture1D(width, target, gl_fmt, gl_internal_fmt,
gl_type, type, &texture, &error, true, data);
break; break;
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER:
h = 1; d = 1; h = 1;
buffer = CreateGLTextureBuffer(width, target, gl_fmt, gl_internal_fmt, d = 1;
gl_type, type, &texture, &glbuf, &error, true, data); buffer = CreateGLTextureBuffer(
width, target, gl_fmt, gl_internal_fmt, gl_type, type, &texture,
&glbuf, &error, true, data);
break; break;
case GL_RENDERBUFFER: case GL_RENDERBUFFER:
case GL_COLOR_ATTACHMENT0: case GL_COLOR_ATTACHMENT0:
d = 1; d = 1;
buffer = CreateGLRenderbuffer(width, height, target, gl_fmt, buffer = CreateGLRenderbuffer(
gl_internal_fmt, gl_type, type, &glFramebuffer, &glRenderbuffer, &error, width, height, target, gl_fmt, gl_internal_fmt, gl_type, type,
data, true); &glFramebuffer, &glRenderbuffer, &error, data, true);
break; break;
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
case GL_TEXTURE_RECTANGLE_EXT: case GL_TEXTURE_RECTANGLE_EXT:
case GL_TEXTURE_CUBE_MAP: case GL_TEXTURE_CUBE_MAP:
d = 1; d = 1;
buffer = CreateGLTexture2D(width, height, target, gl_fmt, gl_internal_fmt, buffer = CreateGLTexture2D(width, height, target, gl_fmt,
gl_type, type, &texture, &error, true, data); gl_internal_fmt, gl_type, type, &texture,
&error, true, data);
break; break;
case GL_TEXTURE_1D_ARRAY: case GL_TEXTURE_1D_ARRAY:
d = 1; d = 1;
buffer = CreateGLTexture1DArray( width, height, target, gl_fmt, buffer = CreateGLTexture1DArray(width, height, target, gl_fmt,
gl_internal_fmt, gl_type, type, &texture, &error, true, data ); gl_internal_fmt, gl_type, type,
&texture, &error, true, data);
break; break;
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
buffer = CreateGLTexture2DArray( width, height, depth, target, gl_fmt, buffer = CreateGLTexture2DArray(width, height, depth, target,
gl_internal_fmt, gl_type, type, &texture, &error, true, data ); gl_fmt, gl_internal_fmt, gl_type,
type, &texture, &error, true, data);
break; break;
case GL_TEXTURE_3D: case GL_TEXTURE_3D:
buffer = CreateGLTexture3D( width, height, depth, target, gl_fmt, buffer = CreateGLTexture3D(width, height, depth, target, gl_fmt,
gl_internal_fmt, gl_type, type, &texture, &error, data, true ); gl_internal_fmt, gl_type, type, &texture,
&error, data, true);
break; break;
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
case GL_TEXTURE_2D_MULTISAMPLE: case GL_TEXTURE_2D_MULTISAMPLE:
d = 1; d = 1;
buffer = CreateGLTexture2DMultisample( width, height, samples, target, gl_fmt, buffer = CreateGLTexture2DMultisample(
gl_internal_fmt, gl_type, type, &texture, &error, true, data, true ); width, height, samples, target, gl_fmt, gl_internal_fmt,
gl_type, type, &texture, &error, true, data, true);
break; break;
case GL_TEXTURE_2D_MULTISAMPLE_ARRAY: case GL_TEXTURE_2D_MULTISAMPLE_ARRAY:
buffer = CreateGLTexture2DArrayMultisample( width, height, depth, samples, target, gl_fmt, buffer = CreateGLTexture2DArrayMultisample(
gl_internal_fmt, gl_type, type, &texture, &error, true, data, true ); width, height, depth, samples, target, gl_fmt, gl_internal_fmt,
gl_type, type, &texture, &error, true, data, true);
break; break;
#endif #endif
default: default: log_error("Unsupported texture target."); return 1;
log_error("Unsupported texture target.");
return 1;
} }
if ( error == -2 ) { if (error == -2)
log_info("OpenGL texture couldn't be created, because a texture is too big. Skipping test.\n"); {
log_info("OpenGL texture couldn't be created, because a texture is too "
"big. Skipping test.\n");
return 0; return 0;
} }
// Check to see if the texture could not be created for some other reason like // Check to see if the texture could not be created for some other reason
// GL_FRAMEBUFFER_UNSUPPORTED // like GL_FRAMEBUFFER_UNSUPPORTED
if (error == GL_FRAMEBUFFER_UNSUPPORTED) { if (error == GL_FRAMEBUFFER_UNSUPPORTED)
{
log_info("Skipping...\n"); log_info("Skipping...\n");
return 0; return 0;
} }
if ( error != 0 ) { if (error != 0)
if ((gl_fmt == GL_RGBA_INTEGER_EXT) && (!CheckGLIntegerExtensionSupport())){ {
if ((gl_fmt == GL_RGBA_INTEGER_EXT)
&& (!CheckGLIntegerExtensionSupport()))
{
log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. " log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. "
"Skipping test.\n"); "Skipping test.\n");
return 0; return 0;
} else { }
else
{
return error; return error;
} }
} }
BufferOwningPtr<char> inputBuffer(buffer); BufferOwningPtr<char> inputBuffer(buffer);
if( inputBuffer == NULL ) if (inputBuffer == NULL) return -1;
return -1;
cl_image_format clFormat; cl_image_format clFormat;
ExplicitType actualType; ExplicitType actualType;
@@ -541,106 +593,110 @@ static int test_image_format_read(cl_context context, cl_command_queue queue,
// Perform the read: // Perform the read:
GLuint globj = texture; GLuint globj = texture;
if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0)
{
globj = glRenderbuffer; globj = glRenderbuffer;
} }
error = test_image_read( context, queue, target, globj, w, h, d, samples, &clFormat, error = test_image_read(context, queue, target, globj, w, h, d, samples,
&actualType, (void **)&outBuffer ); &clFormat, &actualType, (void **)&outBuffer);
if( error != 0 ) if (error != 0) return error;
return error;
BufferOwningPtr<char> actualResults(outBuffer); BufferOwningPtr<char> actualResults(outBuffer);
if( actualResults == NULL ) if (actualResults == NULL) return -1;
return -1;
log_info( "- Read [%4d x %4d x %4d x %4d] : GL Texture : %s : %s : %s => CL Image : %s : %s \n", log_info("- Read [%4d x %4d x %4d x %4d] : GL Texture : %s : %s : %s => CL "
(int)w, (int)h, (int)d, (int)samples, GetGLFormatName( gl_fmt ), GetGLFormatName( gl_internal_fmt ), "Image : %s : %s \n",
GetGLTypeName( gl_type ), GetChannelOrderName( clFormat.image_channel_order ), (int)w, (int)h, (int)d, (int)samples, GetGLFormatName(gl_fmt),
GetChannelTypeName( clFormat.image_channel_data_type )); GetGLFormatName(gl_internal_fmt), GetGLTypeName(gl_type),
GetChannelOrderName(clFormat.image_channel_order),
GetChannelTypeName(clFormat.image_channel_data_type));
BufferOwningPtr<char> convertedInputs; BufferOwningPtr<char> convertedInputs;
// We have to convert our input buffer to the returned type, so we can validate. // We have to convert our input buffer to the returned type, so we can
// This is necessary because OpenCL might not actually pick an internal format // validate. This is necessary because OpenCL might not actually pick an
// that actually matches our input format (for example, if it picks a normalized // internal format that actually matches our input format (for example, if
// format, the results will come out as floats instead of going in as ints). // 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 ) if (gl_type == GL_UNSIGNED_INT_2_10_10_10_REV)
{ {
cl_uint *p = (cl_uint *)buffer; cl_uint *p = (cl_uint *)buffer;
float *inData = (float *)malloc( w * h * d * samples * sizeof(float) ); float *inData = (float *)malloc(w * h * d * samples * sizeof(float));
for( size_t i = 0; i < 4 * w * h * d * samples; i += 4 ) 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 + 0] = (float)((p[0] >> 20) & 0x3ff) / (float)1023;
inData[ i + 1 ] = (float)( ( p[ 0 ] >> 10 ) & 0x3ff ) / (float)1023; inData[i + 1] = (float)((p[0] >> 10) & 0x3ff) / (float)1023;
inData[ i + 2 ] = (float)( p[ 0 ] & 0x3ff ) / (float)1023; inData[i + 2] = (float)(p[0] & 0x3ff) / (float)1023;
p++; p++;
} }
convertedInputs.reset( inData ); convertedInputs.reset(inData);
if( convertedInputs == NULL ) if (convertedInputs == NULL) return -1;
return -1;
} }
else if ( gl_type == GL_DEPTH24_STENCIL8 ) else if (gl_type == GL_DEPTH24_STENCIL8)
{ {
// GL_DEPTH24_STENCIL8 is treated as CL_UNORM_INT24 + CL_DEPTH_STENCIL where // GL_DEPTH24_STENCIL8 is treated as CL_UNORM_INT24 + CL_DEPTH_STENCIL
// the stencil is ignored. // where the stencil is ignored.
cl_uint *p = (cl_uint *)buffer; cl_uint *p = (cl_uint *)buffer;
float *inData = (float *)malloc( w * h * d * samples * sizeof(float) ); float *inData = (float *)malloc(w * h * d * samples * sizeof(float));
for( size_t i = 0; i < w * h * d * samples; i++ ) for (size_t i = 0; i < w * h * d * samples; i++)
{ {
inData[ i ] = (float)((p[i] >> 8) & 0xffffff) / (float)0xfffffe; inData[i] = (float)((p[i] >> 8) & 0xffffff) / (float)0xfffffe;
} }
convertedInputs.reset( inData ); convertedInputs.reset(inData);
if( convertedInputs == NULL ) if (convertedInputs == NULL) return -1;
return -1;
} }
else if ( gl_type == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) 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 + // 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 // unused 24 + CL_DEPTH_STENCIL; we check the float value and ignore the
// second word // second word
float *p = (float *)buffer; float *p = (float *)buffer;
float *inData = (float *)malloc( w * h * d * samples * sizeof(float) ); float *inData = (float *)malloc(w * h * d * samples * sizeof(float));
for( size_t i = 0; i < w * h * d * samples; i++ ) for (size_t i = 0; i < w * h * d * samples; i++)
{ {
inData[ i ] = p[i*2]; inData[i] = p[i * 2];
} }
convertedInputs.reset( inData ); convertedInputs.reset(inData);
if( convertedInputs == NULL ) if (convertedInputs == NULL) return -1;
return -1;
} }
else else
{ {
convertedInputs.reset(convert_to_expected( inputBuffer, convertedInputs.reset(convert_to_expected(
w * h * d * samples, type, actualType, get_channel_order_channel_count(clFormat.image_channel_order) )); inputBuffer, w * h * d * samples, type, actualType,
if( convertedInputs == NULL ) get_channel_order_channel_count(clFormat.image_channel_order)));
return -1; if (convertedInputs == NULL) return -1;
} }
// Now we validate // Now we validate
if( actualType == kFloat ) if (actualType == kFloat)
{ {
if ( clFormat.image_channel_data_type == CL_UNORM_INT_101010 ) if (clFormat.image_channel_data_type == CL_UNORM_INT_101010)
{ {
return validate_float_results_rgb_101010( convertedInputs, actualResults, w, h, d, samples ); return validate_float_results_rgb_101010(
convertedInputs, actualResults, w, h, d, samples);
} }
else else
{ {
return validate_float_results( convertedInputs, actualResults, w, h, d, samples, get_channel_order_channel_count(clFormat.image_channel_order) ); return validate_float_results(
convertedInputs, actualResults, w, h, d, samples,
get_channel_order_channel_count(clFormat.image_channel_order));
} }
} }
else else
{ {
return validate_integer_results( convertedInputs, actualResults, w, h, d, samples, get_explicit_type_size( actualType ) ); return validate_integer_results(convertedInputs, actualResults, w, h, d,
samples,
get_explicit_type_size(actualType));
} }
} }
@@ -654,7 +710,8 @@ int test_images_read_common(cl_device_id device, cl_context context,
// First, ensure this device supports images. // First, ensure this device supports images.
if (checkForImageSupport(device)) { if (checkForImageSupport(device))
{
log_info("Device does not support images. Skipping test.\n"); log_info("Device does not support images. Skipping test.\n");
return 0; return 0;
} }
@@ -663,50 +720,60 @@ int test_images_read_common(cl_device_id device, cl_context context,
// Test each format on every target, every size. // Test each format on every target, every size.
for ( fidx = 0; fidx < nformats; fidx++ ) { for (fidx = 0; fidx < nformats; fidx++)
for ( tidx = 0; tidx < ntargets; tidx++ ) { {
for (tidx = 0; tidx < ntargets; tidx++)
{
// Texture buffer only takes an internal format, so the level data passed // Texture buffer only takes an internal format, so the level data
// by the test and used for verification must match the internal format // passed by the test and used for verification must match the
if ((targets[tidx] == GL_TEXTURE_BUFFER) && (GetGLFormat(formats[ fidx ].internal) != formats[fidx].formattype)) // internal format
if ((targets[tidx] == GL_TEXTURE_BUFFER)
&& (GetGLFormat(formats[fidx].internal)
!= formats[fidx].formattype))
continue; continue;
if ( formats[ fidx ].datatype == GL_UNSIGNED_INT_2_10_10_10_REV ) if (formats[fidx].datatype == GL_UNSIGNED_INT_2_10_10_10_REV)
{ {
// Check if the RGB 101010 format is supported // Check if the RGB 101010 format is supported
if ( is_rgb_101010_supported( context, targets[ tidx ] ) == 0 ) if (is_rgb_101010_supported(context, targets[tidx]) == 0)
break; // skip break; // skip
} }
if (targets[tidx] != GL_TEXTURE_BUFFER) if (targets[tidx] != GL_TEXTURE_BUFFER)
log_info( "Testing image read for GL format %s : %s : %s : %s\n", log_info("Testing image read for GL format %s : %s : %s : %s\n",
GetGLTargetName( targets[ tidx ] ), GetGLTargetName(targets[tidx]),
GetGLFormatName( formats[ fidx ].internal ), GetGLFormatName(formats[fidx].internal),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLTypeName(formats[fidx].datatype));
else else
log_info( "Testing image read for GL format %s : %s\n", log_info("Testing image read for GL format %s : %s\n",
GetGLTargetName( targets[ tidx ] ), GetGLTargetName(targets[tidx]),
GetGLFormatName( formats[ fidx ].internal )); GetGLFormatName(formats[fidx].internal));
for ( sidx = 0; sidx < nsizes; sidx++ ) { for (sidx = 0; sidx < nsizes; sidx++)
{
// Test this format + size: // Test this format + size:
int err; int err;
if ((err = test_image_format_read(context, queue, if ((err = test_image_format_read(
sizes[sidx].width, sizes[sidx].height, sizes[sidx].depth, context, queue, sizes[sidx].width, sizes[sidx].height,
targets[tidx], &formats[fidx], seed) )) sizes[sidx].depth, targets[tidx], &formats[fidx],
seed)))
{
// Negative return values are errors, positive mean the test
// was skipped
if (err < 0)
{ {
// Negative return values are errors, positive mean the test was skipped
if (err < 0) {
// We land here in the event of test failure. // We land here in the event of test failure.
log_error( "ERROR: Image read test failed for %s : %s : %s : %s\n\n", log_error("ERROR: Image read test failed for %s : %s : "
GetGLTargetName( targets[ tidx ] ), "%s : %s\n\n",
GetGLFormatName( formats[ fidx ].internal ), GetGLTargetName(targets[tidx]),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLFormatName(formats[fidx].internal),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName(formats[fidx].datatype));
error++; error++;
} }
@@ -719,12 +786,14 @@ int test_images_read_common(cl_device_id device, cl_context context,
// Note a successful format test, if we passed every size. // Note a successful format test, if we passed every size.
if( sidx == sizeof (sizes) / sizeof( sizes[0] ) ) { 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 ] ), log_info("passed: Image read test for GL format %s : %s : %s "
GetGLFormatName( formats[ fidx ].internal ), ": %s\n\n",
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLTargetName(targets[tidx]),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLFormatName(formats[fidx].internal),
GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName(formats[fidx].datatype));
} }
} }
} }

515
test_conformance/gl/test_images_write_common.cpp
View File

@@ -17,16 +17,17 @@
#include "common.h" #include "common.h"
#include <limits.h> #include <limits.h>
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#pragma mark - #pragma mark -
#pragma mark Write test kernels #pragma mark Write test kernels
// clang-format off
static const char *kernelpattern_image_write_1D = static const char *kernelpattern_image_write_1D =
"__kernel void sample_test( __global %s4 *source, write_only image1d_t dest )\n" "__kernel void sample_test( __global %s4 *source, write_only image1d_t dest )\n"
"{\n" "{\n"
@@ -174,14 +175,17 @@ static const char * kernelpattern_image_write_2D_array_depth =
#endif #endif
// clang-format on
#pragma mark - #pragma mark -
#pragma mark Utility functions #pragma mark Utility functions
static const char* get_appropriate_write_kernel(GLenum target, static const char *get_appropriate_write_kernel(GLenum target,
ExplicitType type, cl_channel_order channel_order) ExplicitType type,
cl_channel_order channel_order)
{ {
switch (get_base_gl_target(target)) { switch (get_base_gl_target(target))
{
case GL_TEXTURE_1D: case GL_TEXTURE_1D:
if (type == kHalf) if (type == kHalf)
@@ -236,8 +240,9 @@ static const char* get_appropriate_write_kernel(GLenum target,
default: default:
log_error("Unsupported GL tex target (%s) passed to write test: " log_error("Unsupported GL tex target (%s) passed to write test: "
"%s (%s):%d", GetGLTargetName(target), __FUNCTION__, "%s (%s):%d",
__FILE__, __LINE__); GetGLTargetName(target), __FUNCTION__, __FILE__,
__LINE__);
return NULL; return NULL;
} }
} }
@@ -245,7 +250,8 @@ static const char* get_appropriate_write_kernel(GLenum target,
void set_dimensions_by_target(GLenum target, size_t *dims, size_t sizes[3], 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)) { switch (get_base_gl_target(target))
{
case GL_TEXTURE_1D: case GL_TEXTURE_1D:
sizes[0] = width; sizes[0] = width;
*dims = 1; *dims = 1;
@@ -289,21 +295,23 @@ void set_dimensions_by_target(GLenum target, size_t *dims, size_t sizes[3],
default: default:
log_error("Unsupported GL tex target (%s) passed to write test: " log_error("Unsupported GL tex target (%s) passed to write test: "
"%s (%s):%d", GetGLTargetName(target), __FUNCTION__, "%s (%s):%d",
__FILE__, __LINE__); GetGLTargetName(target), __FUNCTION__, __FILE__,
__LINE__);
} }
} }
int test_cl_image_write( cl_context context, cl_command_queue queue, 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, GLenum target, cl_mem clImage, size_t width,
cl_image_format *outFormat, ExplicitType *outType, void **outSourceBuffer, size_t height, size_t depth, cl_image_format *outFormat,
MTdata d, bool supports_half ) ExplicitType *outType, void **outSourceBuffer, MTdata d,
bool supports_half)
{ {
size_t global_dims, global_sizes[3]; size_t global_dims, global_sizes[3];
clProgramWrapper program; clProgramWrapper program;
clKernelWrapper kernel; clKernelWrapper kernel;
clMemWrapper inStream; clMemWrapper inStream;
char* programPtr; char *programPtr;
int error; int error;
char kernelSource[2048]; char kernelSource[2048];
@@ -316,25 +324,27 @@ int test_cl_image_write( cl_context context, cl_command_queue queue,
// Create the kernel source. The target and the data type will influence // Create the kernel source. The target and the data type will influence
// which particular kernel we choose. // which particular kernel we choose.
*outType = get_write_kernel_type( outFormat ); *outType = get_write_kernel_type(outFormat);
size_t channelSize = get_explicit_type_size(*outType); size_t channelSize = get_explicit_type_size(*outType);
const char* appropriateKernel = get_appropriate_write_kernel(target, const char *appropriateKernel = get_appropriate_write_kernel(
*outType, outFormat->image_channel_order); target, *outType, outFormat->image_channel_order);
if (*outType == kHalf && !supports_half) { if (*outType == kHalf && !supports_half)
{
log_info("cl_khr_fp16 isn't supported. Skip this test.\n"); log_info("cl_khr_fp16 isn't supported. Skip this test.\n");
return 0; return 0;
} }
const char* suffix = get_kernel_suffix( outFormat ); const char *suffix = get_kernel_suffix(outFormat);
const char* convert = get_write_conversion( outFormat, *outType ); const char *convert = get_write_conversion(outFormat, *outType);
sprintf(kernelSource, appropriateKernel, get_explicit_type_name( *outType ), sprintf(kernelSource, appropriateKernel, get_explicit_type_name(*outType),
get_explicit_type_name( *outType ), suffix, convert); get_explicit_type_name(*outType), suffix, convert);
programPtr = kernelSource; programPtr = kernelSource;
if( create_single_kernel_helper_with_build_options( context, &program, &kernel, 1, if (create_single_kernel_helper_with_build_options(
(const char **)&programPtr, "sample_test", "" ) ) context, &program, &kernel, 1, (const char **)&programPtr,
"sample_test", ""))
{ {
return -1; return -1;
} }
@@ -346,149 +356,172 @@ int test_cl_image_write( cl_context context, cl_command_queue queue,
cl_ulong alloc_size = 0; cl_ulong alloc_size = 0;
cl_device_id device = NULL; cl_device_id device = NULL;
error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device), &device, NULL); error = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(device),
test_error( error, "Unable to query command queue for 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); error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
test_error( error, "Unable to device for 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) { if (bytes > alloc_size)
log_info(" Skipping: Buffer size (%lu) is greater than CL_DEVICE_MAX_MEM_ALLOC_SIZE (%lu)\n", 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; *outSourceBuffer = NULL;
return 0; return 0;
} }
*outSourceBuffer = CreateRandomData(*outType, width * height * depth * 4, d); *outSourceBuffer =
CreateRandomData(*outType, width * height * depth * 4, d);
inStream = clCreateBuffer( context, CL_MEM_COPY_HOST_PTR, inStream = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR,
channelSize * 4 * width * height * depth, *outSourceBuffer, &error ); channelSize * 4 * width * height * depth,
test_error( error, "Unable to create output buffer" ); *outSourceBuffer, &error);
test_error(error, "Unable to create output buffer");
clSamplerWrapper sampler = clCreateSampler(context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error); clSamplerWrapper sampler = clCreateSampler(
test_error( error, "Unable to create sampler" ); context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &error);
test_error(error, "Unable to create sampler");
error = clSetKernelArg( kernel, 0, sizeof( inStream ), &inStream ); error = clSetKernelArg(kernel, 0, sizeof(inStream), &inStream);
test_error( error, "Unable to set kernel arguments" ); test_error(error, "Unable to set kernel arguments");
error = clSetKernelArg( kernel, 1, sizeof( clImage ), &clImage ); error = clSetKernelArg(kernel, 1, sizeof(clImage), &clImage);
test_error( error, "Unable to set kernel arguments" ); test_error(error, "Unable to set kernel arguments");
// Flush and Acquire. // Flush and Acquire.
glFinish(); glFinish();
error = (*clEnqueueAcquireGLObjects_ptr)( queue, 1, &clImage, 0, NULL, NULL); error = (*clEnqueueAcquireGLObjects_ptr)(queue, 1, &clImage, 0, NULL, NULL);
test_error( error, "Unable to acquire GL obejcts"); test_error(error, "Unable to acquire GL obejcts");
// Execute ( letting OpenCL choose the local size ) // Execute ( letting OpenCL choose the local size )
// Setup the global dimensions and sizes based on the target type. // Setup the global dimensions and sizes based on the target type.
set_dimensions_by_target(target, &global_dims, global_sizes, set_dimensions_by_target(target, &global_dims, global_sizes, width, height,
width, height, depth); depth);
error = clEnqueueNDRangeKernel( queue, kernel, global_dims, NULL, error = clEnqueueNDRangeKernel(queue, kernel, global_dims, NULL,
global_sizes, NULL, 0, NULL, NULL ); global_sizes, NULL, 0, NULL, NULL);
test_error( error, "Unable to execute test kernel" ); test_error(error, "Unable to execute test kernel");
clEventWrapper event; clEventWrapper event;
error = (*clEnqueueReleaseGLObjects_ptr)( queue, 1, &clImage, 0, NULL, &event ); error =
(*clEnqueueReleaseGLObjects_ptr)(queue, 1, &clImage, 0, NULL, &event);
test_error(error, "clEnqueueReleaseGLObjects failed"); test_error(error, "clEnqueueReleaseGLObjects failed");
error = clWaitForEvents( 1, &event ); error = clWaitForEvents(1, &event);
test_error(error, "clWaitForEvents failed"); test_error(error, "clWaitForEvents failed");
return 0; return 0;
} }
static int test_image_write( cl_context context, cl_command_queue queue, 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, GLenum glTarget, GLuint glTexture, size_t width,
cl_image_format *outFormat, ExplicitType *outType, void **outSourceBuffer, size_t height, size_t depth,
MTdata d, bool supports_half ) cl_image_format *outFormat, ExplicitType *outType,
void **outSourceBuffer, MTdata d,
bool supports_half)
{ {
int error; int error;
// Create a CL image from the supplied GL texture // Create a CL image from the supplied GL texture
clMemWrapper image = (*clCreateFromGLTexture_ptr)( context, CL_MEM_WRITE_ONLY, clMemWrapper image = (*clCreateFromGLTexture_ptr)(
glTarget, 0, glTexture, &error ); context, CL_MEM_WRITE_ONLY, glTarget, 0, glTexture, &error);
if ( error != CL_SUCCESS ) { if (error != CL_SUCCESS)
print_error( error, "Unable to create CL image from GL texture" ); {
print_error(error, "Unable to create CL image from GL texture");
GLint fmt; GLint fmt;
glGetTexLevelParameteriv( glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt ); glGetTexLevelParameteriv(glTarget, 0, GL_TEXTURE_INTERNAL_FORMAT, &fmt);
log_error( " Supplied GL texture was base format %s and internal " log_error(" Supplied GL texture was base format %s and internal "
"format %s\n", GetGLBaseFormatName( fmt ), GetGLFormatName( fmt ) ); "format %s\n",
GetGLBaseFormatName(fmt), GetGLFormatName(fmt));
return error; return error;
} }
return test_cl_image_write( context, queue, glTarget, image, return test_cl_image_write(context, queue, glTarget, image, width, height,
width, height, depth, outFormat, outType, outSourceBuffer, d, supports_half ); depth, outFormat, outType, outSourceBuffer, d,
supports_half);
} }
int supportsHalf(cl_context context, bool* supports_half) int supportsHalf(cl_context context, bool *supports_half)
{ {
int error; int error;
cl_uint numDev; cl_uint numDev;
error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDev, NULL); error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint),
&numDev, NULL);
test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed");
cl_device_id* devices = new cl_device_id[numDev]; cl_device_id *devices = new cl_device_id[numDev];
error = clGetContextInfo(context, CL_CONTEXT_DEVICES, numDev * sizeof(cl_device_id), devices, NULL); error = clGetContextInfo(context, CL_CONTEXT_DEVICES,
numDev * sizeof(cl_device_id), devices, NULL);
test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed");
*supports_half = is_extension_available(devices[0], "cl_khr_fp16"); *supports_half = is_extension_available(devices[0], "cl_khr_fp16");
delete [] devices; delete[] devices;
return error; return error;
} }
int supportsMsaa(cl_context context, bool* supports_msaa) int supportsMsaa(cl_context context, bool *supports_msaa)
{ {
int error; int error;
cl_uint numDev; cl_uint numDev;
error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDev, NULL); error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint),
&numDev, NULL);
test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed");
cl_device_id* devices = new cl_device_id[numDev]; cl_device_id *devices = new cl_device_id[numDev];
error = clGetContextInfo(context, CL_CONTEXT_DEVICES, numDev * sizeof(cl_device_id), devices, NULL); error = clGetContextInfo(context, CL_CONTEXT_DEVICES,
numDev * sizeof(cl_device_id), devices, NULL);
test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed");
*supports_msaa = is_extension_available(devices[0], "cl_khr_gl_msaa_sharing"); *supports_msaa =
delete [] devices; is_extension_available(devices[0], "cl_khr_gl_msaa_sharing");
delete[] devices;
return error; return error;
} }
int supportsDepth(cl_context context, bool* supports_depth) int supportsDepth(cl_context context, bool *supports_depth)
{ {
int error; int error;
cl_uint numDev; cl_uint numDev;
error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDev, NULL); error = clGetContextInfo(context, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint),
&numDev, NULL);
test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed"); test_error(error, "clGetContextInfo for CL_CONTEXT_NUM_DEVICES failed");
cl_device_id* devices = new cl_device_id[numDev]; cl_device_id *devices = new cl_device_id[numDev];
error = clGetContextInfo(context, CL_CONTEXT_DEVICES, numDev * sizeof(cl_device_id), devices, NULL); error = clGetContextInfo(context, CL_CONTEXT_DEVICES,
numDev * sizeof(cl_device_id), devices, NULL);
test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed"); test_error(error, "clGetContextInfo for CL_CONTEXT_DEVICES failed");
*supports_depth = is_extension_available(devices[0], "cl_khr_gl_depth_images"); *supports_depth =
delete [] devices; is_extension_available(devices[0], "cl_khr_gl_depth_images");
delete[] devices;
return error; return error;
} }
static int test_image_format_write( cl_context context, cl_command_queue queue, 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, size_t width, size_t height, size_t depth,
GLenum internalFormat, GLenum glType, ExplicitType type, MTdata d ) GLenum target, GLenum format,
GLenum internalFormat, GLenum glType,
ExplicitType type, MTdata d)
{ {
int error; int error;
// If we're testing a half float format, then we need to determine the // 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. // rounding mode of this machine. Punt if we fail to do so.
if( type == kHalf ) if (type == kHalf)
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 1;
return 1;
// Create an appropriate GL texture or renderbuffer, given the target. // Create an appropriate GL texture or renderbuffer, given the target.
@@ -496,57 +529,70 @@ static int test_image_format_write( cl_context context, cl_command_queue queue,
glBufferWrapper glBuf; glBufferWrapper glBuf;
glFramebufferWrapper glFramebuffer; glFramebufferWrapper glFramebuffer;
glRenderbufferWrapper glRenderbuffer; glRenderbufferWrapper glRenderbuffer;
switch (get_base_gl_target(target)) { switch (get_base_gl_target(target))
{
case GL_TEXTURE_1D: case GL_TEXTURE_1D:
CreateGLTexture1D( width, target, format, internalFormat, glType, CreateGLTexture1D(width, target, format, internalFormat, glType,
type, &glTexture, &error, false, d ); type, &glTexture, &error, false, d);
break; break;
case GL_TEXTURE_BUFFER: case GL_TEXTURE_BUFFER:
CreateGLTextureBuffer( width, target, format, internalFormat, glType, CreateGLTextureBuffer(width, target, format, internalFormat, glType,
type, &glTexture, &glBuf, &error, false, d ); type, &glTexture, &glBuf, &error, false, d);
break; break;
case GL_TEXTURE_1D_ARRAY: case GL_TEXTURE_1D_ARRAY:
CreateGLTexture1DArray( width, height, target, format, internalFormat, CreateGLTexture1DArray(width, height, target, format,
glType, type, &glTexture, &error, false, d ); internalFormat, glType, type, &glTexture,
&error, false, d);
break; break;
case GL_TEXTURE_RECTANGLE_EXT: case GL_TEXTURE_RECTANGLE_EXT:
case GL_TEXTURE_2D: case GL_TEXTURE_2D:
case GL_TEXTURE_CUBE_MAP: case GL_TEXTURE_CUBE_MAP:
CreateGLTexture2D( width, height, target, format, internalFormat, glType, CreateGLTexture2D(width, height, target, format, internalFormat,
type, &glTexture, &error, false, d ); glType, type, &glTexture, &error, false, d);
break; break;
case GL_COLOR_ATTACHMENT0: case GL_COLOR_ATTACHMENT0:
case GL_RENDERBUFFER: case GL_RENDERBUFFER:
CreateGLRenderbuffer(width, height, target, format, internalFormat, CreateGLRenderbuffer(width, height, target, format, internalFormat,
glType, type, &glFramebuffer, &glRenderbuffer, &error, d, false); glType, type, &glFramebuffer, &glRenderbuffer,
&error, d, false);
case GL_TEXTURE_2D_ARRAY: case GL_TEXTURE_2D_ARRAY:
CreateGLTexture2DArray( width, height, depth, target, format, CreateGLTexture2DArray(width, height, depth, target, format,
internalFormat, glType, type, &glTexture, &error, false, d ); internalFormat, glType, type, &glTexture,
&error, false, d);
break; break;
case GL_TEXTURE_3D: case GL_TEXTURE_3D:
CreateGLTexture3D( width, height, depth, target, format, CreateGLTexture3D(width, height, depth, target, format,
internalFormat, glType, type, &glTexture, &error, d, false ); internalFormat, glType, type, &glTexture, &error,
d, false);
break; break;
default: default:
log_error("Unsupported GL tex target (%s) passed to write test: " log_error("Unsupported GL tex target (%s) passed to write test: "
"%s (%s):%d", GetGLTargetName(target), __FUNCTION__, "%s (%s):%d",
__FILE__, __LINE__); GetGLTargetName(target), __FUNCTION__, __FILE__,
__LINE__);
} }
// If there was a problem during creation, make sure it isn't a known // If there was a problem during creation, make sure it isn't a known
// cause, and then complain. // cause, and then complain.
if ( error == -2 ) { if (error == -2)
log_info("OpenGL texture couldn't be created, because a texture is too big. Skipping test.\n"); {
log_info("OpenGL texture couldn't be created, because a texture is too "
"big. Skipping test.\n");
return 0; return 0;
} }
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. " log_info("OpenGL version does not support GL_RGBA_INTEGER_EXT. "
"Skipping test.\n"); "Skipping test.\n");
return 0; return 0;
} else { }
else
{
return error; return error;
} }
} }
@@ -558,94 +604,101 @@ static int test_image_format_write( cl_context context, cl_command_queue queue,
void *outSourceBuffer = NULL; void *outSourceBuffer = NULL;
GLenum globj = glTexture; GLenum globj = glTexture;
if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0) { if (target == GL_RENDERBUFFER || target == GL_COLOR_ATTACHMENT0)
{
globj = glRenderbuffer; globj = glRenderbuffer;
} }
bool supports_half = false; bool supports_half = false;
error = supportsHalf(context, &supports_half); error = supportsHalf(context, &supports_half);
if( error != 0 ) if (error != 0) return error;
return error;
error = test_image_write( context, queue, target, globj, width, height, error = test_image_write(context, queue, target, globj, width, height,
depth, &clFormat, &sourceType, (void **)&outSourceBuffer, d, supports_half ); depth, &clFormat, &sourceType,
(void **)&outSourceBuffer, d, supports_half);
if( error != 0 || ((sourceType == kHalf ) && !supports_half)) { if (error != 0 || ((sourceType == kHalf) && !supports_half))
if (outSourceBuffer) {
free(outSourceBuffer); if (outSourceBuffer) free(outSourceBuffer);
return error; return error;
} }
if (!outSourceBuffer) if (!outSourceBuffer) return 0;
return 0;
// If actual source type was half, convert to float for validation. // If actual source type was half, convert to float for validation.
if ( sourceType == kHalf ) if (sourceType == kHalf)
validationType = kFloat; validationType = kFloat;
else else
validationType = sourceType; validationType = sourceType;
BufferOwningPtr<char> validationSource; BufferOwningPtr<char> validationSource;
if ( clFormat.image_channel_data_type == CL_UNORM_INT_101010 ) if (clFormat.image_channel_data_type == CL_UNORM_INT_101010)
{ {
validationSource.reset( outSourceBuffer ); validationSource.reset(outSourceBuffer);
} }
else else
{ {
validationSource.reset( convert_to_expected( outSourceBuffer, validationSource.reset(convert_to_expected(
width * height * depth, sourceType, validationType, get_channel_order_channel_count(clFormat.image_channel_order) ) ); outSourceBuffer, width * height * depth, sourceType, validationType,
get_channel_order_channel_count(clFormat.image_channel_order)));
free(outSourceBuffer); free(outSourceBuffer);
} }
log_info( "- Write for %s [%4ld x %4ld x %4ld] : GL Texture : %s : %s : %s =>" log_info(
"- Write for %s [%4ld x %4ld x %4ld] : GL Texture : %s : %s : %s =>"
" CL Image : %s : %s \n", " CL Image : %s : %s \n",
GetGLTargetName(target), GetGLTargetName(target), width, height, depth, GetGLFormatName(format),
width, height, depth, GetGLFormatName(internalFormat), GetGLTypeName(glType),
GetGLFormatName( format ), GetChannelOrderName(clFormat.image_channel_order),
GetGLFormatName( internalFormat ), GetChannelTypeName(clFormat.image_channel_data_type));
GetGLTypeName( glType),
GetChannelOrderName( clFormat.image_channel_order ),
GetChannelTypeName( clFormat.image_channel_data_type ));
// Read the results from the GL texture. // Read the results from the GL texture.
ExplicitType readType = type; ExplicitType readType = type;
BufferOwningPtr<char> glResults( ReadGLTexture( BufferOwningPtr<char> glResults(
target, glTexture, glBuf, width, format, ReadGLTexture(target, glTexture, glBuf, width, format, internalFormat,
internalFormat, glType, readType, /* unused */ 1, 1 ) ); glType, readType, /* unused */ 1, 1));
if( glResults == NULL ) if (glResults == NULL) return -1;
return -1;
// We have to convert our input buffer to the returned type, so we can validate. // We have to convert our input buffer to the returned type, so we can
// validate.
BufferOwningPtr<char> convertedGLResults; BufferOwningPtr<char> convertedGLResults;
if ( clFormat.image_channel_data_type != CL_UNORM_INT_101010 ) if (clFormat.image_channel_data_type != CL_UNORM_INT_101010)
{ {
convertedGLResults.reset( convert_to_expected( convertedGLResults.reset(convert_to_expected(
glResults, width * height * depth, readType, validationType, get_channel_order_channel_count(clFormat.image_channel_order), glType )); glResults, width * height * depth, readType, validationType,
get_channel_order_channel_count(clFormat.image_channel_order),
glType));
} }
// Validate. // Validate.
int valid = 0; int valid = 0;
if (convertedGLResults) { if (convertedGLResults)
if( sourceType == kFloat || sourceType == kHalf )
{ {
if ( clFormat.image_channel_data_type == CL_UNORM_INT_101010 ) if (sourceType == kFloat || sourceType == kHalf)
{ {
valid = validate_float_results_rgb_101010( validationSource, glResults, width, height, depth, 1 ); if (clFormat.image_channel_data_type == CL_UNORM_INT_101010)
{
valid = validate_float_results_rgb_101010(
validationSource, glResults, width, height, depth, 1);
} }
else else
{ {
valid = validate_float_results( validationSource, convertedGLResults, valid =
width, height, depth, 1, get_channel_order_channel_count(clFormat.image_channel_order) ); validate_float_results(validationSource, convertedGLResults,
width, height, depth, 1,
get_channel_order_channel_count(
clFormat.image_channel_order));
} }
} }
else else
{ {
valid = validate_integer_results( validationSource, convertedGLResults, valid = validate_integer_results(
width, height, depth, 1, get_explicit_type_size( readType ) ); validationSource, convertedGLResults, width, height, depth, 1,
get_explicit_type_size(readType));
} }
} }
@@ -670,7 +723,8 @@ int test_images_write_common(cl_device_id device, cl_context context,
// First, ensure this device supports images. // First, ensure this device supports images.
if (checkForImageSupport(device)) { if (checkForImageSupport(device))
{
log_info("Device does not support images. Skipping test.\n"); log_info("Device does not support images. Skipping test.\n");
return 0; return 0;
} }
@@ -680,8 +734,10 @@ int test_images_write_common(cl_device_id device, cl_context context,
err = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, err = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
sizeof(max_individual_allocation_size), sizeof(max_individual_allocation_size),
&max_individual_allocation_size, NULL); &max_individual_allocation_size, NULL);
if (err) { if (err)
log_error("ERROR: clGetDeviceInfo failed for CL_DEVICE_MAX_MEM_ALLOC_SIZE.\n"); {
log_error("ERROR: clGetDeviceInfo failed for "
"CL_DEVICE_MAX_MEM_ALLOC_SIZE.\n");
error++; error++;
return error; return error;
} }
@@ -689,109 +745,132 @@ int test_images_write_common(cl_device_id device, cl_context context,
size_t total_allocation_size; size_t total_allocation_size;
size_t fidx, tidx, sidx; size_t fidx, tidx, sidx;
for ( fidx = 0; fidx < nformats; fidx++ ) { for (fidx = 0; fidx < nformats; fidx++)
for ( tidx = 0; tidx < ntargets; tidx++ ) { {
for (tidx = 0; tidx < ntargets; tidx++)
{
// Texture buffer only takes an internal format, so the level data passed // Texture buffer only takes an internal format, so the level data
// by the test and used for verification must match the internal format // passed by the test and used for verification must match the
if ((targets[tidx] == GL_TEXTURE_BUFFER) && (GetGLFormat(formats[ fidx ].internal) != formats[fidx].formattype)) // internal format
if ((targets[tidx] == GL_TEXTURE_BUFFER)
&& (GetGLFormat(formats[fidx].internal)
!= formats[fidx].formattype))
continue; continue;
if ( formats[ fidx ].datatype == GL_UNSIGNED_INT_2_10_10_10_REV ) if (formats[fidx].datatype == GL_UNSIGNED_INT_2_10_10_10_REV)
{ {
// Check if the RGB 101010 format is supported // Check if the RGB 101010 format is supported
if ( is_rgb_101010_supported( context, targets[ tidx ] ) == 0 ) if (is_rgb_101010_supported(context, targets[tidx]) == 0)
continue; // skip continue; // skip
} }
if (formats[ fidx ].datatype == GL_UNSIGNED_INT_24_8) if (formats[fidx].datatype == GL_UNSIGNED_INT_24_8)
{ {
//check if a implementation supports writing to the depth stencil formats // check if a implementation supports writing to the depth
cl_image_format imageFormat = { CL_DEPTH_STENCIL, CL_UNORM_INT24 }; // stencil formats
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)) 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; continue;
} }
if (formats[ fidx ].datatype == GL_FLOAT_32_UNSIGNED_INT_24_8_REV) if (formats[fidx].datatype == GL_FLOAT_32_UNSIGNED_INT_24_8_REV)
{ {
//check if a implementation supports writing to the depth stencil formats // check if a implementation supports writing to the depth
cl_image_format imageFormat = { CL_DEPTH_STENCIL, CL_FLOAT}; // stencil formats
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)) 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; continue;
} }
if (targets[tidx] != GL_TEXTURE_BUFFER) if (targets[tidx] != GL_TEXTURE_BUFFER)
log_info( "Testing image write for GL format %s : %s : %s : %s\n", log_info(
GetGLTargetName( targets[ tidx ] ), "Testing image write for GL format %s : %s : %s : %s\n",
GetGLFormatName( formats[ fidx ].internal ), GetGLTargetName(targets[tidx]),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLFormatName(formats[fidx].internal),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName(formats[fidx].datatype));
else else
log_info( "Testing image write for GL format %s : %s\n", log_info("Testing image write for GL format %s : %s\n",
GetGLTargetName( targets[ tidx ] ), GetGLTargetName(targets[tidx]),
GetGLFormatName( formats[ fidx ].internal )); GetGLFormatName(formats[fidx].internal));
for (sidx = 0; sidx < nsizes; sidx++) { for (sidx = 0; sidx < nsizes; sidx++)
{
// All tested formats are 4-channel formats // All tested formats are 4-channel formats
total_allocation_size = total_allocation_size = sizes[sidx].width * sizes[sidx].height
sizes[sidx].width * sizes[sidx].height * sizes[sidx].depth * * sizes[sidx].depth * 4
4 * get_explicit_type_size( formats[ fidx ].type ); * get_explicit_type_size(formats[fidx].type);
if (total_allocation_size > max_individual_allocation_size) { if (total_allocation_size > max_individual_allocation_size)
log_info( "The requested allocation size (%gMB) is larger than the " {
log_info("The requested allocation size (%gMB) is larger "
"than the "
"maximum individual allocation size (%gMB)\n", "maximum individual allocation size (%gMB)\n",
total_allocation_size/(1024.0*1024.0), total_allocation_size / (1024.0 * 1024.0),
max_individual_allocation_size/(1024.0*1024.0)); max_individual_allocation_size
log_info( "Skipping write test for %s : %s : %s : %s " / (1024.0 * 1024.0));
log_info("Skipping write test for %s : %s : %s : %s "
" and size (%ld, %ld, %ld)\n", " and size (%ld, %ld, %ld)\n",
GetGLTargetName( targets[ tidx ] ), GetGLTargetName(targets[tidx]),
GetGLFormatName( formats[ fidx ].internal ), GetGLFormatName(formats[fidx].internal),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName( formats[ fidx ].datatype ), GetGLTypeName(formats[fidx].datatype),
sizes[sidx].width, sizes[sidx].width, sizes[sidx].height,
sizes[sidx].height,
sizes[sidx].depth); sizes[sidx].depth);
continue; continue;
} }
#ifdef GL_VERSION_3_2 #ifdef GL_VERSION_3_2
if (get_base_gl_target(targets[ tidx ]) == GL_TEXTURE_2D_MULTISAMPLE || if (get_base_gl_target(targets[tidx])
get_base_gl_target(targets[ tidx ]) == GL_TEXTURE_2D_MULTISAMPLE_ARRAY) == GL_TEXTURE_2D_MULTISAMPLE
|| get_base_gl_target(targets[tidx])
== GL_TEXTURE_2D_MULTISAMPLE_ARRAY)
{ {
bool supports_msaa; bool supports_msaa;
int errorInGetInfo = supportsMsaa(context, &supports_msaa); int errorInGetInfo = supportsMsaa(context, &supports_msaa);
if (errorInGetInfo != 0) return errorInGetInfo; if (errorInGetInfo != 0) return errorInGetInfo;
if (!supports_msaa) return 0; if (!supports_msaa) return 0;
} }
if (formats[ fidx ].formattype == GL_DEPTH_COMPONENT || if (formats[fidx].formattype == GL_DEPTH_COMPONENT
formats[ fidx ].formattype == GL_DEPTH_STENCIL) || formats[fidx].formattype == GL_DEPTH_STENCIL)
{ {
bool supports_depth; bool supports_depth;
int errorInGetInfo = supportsDepth(context, &supports_depth); int errorInGetInfo =
supportsDepth(context, &supports_depth);
if (errorInGetInfo != 0) return errorInGetInfo; if (errorInGetInfo != 0) return errorInGetInfo;
if (!supports_depth) return 0; if (!supports_depth) return 0;
} }
#endif #endif
if( test_image_format_write( context, queue, if (test_image_format_write(
sizes[sidx].width, context, queue, sizes[sidx].width, sizes[sidx].height,
sizes[sidx].height, sizes[sidx].depth, targets[tidx],
sizes[sidx].depth, formats[fidx].formattype, formats[fidx].internal,
targets[ tidx ], formats[fidx].datatype, formats[fidx].type, seed))
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 " log_error(
"ERROR: Image write test failed for %s : %s : %s : %s "
" and size (%ld, %ld, %ld)\n\n", " and size (%ld, %ld, %ld)\n\n",
GetGLTargetName( targets[ tidx ] ), GetGLTargetName(targets[tidx]),
GetGLFormatName( formats[ fidx ].internal ), GetGLFormatName(formats[fidx].internal),
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName( formats[ fidx ].datatype ), GetGLTypeName(formats[fidx].datatype),
sizes[sidx].width, sizes[sidx].width, sizes[sidx].height,
sizes[sidx].height,
sizes[sidx].depth); sizes[sidx].depth);
error++; error++;
@@ -801,12 +880,14 @@ int test_images_write_common(cl_device_id device, cl_context context,
// If we passed all sizes (check versus size loop count): // If we passed all sizes (check versus size loop count):
if (sidx == nsizes) { if (sidx == nsizes)
log_info( "passed: Image write for GL format %s : %s : %s : %s\n\n", {
GetGLTargetName( targets[ tidx ] ), log_info(
GetGLFormatName( formats[ fidx ].internal ), "passed: Image write for GL format %s : %s : %s : %s\n\n",
GetGLBaseFormatName( formats[ fidx ].formattype ), GetGLTargetName(targets[tidx]),
GetGLTypeName( formats[ fidx ].datatype ) ); GetGLFormatName(formats[fidx].internal),
GetGLBaseFormatName(formats[fidx].formattype),
GetGLTypeName(formats[fidx].datatype));
} }
} }
} }

391
test_conformance/gl/test_renderbuffer.cpp
View File

@@ -15,36 +15,33 @@
// //
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
#if defined (__linux__) #if defined(__linux__)
GLboolean GLboolean gluCheckExtension(const GLubyte *extension, const GLubyte *extensions)
gluCheckExtension(const GLubyte *extension, const GLubyte *extensions)
{ {
const GLubyte *start; const GLubyte *start;
GLubyte *where, *terminator; GLubyte *where, *terminator;
/* Extension names should not have spaces. */ /* Extension names should not have spaces. */
where = (GLubyte *) strchr((const char*)extension, ' '); where = (GLubyte *)strchr((const char *)extension, ' ');
if (where || *extension == '\0') if (where || *extension == '\0') return 0;
return 0;
/* It takes a bit of care to be fool-proof about parsing the /* It takes a bit of care to be fool-proof about parsing the
OpenGL extensions string. Don't be fooled by sub-strings, OpenGL extensions string. Don't be fooled by sub-strings,
etc. */ etc. */
start = extensions; start = extensions;
for (;;) { for (;;)
where = (GLubyte *) strstr((const char *) start, (const char*) extension); {
if (!where) where = (GLubyte *)strstr((const char *)start, (const char *)extension);
break; if (!where) break;
terminator = where + strlen((const char*) extension); terminator = where + strlen((const char *)extension);
if (where == start || *(where - 1) == ' ') if (where == start || *(where - 1) == ' ')
if (*terminator == ' ' || *terminator == '\0') if (*terminator == ' ' || *terminator == '\0') return 1;
return 1;
start = terminator; start = terminator;
} }
return 0; return 0;
@@ -53,55 +50,67 @@ gluCheckExtension(const GLubyte *extension, const GLubyte *extensions)
// This is defined in the write common code: // This is defined in the write common code:
extern int test_cl_image_write( cl_context context, cl_command_queue queue, 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, GLenum target, cl_mem clImage, size_t width,
cl_image_format *outFormat, ExplicitType *outType, void **outSourceBuffer, size_t height, size_t depth,
MTdata d, bool supports_half ); 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, 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, GLenum gl_target, cl_mem image, size_t width,
cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ); 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, static int test_attach_renderbuffer_read_image(
size_t imageWidth, size_t imageHeight, cl_image_format *outFormat, ExplicitType *outType, void **outResultBuffer ) 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; int error;
// Create a CL image from the supplied GL renderbuffer // Create a CL image from the supplied GL renderbuffer
cl_mem image = (*clCreateFromGLRenderbuffer_ptr)( context, CL_MEM_READ_ONLY, glRenderbuffer, &error ); cl_mem image = (*clCreateFromGLRenderbuffer_ptr)(context, CL_MEM_READ_ONLY,
if( error != CL_SUCCESS ) 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 error;
} }
return test_cl_image_read( context, queue, glTarget, image, imageWidth, return test_cl_image_read(context, queue, glTarget, image, imageWidth,
imageHeight, 1, 1, outFormat, outType, outResultBuffer ); imageHeight, 1, 1, outFormat, outType,
outResultBuffer);
} }
int test_renderbuffer_read_image( cl_context context, cl_command_queue queue, int test_renderbuffer_read_image(cl_context context, cl_command_queue queue,
GLsizei width, GLsizei height, GLenum attachment, GLsizei width, GLsizei height,
GLenum format, GLenum internalFormat, GLenum attachment, GLenum format,
GLenum glType, ExplicitType type, MTdata d ) GLenum internalFormat, GLenum glType,
ExplicitType type, MTdata d)
{ {
int error; int error;
if( type == kHalf ) if (type == kHalf)
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 1;
return 1;
// Create the GL renderbuffer // Create the GL renderbuffer
glFramebufferWrapper glFramebuffer; glFramebufferWrapper glFramebuffer;
glRenderbufferWrapper glRenderbuffer; 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<char> inputBuffer(tmp); BufferOwningPtr<char> 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; return 0;
} }
else else
@@ -114,14 +123,18 @@ int test_renderbuffer_read_image( cl_context context, cl_command_queue queue,
cl_image_format clFormat; cl_image_format clFormat;
ExplicitType actualType; ExplicitType actualType;
char *outBuffer; char *outBuffer;
error = test_attach_renderbuffer_read_image( context, queue, attachment, glRenderbuffer, width, height, &clFormat, &actualType, (void **)&outBuffer ); error = test_attach_renderbuffer_read_image(
if( error != 0 ) context, queue, attachment, glRenderbuffer, width, height, &clFormat,
return error; &actualType, (void **)&outBuffer);
if (error != 0) return error;
BufferOwningPtr<char> actualResults(outBuffer); BufferOwningPtr<char> actualResults(outBuffer);
log_info( "- Read [%4d x %4d] : GL renderbuffer : %s : %s : %s => CL Image : %s : %s \n", width, height, log_info("- Read [%4d x %4d] : GL renderbuffer : %s : %s : %s => CL Image "
GetGLFormatName( format ), GetGLFormatName( internalFormat ), GetGLTypeName( glType), ": %s : %s \n",
GetChannelOrderName( clFormat.image_channel_order ), GetChannelTypeName( clFormat.image_channel_data_type )); width, height, GetGLFormatName(format),
GetGLFormatName(internalFormat), GetGLTypeName(glType),
GetChannelOrderName(clFormat.image_channel_order),
GetChannelTypeName(clFormat.image_channel_data_type));
#ifdef DEBUG #ifdef DEBUG
log_info("- start read GL data -- \n"); log_info("- start read GL data -- \n");
@@ -129,38 +142,50 @@ int test_renderbuffer_read_image( cl_context context, cl_command_queue queue,
log_info("- end read GL data -- \n"); log_info("- end read GL data -- \n");
#endif #endif
// We have to convert our input buffer to the returned type, so we can validate. // We have to convert our input buffer to the returned type, so we can
BufferOwningPtr<char> convertedInput(convert_to_expected( inputBuffer, width * height, type, actualType, get_channel_order_channel_count(clFormat.image_channel_order) )); // validate.
BufferOwningPtr<char> convertedInput(convert_to_expected(
inputBuffer, width * height, type, actualType,
get_channel_order_channel_count(clFormat.image_channel_order)));
#ifdef DEBUG #ifdef DEBUG
log_info("- start input data -- \n"); log_info("- start input data -- \n");
DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height, convertedInput); DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height,
convertedInput);
log_info("- end input data -- \n"); log_info("- end input data -- \n");
#endif #endif
#ifdef DEBUG #ifdef DEBUG
log_info("- start converted data -- \n"); log_info("- start converted data -- \n");
DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height, actualResults); DumpGLBuffer(GetGLTypeForExplicitType(actualType), width, height,
actualResults);
log_info("- end converted data -- \n"); log_info("- end converted data -- \n");
#endif #endif
// Now we validate // Now we validate
int valid = 0; int valid = 0;
if(convertedInput) { if (convertedInput)
if( actualType == kFloat ) {
valid = validate_float_results( convertedInput, actualResults, width, height, 1, get_channel_order_channel_count(clFormat.image_channel_order) ); if (actualType == kFloat)
valid = validate_float_results(
convertedInput, actualResults, width, height, 1,
get_channel_order_channel_count(clFormat.image_channel_order));
else 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; 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 }; GLenum attachments[] = { GL_COLOR_ATTACHMENT0_EXT };
struct { struct
{
GLenum internal; GLenum internal;
GLenum format; GLenum format;
GLenum datatype; GLenum datatype;
@@ -172,8 +197,9 @@ int test_renderbuffer_read( cl_device_id device, cl_context context, cl_command_
{ GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar },
{ GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort },
// Renderbuffers with integer formats do not seem to work reliably across // Renderbuffers with integer formats do not seem to work reliably across
// platforms/implementations. Disabling this in version 1.0 of CL conformance tests. // platforms/implementations. Disabling this in version 1.0 of CL
// conformance tests.
#ifdef TEST_INTEGER_FORMATS #ifdef TEST_INTEGER_FORMATS
@@ -195,66 +221,70 @@ int test_renderbuffer_read( cl_device_id device, cl_context context, cl_command_
#else #else
size_t iter = 6; size_t iter = 6;
#endif #endif
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Check if images are supported // Check if images are supported
if (checkForImageSupport(device)) { if (checkForImageSupport(device))
{
log_info("Device does not support images. Skipping test.\n"); log_info("Device does not support images. Skipping test.\n");
return 0; return 0;
} }
if( !gluCheckExtension( (const GLubyte *)"GL_EXT_framebuffer_object", glGetString( GL_EXTENSIONS ) ) ) if (!gluCheckExtension((const GLubyte *)"GL_EXT_framebuffer_object",
glGetString(GL_EXTENSIONS)))
{ {
log_info( "Renderbuffers are not supported by this OpenGL implementation; skipping test\n" ); log_info("Renderbuffers are not supported by this OpenGL "
"implementation; skipping test\n");
return 0; return 0;
} }
// Loop through a set of GL formats, testing a set of sizes against each one // 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; size_t i;
log_info( "Testing renderbuffer read for %s : %s : %s : %s\n", log_info("Testing renderbuffer read for %s : %s : %s : %s\n",
GetGLAttachmentName( attachments[ attIdx ] ), GetGLAttachmentName(attachments[attIdx]),
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLFormatName(formats[fmtIdx].internal),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLBaseFormatName(formats[fmtIdx].format),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); 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 width = random_in_range(16, 512, seed);
GLsizei height = random_in_range( 16, 512, seed ); GLsizei height = random_in_range(16, 512, seed);
#ifdef DEBUG #ifdef DEBUG
width = height = 4; width = height = 4;
#endif #endif
if( test_renderbuffer_read_image( context, queue, width, height, if (test_renderbuffer_read_image(
attachments[ attIdx ], context, queue, width, height, attachments[attIdx],
formats[ fmtIdx ].format, formats[fmtIdx].format, formats[fmtIdx].internal,
formats[ fmtIdx ].internal, formats[fmtIdx].datatype, formats[fmtIdx].type, seed))
formats[ fmtIdx ].datatype,
formats[ fmtIdx ].type, seed ) )
{ {
log_error( "ERROR: Renderbuffer read test failed for %s : %s : %s : %s\n\n", log_error("ERROR: Renderbuffer read test failed for %s : "
GetGLAttachmentName( attachments[ attIdx ] ), "%s : %s : %s\n\n",
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLAttachmentName(attachments[attIdx]),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLFormatName(formats[fmtIdx].internal),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLBaseFormatName(formats[fmtIdx].format),
GetGLTypeName(formats[fmtIdx].datatype));
error++; 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", log_info("passed: Renderbuffer read test passed for %s : %s : "
GetGLAttachmentName( attachments[ attIdx ] ), "%s : %s\n\n",
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLAttachmentName(attachments[attIdx]),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLFormatName(formats[fmtIdx].internal),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); 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 ------------------------- #pragma mark -------------------- Write tests -------------------------
int test_attach_renderbuffer_write_to_image( cl_context context, cl_command_queue queue, GLenum glTarget, GLuint glRenderbuffer, int test_attach_renderbuffer_write_to_image(
size_t imageWidth, size_t imageHeight, cl_image_format *outFormat, ExplicitType *outType, MTdata d, void **outSourceBuffer, bool supports_half ) 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; int error;
// Create a CL image from the supplied GL renderbuffer // Create a CL image from the supplied GL renderbuffer
clMemWrapper image = (*clCreateFromGLRenderbuffer_ptr)( context, CL_MEM_WRITE_ONLY, glRenderbuffer, &error ); clMemWrapper image = (*clCreateFromGLRenderbuffer_ptr)(
if( error != CL_SUCCESS ) 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 error;
} }
return test_cl_image_write( context, queue, glTarget, image, imageWidth, return test_cl_image_write(context, queue, glTarget, image, imageWidth,
imageHeight, 1, outFormat, outType, outSourceBuffer, d, supports_half ); imageHeight, 1, outFormat, outType,
outSourceBuffer, d, supports_half);
} }
int test_renderbuffer_image_write( cl_context context, cl_command_queue queue, int test_renderbuffer_image_write(cl_context context, cl_command_queue queue,
GLsizei width, GLsizei height, GLenum attachment, GLsizei width, GLsizei height,
GLenum format, GLenum internalFormat, GLenum attachment, GLenum format,
GLenum glType, ExplicitType type, MTdata d ) GLenum internalFormat, GLenum glType,
ExplicitType type, MTdata d)
{ {
int error; int error;
if( type == kHalf ) if (type == kHalf)
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 1;
return 1;
// Create the GL renderbuffer // Create the GL renderbuffer
glFramebufferWrapper glFramebuffer; glFramebufferWrapper glFramebuffer;
glRenderbufferWrapper glRenderbuffer; glRenderbufferWrapper glRenderbuffer;
CreateGLRenderbuffer( width, height, attachment, format, internalFormat, glType, type, &glFramebuffer, &glRenderbuffer, &error, d, false ); CreateGLRenderbuffer(width, height, attachment, format, internalFormat,
if( error != 0 ) 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; return 0;
} }
else else
@@ -318,27 +357,34 @@ int test_renderbuffer_image_write( cl_context context, cl_command_queue queue,
bool supports_half = false; bool supports_half = false;
error = supportsHalf(context, &supports_half); error = supportsHalf(context, &supports_half);
if( error != 0 ) if (error != 0) return error;
return error;
error = test_attach_renderbuffer_write_to_image( context, queue, attachment, glRenderbuffer, width, height, &clFormat, &sourceType, d, (void **)&outSourceBuffer, supports_half ); error = test_attach_renderbuffer_write_to_image(
if( error != 0 || ((sourceType == kHalf ) && !supports_half)) context, queue, attachment, glRenderbuffer, width, height, &clFormat,
return error; &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 actual source type was half, convert to float for validation.
if( sourceType == kHalf ) if (sourceType == kHalf)
validationType = kFloat; validationType = kFloat;
else else
validationType = sourceType; validationType = sourceType;
BufferOwningPtr<char> validationSource( convert_to_expected( outSourceBuffer, width * height, sourceType, validationType, get_channel_order_channel_count(clFormat.image_channel_order) ) ); BufferOwningPtr<char> 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, log_info("- Write [%4d x %4d] : GL Renderbuffer : %s : %s : %s => CL Image "
GetGLFormatName( format ), GetGLFormatName( internalFormat ), GetGLTypeName( glType), ": %s : %s \n",
GetChannelOrderName( clFormat.image_channel_order ), GetChannelTypeName( clFormat.image_channel_data_type )); 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 // Now read the results from the GL renderbuffer
BufferOwningPtr<char> resultData( ReadGLRenderbuffer( glFramebuffer, glRenderbuffer, attachment, format, internalFormat, glType, type, width, height ) ); BufferOwningPtr<char> resultData(
ReadGLRenderbuffer(glFramebuffer, glRenderbuffer, attachment, format,
internalFormat, glType, type, width, height));
#ifdef DEBUG #ifdef DEBUG
log_info("- start result data -- \n"); log_info("- start result data -- \n");
@@ -346,38 +392,50 @@ int test_renderbuffer_image_write( cl_context context, cl_command_queue queue,
log_info("- end result data -- \n"); log_info("- end result data -- \n");
#endif #endif
// We have to convert our input buffer to the returned type, so we can validate. // We have to convert our input buffer to the returned type, so we can
BufferOwningPtr<char> convertedData( convert_to_expected( resultData, width * height, type, validationType, get_channel_order_channel_count(clFormat.image_channel_order) ) ); // validate.
BufferOwningPtr<char> convertedData(convert_to_expected(
resultData, width * height, type, validationType,
get_channel_order_channel_count(clFormat.image_channel_order)));
#ifdef DEBUG #ifdef DEBUG
log_info("- start input data -- \n"); log_info("- start input data -- \n");
DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height, validationSource); DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height,
validationSource);
log_info("- end input data -- \n"); log_info("- end input data -- \n");
#endif #endif
#ifdef DEBUG #ifdef DEBUG
log_info("- start converted data -- \n"); log_info("- start converted data -- \n");
DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height, convertedData); DumpGLBuffer(GetGLTypeForExplicitType(validationType), width, height,
convertedData);
log_info("- end converted data -- \n"); log_info("- end converted data -- \n");
#endif #endif
// Now we validate // Now we validate
int valid = 0; int valid = 0;
if(convertedData) { 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 (sourceType == kFloat || sourceType == kHalf)
valid = validate_float_results(
validationSource, convertedData, width, height, 1,
get_channel_order_channel_count(clFormat.image_channel_order));
else 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; 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 }; GLenum attachments[] = { GL_COLOR_ATTACHMENT0_EXT };
struct { struct
{
GLenum internal; GLenum internal;
GLenum format; GLenum format;
GLenum datatype; GLenum datatype;
@@ -389,8 +447,9 @@ int test_renderbuffer_write( cl_device_id device, cl_context context, cl_command
{ GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar },
{ GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort },
// Renderbuffers with integer formats do not seem to work reliably across // Renderbuffers with integer formats do not seem to work reliably across
// platforms/implementations. Disabling this in version 1.0 of CL conformance tests. // platforms/implementations. Disabling this in version 1.0 of CL
// conformance tests.
#ifdef TEST_INTEGER_FORMATS #ifdef TEST_INTEGER_FORMATS
@@ -411,64 +470,68 @@ int test_renderbuffer_write( cl_device_id device, cl_context context, cl_command
#ifdef DEBUG #ifdef DEBUG
iter = 1; iter = 1;
#endif #endif
RandomSeed seed( gRandomSeed ); RandomSeed seed(gRandomSeed);
// Check if images are supported // Check if images are supported
if (checkForImageSupport(device)) { if (checkForImageSupport(device))
{
log_info("Device does not support images. Skipping test.\n"); log_info("Device does not support images. Skipping test.\n");
return 0; return 0;
} }
if( !gluCheckExtension( (const GLubyte *)"GL_EXT_framebuffer_object", glGetString( GL_EXTENSIONS ) ) ) if (!gluCheckExtension((const GLubyte *)"GL_EXT_framebuffer_object",
glGetString(GL_EXTENSIONS)))
{ {
log_info( "Renderbuffers are not supported by this OpenGL implementation; skipping test\n" ); log_info("Renderbuffers are not supported by this OpenGL "
"implementation; skipping test\n");
return 0; return 0;
} }
// Loop through a set of GL formats, testing a set of sizes against each one // 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", log_info("Testing Renderbuffer write test for %s : %s : %s : %s\n",
GetGLAttachmentName( attachments[ attIdx ] ), GetGLAttachmentName(attachments[attIdx]),
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLFormatName(formats[fmtIdx].internal),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLBaseFormatName(formats[fmtIdx].format),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLTypeName(formats[fmtIdx].datatype));
size_t i; size_t i;
for( i = 0; i < iter; i++ ) for (i = 0; i < iter; i++)
{ {
GLsizei width = random_in_range( 16, 512, seed ); GLsizei width = random_in_range(16, 512, seed);
GLsizei height = random_in_range( 16, 512, seed ); GLsizei height = random_in_range(16, 512, seed);
#ifdef DEBUG #ifdef DEBUG
width = height = 4; width = height = 4;
#endif #endif
if( test_renderbuffer_image_write( context, queue, width, height, if (test_renderbuffer_image_write(
attachments[ attIdx ], context, queue, width, height, attachments[attIdx],
formats[ fmtIdx ].format, formats[fmtIdx].format, formats[fmtIdx].internal,
formats[ fmtIdx ].internal, formats[fmtIdx].datatype, formats[fmtIdx].type, seed))
formats[ fmtIdx ].datatype,
formats[ fmtIdx ].type, seed ) )
{ {
log_error( "ERROR: Renderbuffer write test failed for %s : %s : %s : %s\n\n", log_error("ERROR: Renderbuffer write test failed for %s : "
GetGLAttachmentName( attachments[ attIdx ] ), "%s : %s : %s\n\n",
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLAttachmentName(attachments[attIdx]),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLFormatName(formats[fmtIdx].internal),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLBaseFormatName(formats[fmtIdx].format),
GetGLTypeName(formats[fmtIdx].datatype));
error++; 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", log_info("passed: Renderbuffer write test passed for %s : %s : "
GetGLAttachmentName( attachments[ attIdx ] ), "%s : %s\n\n",
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLAttachmentName(attachments[attIdx]),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLFormatName(formats[fmtIdx].internal),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLBaseFormatName(formats[fmtIdx].format),
GetGLTypeName(formats[fmtIdx].datatype));
} }
} }
} }

103
test_conformance/gl/test_renderbuffer_info.cpp
View File

@@ -14,59 +14,63 @@
// limitations under the License. // limitations under the License.
// //
#include "testBase.h" #include "testBase.h"
#if defined( __APPLE__ ) #if defined(__APPLE__)
#include <OpenGL/glu.h> #include <OpenGL/glu.h>
#else #else
#include <GL/glu.h> #include <GL/glu.h>
#include <CL/cl_gl.h> #include <CL/cl_gl.h>
#endif #endif
static int test_renderbuffer_object_info( cl_context context, cl_command_queue queue, static int test_renderbuffer_object_info(cl_context context,
GLsizei width, GLsizei height, GLenum attachment, cl_command_queue queue, GLsizei width,
GLsizei height, GLenum attachment,
GLenum format, GLenum internalFormat, GLenum format, GLenum internalFormat,
GLenum glType, ExplicitType type, MTdata d ) GLenum glType, ExplicitType type,
MTdata d)
{ {
int error; int error;
if( type == kHalf ) if (type == kHalf)
if( DetectFloatToHalfRoundingMode(queue) ) if (DetectFloatToHalfRoundingMode(queue)) return 1;
return 1;
// Create the GL render buffer // Create the GL render buffer
glFramebufferWrapper glFramebuffer; glFramebufferWrapper glFramebuffer;
glRenderbufferWrapper glRenderbuffer; glRenderbufferWrapper glRenderbuffer;
BufferOwningPtr<char> inputBuffer(CreateGLRenderbuffer( width, height, attachment, format, internalFormat, glType, type, &glFramebuffer, &glRenderbuffer, &error, d, true )); BufferOwningPtr<char> inputBuffer(CreateGLRenderbuffer(
if( error != 0 ) width, height, attachment, format, internalFormat, glType, type,
return error; &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"); test_error(error, "clCreateFromGLRenderbuffer failed");
log_info( "- Given a GL format of %s, input type was %s, size was %d x %d\n", log_info("- Given a GL format of %s, input type was %s, size was %d x %d\n",
GetGLFormatName( internalFormat ), GetGLFormatName(internalFormat), get_explicit_type_name(type),
get_explicit_type_name( type ), (int)width, (int)height ); (int)width, (int)height);
// Verify the expected information here. // 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 }; GLenum attachments[] = { GL_COLOR_ATTACHMENT0_EXT };
struct { struct
{
GLenum internal; GLenum internal;
GLenum format; GLenum format;
GLenum datatype; GLenum datatype;
ExplicitType type; ExplicitType type;
} formats[] = { } formats[] = { { GL_RGBA, GL_RGBA, GL_UNSIGNED_BYTE, kUChar },
{ GL_RGBA, GL_RGBA, GL_UNSIGNED_BYTE, kUChar },
{ GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar }, { GL_RGBA8, GL_RGBA, GL_UNSIGNED_BYTE, kUChar },
{ GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort }, { GL_RGBA16, GL_RGBA, GL_UNSIGNED_SHORT, kUShort },
{ GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat }, { GL_RGBA32F_ARB, GL_RGBA, GL_FLOAT, kFloat },
{ GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } { GL_RGBA16F_ARB, GL_RGBA, GL_HALF_FLOAT, kHalf } };
};
size_t fmtIdx, tgtIdx; size_t fmtIdx, tgtIdx;
int error = 0; int error = 0;
@@ -74,58 +78,61 @@ int test_renderbuffer_getinfo( cl_device_id device, cl_context context, cl_comma
RandomSeed seed(gRandomSeed); RandomSeed seed(gRandomSeed);
// Check if images are supported // Check if images are supported
if (checkForImageSupport(device)) { if (checkForImageSupport(device))
{
log_info("Device does not support images. Skipping test.\n"); log_info("Device does not support images. Skipping test.\n");
return 0; return 0;
} }
if( !gluCheckExtension( (const GLubyte *)"GL_EXT_framebuffer_object", glGetString( GL_EXTENSIONS ) ) ) if (!gluCheckExtension((const GLubyte *)"GL_EXT_framebuffer_object",
glGetString(GL_EXTENSIONS)))
{ {
log_info( "Renderbuffers are not supported by this OpenGL implementation; skipping test\n" ); log_info("Renderbuffers are not supported by this OpenGL "
"implementation; skipping test\n");
return 0; return 0;
} }
// Loop through a set of GL formats, testing a set of sizes against each one // 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", log_info("Testing Renderbuffer object info for %s : %s : %s\n",
GetGLFormatName( formats[ fmtIdx ].internal ), GetGLFormatName(formats[fmtIdx].internal),
GetGLBaseFormatName( formats[ fmtIdx ].format ), GetGLBaseFormatName(formats[fmtIdx].format),
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLTypeName(formats[fmtIdx].datatype));
size_t i; size_t i;
for( i = 0; i < iter; i++ ) for (i = 0; i < iter; i++)
{ {
GLsizei width = random_in_range( 16, 512, seed ); GLsizei width = random_in_range(16, 512, seed);
GLsizei height = 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, if (test_renderbuffer_object_info(
attachments[ tgtIdx ], context, queue, (int)width, (int)height,
formats[ fmtIdx ].format, attachments[tgtIdx], formats[fmtIdx].format,
formats[ fmtIdx ].internal, formats[fmtIdx].internal, formats[fmtIdx].datatype,
formats[ fmtIdx ].datatype, formats[fmtIdx].type, seed))
formats[ fmtIdx ].type, seed ) )
{ {
log_error( "ERROR: Renderbuffer write test failed for GL format %s : %s\n\n", log_error("ERROR: Renderbuffer write test failed for GL "
GetGLFormatName( formats[ fmtIdx ].internal ), "format %s : %s\n\n",
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLFormatName(formats[fmtIdx].internal),
GetGLTypeName(formats[fmtIdx].datatype));
error++; 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", log_info("passed: Renderbuffer write test passed for GL format "
GetGLFormatName( formats[ fmtIdx ].internal ), "%s : %s\n\n",
GetGLTypeName( formats[ fmtIdx ].datatype ) ); GetGLFormatName(formats[fmtIdx].internal),
GetGLTypeName(formats[fmtIdx].datatype));
} }
} }
} }
return error; return error;
} }