| // |
| // Copyright (c) 2017 The Khronos Group Inc. |
| // |
| // Licensed under the Apache License, Version 2.0 (the "License"); |
| // you may not use this file except in compliance with the License. |
| // You may obtain a copy of the License at |
| // |
| // http://www.apache.org/licenses/LICENSE-2.0 |
| // |
| // Unless required by applicable law or agreed to in writing, software |
| // distributed under the License is distributed on an "AS IS" BASIS, |
| // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. |
| // See the License for the specific language governing permissions and |
| // limitations under the License. |
| // |
| #ifndef _kernelHelpers_h |
| #define _kernelHelpers_h |
| |
| // Configuration |
| #include "../config.hpp" |
| |
| #include "compat.h" |
| #include "testHarness.h" |
| |
| #include <stdio.h> |
| #include <stdlib.h> |
| |
| #if defined(__MINGW32__) |
| #include <malloc.h> |
| #endif |
| |
| #include <string.h> |
| |
| #ifdef __APPLE__ |
| #include <OpenCL/opencl.h> |
| #else |
| #include <CL/opencl.h> |
| #endif |
| |
| #include "deviceInfo.h" |
| #include "harness/alloc.h" |
| |
| #include <functional> |
| |
| /* |
| * The below code is intended to be used at the top of kernels that appear |
| * inline in files to set line and file info for the kernel: |
| * |
| * const char *source = { |
| * INIT_OPENCL_DEBUG_INFO |
| * "__kernel void foo( int x )\n" |
| * "{\n" |
| * " ...\n" |
| * "}\n" |
| * }; |
| */ |
| #define INIT_OPENCL_DEBUG_INFO SET_OPENCL_LINE_INFO(__LINE__, __FILE__) |
| #define SET_OPENCL_LINE_INFO(_line, _file) \ |
| "#line " STRINGIFY(_line) " " STRINGIFY(_file) "\n" |
| #ifndef STRINGIFY_VALUE |
| #define STRINGIFY_VALUE(_x) STRINGIFY(_x) |
| #endif |
| #ifndef STRINGIFY |
| #define STRINGIFY(_x) #_x |
| #endif |
| |
| const int MAX_LEN_FOR_KERNEL_LIST = 20; |
| |
| /* Helper that creates a single program and kernel from a single-kernel program |
| * source */ |
| extern int |
| create_single_kernel_helper(cl_context context, cl_program *outProgram, |
| cl_kernel *outKernel, unsigned int numKernelLines, |
| const char **kernelProgram, const char *kernelName, |
| const char *buildOptions = NULL, |
| const bool openclCXX = false); |
| |
| extern int create_single_kernel_helper_with_build_options( |
| cl_context context, cl_program *outProgram, cl_kernel *outKernel, |
| unsigned int numKernelLines, const char **kernelProgram, |
| const char *kernelName, const char *buildOptions, |
| const bool openclCXX = false); |
| |
| extern int create_single_kernel_helper_create_program( |
| cl_context context, cl_program *outProgram, unsigned int numKernelLines, |
| const char **kernelProgram, const char *buildOptions = NULL, |
| const bool openclCXX = false); |
| |
| extern int create_single_kernel_helper_create_program_for_device( |
| cl_context context, cl_device_id device, cl_program *outProgram, |
| unsigned int numKernelLines, const char **kernelProgram, |
| const char *buildOptions = NULL, const bool openclCXX = false); |
| |
| /* Creates OpenCL C++ program. This one must be used for creating OpenCL C++ |
| * program. */ |
| extern int create_openclcpp_program(cl_context context, cl_program *outProgram, |
| unsigned int numKernelLines, |
| const char **kernelProgram, |
| const char *buildOptions = NULL); |
| |
| /* Builds program (outProgram) and creates one kernel */ |
| int build_program_create_kernel_helper( |
| cl_context context, cl_program *outProgram, cl_kernel *outKernel, |
| unsigned int numKernelLines, const char **kernelProgram, |
| const char *kernelName, const char *buildOptions = NULL); |
| |
| /* Helper to obtain the biggest fit work group size for all the devices in a |
| * given group and for the given global thread size */ |
| extern int get_max_common_work_group_size(cl_context context, cl_kernel kernel, |
| size_t globalThreadSize, |
| size_t *outSize); |
| |
| /* Helper to obtain the biggest fit work group size for all the devices in a |
| * given group and for the given global thread size */ |
| extern int get_max_common_2D_work_group_size(cl_context context, |
| cl_kernel kernel, |
| size_t *globalThreadSize, |
| size_t *outSizes); |
| |
| /* Helper to obtain the biggest fit work group size for all the devices in a |
| * given group and for the given global thread size */ |
| extern int get_max_common_3D_work_group_size(cl_context context, |
| cl_kernel kernel, |
| size_t *globalThreadSize, |
| size_t *outSizes); |
| |
| /* Helper to obtain the biggest allowed work group size for all the devices in a |
| * given group */ |
| extern int get_max_allowed_work_group_size(cl_context context, cl_kernel kernel, |
| size_t *outSize, size_t *outLimits); |
| |
| /* Helper to obtain the biggest allowed 1D work group size on a given device */ |
| extern int get_max_allowed_1d_work_group_size_on_device(cl_device_id device, |
| cl_kernel kernel, |
| size_t *outSize); |
| |
| /* Helper to determine if a device supports an image format */ |
| extern int is_image_format_supported(cl_context context, cl_mem_flags flags, |
| cl_mem_object_type image_type, |
| const cl_image_format *fmt); |
| |
| /* Helper to get pixel size for a pixel format */ |
| size_t get_pixel_bytes(const cl_image_format *fmt); |
| |
| /* Verify the given device supports images. */ |
| extern test_status verifyImageSupport(cl_device_id device); |
| |
| /* Checks that the given device supports images. Same as verify, but doesn't |
| * print an error */ |
| extern int checkForImageSupport(cl_device_id device); |
| extern int checkFor3DImageSupport(cl_device_id device); |
| extern int checkForReadWriteImageSupport(cl_device_id device); |
| |
| /* Checks that a given queue property is supported on the specified device. |
| * Returns 1 if supported, 0 if not or an error. */ |
| extern int checkDeviceForQueueSupport(cl_device_id device, |
| cl_command_queue_properties prop); |
| |
| /* Helper to obtain the min alignment for a given context, i.e the max of all |
| * min alignments for devices attached to the context*/ |
| size_t get_min_alignment(cl_context context); |
| |
| /* Helper to obtain the default rounding mode for single precision computation. |
| * (Double is always CL_FP_ROUND_TO_NEAREST.) Returns 0 on error. */ |
| cl_device_fp_config get_default_rounding_mode(cl_device_id device); |
| |
| #define PASSIVE_REQUIRE_IMAGE_SUPPORT(device) \ |
| if (checkForImageSupport(device)) \ |
| { \ |
| log_info( \ |
| "\n\tNote: device does not support images. Skipping test...\n"); \ |
| return TEST_SKIPPED_ITSELF; \ |
| } |
| |
| #define PASSIVE_REQUIRE_3D_IMAGE_SUPPORT(device) \ |
| if (checkFor3DImageSupport(device)) \ |
| { \ |
| log_info("\n\tNote: device does not support 3D images. Skipping " \ |
| "test...\n"); \ |
| return TEST_SKIPPED_ITSELF; \ |
| } |
| |
| #define PASSIVE_REQUIRE_FP16_SUPPORT(device) \ |
| if (!is_extension_available(device, "cl_khr_fp16")) \ |
| { \ |
| log_info( \ |
| "\n\tNote: device does not support fp16. Skipping test...\n"); \ |
| return TEST_SKIPPED_ITSELF; \ |
| } |
| |
| /* Prints out the standard device header for all tests given the device to print |
| * for */ |
| extern int printDeviceHeader(cl_device_id device); |
| |
| // Execute the CL_DEVICE_OPENCL_C_VERSION query and return the OpenCL C version |
| // is supported by the device. |
| Version get_device_cl_c_version(cl_device_id device); |
| |
| // Gets the latest (potentially non-backward compatible) OpenCL C version |
| // supported by the device. |
| Version get_device_latest_cl_c_version(cl_device_id device); |
| |
| // Gets the maximum universally supported OpenCL C version in a context, i.e. |
| // the OpenCL C version supported by all devices in a context. |
| Version get_max_OpenCL_C_for_context(cl_context context); |
| |
| // Checks whether a particular OpenCL C version is supported by the device. |
| bool device_supports_cl_c_version(cl_device_id device, Version version); |
| |
| // Poll fn every interval_ms until timeout_ms or it returns true |
| bool poll_until(unsigned timeout_ms, unsigned interval_ms, |
| std::function<bool()> fn); |
| |
| #endif // _kernelHelpers_h |
