Mode 2 :hyPACK-2013 OpenCL (Open Computing Language)

The OpenCL Architecture

OpenCL provides a uniform programming environment for software developers to write efficient, portable code for high-performance compute servers, desktop computer systems and hand-held devices using a diverse mix of multi-core CPUs, GPUs, Cell-type architectures and other parallel processors such as DSPs. OpenCL is a framework for parallel programming and includes a language, API, libraries, and runtime system to support software development. Using OpenCL, a programmer can write general purpose programs that execute on GPUs without need to map their algorithms onto a 3D graphics API such as OpenGL.

The OpenCL architecture model use hierarchy of models such as platform Model, Memory Model, Execution Model, and Programming model. The important points are described below.

• The Platform model consists of a host connected to one or more OpenCL devices.

• An OpenCL device is divided into one or more compute units (CUs), which are further divided into one or more processing elements (PEs). Computations on a device occur within the processing elements.

• An OpenCL application runs on a host according to the models, particular to the host platform.

• The OpenCL application submits commands from the host to execute computations on the processing elements within a device.

• The processing elements within a compute unit execute a single stream of instructions as SIMD units or SPMD units.

• Execution of an OpenCL program occurs in two parts: a host program that executes on the particular host platform and kernels that execute on one or more OpenCL devices.

• The core of the OpenCL execution model is defined by the kernels execute. The concepts of kernel instance are called a work-item and these work-items are organized into Work-groups.

• Execution Model: Context and Command Queues - The host defines a context for the execution of the kernels. The context includes

Devices : The collection of OpenCL devices to be used by the host;
Kernels The OpenCL functions that run on OpenCL devices);
Program Objects: The program source and executable that implement the kernels);
Memory Objects : A set of memory objects visible to the host and the OpenCL devices Memory objects contain values that can be operated on by instances of a kernel.

• The OpenCL execution model supports two categories of kernels: OpenCL Kernels & Native Kernels.

• Work-Item(s) executing a kernel have access to four distinct memory regions. Global Memory; Constant Memory; Local Memory; & Private Memory. In OpenCL, The kernel or the host can allocate from a memory region, which depends upon the type of allocation and the type of access allowed.

• Table describes whether the kernel or the host from a memory region, the type of allocation (static i.e. compile time versus dynamic time i.e. runtime) and the type of access allowed i.e. whether the kernel or the host can read and/or write to a memory region.

• The application running on the host uses the OpenCL API to create memory objects in global memory, and to enqueue memory commands (Refer OpenCL API specification) that operate on those memory objects.

• OpenCL uses a relaxed consistency memory model: i.e the state of memory visible to a work item is not guaranteed to be consistent across the collection of work-items at all times.

• The OpenCL Framework allows applications to use host and one or more OpenCL devices as a single heterogeneous parallel computer system. The framework contains the components OpenCL Platform layer; OpenCL Runtime; & OpenCL Compiler

The OpenCL Platform Layer

• The OpenCL platform layer which implements platform specific features that allow applications to query OpenCL device configuration information and to create OpenCL contexts using one or more devices.

Querying Platform Info

Querying Devices

Contexts

The OpenCL Runtime

Command Queues

Memory Objects


- Creating Buffer Objects,
- Reading, Writing and copying Buffer Objects
- Retaining and Releasing Memory Objects
- Creating Image Objects,
- Querying List of Supported Image formats,
- Reading, Writing and Copying Image objects
- Copying between Image and Buffer Objects
- Mapping and Unmapping Memory Objects
- Memory Object Queries

Sampler Objects


Program Objects


- Creating Program Objects
- Building Program Executables


Build Options


- Options (Preprocessor, Math Intrinsic, Optimization)
- Uploading the OpenCL compiler
- Program Object Queries


Kernel Objects


- Creating Kernel Objects
- Setting Kernel Arguments
- Kernel Object Queries


Executing Kernels


Event Objects


Profiling Operations on Memory Objects and Kernels


Flush and Finish

The OpenCL Compiler : Building & Running Programs

The Compiler tool-chain provides a common framework for both CPUs & GPUs, sharing the front-end and some high-level compiler transformations. The back-ends are optimized for the device type (CPU or GPU). Most of the application remains same, but OpenCL APIs are included at various parts of the code. The kernels are compiled by the OpenCL compiler to either CPU binaries or GPU binaries, depending on that target device.

• CPU Processing :For CPU processing, the OpenCL runtime uses the LLVM AS ( Low-level virtual Machine ) to generate x86 binaries. The OpenCL runtime automatically determines the number of processing elements or cores, present in the CPU and distributes the OpenCL kernel between them.

• GPU Processing : For GPU processing, the OpenCL runtime layer generates GPU specific AMD -ATI binaries with CAL or CUDA enabled NVIDIA architecture GPU binaries.

Compiling Program

An OpenCL application consists of a host program (C/C++) and an optional kernel program (.cl). To compile an OpenCL application, the host program must be compiled and this can be done using the off-the-shelf compiler such as g++ or MSV++. The application kernels are compiled into device-specific binaries using the OpenCL compiler. The compiler uses a standard C front-end as well as the LLVM framework, with extensions for OpenCL.

To compile OpenCL applications on Windows requires that Visual Studio 2008 Professional Edition and the Intel C compiler and all C++ files must be added with appropriate settings. To compile OpenCL applications on Linux requires that the gcc or the Intel C compiler is installed and all C++ files must be compiled with appropriate settings on 32-bit /64-bit systems.

The OpenCL Library and runtime environment depends upon the target GPU (i.e CUDA enabled NVIDIA or AMD ATI - Stream DSK).

Running Program

An OpenCL application is compiled on the target system, the runtime system assigns the work in the command queues to the underling devices. Commands are placed into the queue using the clEnqueue commands shown below. The commands can be broadly classified into three categories.

• kernel commands (for example, clEnqueueNDRangeKernel(), etc.).

• Memory commands (for example, clEnqueueNDReadBuffer(), etc.), and

• Memory commands (for example, clEnqueueWaitForEvents(),etc.

An OpenCL application can create multiple command queues and please refer OpenCL specification or OpenCL Programming Guide for the CUDA Architecture or AMD ATI Stream computing OpenCL Programming Guide.

How an OpenCL application is built ? :

• First querying the runtime to determine which platforms are present. There can be any number of different OpenCL implementations are present.

• Create a context (The OpenCL Context has associated with it a number of compute devices such as CPU or GPU devices)

  Within a context, OpenCL guarantees a relaxed consistency between these devices. This means that memory objects, such as buffers or images, are allocated per context, but changes made by one device are only guaranteed to be visible by another device at well-defined synchronization points.

• OpenCL provides events, with the ability to synchronization on a given event to enforce the correct order of execution,

• Many operations are performed with respect to a given context: there also are many operations there are specific to a device. For example, program compilation and kernel execution are done on a peer-device basis.

• Performing work with a device, such as executing kernels or moving data to end from the device's local memory is done using a corresponding a command queue.

• A command queue is associated with a single device and a given context. : all work for a specific device is done through this interface. Note that while a single command queue can be associated with only a single device. For example, it is possible to have one command queue for executing kernels and a command kernel for managing data transfers between the host and the device.

Most OpenCL program follows the same pattern. Given a specific platform, select a device or devices to create a context, allocate memory, create device-specific command queues to create a context, allocate memory, create device-specific command queues, and perform data transfers & computations.

Generally, the platform is the gateway to accessing specific devices, given these devices and a corresponding context the application is independent of the platform. Given a context, the application can:

• Create a command queues

• Create programs to run on one or more associated devices

• Create kernels within those programs

• Allocate memory buffers or image, either on the host or on the device(s) (memory can be copied between the host and device)

• Write data to the device

• Submit the kernel (with appropriate arguments) to the command queue for execution.

• Read data back to the host form the device.

The relationship between context(s), device(s), buffer(s), program(s), kernel(s), and command queue(S) is best seen by looking at simple code.

An Overview of Basic Programming Steps :

Given below, illustrate the basic programming steps required for a minimum amount of code. Many test programs might require similar steps and these steps do not include error checks.