def gen_checks(type_str, size):
vals_list = append_fp_postfix(type_str, Data.vals_list_dict[size])
if 'double' in type_str or 'half' in type_str or 'float' in type_str:
vals_list = Data.vals_list_dict_float[size]
reverse_vals_list = vals_list[::-1]
kernel_name = 'KERNEL_API_' + type_str + str(size)
test_string = vector_api_template.substitute(
type=type_str,
size=size,
vals=', '.join(vals_list),
reversed_vals=', '.join(reverse_vals_list),
swizIndexes=', '.join(Data.swizzle_elem_list_dict[size][::-1]))
if 'double' in type_str:
test_string += 'check_convert_as_all_dims<' + type_str + ',' + str(
size) + ', double>(inputVec);\n'
if 'half' in type_str:
test_string += 'check_convert_as_all_dims<' + type_str + ',' + str(
size) + ', sycl::half>(inputVec);\n'
if size != 1:
test_string += lo_hi_odd_even_template.substitute(
type=type_str, vals=', '.join(vals_list))
string = wrap_with_kernel(
type_str, kernel_name,
'API test for sycl::vec<' + type_str + ', ' + str(size) + '>',
test_string)
return wrap_with_test_func(TEST_NAME, type_str, string, str(size))
def gen_xyzw_str(type_str, size):
"""Generates tests for each combination of xyzw up to the size given.
For example: size=3 will generate all permutations of xyz"""
xyzw_string = ''
for length in range(size, size + 1):
for index_subset, value_subset in zip(
permutations(Data.swizzle_xyzw_list_dict[size][:size], length),
permutations(Data.vals_list_dict[size][:size], length)):
index_string = ''
for index in index_subset:
index_string += index
val_list = append_fp_postfix(type_str,
gen_ordered_values(index_string, type_str))
val_string = ', '.join(val_list)
test_string = swizzle_xyzw_rgba_assignment_template.substitute(
type=type_str,
size=size,
testVecValues=generate_value_list(type_str, size),
indexes=index_string,
vecName=Data.vec_name_dict[size],
orderedValues=val_string)
xyzw_string += wrap_with_kernel(
type_str,
'XYZW_KERNEL_' + type_str + str(size) + index_string,
'Swizzle assignment: vec<' + type_str + ', ' + str(size) + '>.'
+ index_string, test_string)
return xyzw_string
def generate_default(type_str, size):
"""Generates test for vec()"""
test_string = default_constructor_vec_template.substitute(type=type_str,
size=size)
return wrap_with_kernel(
type_str, 'VEC_DEFAULT_CONSTRUCTOR_KERNEL_' + type_str + str(size),
'Default constructor, sycl::vec<' + type_str + ', ' + str(size) + '>',
test_string)
def generate_opencl(type_str, size):
"""Generates test for vec(vector_t openclVector)"""
test_string = opencl_constructor_vec_template.substitute(type=type_str,
size=size)
return '#ifdef __SYCL_DEVICE_ONLY__\n' + wrap_with_kernel(
type_str, 'VEC_OPENCL_CONSTRUCTOR_KERNEL_' + type_str + str(size),
'vec(vector_t openclVector), sycl::vec<' + type_str + ', ' +
str(size) + '>', test_string) + '#endif // __SYCL_DEVICE_ONLY__\n'
def gen_alias_test(type_str, size):
alias_vec_name = Data.alias_dict[type_str] + str(size)
test_string = alias_test_template.substitute(
aliasVecName=alias_vec_name, typeName=type_str, size=size)
string = wrap_with_kernel(
type_str, 'KERNEL_alias_' + alias_vec_name.replace('sycl::', ''),
'Alias vector test: ' + alias_vec_name, test_string)
return wrap_with_test_func(TEST_NAME, type_str, string, str(size))
def generate_all_type_test(type_str, size):
test_string = all_type_test_template.substitute(type=type_str,
size=str(size),
swizzle=get_swizzle(size),
test_value_1=1,
test_value_2=2)
return wrap_with_kernel(
type_str, 'VEC_ALL_TYPE_OPERATOR_KERNEL_' +
type_str.replace('cl::sycl::', '').replace(' ', '') + str(size),
'All types operator test, cl::sycl::vec<' + type_str + ', ' +
str(size) + '>', test_string)
def generate_non_fp_arithmetic_test(type_str, size):
test_string = non_fp_arithmetic_test_template.substitute(
type=type_str,
size=str(size),
swizzle=get_swizzle(size),
test_value_1=1,
test_value_2=2)
return wrap_with_kernel(
type_str, 'VEC_NON_FP_ARITHMETIC_OPERATOR_KERNEL_' +
type_str.replace('cl::sycl::', '').replace(' ', '') + str(size),
'Non FP arithmetic operator test, cl::sycl::vec<' + type_str + ', ' +
str(size) + '>', test_string)
def gen_optional_checks(type_str, size, dest, dest_type, TEST_NAME_OP):
kernel_name = 'KERNEL_CONVERT_AS_' + type_str + str(size) + dest
test_string = as_convert_call_template.substitute(
type=type_str,
size=size,
vals=', '.join(type_str, Data.vals_list_dict_float[size]),
dest_type=dest_type)
string = wrap_with_kernel(
dest_type, kernel_name, 'convert() as() test for sycl::vec<' +
type_str + ', ' + str(size) + '> to ' + dest, test_string)
return wrap_with_test_func(TEST_NAME_OP, type_str, string, str(size))
def generate_all_types_specific_return_type_test(type_str, size):
test_string = specific_return_type_test_template.substitute(
type=type_str,
size=str(size),
swizzle=get_swizzle(size),
ret_type=Data.opencl_sized_return_type_dict[type_str],
test_value_1=1,
test_value_2=2)
return wrap_with_kernel(
type_str, 'VEC_SPECIFIC_RETURN_TYPE_OPERATOR_KERNEL_' +
type_str.replace('cl::sycl::', '').replace(' ', '') + str(size),
'Specific return type operator test, cl::sycl::vec<' + type_str +
', ' + str(size) + '>', test_string)
def generate_vec(type_str, size):
"""Generates test for vec<T, dims>(const &vec<T, dims>)"""
val_list = []
for _ in itertools.repeat(None, size):
val_list.append(Data.value_default_dict[type_str])
test_string = vec_constructor_vec_template.substitute(
type=type_str,
size=size,
val=Data.value_default_dict[type_str],
vals=', '.join(val_list))
return wrap_with_kernel(
type_str, 'VEC_VEC_CONSTRUCTOR_KERNEL_' + type_str + str(size),
'const &vec constructor, sycl::vec<' + type_str + ', ' + str(size) +
'>', test_string)
def generate_explicit(type_str, size):
"""Generates test for vec(const T &arg)"""
val_list = []
for _ in itertools.repeat(None, size):
val_list.append(Data.value_default_dict[type_str])
test_string = explicit_constructor_vec_template.substitute(
type=type_str,
size=size,
val=Data.value_default_dict[type_str],
vals=', '.join(val_list))
return wrap_with_kernel(
type_str, 'VEC_EXPLICIT_CONSTRUCTOR_KERNEL_' + type_str.replace(
'cl::sycl::', '').replace(' ', '') + str(size),
'Explicit constructor, cl::sycl::vec<' + type_str + ', ' + str(size) +
'>', test_string)
def gen_interop_checks(type_str, reverse_type_str, size):
vals_list = append_fp_postfix(type_str, Data.vals_list_dict[size])
reverse_vals_list = vals_list[::-1]
kernel_name = make_interop_kernel_name(type_str, size)
test_string = vector_api_template.substitute(
type=type_str,
size=size,
vals=', '.join(vals_list),
reversed_vals=', '.join(reverse_vals_list),
convertType=reverse_type_str,
asType=reverse_type_str,
swizIndexes=', '.join(Data.swizzle_elem_list_dict[size][::-1]))
if size != 1:
test_string += lo_hi_odd_even_template.substitute(type=type_str)
string = wrap_with_kernel(
type_str, kernel_name,
'API test for cl::sycl::vec<' + type_str + ', ' + str(size) + '>',
test_string)
return wrap_with_test_func(TEST_NAME, type_str, string, str(size))
def gen_host_checks(type_str, reverse_type_str, size):
"""Uses the above string templates to generate tests for each vec api function except load and store.
Load and store are handled separately.
lo() hi() odd() and even() are handled with a separate function and template to other api functions
as they can only be performed on vectors of size 2 or greater."""
vals_list = append_fp_postfix(type_str, Data.vals_list_dict[size])
reverse_vals_list = vals_list[::-1]
kernel_name = make_host_kernel_name(type_str, size)
test_string = vector_api_template.substitute(
type=type_str,
size=size,
vals=', '.join(vals_list),
reversed_vals=', '.join(reverse_vals_list),
convertType=reverse_type_str,
asType=reverse_type_str,
swizIndexes=', '.join(Data.swizzle_elem_list_dict[size][::-1]))
if size != 1:
test_string += lo_hi_odd_even_template.substitute(type=type_str)
string = wrap_with_kernel(
type_str, kernel_name,
'API test for cl::sycl::vec<' + type_str + ', ' + str(size) + '>',
test_string)
return wrap_with_test_func(TEST_NAME, type_str, string, str(size))
def gen_elem_str(type_str, size):
"""Generates in order, reverse order, and swapped pairs
of elements up to the given size.
For example: size 4 will generate element sequences
0, 1, 2, 3
1, 0, 3, 2
2, 3, 1, 0
3, 2, 1, 0"""
index_list = Data.swizzle_elem_list_dict[size][:size]
index_string = ', '.join(index_list)
test_string = swizzle_elem_assignment_template.substitute(
type=type_str,
size=size,
testVecValues=generate_value_list(type_str, size),
indexes=index_string,
vecName=Data.vec_name_dict[size],
orderedValues=', '.join(
append_fp_postfix(type_str, gen_ordered_values(index_list, type_str))))
if size == 1:
return wrap_with_kernel(
test_string, 'ELEM_KERNEL_' + type_str + str(size) +
''.join(Data.swizzle_elem_list_dict[size][:size]).replace(
'sycl::elem::', ''), 'Swizzle assignment test for vec<' +
type_str + ', ' + str(size) + '> .swizzle<' +
', '.join(Data.swizzle_elem_list_dict[size][:size]) + '>',
test_string)
index_list = swap_pairs(Data.swizzle_elem_list_dict[size][:size])
index_string = ', '.join(index_list)
test_string += swizzle_elem_assignment_template.substitute(
type=type_str,
size=size,
testVecValues=generate_value_list(type_str, size),
indexes=', '.join(
swap_pairs(Data.swizzle_elem_list_dict[size][:size])),
vecName=Data.vec_name_dict[size],
orderedValues=', '.join(
append_fp_postfix(type_str, gen_ordered_values(index_list, type_str))))
index_list = swap_pairs(Data.swizzle_elem_list_dict[size][::-1])
index_string = ', '.join(index_list)
test_string += swizzle_elem_assignment_template.substitute(
type=type_str,
size=size,
testVecValues=generate_value_list(type_str, size),
indexes=', '.join(swap_pairs(Data.swizzle_elem_list_dict[size][::-1])),
vecName=Data.vec_name_dict[size],
orderedValues=', '.join(
append_fp_postfix(type_str, gen_ordered_values(index_list, type_str))))
index_list = Data.swizzle_elem_list_dict[size][::-1]
index_string = ', '.join(index_list)
test_string += swizzle_elem_assignment_template.substitute(
type=type_str,
size=size,
testVecValues=generate_value_list(type_str, size),
indexes=', '.join(Data.swizzle_elem_list_dict[size][::-1]),
vecName=Data.vec_name_dict[size],
orderedValues=', '.join(
append_fp_postfix(type_str, gen_ordered_values(index_list, type_str))))
return wrap_with_kernel(
test_string, 'KERNEL_SWIZZLE_ASSIGNMENT_ELEM_' + type_str + str(size) +
''.join(Data.swizzle_elem_list_dict[size][:size]).replace(
'sycl::elem::', ''), 'Swizzle assignment test for vec<' +
type_str + ', ' + str(size) + '> .swizzle<' +
', '.join(Data.swizzle_elem_list_dict[size][:size]) + '>', test_string)
def gen_swizzle_test(type_str, convert_type_str, as_type_str, size):
string = ''
if size <= 4:
for length in range(size, size + 1):
for index_subset, value_subset in zip(
product(Data.swizzle_xyzw_list_dict[size][:size],
repeat=length),
product(Data.vals_list_dict[size][:size], repeat=length)):
index_list = []
val_list = []
for index, value in zip(index_subset, value_subset):
index_list.append(index)
val_list.append(value)
val_list = append_fp_postfix(type_str, val_list)
index_string = ''.join(index_list)
test_string = swizzle_template.substitute(
name=Data.vec_name_dict[size],
indexes=index_string,
type=type_str,
testVecValues=generate_value_list(type_str, size),
in_order_vals=', '.join(val_list),
reversed_vals=', '.join(val_list[::-1]),
in_order_pair_vals=', '.join(swap_pairs(val_list)),
reverse_order_pair_vals=', '.join(
swap_pairs(val_list[::-1])),
in_order_positions=''.join(
Data.swizzle_xyzw_list_dict[size][:size]),
size=size,
swiz_vals=Data.swizzle_elem_list_dict[size][::-1],
convert_type=convert_type_str,
as_type=as_type_str)
if size > 1:
test_string += lo_hi_odd_even_template.substitute(
type=type_str, size=size)
test_string += swizzle_elem_template.substitute(
type=type_str,
size=size,
in_order_swiz_indexes=', '.join(
Data.swizzle_elem_list_dict[size]),
reverse_order_swiz_indexes=', '.join(
Data.swizzle_elem_list_dict[size][::-1]),
in_order_reversed_pair_swiz_indexes=', '.join(
swap_pairs(Data.swizzle_elem_list_dict[size])),
reverse_order_reversed_pair_swiz_indexes=', '.join(
swap_pairs(Data.swizzle_elem_list_dict[size][::-1])))
string += wrap_with_kernel(
type_str, 'KERNEL_' +
type_str.replace('cl::sycl::', '').replace(' ', '') +
str(size) + index_string,
'vec<' + type_str + ', ' + str(size) + '>.' + index_string,
test_string)
if size == 4:
for length in range(size, size + 1):
for index_subset, value_subset in zip(
product(Data.swizzle_rgba_list_dict[size][:size],
repeat=length),
product(Data.vals_list_dict[size][:size],
repeat=length)):
index_list = []
val_list = []
for index, value in zip(index_subset, value_subset):
index_list.append(index)
val_list.append(value)
index_string = ''.join(index_list)
test_string = swizzle_template.substitute(
name=Data.vec_name_dict[size],
indexes=index_string,
type=type_str,
testVecValues=generate_value_list(type_str, size),
in_order_vals=', '.join(val_list),
reversed_vals=', '.join(val_list[::-1]),
in_order_pair_vals=', '.join(swap_pairs(val_list)),
reverse_order_pair_vals=', '.join(
swap_pairs(val_list[::-1])),
in_order_positions=''.join(
Data.swizzle_rgba_list_dict[size][:size]),
size=size,
swiz_vals=Data.swizzle_elem_list_dict[size][::-1],
convert_type=convert_type_str,
as_type=as_type_str)
test_string += lo_hi_odd_even_template.substitute(
type=type_str, size=size)
test_string += swizzle_elem_template.substitute(
type=type_str,
size=size,
in_order_swiz_indexes=', '.join(
Data.swizzle_elem_list_dict[size]),
reverse_order_swiz_indexes=', '.join(
Data.swizzle_elem_list_dict[size][::-1]),
in_order_reversed_pair_swiz_indexes=', '.join(
swap_pairs(Data.swizzle_elem_list_dict[size])),
reverse_order_reversed_pair_swiz_indexes=', '.join(
swap_pairs(
Data.swizzle_elem_list_dict[size][::-1])))
string += wrap_with_kernel(
type_str, 'KERNEL_' +
type_str.replace('cl::sycl::', '').replace(' ', '') +
str(size) + index_string, 'vec<' + type_str + ', ' +
str(size) + '>.' + index_string, test_string)
else:
test_string = swizzle_full_test_template.substitute(
name=Data.vec_name_dict[size],
type=type_str,
size=size,
testVecValues=generate_value_list(type_str, size),
convert_type=convert_type_str,
as_type=as_type_str,
in_order_swiz_indexes=', '.join(Data.swizzle_elem_list_dict[size]),
reverse_order_swiz_indexes=', '.join(
Data.swizzle_elem_list_dict[size][::-1]),
in_order_reversed_pair_swiz_indexes=', '.join(
swap_pairs(Data.swizzle_elem_list_dict[size])),
reverse_order_reversed_pair_swiz_indexes=', '.join(
swap_pairs(Data.swizzle_elem_list_dict[size][::-1])),
in_order_vals=', '.join(Data.vals_list_dict[size]),
reversed_vals=', '.join(Data.vals_list_dict[size][::-1]),
in_order_pair_vals=', '.join(swap_pairs(
Data.vals_list_dict[size])),
reverse_order_pair_vals=', '.join(
swap_pairs(Data.vals_list_dict[size][::-1])))
string += wrap_with_kernel(
type_str, 'ELEM_KERNEL_' +
type_str.replace('cl::sycl::', '').replace(' ', '') + str(size) +
''.join(Data.swizzle_elem_list_dict[size][:size]).replace(
'cl::sycl::elem::', ''),
'vec<' + type_str + ', ' + str(size) + '> .swizzle<' +
', '.join(Data.swizzle_elem_list_dict[size][:size]) + '>',
test_string)
return string
