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

Use the same kernel name for all tests (#1194)

Browse Source 
Make signature of BuildKernel more consistent across tests: now only two
variants exist:

  int BuildKernel(const char *name, int vectorSize, cl_kernel *k,
                  cl_program *p, bool relaxedMode)

or

  int BuildKernel(const char *name/symbol, int vectorSize,
                  cl_uint kernel_count, cl_kernel *k, cl_program *p,
                  bool relaxedMode)

Signed-off-by: Marco Antognini <marco.antognini@arm.com>
This commit is contained in:
Marco Antognini
2021-03-24 16:20:05 +00:00
committed by GitHub
parent ef19796590
commit 8488d4b2de
2 changed files with 23 additions and 39 deletions
Download Patch File Download Diff File Expand all files Collapse all files

32
test_conformance/math_brute_force/binary_operator_double.cpp
View File

@@ -20,14 +20,12 @@
#include <cstring> #include <cstring>
static int BuildKernel(const char *name, const char *operator_symbol, static int BuildKernel(const char *operator_symbol, int vectorSize,
int vectorSize, cl_uint kernel_count, cl_kernel *k, cl_uint kernel_count, cl_kernel *k, cl_program *p,
cl_program *p, bool relaxedMode) bool relaxedMode)
{ {
const char *c[] = { "#pragma OPENCL EXTENSION cl_khr_fp64 : enable\n", const char *c[] = { "#pragma OPENCL EXTENSION cl_khr_fp64 : enable\n",
"__kernel void ", "__kernel void math_kernel",
name,
"_kernel",
sizeNames[vectorSize], sizeNames[vectorSize],
"( __global double", "( __global double",
sizeNames[vectorSize], sizeNames[vectorSize],
@@ -38,16 +36,14 @@ static int BuildKernel(const char *name, const char *operator_symbol,
"* in2 )\n" "* in2 )\n"
"{\n" "{\n"
" size_t i = get_global_id(0);\n" " size_t i = get_global_id(0);\n"
" out[i] = in1[i] ", " out[i] = in1[i] ",
operator_symbol, operator_symbol,
" in2[i];\n" " in2[i];\n"
"}\n" }; "}\n" };
const char *c3[] = { const char *c3[] = {
"#pragma OPENCL EXTENSION cl_khr_fp64 : enable\n", "#pragma OPENCL EXTENSION cl_khr_fp64 : enable\n",
"__kernel void ", "__kernel void math_kernel",
name,
"_kernel",
sizeNames[vectorSize], sizeNames[vectorSize],
"( __global double* out, __global double* in, __global double* in2)\n" "( __global double* out, __global double* in, __global double* in2)\n"
"{\n" "{\n"
@@ -105,7 +101,7 @@ static int BuildKernel(const char *name, const char *operator_symbol,
} }
char testName[32]; char testName[32];
snprintf(testName, sizeof(testName) - 1, "%s_kernel%s", name, snprintf(testName, sizeof(testName) - 1, "math_kernel%s",
sizeNames[vectorSize]); sizeNames[vectorSize]);
return MakeKernels(kern, (cl_uint)kernSize, testName, kernel_count, k, p, return MakeKernels(kern, (cl_uint)kernSize, testName, kernel_count, k, p,
@@ -118,7 +114,6 @@ typedef struct BuildKernelInfo
cl_uint kernel_count; cl_uint kernel_count;
cl_kernel **kernels; cl_kernel **kernels;
cl_program *programs; cl_program *programs;
const char *name;
const char *operator_symbol; const char *operator_symbol;
bool relaxedMode; // Whether to build with -cl-fast-relaxed-math. bool relaxedMode; // Whether to build with -cl-fast-relaxed-math.
} BuildKernelInfo; } BuildKernelInfo;
@@ -127,7 +122,7 @@ static cl_int BuildKernelFn(cl_uint job_id, cl_uint thread_id UNUSED, void *p)
{ {
BuildKernelInfo *info = (BuildKernelInfo *)p; BuildKernelInfo *info = (BuildKernelInfo *)p;
cl_uint i = info->offset + job_id; cl_uint i = info->offset + job_id;
return BuildKernel(info->name, info->operator_symbol, i, info->kernel_count, return BuildKernel(info->operator_symbol, i, info->kernel_count,
info->kernels[i], info->programs + i, info->relaxedMode); info->kernels[i], info->programs + i, info->relaxedMode);
} }
@@ -403,13 +398,10 @@ int TestFunc_Double_Double_Double_Operator(const Func *f, MTdata d,
// Init the kernels // Init the kernels
{ {
BuildKernelInfo build_info = { gMinVectorSizeIndex, BuildKernelInfo build_info = {
test_info.threadCount, gMinVectorSizeIndex, test_info.threadCount, test_info.k,
test_info.k, test_info.programs, f->nameInCode, relaxedMode
test_info.programs, };
f->name,
f->nameInCode,
relaxedMode };
if ((error = ThreadPool_Do(BuildKernelFn, if ((error = ThreadPool_Do(BuildKernelFn,
gMaxVectorSizeIndex - gMinVectorSizeIndex, gMaxVectorSizeIndex - gMinVectorSizeIndex,
&build_info))) &build_info)))

30
test_conformance/math_brute_force/binary_operator_float.cpp
View File

@@ -20,13 +20,11 @@
#include <cstring> #include <cstring>
static int BuildKernel(const char *name, const char *operator_symbol, static int BuildKernel(const char *operator_symbol, int vectorSize,
int vectorSize, cl_uint kernel_count, cl_kernel *k, cl_uint kernel_count, cl_kernel *k, cl_program *p,
cl_program *p, bool relaxedMode) bool relaxedMode)
{ {
const char *c[] = { "__kernel void ", const char *c[] = { "__kernel void math_kernel",
name,
"_kernel",
sizeNames[vectorSize], sizeNames[vectorSize],
"( __global float", "( __global float",
sizeNames[vectorSize], sizeNames[vectorSize],
@@ -43,9 +41,7 @@ static int BuildKernel(const char *name, const char *operator_symbol,
"}\n" }; "}\n" };
const char *c3[] = { const char *c3[] = {
"__kernel void ", "__kernel void math_kernel",
name,
"_kernel",
sizeNames[vectorSize], sizeNames[vectorSize],
"( __global float* out, __global float* in, __global float* in2)\n" "( __global float* out, __global float* in, __global float* in2)\n"
"{\n" "{\n"
@@ -103,7 +99,7 @@ static int BuildKernel(const char *name, const char *operator_symbol,
} }
char testName[32]; char testName[32];
snprintf(testName, sizeof(testName) - 1, "%s_kernel%s", name, snprintf(testName, sizeof(testName) - 1, "math_kernel%s",
sizeNames[vectorSize]); sizeNames[vectorSize]);
return MakeKernels(kern, (cl_uint)kernSize, testName, kernel_count, k, p, return MakeKernels(kern, (cl_uint)kernSize, testName, kernel_count, k, p,
@@ -116,7 +112,6 @@ typedef struct BuildKernelInfo
cl_uint kernel_count; cl_uint kernel_count;
cl_kernel **kernels; cl_kernel **kernels;
cl_program *programs; cl_program *programs;
const char *name;
const char *operator_symbol; const char *operator_symbol;
bool relaxedMode; // Whether to build with -cl-fast-relaxed-math. bool relaxedMode; // Whether to build with -cl-fast-relaxed-math.
} BuildKernelInfo; } BuildKernelInfo;
@@ -125,7 +120,7 @@ static cl_int BuildKernelFn(cl_uint job_id, cl_uint thread_id UNUSED, void *p)
{ {
BuildKernelInfo *info = (BuildKernelInfo *)p; BuildKernelInfo *info = (BuildKernelInfo *)p;
cl_uint i = info->offset + job_id; cl_uint i = info->offset + job_id;
return BuildKernel(info->name, info->operator_symbol, i, info->kernel_count, return BuildKernel(info->operator_symbol, i, info->kernel_count,
info->kernels[i], info->programs + i, info->relaxedMode); info->kernels[i], info->programs + i, info->relaxedMode);
} }
@@ -395,13 +390,10 @@ int TestFunc_Float_Float_Float_Operator(const Func *f, MTdata d,
// Init the kernels // Init the kernels
{ {
BuildKernelInfo build_info = { gMinVectorSizeIndex, BuildKernelInfo build_info = {
test_info.threadCount, gMinVectorSizeIndex, test_info.threadCount, test_info.k,
test_info.k, test_info.programs, f->nameInCode, relaxedMode
test_info.programs, };
f->name,
f->nameInCode,
relaxedMode };
if ((error = ThreadPool_Do(BuildKernelFn, if ((error = ThreadPool_Do(BuildKernelFn,
gMaxVectorSizeIndex - gMinVectorSizeIndex, gMaxVectorSizeIndex - gMinVectorSizeIndex,
&build_info))) &build_info)))