diff --git a/test_common/harness/compat.h b/test_common/harness/compat.h index 3b557852..4053b7ee 100644 --- a/test_common/harness/compat.h +++ b/test_common/harness/compat.h @@ -309,13 +309,6 @@ EXTERN_C int __builtin_clz(unsigned int pattern); #endif -#ifndef MIN -#define MIN(x, y) (((x) < (y)) ? (x) : (y)) -#endif -#ifndef MAX -#define MAX(x, y) (((x) > (y)) ? (x) : (y)) -#endif - /*----------------------------------------------------------------------------- WARNING: DO NOT USE THESE MACROS: diff --git a/test_common/harness/errorHelpers.cpp b/test_common/harness/errorHelpers.cpp index ea928bc3..eaccf641 100644 --- a/test_common/harness/errorHelpers.cpp +++ b/test_common/harness/errorHelpers.cpp @@ -18,6 +18,8 @@ #include #include +#include + #include "errorHelpers.h" #include "parseParameters.h" @@ -301,10 +303,6 @@ const char *GetQueuePropertyName(cl_command_queue_properties property) } } -#ifndef MAX -#define MAX(_a, _b) ((_a) > (_b) ? (_a) : (_b)) -#endif - #if defined(_MSC_VER) #define scalbnf(_a, _i) ldexpf(_a, _i) #define scalbn(_a, _i) ldexp(_a, _i) @@ -357,7 +355,7 @@ static float Ulp_Error_Half_Float(float test, double reference) // The unbiased exponent of the ulp unit place int ulp_exp = - HALF_MANT_DIG - 1 - MAX(ilogb(reference), HALF_MIN_EXP - 1); + HALF_MANT_DIG - 1 - std::max(ilogb(reference), HALF_MIN_EXP - 1); // Scale the exponent of the error return (float)scalbn(testVal - reference, ulp_exp); @@ -365,7 +363,7 @@ static float Ulp_Error_Half_Float(float test, double reference) // reference is a normal power of two or a zero int ulp_exp = - HALF_MANT_DIG - 1 - MAX(ilogb(reference) - 1, HALF_MIN_EXP - 1); + HALF_MANT_DIG - 1 - std::max(ilogb(reference) - 1, HALF_MIN_EXP - 1); // Scale the exponent of the error return (float)scalbn(testVal - reference, ulp_exp); @@ -437,7 +435,8 @@ float Ulp_Error(float test, double reference) return 0.0f; // if we are expecting a NaN, any NaN is fine // The unbiased exponent of the ulp unit place - int ulp_exp = FLT_MANT_DIG - 1 - MAX(ilogb(reference), FLT_MIN_EXP - 1); + int ulp_exp = + FLT_MANT_DIG - 1 - std::max(ilogb(reference), FLT_MIN_EXP - 1); // Scale the exponent of the error return (float)scalbn(testVal - reference, ulp_exp); @@ -445,7 +444,8 @@ float Ulp_Error(float test, double reference) // reference is a normal power of two or a zero // The unbiased exponent of the ulp unit place - int ulp_exp = FLT_MANT_DIG - 1 - MAX(ilogb(reference) - 1, FLT_MIN_EXP - 1); + int ulp_exp = + FLT_MANT_DIG - 1 - std::max(ilogb(reference) - 1, FLT_MIN_EXP - 1); // Scale the exponent of the error return (float)scalbn(testVal - reference, ulp_exp); @@ -513,7 +513,7 @@ float Ulp_Error_Double(double test, long double reference) // The unbiased exponent of the ulp unit place int ulp_exp = - DBL_MANT_DIG - 1 - MAX(ilogbl(reference), DBL_MIN_EXP - 1); + DBL_MANT_DIG - 1 - std::max(ilogbl(reference), DBL_MIN_EXP - 1); // Scale the exponent of the error float result = (float)scalbnl(testVal - reference, ulp_exp); @@ -529,7 +529,7 @@ float Ulp_Error_Double(double test, long double reference) // reference is a normal power of two or a zero // The unbiased exponent of the ulp unit place int ulp_exp = - DBL_MANT_DIG - 1 - MAX(ilogbl(reference) - 1, DBL_MIN_EXP - 1); + DBL_MANT_DIG - 1 - std::max(ilogbl(reference) - 1, DBL_MIN_EXP - 1); // Scale the exponent of the error float result = (float)scalbnl(testVal - reference, ulp_exp); diff --git a/test_common/harness/imageHelpers.cpp b/test_common/harness/imageHelpers.cpp index 314709f8..3a5c5533 100644 --- a/test_common/harness/imageHelpers.cpp +++ b/test_common/harness/imageHelpers.cpp @@ -690,9 +690,6 @@ int has_alpha(const cl_image_format *format) _b ^= _a; \ _a ^= _b; \ } while (0) -#ifndef MAX -#define MAX(_a, _b) ((_a) > (_b) ? (_a) : (_b)) -#endif void get_max_sizes( size_t *numberOfSizes, const int maxNumberOfSizes, size_t sizes[][3], diff --git a/test_conformance/basic/test_enqueued_local_size.cpp b/test_conformance/basic/test_enqueued_local_size.cpp index f52162a8..91fe1434 100644 --- a/test_conformance/basic/test_enqueued_local_size.cpp +++ b/test_conformance/basic/test_enqueued_local_size.cpp @@ -14,13 +14,15 @@ // limitations under the License. // #include "harness/compat.h" +#include "harness/rounding_mode.h" #include #include #include #include #include -#include "harness/rounding_mode.h" + +#include #include "procs.h" @@ -124,8 +126,8 @@ test_enqueued_local_size(cl_device_id device, cl_context context, cl_command_que err = clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(max_wgs), &max_wgs, NULL); test_error( err, "clGetDeviceInfo failed."); - localsize[0] = MIN(16, max_wgs); - localsize[1] = MIN(11, max_wgs / localsize[0]); + localsize[0] = std::min(16, max_wgs); + localsize[1] = std::min(11, max_wgs / localsize[0]); // If we need to use uniform workgroups because non-uniform workgroups are // not supported, round up to the next global size that is divisible by the // local size. diff --git a/test_conformance/buffers/test_sub_buffers.cpp b/test_conformance/buffers/test_sub_buffers.cpp index 691509fd..d6ab111e 100644 --- a/test_conformance/buffers/test_sub_buffers.cpp +++ b/test_conformance/buffers/test_sub_buffers.cpp @@ -15,6 +15,8 @@ // #include "procs.h" +#include + // Design: // To test sub buffers, we first create one main buffer. We then create several sub-buffers and // queue Actions on each one. Each Action is encapsulated in a class so it can keep track of @@ -101,13 +103,6 @@ public: } }; -#ifndef MAX -#define MAX( _a, _b ) ( (_a) > (_b) ? (_a) : (_b) ) -#endif -#ifndef MIN -#define MIN( _a, _b ) ( (_a) < (_b) ? (_a) : (_b) ) -#endif - class CopyAction : public Action { public: @@ -117,7 +112,8 @@ public: virtual cl_int Execute( cl_context context, cl_command_queue queue, cl_char tag, SubBufferWrapper &buffer1, SubBufferWrapper &buffer2, cl_char *parentBufferState ) { // Copy from sub-buffer 1 to sub-buffer 2 - size_t size = get_random_size_t( 0, MIN( buffer1.mSize, buffer2.mSize ), GetRandSeed() ); + size_t size = get_random_size_t( + 0, std::min(buffer1.mSize, buffer2.mSize), GetRandSeed()); size_t startOffset = get_random_size_t( 0, buffer1.mSize - size, GetRandSeed() ); size_t endOffset = get_random_size_t( 0, buffer2.mSize - size, GetRandSeed() ); @@ -266,7 +262,11 @@ int test_sub_buffers_read_write_core( cl_context context, cl_command_queue queue endRange = mainSize; size_t offset = get_random_size_t( toStartFrom / addressAlign, endRange / addressAlign, Action::GetRandSeed() ) * addressAlign; - size_t size = get_random_size_t( 1, ( MIN( mainSize / 8, mainSize - offset ) ) / addressAlign, Action::GetRandSeed() ) * addressAlign; + size_t size = + get_random_size_t( + 1, (std::min(mainSize / 8, mainSize - offset)) / addressAlign, + Action::GetRandSeed()) + * addressAlign; error = subBuffers[ numSubBuffers ].Allocate( mainBuffer, CL_MEM_READ_WRITE, offset, size ); test_error( error, "Unable to allocate sub buffer" ); @@ -443,7 +443,7 @@ int test_sub_buffers_read_write_dual_devices( cl_device_id deviceID, cl_context error = get_reasonable_buffer_size( otherDevice, maxBuffer2 ); test_error( error, "Unable to get buffer size for secondary device" ); - maxBuffer1 = MIN( maxBuffer1, maxBuffer2 ); + maxBuffer1 = std::min(maxBuffer1, maxBuffer2); cl_uint addressAlign1Bits, addressAlign2Bits; error = clGetDeviceInfo( deviceID, CL_DEVICE_MEM_BASE_ADDR_ALIGN, sizeof( addressAlign1Bits ), &addressAlign1Bits, NULL ); @@ -452,7 +452,7 @@ int test_sub_buffers_read_write_dual_devices( cl_device_id deviceID, cl_context error = clGetDeviceInfo( otherDevice, CL_DEVICE_MEM_BASE_ADDR_ALIGN, sizeof( addressAlign2Bits ), &addressAlign2Bits, NULL ); test_error( error, "Unable to get secondary device's address alignment" ); - cl_uint addressAlign1 = MAX( addressAlign1Bits, addressAlign2Bits ) / 8; + cl_uint addressAlign1 = std::max(addressAlign1Bits, addressAlign2Bits) / 8; // Finally time to run! return test_sub_buffers_read_write_core( testingContext, queue1, queue2, maxBuffer1, addressAlign1 ); diff --git a/test_conformance/conversions/test_conversions.cpp b/test_conformance/conversions/test_conversions.cpp index 87b8ead7..e8e572e6 100644 --- a/test_conformance/conversions/test_conversions.cpp +++ b/test_conformance/conversions/test_conversions.cpp @@ -47,6 +47,8 @@ #endif #include +#include + #include "Sleep.h" #include "basic_test_conversions.h" @@ -1003,7 +1005,8 @@ static int DoTest( cl_device_id device, Type outType, Type inType, SaturationMod uint64_t i; gTestCount++; - size_t blockCount = BUFFER_SIZE / MAX( gTypeSizes[ inType ], gTypeSizes[ outType ] ); + size_t blockCount = + BUFFER_SIZE / std::max(gTypeSizes[inType], gTypeSizes[outType]); size_t step = blockCount; uint64_t lastCase = 1ULL << (8*gTypeSizes[ inType ]); cl_event writeInputBuffer = NULL; @@ -1078,7 +1081,7 @@ static int DoTest( cl_device_id device, Type outType, Type inType, SaturationMod fflush(stdout); } - cl_uint count = (uint32_t) MIN( blockCount, lastCase - i ); + cl_uint count = (uint32_t)std::min((uint64_t)blockCount, lastCase - i); writeInputBufferInfo.count = count; // Crate a user event to represent the status of the reference value computation completion diff --git a/test_conformance/device_execution/enqueue_ndrange.cpp b/test_conformance/device_execution/enqueue_ndrange.cpp index 8ced6629..f228f063 100644 --- a/test_conformance/device_execution/enqueue_ndrange.cpp +++ b/test_conformance/device_execution/enqueue_ndrange.cpp @@ -18,6 +18,7 @@ #include "harness/testHarness.h" #include "harness/typeWrappers.h" +#include #include #include "procs.h" @@ -645,7 +646,7 @@ int test_enqueue_ndrange(cl_device_id device, cl_context context, cl_command_que max_local_size = (max_local_size > MAX_GWS)? MAX_GWS: max_local_size; if(gWimpyMode) { - max_local_size = MIN(8, max_local_size); + max_local_size = std::min((size_t)8, max_local_size); } cl_uint num = 10; diff --git a/test_conformance/device_execution/host_queue_order.cpp b/test_conformance/device_execution/host_queue_order.cpp index 2b5688d1..5376ea40 100644 --- a/test_conformance/device_execution/host_queue_order.cpp +++ b/test_conformance/device_execution/host_queue_order.cpp @@ -18,6 +18,7 @@ #include "harness/testHarness.h" #include "harness/typeWrappers.h" +#include #include #include "procs.h" @@ -124,7 +125,7 @@ int test_host_queue_order(cl_device_id device, cl_context context, cl_command_qu cl_uint num = arr_size(result); if( gWimpyMode ) { - num = MAX(num / 16, 4); + num = std::max(num / 16, 4U); } clMemWrapper res_mem; diff --git a/test_conformance/half/Test_roundTrip.cpp b/test_conformance/half/Test_roundTrip.cpp index 69fc7e41..1ab40937 100644 --- a/test_conformance/half/Test_roundTrip.cpp +++ b/test_conformance/half/Test_roundTrip.cpp @@ -14,6 +14,9 @@ // limitations under the License. // #include + +#include + #include "cl_utils.h" #include "tests.h" #include "harness/testHarness.h" @@ -156,7 +159,7 @@ int test_roundTrip( cl_device_id device, cl_context context, cl_command_queue qu } // Figure out how many elements are in a work block - size_t elementSize = MAX( sizeof(cl_half), sizeof(cl_float)); + size_t elementSize = std::max(sizeof(cl_half), sizeof(cl_float)); size_t blockCount = (size_t)getBufferSize(device) / elementSize; //elementSize is a power of two uint64_t lastCase = 1ULL << (8*sizeof(cl_half)); // number of cl_half size_t stride = blockCount; @@ -168,7 +171,7 @@ int test_roundTrip( cl_device_id device, cl_context context, cl_command_queue qu for( i = 0; i < (uint64_t)lastCase; i += stride ) { - count = (uint32_t) MIN( blockCount, lastCase - i ); + count = (uint32_t)std::min((uint64_t)blockCount, lastCase - i); //Init the input stream uint16_t *p = (uint16_t *)gIn_half; diff --git a/test_conformance/half/Test_vLoadHalf.cpp b/test_conformance/half/Test_vLoadHalf.cpp index 5dfac7a3..e9354019 100644 --- a/test_conformance/half/Test_vLoadHalf.cpp +++ b/test_conformance/half/Test_vLoadHalf.cpp @@ -17,6 +17,9 @@ #include "harness/testHarness.h" #include + +#include + #include "cl_utils.h" #include "tests.h" @@ -429,7 +432,7 @@ int Test_vLoadHalf_private( cl_device_id device, bool aligned ) } // Figure out how many elements are in a work block - size_t elementSize = MAX( sizeof(cl_half), sizeof(cl_float)); + size_t elementSize = std::max(sizeof(cl_half), sizeof(cl_float)); size_t blockCount = getBufferSize(device) / elementSize; // elementSize is power of 2 uint64_t lastCase = 1ULL << (8*sizeof(cl_half)); // number of things of size cl_half @@ -447,7 +450,7 @@ int Test_vLoadHalf_private( cl_device_id device, bool aligned ) for( i = 0; i < (uint64_t)lastCase; i += blockCount ) { - count = (uint32_t) MIN( blockCount, lastCase - i ); + count = (uint32_t)std::min((uint64_t)blockCount, lastCase - i); //Init the input stream uint16_t *p = (uint16_t *)gIn_half; diff --git a/test_conformance/half/Test_vStoreHalf.cpp b/test_conformance/half/Test_vStoreHalf.cpp index c3a328ad..85824a9f 100644 --- a/test_conformance/half/Test_vStoreHalf.cpp +++ b/test_conformance/half/Test_vStoreHalf.cpp @@ -18,6 +18,9 @@ #include "harness/testHarness.h" #include + +#include + #include "cl_utils.h" #include "tests.h" @@ -674,7 +677,7 @@ int Test_vStoreHalf_private( cl_device_id device, f2h referenceFunc, d2h doubleR } // end for vector size // Figure out how many elements are in a work block - size_t elementSize = MAX( sizeof(cl_ushort), sizeof(float)); + size_t elementSize = std::max(sizeof(cl_ushort), sizeof(float)); size_t blockCount = BUFFER_SIZE / elementSize; // elementSize is power of 2 uint64_t lastCase = 1ULL << (8*sizeof(float)); // number of floats. size_t stride = blockCount; @@ -726,7 +729,7 @@ int Test_vStoreHalf_private( cl_device_id device, f2h referenceFunc, d2h doubleR for( i = 0; i < lastCase; i += stride ) { - count = (cl_uint) MIN( blockCount, lastCase - i ); + count = (cl_uint)std::min((uint64_t)blockCount, lastCase - i); fref.i = i; dref.i = i; @@ -1272,7 +1275,7 @@ int Test_vStoreaHalf_private( cl_device_id device, f2h referenceFunc, d2h double } // Figure out how many elements are in a work block - size_t elementSize = MAX( sizeof(cl_ushort), sizeof(float)); + size_t elementSize = std::max(sizeof(cl_ushort), sizeof(float)); size_t blockCount = BUFFER_SIZE / elementSize; uint64_t lastCase = 1ULL << (8*sizeof(float)); size_t stride = blockCount; @@ -1323,7 +1326,7 @@ int Test_vStoreaHalf_private( cl_device_id device, f2h referenceFunc, d2h double for( i = 0; i < (uint64_t)lastCase; i += stride ) { - count = (cl_uint) MIN( blockCount, lastCase - i ); + count = (cl_uint)std::min((uint64_t)blockCount, lastCase - i); fref.i = i; dref.i = i; diff --git a/test_conformance/images/kernel_read_write/test_common.cpp b/test_conformance/images/kernel_read_write/test_common.cpp index 375ee587..6b3cf849 100644 --- a/test_conformance/images/kernel_read_write/test_common.cpp +++ b/test_conformance/images/kernel_read_write/test_common.cpp @@ -16,6 +16,7 @@ #include "test_common.h" +#include cl_sampler create_sampler(cl_context context, image_sampler_data *sdata, bool test_mipmaps, cl_int *error) { cl_sampler sampler = nullptr; @@ -934,13 +935,13 @@ int test_read_image(cl_context context, cl_command_queue queue, { err4 = 0.0f; } - float maxErr1 = MAX( + float maxErr1 = std::max( maxErr * maxPixel.p[0], FLT_MIN); - float maxErr2 = MAX( + float maxErr2 = std::max( maxErr * maxPixel.p[1], FLT_MIN); - float maxErr3 = MAX( + float maxErr3 = std::max( maxErr * maxPixel.p[2], FLT_MIN); - float maxErr4 = MAX( + float maxErr4 = std::max( maxErr * maxPixel.p[3], FLT_MIN); if (!(err1 <= maxErr1) @@ -1039,17 +1040,17 @@ int test_read_image(cl_context context, cl_command_queue queue, float err4 = ABS_ERROR(resultPtr[3], expected[3]); float maxErr1 = - MAX(maxErr * maxPixel.p[0], - FLT_MIN); + std::max(maxErr * maxPixel.p[0], + FLT_MIN); float maxErr2 = - MAX(maxErr * maxPixel.p[1], - FLT_MIN); + std::max(maxErr * maxPixel.p[1], + FLT_MIN); float maxErr3 = - MAX(maxErr * maxPixel.p[2], - FLT_MIN); + std::max(maxErr * maxPixel.p[2], + FLT_MIN); float maxErr4 = - MAX(maxErr * maxPixel.p[3], - FLT_MIN); + std::max(maxErr * maxPixel.p[3], + FLT_MIN); if (!(err1 <= maxErr1) diff --git a/test_conformance/images/kernel_read_write/test_iterations.cpp b/test_conformance/images/kernel_read_write/test_iterations.cpp index 03ca9595..3b779fab 100644 --- a/test_conformance/images/kernel_read_write/test_iterations.cpp +++ b/test_conformance/images/kernel_read_write/test_iterations.cpp @@ -16,6 +16,8 @@ #include "test_common.h" #include +#include + #if defined( __APPLE__ ) #include #include @@ -434,7 +436,8 @@ int validate_image_2D_depth_results(void *imageValues, void *resultValues, doubl float err1 = ABS_ERROR(resultPtr[0], expected[0]); // Clamp to the minimum absolute error for the format if (err1 > 0 && err1 < formatAbsoluteError) { err1 = 0.0f; } - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); // Check if the result matches. if( ! (err1 <= maxErr1) ) @@ -484,7 +487,8 @@ int validate_image_2D_depth_results(void *imageValues, void *resultValues, doubl imageSampler, expected, 0, &containsDenormals ); float err1 = ABS_ERROR(resultPtr[0], expected[0]); - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); if( ! (err1 <= maxErr1) ) @@ -598,10 +602,14 @@ int validate_image_2D_results(void *imageValues, void *resultValues, double form if (err2 > 0 && err2 < formatAbsoluteError) { err2 = 0.0f; } if (err3 > 0 && err3 < formatAbsoluteError) { err3 = 0.0f; } if (err4 > 0 && err4 < formatAbsoluteError) { err4 = 0.0f; } - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = + std::max(maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = + std::max(maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = + std::max(maxErr * maxPixel.p[3], FLT_MIN); // Check if the result matches. if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || @@ -671,10 +679,14 @@ int validate_image_2D_results(void *imageValues, void *resultValues, double form float err2 = ABS_ERROR(resultPtr[1], expected[1]); float err3 = ABS_ERROR(resultPtr[2], expected[2]); float err4 = ABS_ERROR(resultPtr[3], expected[3]); - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = + std::max(maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = + std::max(maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = + std::max(maxErr * maxPixel.p[3], FLT_MIN); if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || diff --git a/test_conformance/images/kernel_read_write/test_read_1D.cpp b/test_conformance/images/kernel_read_write/test_read_1D.cpp index c9ba4e84..68113f9a 100644 --- a/test_conformance/images/kernel_read_write/test_read_1D.cpp +++ b/test_conformance/images/kernel_read_write/test_read_1D.cpp @@ -17,6 +17,8 @@ #include "test_common.h" #include +#include + #if defined( __APPLE__ ) #include #include @@ -669,10 +671,14 @@ int test_read_image_1D( cl_context context, cl_command_queue queue, cl_kernel ke if (err2 > 0 && err2 < formatAbsoluteError) { err2 = 0.0f; } if (err3 > 0 && err3 < formatAbsoluteError) { err3 = 0.0f; } if (err4 > 0 && err4 < formatAbsoluteError) { err4 = 0.0f; } - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = + std::max(maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = + std::max(maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = + std::max(maxErr * maxPixel.p[3], FLT_MIN); // Check if the result matches. if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || @@ -732,10 +738,14 @@ int test_read_image_1D( cl_context context, cl_command_queue queue, cl_kernel ke ABS_ERROR(resultPtr[2], expected[2]); float err4 = ABS_ERROR(resultPtr[3], expected[3]); - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = + std::max(maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = + std::max(maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = + std::max(maxErr * maxPixel.p[3], FLT_MIN); if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || diff --git a/test_conformance/images/kernel_read_write/test_read_1D_array.cpp b/test_conformance/images/kernel_read_write/test_read_1D_array.cpp index b3287ded..ac266ad7 100644 --- a/test_conformance/images/kernel_read_write/test_read_1D_array.cpp +++ b/test_conformance/images/kernel_read_write/test_read_1D_array.cpp @@ -16,13 +16,14 @@ #include "test_common.h" #include +#include + #if defined( __APPLE__ ) #include #include #include #endif - const char *read1DArrayKernelSourcePattern = "__kernel void sample_kernel( read_only image1d_array_t input,%s __global float *xOffsets, __global float *yOffsets, __global %s4 *results %s)\n" "{\n" @@ -772,10 +773,14 @@ int test_read_image_1D_array( cl_context context, cl_command_queue queue, cl_ker if (err2 > 0 && err2 < formatAbsoluteError) { err2 = 0.0f; } if (err3 > 0 && err3 < formatAbsoluteError) { err3 = 0.0f; } if (err4 > 0 && err4 < formatAbsoluteError) { err4 = 0.0f; } - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = + std::max(maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = + std::max(maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = + std::max(maxErr * maxPixel.p[3], FLT_MIN); // Check if the result matches. if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || @@ -838,10 +843,14 @@ int test_read_image_1D_array( cl_context context, cl_command_queue queue, cl_ker ABS_ERROR(resultPtr[2], expected[2]); float err4 = ABS_ERROR(resultPtr[3], expected[3]); - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = + std::max(maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = + std::max(maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = + std::max(maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = + std::max(maxErr * maxPixel.p[3], FLT_MIN); if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || diff --git a/test_conformance/images/kernel_read_write/test_read_2D_array.cpp b/test_conformance/images/kernel_read_write/test_read_2D_array.cpp index 7cb334b2..11b78814 100644 --- a/test_conformance/images/kernel_read_write/test_read_2D_array.cpp +++ b/test_conformance/images/kernel_read_write/test_read_2D_array.cpp @@ -16,6 +16,8 @@ #include "test_common.h" #include +#include + // Utility function to clamp down image sizes for certain tests to avoid // using too much memory. static size_t reduceImageSizeRange(size_t maxDimSize) { @@ -617,7 +619,8 @@ int test_read_image_2D_array( cl_context context, cl_command_queue queue, cl_ker ABS_ERROR(resultPtr[0], expected[0]); // Clamp to the minimum absolute error for the format if (err1 > 0 && err1 < formatAbsoluteError) { err1 = 0.0f; } - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); + float maxErr1 = std::max( + maxErr * maxPixel.p[0], FLT_MIN); if( ! (err1 <= maxErr1) ) { @@ -661,7 +664,8 @@ int test_read_image_2D_array( cl_context context, cl_command_queue queue, cl_ker float err1 = ABS_ERROR(resultPtr[0], expected[0]); - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); + float maxErr1 = std::max( + maxErr * maxPixel.p[0], FLT_MIN); if( ! (err1 <= maxErr1) ) @@ -942,10 +946,14 @@ int test_read_image_2D_array( cl_context context, cl_command_queue queue, cl_ker if (err2 > 0 && err2 < formatAbsoluteError) { err2 = 0.0f; } if (err3 > 0 && err3 < formatAbsoluteError) { err3 = 0.0f; } if (err4 > 0 && err4 < formatAbsoluteError) { err4 = 0.0f; } - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = std::max( + maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = std::max( + maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = std::max( + maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = std::max( + maxErr * maxPixel.p[3], FLT_MIN); if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || ! (err3 <= maxErr3) || ! (err4 <= maxErr4) ) { @@ -1004,10 +1012,14 @@ int test_read_image_2D_array( cl_context context, cl_command_queue queue, cl_ker expected[2]); float err4 = ABS_ERROR(resultPtr[3], expected[3]); - float maxErr1 = MAX( maxErr * maxPixel.p[0], FLT_MIN ); - float maxErr2 = MAX( maxErr * maxPixel.p[1], FLT_MIN ); - float maxErr3 = MAX( maxErr * maxPixel.p[2], FLT_MIN ); - float maxErr4 = MAX( maxErr * maxPixel.p[3], FLT_MIN ); + float maxErr1 = std::max( + maxErr * maxPixel.p[0], FLT_MIN); + float maxErr2 = std::max( + maxErr * maxPixel.p[1], FLT_MIN); + float maxErr3 = std::max( + maxErr * maxPixel.p[2], FLT_MIN); + float maxErr4 = std::max( + maxErr * maxPixel.p[3], FLT_MIN); if( ! (err1 <= maxErr1) || ! (err2 <= maxErr2) || ! (err3 <= maxErr3) || ! (err4 <= maxErr4) ) diff --git a/test_conformance/integer_ops/test_add_sat.cpp b/test_conformance/integer_ops/test_add_sat.cpp index c0e45d11..e33f5c67 100644 --- a/test_conformance/integer_ops/test_add_sat.cpp +++ b/test_conformance/integer_ops/test_add_sat.cpp @@ -21,27 +21,18 @@ #include #include +#include + #include "procs.h" -#define UCHAR_MIN 0 -#define USHRT_MIN 0 -#define UINT_MIN 0 - -#ifndef MAX -#define MAX( _a, _b ) ( (_a) > (_b) ? (_a) : (_b) ) -#endif -#ifndef MIN -#define MIN( _a, _b ) ( (_a) < (_b) ? (_a) : (_b) ) -#endif - static int verify_addsat_char( const cl_char *inA, const cl_char *inB, const cl_char *outptr, int n, const char *sizeName, int vecSize ) { int i; for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] + (cl_int) inB[i]; - r = MAX( r, CL_CHAR_MIN ); - r = MIN( r, CL_CHAR_MAX ); + r = std::max(r, CL_CHAR_MIN); + r = std::min(r, CL_CHAR_MAX); if( r != outptr[i] ) { log_info( "\n%d) Failure for add_sat( (char%s) 0x%2.2x, (char%s) 0x%2.2x) = *0x%2.2x vs 0x%2.2x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } @@ -55,9 +46,9 @@ static int verify_addsat_uchar( const cl_uchar *inA, const cl_uchar *inB, const for( i = 0; i < n; i++ ) { cl_int r = (int) inA[i] + (int) inB[i]; - r = MAX( r, 0 ); - r = MIN( r, CL_UCHAR_MAX ); - if( r != outptr[i] ) + r = std::max(r, 0); + r = std::min(r, CL_UCHAR_MAX); + if (r != outptr[i]) { log_info( "\n%d) Failure for add_sat( (uchar%s) 0x%2.2x, (uchar%s) 0x%2.2x) = *0x%2.2x vs 0x%2.2x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } } return 0; @@ -69,8 +60,8 @@ static int verify_addsat_short( const cl_short *inA, const cl_short *inB, const for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] + (cl_int) inB[i]; - r = MAX( r, CL_SHRT_MIN ); - r = MIN( r, CL_SHRT_MAX ); + r = std::max(r, CL_SHRT_MIN); + r = std::min(r, CL_SHRT_MAX); if( r != outptr[i] ) { log_info( "\n%d) Failure for add_sat( (short%s) 0x%4.4x, (short%s) 0x%4.4x) = *0x%4.4x vs 0x%4.4x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } @@ -84,8 +75,8 @@ static int verify_addsat_ushort( const cl_ushort *inA, const cl_ushort *inB, con for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] + (cl_int) inB[i]; - r = MAX( r, 0 ); - r = MIN( r, CL_USHRT_MAX ); + r = std::max(r, 0); + r = std::min(r, CL_USHRT_MAX); if( r != outptr[i] ) { log_info( "\n%d) Failure for add_sat( (ushort%s) 0x%4.4x, (ushort%s) 0x%4.4x) = *0x%4.4x vs 0x%4.4x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } diff --git a/test_conformance/integer_ops/test_integers.cpp b/test_conformance/integer_ops/test_integers.cpp index 8d77b24b..6fa18e1e 100644 --- a/test_conformance/integer_ops/test_integers.cpp +++ b/test_conformance/integer_ops/test_integers.cpp @@ -16,14 +16,9 @@ #include "testBase.h" #include "harness/conversions.h" -#define TEST_SIZE 512 +#include -#ifndef MIN - #define MIN( _a, _b ) ((_a) < (_b) ? (_a) : (_b)) -#endif -#ifndef MAX - #define MAX( _a, _b ) ((_a) > (_b) ? (_a) : (_b)) -#endif +#define TEST_SIZE 512 const char *singleParamIntegerKernelSourcePattern = "__kernel void sample_test(__global %s *sourceA, __global %s *destValues)\n" @@ -1512,19 +1507,20 @@ bool verify_integer_clamp( void *sourceA, void *sourceB, void *sourceC, void *de switch( vecAType ) { case kULong: - ((cl_ulong*) destination)[0] = MAX(MIN(valueA, valueC), valueB); + ((cl_ulong *)destination)[0] = + std::max(std::min(valueA, valueC), valueB); break; case kUInt: - ((cl_uint*) destination)[0] = (cl_uint) - (MAX(MIN(valueA, valueC), valueB)); + ((cl_uint *)destination)[0] = + (cl_uint)(std::max(std::min(valueA, valueC), valueB)); break; case kUShort: - ((cl_ushort*) destination)[0] = (cl_ushort) - (MAX(MIN(valueA, valueC), valueB)); + ((cl_ushort *)destination)[0] = + (cl_ushort)(std::max(std::min(valueA, valueC), valueB)); break; case kUChar: - ((cl_uchar*) destination)[0] = (cl_uchar) - (MAX(MIN(valueA, valueC), valueB)); + ((cl_uchar *)destination)[0] = + (cl_uchar)(std::max(std::min(valueA, valueC), valueB)); break; default: //error -- should never get here @@ -1576,19 +1572,20 @@ bool verify_integer_clamp( void *sourceA, void *sourceB, void *sourceC, void *de switch( vecAType ) { case kLong: - ((cl_long*) destination)[0] = MAX(MIN(valueA, valueC), valueB); + ((cl_long *)destination)[0] = + std::max(std::min(valueA, valueC), valueB); break; case kInt: - ((cl_int*) destination)[0] = (cl_int) - (MAX(MIN(valueA, valueC), valueB)); + ((cl_int *)destination)[0] = + (cl_int)(std::max(std::min(valueA, valueC), valueB)); break; case kShort: - ((cl_short*) destination)[0] = (cl_short) - (MAX(MIN(valueA, valueC), valueB)); + ((cl_short *)destination)[0] = + (cl_short)(std::max(std::min(valueA, valueC), valueB)); break; case kChar: - ((cl_char*) destination)[0] = (cl_char) - (MAX(MIN(valueA, valueC), valueB)); + ((cl_char *)destination)[0] = + (cl_char)(std::max(std::min(valueA, valueC), valueB)); break; default: //error -- should never get here @@ -1654,13 +1651,16 @@ bool verify_integer_mad_sat( void *sourceA, void *sourceB, void *sourceC, void * ((cl_ulong*) destination)[0] = multLo; break; case kUInt: - ((cl_uint*) destination)[0] = (cl_uint) MIN( multLo, (cl_ulong) CL_UINT_MAX ); + ((cl_uint *)destination)[0] = + (cl_uint)std::min(multLo, (cl_ulong)CL_UINT_MAX); break; case kUShort: - ((cl_ushort*) destination)[0] = (cl_ushort) MIN( multLo, (cl_ulong) CL_USHRT_MAX ); + ((cl_ushort *)destination)[0] = + (cl_ushort)std::min(multLo, (cl_ulong)CL_USHRT_MAX); break; case kUChar: - ((cl_uchar*) destination)[0] = (cl_uchar) MIN( multLo, (cl_ulong) CL_UCHAR_MAX ); + ((cl_uchar *)destination)[0] = + (cl_uchar)std::min(multLo, (cl_ulong)CL_UCHAR_MAX); break; default: //error -- should never get here @@ -1744,18 +1744,18 @@ bool verify_integer_mad_sat( void *sourceA, void *sourceB, void *sourceC, void * ((cl_long*) destination)[0] = result; break; case kInt: - result = MIN( result, (cl_long) CL_INT_MAX ); - result = MAX( result, (cl_long) CL_INT_MIN ); + result = std::min(result, (cl_long)CL_INT_MAX); + result = std::max(result, (cl_long)CL_INT_MIN); ((cl_int*) destination)[0] = (cl_int) result; break; case kShort: - result = MIN( result, (cl_long) CL_SHRT_MAX ); - result = MAX( result, (cl_long) CL_SHRT_MIN ); + result = std::min(result, (cl_long)CL_SHRT_MAX); + result = std::max(result, (cl_long)CL_SHRT_MIN); ((cl_short*) destination)[0] = (cl_short) result; break; case kChar: - result = MIN( result, (cl_long) CL_CHAR_MAX ); - result = MAX( result, (cl_long) CL_CHAR_MIN ); + result = std::min(result, (cl_long)CL_CHAR_MAX); + result = std::max(result, (cl_long)CL_CHAR_MIN); ((cl_char*) destination)[0] = (cl_char) result; break; default: diff --git a/test_conformance/integer_ops/test_sub_sat.cpp b/test_conformance/integer_ops/test_sub_sat.cpp index 845d1064..2a88ee0d 100644 --- a/test_conformance/integer_ops/test_sub_sat.cpp +++ b/test_conformance/integer_ops/test_sub_sat.cpp @@ -21,28 +21,18 @@ #include #include +#include + #include "procs.h" -#define UCHAR_MIN 0 -#define USHRT_MIN 0 -#define UINT_MIN 0 - -#ifndef MAX -#define MAX( _a, _b ) ( (_a) > (_b) ? (_a) : (_b) ) -#endif -#ifndef MIN -#define MIN( _a, _b ) ( (_a) < (_b) ? (_a) : (_b) ) -#endif - - static int verify_subsat_char( const cl_char *inA, const cl_char *inB, const cl_char *outptr, int n, const char *sizeName, int vecSize ) { int i; for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] - (cl_int) inB[i]; - r = MAX( r, CL_CHAR_MIN ); - r = MIN( r, CL_CHAR_MAX ); + r = std::max(r, CL_CHAR_MIN); + r = std::min(r, CL_CHAR_MAX); if( r != outptr[i] ) { log_info( "\n%d) Failure for sub_sat( (char%s) 0x%2.2x, (char%s) 0x%2.2x) = *0x%2.2x vs 0x%2.2x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } @@ -56,9 +46,9 @@ static int verify_subsat_uchar( const cl_uchar *inA, const cl_uchar *inB, const for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] - (cl_int) inB[i]; - r = MAX( r, 0 ); - r = MIN( r, CL_UCHAR_MAX ); - if( r != outptr[i] ) + r = std::max(r, 0); + r = std::min(r, CL_UCHAR_MAX); + if (r != outptr[i]) { log_info( "\n%d) Failure for sub_sat( (uchar%s) 0x%2.2x, (uchar%s) 0x%2.2x) = *0x%2.2x vs 0x%2.2x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } } return 0; @@ -70,8 +60,8 @@ static int verify_subsat_short( const cl_short *inA, const cl_short *inB, const for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] - (cl_int) inB[i]; - r = MAX( r, CL_SHRT_MIN ); - r = MIN( r, CL_SHRT_MAX ); + r = std::max(r, CL_SHRT_MIN); + r = std::min(r, CL_SHRT_MAX); if( r != outptr[i] ) { log_info( "\n%d) Failure for sub_sat( (short%s) 0x%4.4x, (short%s) 0x%4.4x) = *0x%4.4x vs 0x%4.4x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } @@ -85,8 +75,8 @@ static int verify_subsat_ushort( const cl_ushort *inA, const cl_ushort *inB, con for( i = 0; i < n; i++ ) { cl_int r = (cl_int) inA[i] - (cl_int) inB[i]; - r = MAX( r, 0 ); - r = MIN( r, CL_USHRT_MAX ); + r = std::max(r, 0); + r = std::min(r, CL_USHRT_MAX); if( r != outptr[i] ) { log_info( "\n%d) Failure for sub_sat( (ushort%s) 0x%4.4x, (ushort%s) 0x%4.4x) = *0x%4.4x vs 0x%4.4x\n", i, sizeName, inA[i], sizeName, inB[i], r, outptr[i] ); return -1; } diff --git a/test_conformance/integer_ops/test_unary_ops.cpp b/test_conformance/integer_ops/test_unary_ops.cpp index 72940eaa..c91c85ae 100644 --- a/test_conformance/integer_ops/test_unary_ops.cpp +++ b/test_conformance/integer_ops/test_unary_ops.cpp @@ -107,7 +107,7 @@ int test_unary_op( cl_command_queue queue, cl_context context, OpKonstants which // For sub ops, the min control value is 2. Otherwise, it's 0 controlData[ i ] |= 0x02; else if( whichOp == kIncrement ) - // For addition ops, the MAX control value is 1. Otherwise, it's 3 + // For addition ops, the max control value is 1. Otherwise, it's 3 controlData[ i ] &= ~0x02; } streams[1] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, diff --git a/test_conformance/math_brute_force/macro_binary_double.cpp b/test_conformance/math_brute_force/macro_binary_double.cpp index 6db6aa56..d3e8071f 100644 --- a/test_conformance/math_brute_force/macro_binary_double.cpp +++ b/test_conformance/math_brute_force/macro_binary_double.cpp @@ -496,7 +496,8 @@ cl_int Test(cl_uint job_id, cl_uint thread_id, void *data) } - for (auto k = MAX(1, gMinVectorSizeIndex); k < gMaxVectorSizeIndex; k++) + for (auto k = std::max(1U, gMinVectorSizeIndex); + k < gMaxVectorSizeIndex; k++) { q = (cl_long *)out[k]; // If we aren't getting the correctly rounded result diff --git a/test_conformance/math_brute_force/macro_binary_float.cpp b/test_conformance/math_brute_force/macro_binary_float.cpp index d6d5c8eb..6c7c8c05 100644 --- a/test_conformance/math_brute_force/macro_binary_float.cpp +++ b/test_conformance/math_brute_force/macro_binary_float.cpp @@ -485,7 +485,8 @@ cl_int Test(cl_uint job_id, cl_uint thread_id, void *data) goto exit; } - for (auto k = MAX(1, gMinVectorSizeIndex); k < gMaxVectorSizeIndex; k++) + for (auto k = std::max(1U, gMinVectorSizeIndex); + k < gMaxVectorSizeIndex; k++) { q = out[k]; // If we aren't getting the correctly rounded result diff --git a/test_conformance/math_brute_force/macro_unary_double.cpp b/test_conformance/math_brute_force/macro_unary_double.cpp index 1978c185..7f3521c6 100644 --- a/test_conformance/math_brute_force/macro_unary_double.cpp +++ b/test_conformance/math_brute_force/macro_unary_double.cpp @@ -304,7 +304,8 @@ cl_int Test(cl_uint job_id, cl_uint thread_id, void *data) } - for (auto k = MAX(1, gMinVectorSizeIndex); k < gMaxVectorSizeIndex; k++) + for (auto k = std::max(1U, gMinVectorSizeIndex); + k < gMaxVectorSizeIndex; k++) { q = out[k]; // If we aren't getting the correctly rounded result diff --git a/test_conformance/math_brute_force/macro_unary_float.cpp b/test_conformance/math_brute_force/macro_unary_float.cpp index ece5e9b6..0cd54de4 100644 --- a/test_conformance/math_brute_force/macro_unary_float.cpp +++ b/test_conformance/math_brute_force/macro_unary_float.cpp @@ -309,8 +309,8 @@ cl_int Test(cl_uint job_id, cl_uint thread_id, void *data) } - for (auto k = MAX(1, gMinVectorSizeIndex); k < gMaxVectorSizeIndex; - k++) + for (auto k = std::max(1U, gMinVectorSizeIndex); + k < gMaxVectorSizeIndex; k++) { q = out[k]; // If we aren't getting the correctly rounded result diff --git a/test_conformance/math_brute_force/main.cpp b/test_conformance/math_brute_force/main.cpp index 6691f462..1a6e0c4e 100644 --- a/test_conformance/math_brute_force/main.cpp +++ b/test_conformance/math_brute_force/main.cpp @@ -18,6 +18,7 @@ #include "sleep.h" #include "utility.h" +#include #include #include #include @@ -1239,7 +1240,7 @@ float Bruteforce_Ulp_Error_Double(double test, long double reference) // The unbiased exponent of the ulp unit place int ulp_exp = - DBL_MANT_DIG - 1 - MAX(ilogbl(reference), DBL_MIN_EXP - 1); + DBL_MANT_DIG - 1 - std::max(ilogbl(reference), DBL_MIN_EXP - 1); // Scale the exponent of the error float result = (float)scalbnl(testVal - reference, ulp_exp); @@ -1255,7 +1256,7 @@ float Bruteforce_Ulp_Error_Double(double test, long double reference) // reference is a normal power of two or a zero // The unbiased exponent of the ulp unit place int ulp_exp = - DBL_MANT_DIG - 1 - MAX(ilogbl(reference) - 1, DBL_MIN_EXP - 1); + DBL_MANT_DIG - 1 - std::max(ilogbl(reference) - 1, DBL_MIN_EXP - 1); // allow correctly rounded results to pass through unmolested. (We might add // error to it below.) There is something of a performance optimization here diff --git a/test_conformance/profiling/execute.cpp b/test_conformance/profiling/execute.cpp index edfc043c..0541bfa5 100644 --- a/test_conformance/profiling/execute.cpp +++ b/test_conformance/profiling/execute.cpp @@ -21,6 +21,8 @@ #include #include +#include + #include "procs.h" #include "harness/testHarness.h" #include "harness/errorHelpers.h" @@ -29,12 +31,6 @@ typedef unsigned char uchar; #endif -#undef MIN -#define MIN(x,y) ( (x) < (y) ? (x) : (y) ) - -#undef MAX -#define MAX(x,y) ( (x) > (y) ? (x) : (y) ) - //#define CREATE_OUTPUT 1 extern int writePPM( const char *filename, uchar *buf, int xsize, int ysize ); @@ -73,8 +69,8 @@ static const char *image_filter_src = static void read_imagef( int x, int y, int w, int h, int nChannels, uchar *src, float *srcRgb ) { // clamp the coords - int x0 = MIN( MAX( x, 0 ), w - 1 ); - int y0 = MIN( MAX( y, 0 ), h - 1 ); + int x0 = std::min(std::max(x, 0), w - 1); + int y0 = std::min(std::max(y, 0), h - 1); // get tine index int indx = ( y0 * w + x0 ) * nChannels; diff --git a/test_conformance/workgroups/test_wg_broadcast.cpp b/test_conformance/workgroups/test_wg_broadcast.cpp index 35559476..29380211 100644 --- a/test_conformance/workgroups/test_wg_broadcast.cpp +++ b/test_conformance/workgroups/test_wg_broadcast.cpp @@ -20,6 +20,8 @@ #include #include +#include + #include "procs.h" @@ -310,7 +312,7 @@ test_work_group_broadcast_2D(cl_device_id device, cl_context context, cl_command localsize[0] = localsize[1] = 1; } - num_workgroups = MAX(n_elems/wg_size[0], 16); + num_workgroups = std::max(n_elems / wg_size[0], (size_t)16); globalsize[0] = num_workgroups * localsize[0]; globalsize[1] = num_workgroups * localsize[1]; num_elements = globalsize[0] * globalsize[1]; @@ -437,7 +439,7 @@ test_work_group_broadcast_3D(cl_device_id device, cl_context context, cl_command localsize[0] = localsize[1] = localsize[2] = 1; } - num_workgroups = MAX(n_elems/wg_size[0], 8); + num_workgroups = std::max(n_elems / wg_size[0], (size_t)8); globalsize[0] = num_workgroups * localsize[0]; globalsize[1] = num_workgroups * localsize[1]; globalsize[2] = num_workgroups * localsize[2]; diff --git a/test_conformance/workgroups/test_wg_scan_exclusive_max.cpp b/test_conformance/workgroups/test_wg_scan_exclusive_max.cpp index 12338b68..644b3ccf 100644 --- a/test_conformance/workgroups/test_wg_scan_exclusive_max.cpp +++ b/test_conformance/workgroups/test_wg_scan_exclusive_max.cpp @@ -20,8 +20,9 @@ #include #include -#include "procs.h" +#include +#include "procs.h" const char *wg_scan_exclusive_max_kernel_code_int = "__kernel void test_wg_scan_exclusive_max_int(global int *input, global int *output)\n" @@ -79,7 +80,7 @@ verify_wg_scan_exclusive_max_int(int *inptr, int *outptr, size_t n, size_t wg_si log_info("work_group_scan_exclusive_max int: Error at %u: expected = %d, got = %d\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; } - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); } } @@ -103,7 +104,7 @@ verify_wg_scan_exclusive_max_uint(unsigned int *inptr, unsigned int *outptr, siz log_info("work_group_scan_exclusive_max int: Error at %u: expected = %u, got = %u\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; } - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); } } @@ -127,7 +128,7 @@ verify_wg_scan_exclusive_max_long(cl_long *inptr, cl_long *outptr, size_t n, siz log_info("work_group_scan_exclusive_max long: Error at %u: expected = %lld, got = %lld\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; } - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); } } @@ -151,7 +152,7 @@ verify_wg_scan_exclusive_max_ulong(cl_ulong *inptr, cl_ulong *outptr, size_t n, log_info("work_group_scan_exclusive_max ulong: Error at %u: expected = %llu, got = %llu\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; } - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); } } diff --git a/test_conformance/workgroups/test_wg_scan_exclusive_min.cpp b/test_conformance/workgroups/test_wg_scan_exclusive_min.cpp index f4e6bf97..3c6dfc87 100644 --- a/test_conformance/workgroups/test_wg_scan_exclusive_min.cpp +++ b/test_conformance/workgroups/test_wg_scan_exclusive_min.cpp @@ -20,8 +20,9 @@ #include #include -#include "procs.h" +#include +#include "procs.h" const char *wg_scan_exclusive_min_kernel_code_int = "__kernel void test_wg_scan_exclusive_min_int(global int *input, global int *output)\n" @@ -80,7 +81,7 @@ verify_wg_scan_exclusive_min_int(int *inptr, int *outptr, size_t n, size_t wg_si log_info("work_group_scan_exclusive_min int: Error at %u: expected = %d, got = %d\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1; } - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); } } @@ -104,7 +105,7 @@ verify_wg_scan_exclusive_min_uint(unsigned int *inptr, unsigned int *outptr, siz log_info("work_group_scan_exclusive_min int: Error at %u: expected = %u, got = %u\n", j+i, min_, outptr[j+i]); return -1; } - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); } } @@ -128,7 +129,7 @@ verify_wg_scan_exclusive_min_long(cl_long *inptr, cl_long *outptr, size_t n, siz log_info("work_group_scan_exclusive_min long: Error at %u: expected = %lld, got = %lld\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1; } - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); } } @@ -152,7 +153,7 @@ verify_wg_scan_exclusive_min_ulong(cl_ulong *inptr, cl_ulong *outptr, size_t n, log_info("work_group_scan_exclusive_min ulong: Error at %u: expected = %llu, got = %llu\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1; } - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); } } diff --git a/test_conformance/workgroups/test_wg_scan_inclusive_max.cpp b/test_conformance/workgroups/test_wg_scan_inclusive_max.cpp index 44ebf805..2a2e230e 100644 --- a/test_conformance/workgroups/test_wg_scan_inclusive_max.cpp +++ b/test_conformance/workgroups/test_wg_scan_inclusive_max.cpp @@ -20,6 +20,8 @@ #include #include +#include + #include "procs.h" @@ -75,7 +77,7 @@ verify_wg_scan_inclusive_max_int(int *inptr, int *outptr, size_t n, size_t wg_si m = wg_size; for (i = 0; i < m; ++i) { - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); if (outptr[j+i] != max_) { log_info("work_group_scan_inclusive_max int: Error at %u: expected = %d, got = %d\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; @@ -99,7 +101,7 @@ verify_wg_scan_inclusive_max_uint(unsigned int *inptr, unsigned int *outptr, siz m = wg_size; for (i = 0; i < m; ++i) { - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); if (outptr[j+i] != max_) { log_info("work_group_scan_inclusive_max int: Error at %lu: expected = %u, got = %u\n", (unsigned long)(j+i), max_, outptr[j+i]); return -1; @@ -123,7 +125,7 @@ verify_wg_scan_inclusive_max_long(cl_long *inptr, cl_long *outptr, size_t n, siz m = wg_size; for (i = 0; i < m; ++i) { - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); if (outptr[j+i] != max_) { log_info("work_group_scan_inclusive_max long: Error at %u: expected = %lld, got = %lld\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; @@ -147,7 +149,7 @@ verify_wg_scan_inclusive_max_ulong(cl_ulong *inptr, cl_ulong *outptr, size_t n, m = wg_size; for (i = 0; i < m; ++i) { - max_ = MAX(inptr[j+i], max_); + max_ = std::max(inptr[j + i], max_); if (outptr[j+i] != max_) { log_info("work_group_scan_inclusive_max ulong: Error at %u: expected = %llu, got = %llu\n", (unsigned int)(j+i), max_, outptr[j+i]); return -1; diff --git a/test_conformance/workgroups/test_wg_scan_inclusive_min.cpp b/test_conformance/workgroups/test_wg_scan_inclusive_min.cpp index f2f05788..adbdad56 100644 --- a/test_conformance/workgroups/test_wg_scan_inclusive_min.cpp +++ b/test_conformance/workgroups/test_wg_scan_inclusive_min.cpp @@ -20,6 +20,8 @@ #include #include +#include + #include "procs.h" @@ -75,7 +77,7 @@ verify_wg_scan_inclusive_min_int(int *inptr, int *outptr, size_t n, size_t wg_si m = wg_size; for (i = 0; i < m; ++i) { - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); if (outptr[j+i] != min_) { log_info("work_group_scan_inclusive_min int: Error at %u: expected = %d, got = %d\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1; @@ -99,7 +101,7 @@ verify_wg_scan_inclusive_min_uint(unsigned int *inptr, unsigned int *outptr, siz m = wg_size; for (i = 0; i < m; ++i) { - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); if (outptr[j+i] != min_) { log_info("work_group_scan_inclusive_min int: Error at %u: expected = %u, got = %u\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1; @@ -123,7 +125,7 @@ verify_wg_scan_inclusive_min_long(cl_long *inptr, cl_long *outptr, size_t n, siz m = wg_size; for (i = 0; i < m; ++i) { - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); if (outptr[j+i] != min_) { log_info("work_group_scan_inclusive_min long: Error at %u: expected = %lld, got = %lld\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1; @@ -147,7 +149,7 @@ verify_wg_scan_inclusive_min_ulong(cl_ulong *inptr, cl_ulong *outptr, size_t n, m = wg_size; for (i = 0; i < m; ++i) { - min_ = MIN(inptr[j+i], min_); + min_ = std::min(inptr[j + i], min_); if (outptr[j+i] != min_) { log_info("work_group_scan_inclusive_min ulong: Error at %u: expected = %llu, got = %llu\n", (unsigned int)(j+i), min_, outptr[j+i]); return -1;