Multi-threaded Programming
Contents
Introduction
OpenCL is a standard which defines a framework, an API and a programming language for parallel computation on heterogeneous systems like client computer systems, high- performance computing servers as well as hand-held devices.
The standard is maintained by the Khronos Group and supported by a large consortium of industry leaders including Apple, Intel, AMD, NVIDIA and ARM. Influenced by NVIDIA’s CUDA from the GPU side and by OpenMP which originates from the classical CPU side, the open OpenCL standard is characterized by a formulation which is abstract enough to support both CPU and GPU computing resources.
This is an ambitious goal, since providing an abstract interface together with a peak performance is a challenging task. OpenCL employs a strict isolation of the computation work into fundamental units, the kernels. These kernels can be developed in the OpenCL C programming language, a subset of the C99 language, with some additional OpenCL specific keywords.
In general, these kernels are hardware independent and compiled by the OpenCL runtime when they are loaded. To be able to fully exploit the parallel execution of the kernel code, several kernel instances, the work items, are started to process a set of input values. The actual number of concurrently running work items is determined by the OpenCL system. How a concrete algorithm can be partitioned into work items has to be decided by the programmer.
Reference Material
Khronos Group OpenCL
OpenCL 1.2 Quick Reference Card
http://www.khronos.org/files/opencl-1-2-quick-reference-card.pdf
OpenCL 1.2 Full Documentation
Intel SDK for OpenCL
http://software.intel.com/en-us/articles/vcsource-tools-opencl-sdk/
AMD OpenCL Zone
NVIDIA OpenCL
Programming Environment and Libraries
c++ interface / blah
Project: Boostraping OpenCL and Vector Addition
TODO: give overview of what to do
Compiling and running the test program
Open the folder project_vectoradd, create the build files using CMake and compile the application.
$ cd project_vectoradd/ [hauth@vdt-corei7avx project_vectoradd]$ cmake . [hauth@vdt-corei7avx project_vectoradd]$ make [100%] Built target vectoradd [hauth@vdt-corei7avx project_vectoradd]$ ./vectoradd
Once the application was compiled successully run it. The output should be along the following lines:
$ ./vectoradd Testing Platform : Intel(R) OpenCL > Selected Compute Device : Intel(R) Core(TM) i7-3930K CPU @ 3.20GHz Transferring data to device memory took 2e-06 s Running vectorAdd kernel took 2e-06 s Transferring data to host memory took 0 s All done
Task: Understand the existing code
The first task is to read and understand the existing source code. To do so, you have to edit the file vectoradd.cpp in your favorite text editor, use nano if you are not sure which tool to use.
$ nano vectoradd.cpp
Take your time to familiarize yourself with the sourcecode which is already in the file. Some of the intial steps of setting up the OpenCL system are already provided:
OpenCL compute context
An OpenCL platform is automaticaly selected, depending on the required device type (CPU/GPU) and a compute context is created. You can change the required device type by modifying the constant devType:
// Desired Device type. // can be CL_DEVICE_TYPE_GPU or CL_DEVICE_TYPE_CPU in this example const cl_device_type devType = CL_DEVICE_TYPE_CPU;
OpenCL command queue
A command queue is created on the context
Kernel compile
The variable contains kernelSourceAdd the source code of the OpenCL kernel. This string ist used to compile and register the kernel with the OpenCL runtime system. If you want to extend the kernel code, you have to add to the string contained in the variable kernelSourceAdd.
Host data buffers
The variables hostVectorA, hostVectorB and hostVectorC hold buffers which will be used later to transfer data to and from the OpenCL kernel. The underlying C++ type is std::vector which can be conviniently used to acquire a contingious space of memory in host’s RAM.
FloatingPoint vp = 23.0f; FloatingPointVector hostVectorA(itemCount, vp); FloatingPointVector hostVectorB(itemCount, vp); FloatingPointVector hostVectorC(itemCount, vp);
Various timing measurements
To quantify the runtime of the OpenCL operations, various timing objects are present. For the time measurements to be correct, only the operations indicated by the TODO comments must be between the creation of the Timing object and the call to EndWithReport() of the respective timing object.
Timing t_transfer_input("Transferring data to device memory"); // TODO: transfer input buffer here t_transfer_input.EndWithReport();
Task: Running the OpenCL Kernel
The OpenCL kernel is already created and associated with the compiled program:
cl::Kernel kernelAdd(pProgram, "vectorAdd");
Use the enqueueNDRangeKernel(…) method on the command queue object to add a kernel instance to the queue <ref>http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueNDRangeKernel.html </ref>
Call the finish() method on the command queue object to ensure the kernel is done processing before the host program continues execution. <ref>http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clFinish.html </ref>
Modifications to play around
- switch to double , how does the runtime change for CPU/GPU ?
- switch to float4 vector type. can you perform the same addition operations ?
Project: N-Body Simulation
<references />