OpenCL
OpenCL (Open Computing Language) é uma arquitetura para escrever programas que funcionam em plataformas heterogêneas, consistindo em CPUs, GPUs e outros processadores. OpenCL inclui uma linguagem (baseada em C99) para escrever kernels (funções executadas em dispositivos OpenCL), além de APIs que são usadas para definir e depois controlar as plataformas heterogênea. OpenCL permite programação paralela usando, tanto o paralelismo de tarefas, como de dados.
Ela foi adotada para controladores de placas gráficas pela AMD/ATI, que a tornou na sua única oferta de GPU como Stream SDK, e pela Nvidia, que oferece também OpenCL como a escolha para o seu Compute Unified Device Architecture (CUDA) nos seus controladores.
A arquitetura OpenCL partilha uma série de interfaces computacionais, tanto com CUDA, como com a concorrente DirectCompute da Microsoft.
A proposta OpenCL é similar às propostas OpenGL e OpenAL, que são padrões abertos da indústria para gráficos 3D e áudio, respectivamente. OpenCL estende o poder da GPU além do uso gráfico (GPGPU). OpenCL é gerido pelo consórcio tecnológico Khronos Group.
Histórico
[editar | editar código-fonte]OpenCL foi inicialmente desenvolvida pela Apple Inc., que tem direitos sobre a marca, e produziu uma proposta inicial em colaboração com equipes técnicas na AMD, IBM, Intel e Nvidia. Apple submeteu esta proposta inicial ao Khronos Group. Em 16 de junho de 2008, o Khronos Compute Working Group foi formado[1] com representantes de companhias fabricantes de CPU, GPU, embedded-processor e software. Este grupo trabalhou por cinco meses para finalizar os detalhes técnicos da especificação OpenCL 1.0 em 18 de novembro de 2008.[2]
Esta especificação técnica foi revisada por membros da Khronos e aprovada como release pública em 8 de dezembro de 2008.[3]
OpenCL 1.0 foi introduzida no Mac OS X v10.6 ('Snow Leopard') em 28 de agosto de 2008. De acordo com a nota de imprensa:[4]
Snow Leopard terá extensões para suporte à hardware moderno com Open Computing Language (OpenCL), que permite qualquer aplicação explorar os vastos gigaflops do poder de computação da GPU antes disponíveis somente para aplicações gráficas. OpenCL é baseado na linguagem de programação C e está sendo proposto como um padrão aberto.
AMD decidiu oferecer suporte a OpenCL (e DirectX 11) ao invés do agora obsoleto Close to Metal em seu Stream framework.[5][6]
RapidMind anunciou a adoção de OpenCL sob sua plataforma de desenvolvimento, para permitir suporte a GPUs de múltiplos fabricantes com uma única interface.[7]
Nvidia anunciou em 9 de dezembro de 2008 agregar suporte completo à especificação OpenCL 1.0 ao seu GPU Computing Toolkit.[8]
Nvidia anunciou em 13 de maio de 2009 aos desenvolvedores o "OpenCL 1.0 Conformance Candidate Release", é a primeira solução de compilação de códigos desenvolvidos em OpenCL.
Exemplo
[editar | editar código-fonte]O modelo de programação OpenCL é bastante semelhante ao modelo de programação CUDA de Nvidia. Abaixo temos os códigos fontes de dois núcleos que fazem a adição de dois vetores:
Código C para um núcleo CUDA:
__global__ void
vectorAdd(const float * a, const float * b, float * c)
{
// Indice do vetor
int nIndex = blockIdx.x * blockDim.x + threadIdx.x;
c[nIndex] = a[nIndex] + b[nIndex];
}
Código C para um núcleo OpenCL:
__kernel void
vectorAdd(__global const float * a,
__global const float * b,
__global float * c)
{
// Indice do vetor
int nIndex = get_global_id(0);
c[nIndex] = a[nIndex] + b[nIndex];
}
Código[9] C para o HOST CUDA:
Não existe ainda um compilador OpenCL
/* main routine that executes on the host */
int main(void)
{
float *a_h, *a_d; // Pointer to host & device arrays
const int N = 10; // Number of elements in arrays
size_t size = N * sizeof(float);
a_h = (float *)malloc(size); // Allocate array on host
cudaMalloc((void **) &a_d, size); // Allocate array on device
// Initialize host array and copy it to CUDA device
for (int i=0; i<N; i++) a_h[i] = (float)i;
cudaMemcpy(a_d, a_h, size, cudaMemcpyHostToDevice);
// Do calculation on device:
int block_size = 3;
int n_blocks = N/block_size + (N % block_size == 0 ? 0:1);
square_array <<< n_blocks, block_size >>> (a_d, N);
// Retrieve result from device and store it in host array
cudaMemcpy(a_h, a_d, sizeof(float)*N, cudaMemcpyDeviceToHost);
// Print results
for (int i=0; i<N; i++) printf("%d %f\n", i, a_h[i]);
// Cleanup
free(a_h); cudaFree(a_d);
}
Código[9] C para o HOST OpenCL:
# include <stdio.h>
# include <cl.h>
int main(void){ // main routine that executes on the host
float *a_h; // Pointer to host & device arrays
const int N = 10; // Number of elements in arrays
size_t size = N * sizeof(float);
a_h = (float *)malloc(size); // Allocate array on host
// Initialize host array and copy it to CUDA device
for (int i=0; i<N; i++) a_h[i] = (float)i;
// create the OpenCL context on a GPU device
cl_context context = clCreateContextFromType(0, // (must be 0)
CL_DEVICE_TYPE_ GPU, NULL, // error callback
NULL, // user data
NULL); // error code
// get the list of GPU devices associated with context
size_t cb;
clGetContextInfo(context, CL_CONTEXT_DEVICES, 0, NULL, &cb);
cl_device_id *devices = malloc(cb);
clGetContextInfo(context, CL_CONTEXT_DEVICES, cb, devices, NULL);
// create a command-queue
cl_cmd_queue cmd_queue = clCreateCommandQueue(context,
devices[0], 0, // default options
NULL); // error code
cl_mem memobjs[1];
// allocate input buffer memory objects
memobjs[0] = clCreateBuffer(context,
CL_MEM_READ_ONLY | // flags
CL_MEM_COPY_HOST_PTR,
sizeof(cl_float)*n, // size
a_h, // host pointer
NULL); // error code
// create the program
cl_program program = clCreateProgramWithSource(
context,
1, // string count
&program_source, // program strings
NULL, // string lengths
NULL); // error code
// build the program
cl_int err = clBuildProgram(program,
0, // num devices in device list
NULL, // device list
NULL, // options
NULL, // notifier callback function ptr
NULL); // error code
// create the kernel
cl_kernel kernel = clCreateKernel(program, “square_array”, NULL);
// set “a” vector argument
err = clSetKernelArg(kernel,
0, // argument index
(void *)&memobjs[0], // argument data
sizeof(cl_mem)); // argument data size
size_t global_work_size[1] = n; // set work-item dimensions
// execute kernel
err = clEnqueueNDRangeKernel(cmd_queue, kernel,
1, // Work dimensions
NULL, // must be NULL (work offset)
global_work_size,
NULL, // automatic local work size
0, // no events to wait on
NULL, // event list
NULL); // event for this kernel
// read output array
err = clEnqueueReadBuffer( context, memobjs[0],
CL_TRUE, // blocking
0, // offset
n*sizeof(cl_float), // size
a_h, // pointer
0, NULL, NULL); // events
// Print results
for (int i=0; i<N; i++) printf("%d %f\n", i, a_h[i]);
// Cleanup Host Memory
free(a_h);
}
- ↑ «Khronos Launches Heterogeneous Computing Initiative» (Nota de imprensa). Khronos Group. 16 de junho de 2008
- ↑ «OpenCL gets touted in Texas». MacWorld. 20 de novembro de 2008
- ↑ «The Khronos Group Releases OpenCL 1.0 Specification» (Nota de imprensa). Khronos Group. 8 de dezembro de 2008
- ↑ «Apple Previews Mac OS X Snow Leopard to Developers» (Nota de imprensa). Apple Inc. 9 de junho de 2008
- ↑ «AMD Drives Adoption of Industry Standards in GPGPU Software Development» (Nota de imprensa). AMD
- ↑ «AMD Backs OpenCL, Microsoft DirectX 11». eWeek
- ↑ «HPCWire: RapidMind Embraces Open Source and Standards Projects». HPCWire. Consultado em 10 de abril de 2009. Arquivado do original em 18 de dezembro de 2008
- ↑ «NVIDIA Adds OpenCL To Its Industry Leading GPU Computing Toolkit» (Nota de imprensa). Nvidia
- ↑ a b «GPU Modeling». Blogspot. 21 de abril de 2009