• Home
  • Line#
  • Scopes#
  • Navigate#
  • Raw
  • Download
1 //
2 // Copyright (c) 2017 The Khronos Group Inc.
3 //
4 // Licensed under the Apache License, Version 2.0 (the "License");
5 // you may not use this file except in compliance with the License.
6 // You may obtain a copy of the License at
7 //
8 //    http://www.apache.org/licenses/LICENSE-2.0
9 //
10 // Unless required by applicable law or agreed to in writing, software
11 // distributed under the License is distributed on an "AS IS" BASIS,
12 // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13 // See the License for the specific language governing permissions and
14 // limitations under the License.
15 //
16 #ifndef _kernelHelpers_h
17 #define _kernelHelpers_h
18 
19 // Configuration
20 #include "../config.hpp"
21 
22 #include "compat.h"
23 #include "testHarness.h"
24 
25 #include <stdio.h>
26 #include <stdlib.h>
27 
28 #if defined(__MINGW32__)
29 #include <malloc.h>
30 #endif
31 
32 #include <string.h>
33 
34 #ifdef __APPLE__
35 #include <OpenCL/opencl.h>
36 #else
37 #include <CL/opencl.h>
38 #endif
39 
40 #include "deviceInfo.h"
41 #include "harness/alloc.h"
42 
43 #include <functional>
44 
45 /*
46  *  The below code is intended to be used at the top of kernels that appear
47  * inline in files to set line and file info for the kernel:
48  *
49  *  const char *source = {
50  *      INIT_OPENCL_DEBUG_INFO
51  *      "__kernel void foo( int x )\n"
52  *      "{\n"
53  *      "   ...\n"
54  *      "}\n"
55  *  };
56  */
57 #define INIT_OPENCL_DEBUG_INFO SET_OPENCL_LINE_INFO(__LINE__, __FILE__)
58 #define SET_OPENCL_LINE_INFO(_line, _file)                                     \
59     "#line " STRINGIFY(_line) " " STRINGIFY(_file) "\n"
60 #ifndef STRINGIFY_VALUE
61 #define STRINGIFY_VALUE(_x) STRINGIFY(_x)
62 #endif
63 #ifndef STRINGIFY
64 #define STRINGIFY(_x) #_x
65 #endif
66 
67 const int MAX_LEN_FOR_KERNEL_LIST = 20;
68 
69 /* Helper that creates a single program and kernel from a single-kernel program
70  * source */
71 extern int
72 create_single_kernel_helper(cl_context context, cl_program *outProgram,
73                             cl_kernel *outKernel, unsigned int numKernelLines,
74                             const char **kernelProgram, const char *kernelName,
75                             const char *buildOptions = NULL);
76 
77 extern int create_single_kernel_helper_with_build_options(
78     cl_context context, cl_program *outProgram, cl_kernel *outKernel,
79     unsigned int numKernelLines, const char **kernelProgram,
80     const char *kernelName, const char *buildOptions);
81 
82 extern int create_single_kernel_helper_create_program(
83     cl_context context, cl_program *outProgram, unsigned int numKernelLines,
84     const char **kernelProgram, const char *buildOptions = NULL);
85 
86 extern int create_single_kernel_helper_create_program_for_device(
87     cl_context context, cl_device_id device, cl_program *outProgram,
88     unsigned int numKernelLines, const char **kernelProgram,
89     const char *buildOptions = NULL);
90 
91 /* Creates OpenCL C++ program. This one must be used for creating OpenCL C++
92  * program. */
93 extern int create_openclcpp_program(cl_context context, cl_program *outProgram,
94                                     unsigned int numKernelLines,
95                                     const char **kernelProgram,
96                                     const char *buildOptions = NULL);
97 
98 /* Builds program (outProgram) and creates one kernel */
99 int build_program_create_kernel_helper(
100     cl_context context, cl_program *outProgram, cl_kernel *outKernel,
101     unsigned int numKernelLines, const char **kernelProgram,
102     const char *kernelName, const char *buildOptions = NULL);
103 
104 /* Helper to obtain the biggest fit work group size for all the devices in a
105  * given group and for the given global thread size */
106 extern int get_max_common_work_group_size(cl_context context, cl_kernel kernel,
107                                           size_t globalThreadSize,
108                                           size_t *outSize);
109 
110 /* Helper to obtain the biggest fit work group size for all the devices in a
111  * given group and for the given global thread size */
112 extern int get_max_common_2D_work_group_size(cl_context context,
113                                              cl_kernel kernel,
114                                              size_t *globalThreadSize,
115                                              size_t *outSizes);
116 
117 /* Helper to obtain the biggest fit work group size for all the devices in a
118  * given group and for the given global thread size */
119 extern int get_max_common_3D_work_group_size(cl_context context,
120                                              cl_kernel kernel,
121                                              size_t *globalThreadSize,
122                                              size_t *outSizes);
123 
124 /* Helper to obtain the biggest allowed work group size for all the devices in a
125  * given group */
126 extern int get_max_allowed_work_group_size(cl_context context, cl_kernel kernel,
127                                            size_t *outSize, size_t *outLimits);
128 
129 /* Helper to obtain the biggest allowed 1D work group size on a given device */
130 extern int get_max_allowed_1d_work_group_size_on_device(cl_device_id device,
131                                                         cl_kernel kernel,
132                                                         size_t *outSize);
133 
134 /* Helper to determine if a device supports an image format */
135 extern int is_image_format_supported(cl_context context, cl_mem_flags flags,
136                                      cl_mem_object_type image_type,
137                                      const cl_image_format *fmt);
138 
139 /* Helper to get pixel size for a pixel format */
140 size_t get_pixel_bytes(const cl_image_format *fmt);
141 
142 /* Verify the given device supports images. */
143 extern test_status verifyImageSupport(cl_device_id device);
144 
145 /* Checks that the given device supports images. Same as verify, but doesn't
146  * print an error */
147 extern int checkForImageSupport(cl_device_id device);
148 extern int checkFor3DImageSupport(cl_device_id device);
149 extern int checkForReadWriteImageSupport(cl_device_id device);
150 
151 /* Checks that a given queue property is supported on the specified device.
152  * Returns 1 if supported, 0 if not or an error. */
153 extern int checkDeviceForQueueSupport(cl_device_id device,
154                                       cl_command_queue_properties prop);
155 
156 /* Helper to obtain the min alignment for a given context, i.e the max of all
157  * min alignments for devices attached to the context*/
158 size_t get_min_alignment(cl_context context);
159 
160 /* Helper to obtain the default rounding mode for single precision computation.
161  * (Double is always CL_FP_ROUND_TO_NEAREST.) Returns 0 on error. */
162 cl_device_fp_config get_default_rounding_mode(cl_device_id device);
163 
164 #define PASSIVE_REQUIRE_IMAGE_SUPPORT(device)                                  \
165     if (checkForImageSupport(device))                                          \
166     {                                                                          \
167         log_info(                                                              \
168             "\n\tNote: device does not support images. Skipping test...\n");   \
169         return TEST_SKIPPED_ITSELF;                                            \
170     }
171 
172 #define PASSIVE_REQUIRE_3D_IMAGE_SUPPORT(device)                               \
173     if (checkFor3DImageSupport(device))                                        \
174     {                                                                          \
175         log_info("\n\tNote: device does not support 3D images. Skipping "      \
176                  "test...\n");                                                 \
177         return TEST_SKIPPED_ITSELF;                                            \
178     }
179 
180 #define PASSIVE_REQUIRE_FP16_SUPPORT(device)                                   \
181     if (!device_supports_half(device))                                         \
182     {                                                                          \
183         log_info(                                                              \
184             "\n\tNote: device does not support fp16. Skipping test...\n");     \
185         return TEST_SKIPPED_ITSELF;                                            \
186     }
187 
188 /* Prints out the standard device header for all tests given the device to print
189  * for */
190 extern int printDeviceHeader(cl_device_id device);
191 
192 // Execute the CL_DEVICE_OPENCL_C_VERSION query and return the OpenCL C version
193 // is supported by the device.
194 Version get_device_cl_c_version(cl_device_id device);
195 
196 // Gets the latest (potentially non-backward compatible) OpenCL C version
197 // supported by the device.
198 Version get_device_latest_cl_c_version(cl_device_id device);
199 
200 // Gets the maximum universally supported OpenCL C version in a context, i.e.
201 // the OpenCL C version supported by all devices in a context.
202 Version get_max_OpenCL_C_for_context(cl_context context);
203 
204 // Checks whether a particular OpenCL C version is supported by the device.
205 bool device_supports_cl_c_version(cl_device_id device, Version version);
206 
207 // Poll fn every interval_ms until timeout_ms or it returns true
208 bool poll_until(unsigned timeout_ms, unsigned interval_ms,
209                 std::function<bool()> fn);
210 
211 // Checks whether the device supports double data types
212 bool device_supports_double(cl_device_id device);
213 
214 // Checks whether the device supports half data types
215 bool device_supports_half(cl_device_id device);
216 
217 #endif // _kernelHelpers_h
218