diff options
Diffstat (limited to 'include/CL/cl2.hpp')
| -rw-r--r-- | include/CL/cl2.hpp | 238 |
1 files changed, 141 insertions, 97 deletions
diff --git a/include/CL/cl2.hpp b/include/CL/cl2.hpp index c6cd8a716c..0d6e805a0b 100644 --- a/include/CL/cl2.hpp +++ b/include/CL/cl2.hpp @@ -1,5 +1,5 @@ /******************************************************************************* - * Copyright (c) 2008-2015 The Khronos Group Inc. + * Copyright (c) 2008-2016 The Khronos Group Inc. * * Permission is hereby granted, free of charge, to any person obtaining a * copy of this software and/or associated documentation files (the @@ -12,6 +12,11 @@ * The above copyright notice and this permission notice shall be included * in all copies or substantial portions of the Materials. * + * MODIFICATIONS TO THIS FILE MAY MEAN IT NO LONGER ACCURATELY REFLECTS + * KHRONOS STANDARDS. THE UNMODIFIED, NORMATIVE VERSIONS OF KHRONOS + * SPECIFICATIONS AND HEADER INFORMATION ARE LOCATED AT + * https://www.khronos.org/registry/ + * * THE MATERIALS ARE PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, * EXPRESS OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF * MERCHANTABILITY, FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. @@ -23,21 +28,21 @@ /*! \file * - * \brief C++ bindings for OpenCL 1.0 (rev 48), OpenCL 1.1 (rev 33), + * \brief C++ bindings for OpenCL 1.0 (rev 48), OpenCL 1.1 (rev 33), * OpenCL 1.2 (rev 15) and OpenCL 2.0 (rev 29) * \author Lee Howes and Bruce Merry - * - * Derived from the OpenCL 1.x C++ bindings written by + * + * Derived from the OpenCL 1.x C++ bindings written by * Benedict R. Gaster, Laurent Morichetti and Lee Howes * With additions and fixes from: * Brian Cole, March 3rd 2010 and April 2012 * Matt Gruenke, April 2012. * Bruce Merry, February 2013. * Tom Deakin and Simon McIntosh-Smith, July 2013 - * James Price, June-November 2015 + * James Price, 2015- * - * \version 2.0.8 - * \date 2015-11-03 + * \version 2.0.10 + * \date 2016-07-20 * * Optional extension support * @@ -47,6 +52,18 @@ * #define CL_HPP_USE_DX_INTEROP * cl_khr_sub_groups * #define CL_HPP_USE_CL_SUB_GROUPS_KHR + * + * Doxygen documentation for this header is available here: + * + * http://khronosgroup.github.io/OpenCL-CLHPP/ + * + * The latest version of this header can be found on the GitHub releases page: + * + * https://github.com/KhronosGroup/OpenCL-CLHPP/releases + * + * Bugs and patches can be submitted to the GitHub repository: + * + * https://github.com/KhronosGroup/OpenCL-CLHPP */ /*! \mainpage @@ -57,7 +74,7 @@ * The interface is contained with a single C++ header file \em cl2.hpp and all * definitions are contained within the namespace \em cl. There is no additional * requirement to include \em cl.h and to use either the C++ or original C - * bindings; it is enough to simply include \em cl.hpp. + * bindings; it is enough to simply include \em cl2.hpp. * * The bindings themselves are lightweight and correspond closely to the * underlying C API. Using the C++ bindings introduces no additional execution @@ -129,41 +146,64 @@ * * \section parameterization Parameters * This header may be parameterized by a set of preprocessor macros. - * CL_HPP_TARGET_OPENCL_VERSION - * - Defines the target OpenCL runtime version to build the header against. - * Defaults to 200, representing OpenCL 2.0. - * CL_HPP_NO_STD_STRING - * - Do not use the standard library string class. - * cl::string is not defined and may be defined by the user before - * cl2.hpp is included. - * CL_HPP_NO_STD_VECTOR - * - Do not use the standard library vector class. - * cl::vector is not defined and may be defined by the user before - * cl2.hpp is included. - * CL_HPP_NO_STD_ARRAY - * - Do not use the standard library array class. - * cl::array is not defined and may be defined by the user before - * cl2.hpp is included. - * CL_HPP_NO_STD_UNIQUE_PTR - * - Do not use the standard library unique_ptr class. - * cl::pointer and the cl::allocate_pointer function are not defined - * and may be defined by the user before cl2.hpp is included. - * CL_HPP_ENABLE_DEVICE_FISSION - * - Enables device fission for OpenCL 1.2 platforms - * CL_HPP_ENABLE_EXCEPTIONS - * - Enable exceptions for use in the C++ bindings header. - * This is the preferred error handling mechanism but is not required. - * CL_HPP_ENABLE_SIZE_T_COMPATIBILITY - * - Backward compatibility option to support cl.hpp-style size_t class. - * Replaces the updated std::array derived version and removal of size_t - * from the namespace. Note that in this case the new size_t class - * is placed in the cl::compatibility namespace and thus requires - * an additional using declaration for direct backward compatibility. - * CL_HPP_ENABLE_PROGRAM_CONSTRUCTION_FROM_ARRAY_COMPATIBILITY - * - Enable older vector of pairs interface for construction of programs. - * CL_HPP_CL_1_2_DEFAULT_BUILD - * - Default to OpenCL C 1.2 compilation rather than OpenCL C 2.0 - * - applies to use of cl::Program construction and other program build variants. + * + * - CL_HPP_TARGET_OPENCL_VERSION + * + * Defines the target OpenCL runtime version to build the header + * against. Defaults to 200, representing OpenCL 2.0. + * + * - CL_HPP_NO_STD_STRING + * + * Do not use the standard library string class. cl::string is not + * defined and may be defined by the user before cl2.hpp is + * included. + * + * - CL_HPP_NO_STD_VECTOR + * + * Do not use the standard library vector class. cl::vector is not + * defined and may be defined by the user before cl2.hpp is + * included. + * + * - CL_HPP_NO_STD_ARRAY + * + * Do not use the standard library array class. cl::array is not + * defined and may be defined by the user before cl2.hpp is + * included. + * + * - CL_HPP_NO_STD_UNIQUE_PTR + * + * Do not use the standard library unique_ptr class. cl::pointer and + * the cl::allocate_pointer functions are not defined and may be + * defined by the user before cl2.hpp is included. + * + * - CL_HPP_ENABLE_DEVICE_FISSION + * + * Enables device fission for OpenCL 1.2 platforms. + * + * - CL_HPP_ENABLE_EXCEPTIONS + * + * Enable exceptions for use in the C++ bindings header. This is the + * preferred error handling mechanism but is not required. + * + * - CL_HPP_ENABLE_SIZE_T_COMPATIBILITY + * + * Backward compatibility option to support cl.hpp-style size_t + * class. Replaces the updated std::array derived version and + * removal of size_t from the namespace. Note that in this case the + * new size_t class is placed in the cl::compatibility namespace and + * thus requires an additional using declaration for direct backward + * compatibility. + * + * - CL_HPP_ENABLE_PROGRAM_CONSTRUCTION_FROM_ARRAY_COMPATIBILITY + * + * Enable older vector of pairs interface for construction of + * programs. + * + * - CL_HPP_CL_1_2_DEFAULT_BUILD + * + * Default to OpenCL C 1.2 compilation rather than OpenCL C 2.0 + * applies to use of cl::Program construction and other program + * build variants. * * * \section example Example @@ -172,19 +212,19 @@ * bindings, including support for the optional exception feature and * also the supplied vector and string classes, see following sections for * decriptions of these features. - * + * * \code #define CL_HPP_ENABLE_EXCEPTIONS #define CL_HPP_TARGET_OPENCL_VERSION 200 - + #include <CL/cl2.hpp> #include <iostream> #include <vector> #include <memory> #include <algorithm> - + const int numElements = 32; - + int main(void) { // Filter for a 2.0 platform and set it as the default @@ -207,35 +247,45 @@ std::cout << "Error setting default platform."; return -1; } - - std::string kernel1{ - "global int globalA;" - "kernel void updateGlobal(){" - " globalA = 75;" - "}"}; - std::string kernel2{ - "typedef struct { global int *bar; } Foo; kernel void vectorAdd(global const Foo* aNum, global const int *inputA, global const int *inputB, global int *output, int val, write_only pipe int outPipe, queue_t childQueue){" - " output[get_global_id(0)] = inputA[get_global_id(0)] + inputB[get_global_id(0)] + val + *(aNum->bar);" - " write_pipe(outPipe, &val);" - " queue_t default_queue = get_default_queue(); " - " ndrange_t ndrange = ndrange_1D(get_global_size(0)/2, get_global_size(0)/2); " - // Have a child kernel write into third quarter of output - " enqueue_kernel(default_queue, CLK_ENQUEUE_FLAGS_WAIT_KERNEL, ndrange, " - " ^{" - " output[get_global_size(0)*2 + get_global_id(0)] = inputA[get_global_size(0)*2+get_global_id(0)] + inputB[get_global_size(0)*2+get_global_id(0)] + globalA;" - " });" - // Have a child kernel write into last quarter of output - " enqueue_kernel(childQueue, CLK_ENQUEUE_FLAGS_WAIT_KERNEL, ndrange, " - " ^{" - " output[get_global_size(0)*3 + get_global_id(0)] = inputA[get_global_size(0)*3 + get_global_id(0)] + inputB[get_global_size(0)*3 + get_global_id(0)] + globalA + 2;" - " });" - "}" }; + + // Use C++11 raw string literals for kernel source code + std::string kernel1{R"CLC( + global int globalA; + kernel void updateGlobal() + { + globalA = 75; + } + )CLC"}; + std::string kernel2{R"CLC( + typedef struct { global int *bar; } Foo; + kernel void vectorAdd(global const Foo* aNum, global const int *inputA, global const int *inputB, + global int *output, int val, write_only pipe int outPipe, queue_t childQueue) + { + output[get_global_id(0)] = inputA[get_global_id(0)] + inputB[get_global_id(0)] + val + *(aNum->bar); + write_pipe(outPipe, &val); + queue_t default_queue = get_default_queue(); + ndrange_t ndrange = ndrange_1D(get_global_size(0)/2, get_global_size(0)/2); + + // Have a child kernel write into third quarter of output + enqueue_kernel(default_queue, CLK_ENQUEUE_FLAGS_WAIT_KERNEL, ndrange, + ^{ + output[get_global_size(0)*2 + get_global_id(0)] = + inputA[get_global_size(0)*2 + get_global_id(0)] + inputB[get_global_size(0)*2 + get_global_id(0)] + globalA; + }); + + // Have a child kernel write into last quarter of output + enqueue_kernel(childQueue, CLK_ENQUEUE_FLAGS_WAIT_KERNEL, ndrange, + ^{ + output[get_global_size(0)*3 + get_global_id(0)] = + inputA[get_global_size(0)*3 + get_global_id(0)] + inputB[get_global_size(0)*3 + get_global_id(0)] + globalA + 2; + }); + } + )CLC"}; // New simpler string interface style std::vector<std::string> programStrings {kernel1, kernel2}; - cl::Program vectorAddProgram( - programStrings); + cl::Program vectorAddProgram(programStrings); try { vectorAddProgram.build("-cl-std=CL2.0"); } @@ -246,7 +296,7 @@ for (auto &pair : buildInfo) { std::cerr << pair.second << std::endl << std::endl; } - + return 1; } @@ -259,17 +309,17 @@ program2Kernel( cl::EnqueueArgs( cl::NDRange(1))); - + ////////////////// // SVM allocations - - cl::pointer<int> anSVMInt = cl::allocate_svm<int, cl::SVMTraitCoarse<>>(); + + auto anSVMInt = cl::allocate_svm<int, cl::SVMTraitCoarse<>>(); *anSVMInt = 5; - cl::SVMAllocator<int, cl::SVMTraitCoarse<cl::SVMTraitReadOnly<>>> svmAllocReadOnly; + cl::SVMAllocator<Foo, cl::SVMTraitCoarse<cl::SVMTraitReadOnly<>>> svmAllocReadOnly; auto fooPointer = cl::allocate_pointer<Foo>(svmAllocReadOnly); fooPointer->bar = anSVMInt.get(); cl::SVMAllocator<int, cl::SVMTraitCoarse<>> svmAlloc; - std::vector<int, cl::SVMAllocator<int, cl::SVMTraitCoarse<>>> inputA(numElements, 1, svmAlloc); + std::vector<int, cl::SVMAllocator<int, cl::SVMTraitCoarse<>>> inputA(numElements, 1, svmAlloc); cl::coarse_svm_vector<int> inputB(numElements, 2, svmAlloc); // @@ -279,7 +329,7 @@ std::vector<int> output(numElements, 0xdeadbeef); cl::Buffer outputBuffer(begin(output), end(output), false); cl::Pipe aPipe(sizeof(cl_int), numElements / 2); - + // Default command queue, also passed in as a parameter cl::DeviceCommandQueue defaultDeviceQueue = cl::DeviceCommandQueue::makeDefault( cl::Context::getDefault(), cl::Device::getDefault()); @@ -334,17 +384,13 @@ return 0; } - * + * * \endcode * */ #ifndef CL_HPP_ #define CL_HPP_ -#ifdef __GNUC__ -#pragma GCC system_header -#endif - /* Handle deprecated preprocessor definitions. In each case, we only check for * the old name if the new name is not defined, so that user code can define * both and hence work with either version of the bindings. @@ -894,14 +940,12 @@ inline cl_int getInfoHelper(Func f, cl_uint name, vector<vector<unsigned char>>* size_type numBinaries = param->size(); vector<unsigned char*> binariesPointers(numBinaries); - size_type totalSize = 0; for (size_type i = 0; i < numBinaries; ++i) { binariesPointers[i] = (*param)[i].data(); - totalSize += (*param)[i].size(); } - cl_int err = f(name, totalSize, binariesPointers.data(), NULL); + cl_int err = f(name, numBinaries * sizeof(unsigned char*), binariesPointers.data(), NULL); if (err != CL_SUCCESS) { return err; @@ -1534,6 +1578,7 @@ struct ReferenceHandler<cl_event> }; +#if CL_HPP_TARGET_OPENCL_VERSION >= 120 && CL_HPP_MINIMUM_OPENCL_VERSION < 120 // Extracts version number with major in the upper 16 bits, minor in the lower 16 static cl_uint getVersion(const vector<char> &versionInfo) { @@ -1554,12 +1599,11 @@ static cl_uint getVersion(const vector<char> &versionInfo) return (highVersion << 16) | lowVersion; } -#if CL_HPP_TARGET_OPENCL_VERSION >= 120 && CL_HPP_MINIMUM_OPENCL_VERSION < 120 static cl_uint getPlatformVersion(cl_platform_id platform) { size_type size = 0; clGetPlatformInfo(platform, CL_PLATFORM_VERSION, 0, NULL, &size); - + vector<char> versionInfo(size); clGetPlatformInfo(platform, CL_PLATFORM_VERSION, size, versionInfo.data(), &size); return getVersion(versionInfo); @@ -3414,7 +3458,7 @@ public: context_(), SVMTrait::getSVMMemFlags(), size*sizeof(T), - sizeof(T)); + 0); pointer retValue = reinterpret_cast<pointer>( voidPointer); #if defined(CL_HPP_ENABLE_EXCEPTIONS) @@ -3537,7 +3581,7 @@ template <class T, class Alloc, class... Args> cl::pointer<T, detail::Deleter<Alloc>> allocate_pointer(const Alloc &alloc_, Args&&... args) { Alloc alloc(alloc_); - static const size_t copies = 1; + static const size_type copies = 1; // Ensure that creation of the management block and the // object are dealt with separately such that we only provide a deleter @@ -6519,7 +6563,7 @@ inline cl_int cl::Program::getInfo(cl_program_info name, vector<vector<unsigned // Resize the parameter array and constituent arrays param->resize(numBinaries); - for (int i = 0; i < numBinaries; ++i) { + for (size_type i = 0; i < numBinaries; ++i) { (*param)[i].resize(sizes[i]); } @@ -7106,7 +7150,7 @@ public: size_type buffer_slice_pitch, size_type host_row_pitch, size_type host_slice_pitch, - void *ptr, + const void *ptr, const vector<Event>* events = NULL, Event* event = NULL) const { @@ -7254,7 +7298,7 @@ public: const array<size_type, 3>& region, size_type row_pitch, size_type slice_pitch, - void* ptr, + const void* ptr, const vector<Event>* events = NULL, Event* event = NULL) const { @@ -7844,7 +7888,7 @@ public: CL_EXT_PREFIX__VERSION_1_2_DEPRECATED cl_int enqueueTask( const Kernel& kernel, const vector<Event>* events = NULL, - Event* event = NULL) CL_EXT_SUFFIX__VERSION_1_2_DEPRECATED const + Event* event = NULL) const CL_EXT_SUFFIX__VERSION_1_2_DEPRECATED { cl_event tmp; cl_int err = detail::errHandler( @@ -8872,7 +8916,7 @@ inline cl_int enqueueWriteBufferRect( size_type buffer_slice_pitch, size_type host_row_pitch, size_type host_slice_pitch, - void *ptr, + const void *ptr, const vector<Event>* events = NULL, Event* event = NULL) { @@ -8970,7 +9014,7 @@ inline cl_int enqueueWriteImage( const array<size_type, 3>& region, size_type row_pitch, size_type slice_pitch, - void* ptr, + const void* ptr, const vector<Event>* events = NULL, Event* event = NULL) { |