OpenCL

OpenCL
Parte de Computación heterogénea
Información general
Tipo de programa	Interfaz de programación de aplicaciones, GPGPU
Autor	Apple
Desarrollador	Grupo Khronos
Lanzamiento inicial	28 de agosto de 2009
Licencia	Libre de derechos
Información técnica
Programado en	C++ C
Plataformas admitidas	x86-64 IA-32 ARM
Versiones
Última versión estable	1.2 (info) ( 15 de noviembre de 2011 (13 años y 17 días))
Enlaces
Sitio web oficial

OpenCL (Open Computing Language, en español lenguaje de computación abierto) consta de una interfaz de programación de aplicaciones y de un lenguaje de programación. Juntos permiten crear aplicaciones con paralelismo a nivel de datos y de tareas que pueden ejecutarse tanto en unidades centrales de procesamiento como unidades de procesamiento gráfico. El lenguaje está basado en C99, eliminando cierta funcionalidad y extendiéndolo con operaciones vectoriales.​

Apple creó la especificación original y la propuso al Grupo Khronos para convertirla en un estándar abierto y libre de derechos. El 16 de junio de 2008 Khronos creó el Compute Working Group​ para llevar a cabo el proceso de estandarización.

OpenCL forma parte de Mac OS X v10.6 ('Snow Leopard').​ AMD ha decidido apoyar OpenCL en lugar de su antigua API Close to Metal. ​ ​

Ejemplo

Este ejemplo calcula una Transformada rápida de Fourier.​ Las llamadas a la API son las siguientes:

// create a compute context with GPU device
context = clCreateContextFromType(CL_DEVICE_TYPE_GPU);

// create a work-queue
queue = clCreateWorkQueue(context, NULL, NULL, 0);

// allocate the buffer memory objects
memobjs[0] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(float)*2*num_entries, srcA);
memobjs[1] = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float)*2*num_entries, NULL);

// create the compute program
program = clCreateProgramFromSource(context, 1, &fft1D_1024_kernel_src, NULL);

// build the compute program executable
clBuildProgramExecutable(program, false, NULL, NULL);

// create the compute kernel
kernel = clCreateKernel(program, “fft1D_1024”);

// create N-D range object with work-item dimensions
global_work_size[0] = n;
local_work_size[0] = 64;
range = clCreateNDRangeContainer(context, 0, 1, global_work_size, local_work_size);

// set the args values
clSetKernelArg(kernel, 0, (void *)&memobjs[0], sizeof(cl_mem), NULL);
clSetKernelArg(kernel, 1, (void *)&memobjs[1], sizeof(cl_mem), NULL);
clSetKernelArg(kernel, 2, NULL, sizeof(float)*(local_work_size[0]+1)*16, NULL);
clSetKernelArg(kernel, 3, NULL, sizeof(float)*(local_work_size[0]+1)*16, NULL);

 // execute kernel
clExecuteKernel(queue, kernel, NULL, range, NULL, 0, NULL);

El cómputo en sí es éste:

// This kernel computes FFT of length 1024. The 1024 length FFT is decomposed into 
// calls to a radix 16 function, another radix 16 function and then a radix 4 function 
__kernel void fft1D_1024 (__global float2 *in, __global float2 *out, 
                          __local float *sMemx, __local float *sMemy) { 
  int tid = get_local_id(0); 
  int blockIdx = get_group_id(0) * 1024 + tid; 
  float2 data[16]; 
  // starting index of data to/from global memory 
  in = in + blockIdx;  out = out + blockIdx; 
  globalLoads(data, in, 64); // coalesced global reads 
  fftRadix16Pass(data);      // in-place radix-16 pass 
  twiddleFactorMul(data, tid, 1024, 0); 
  // local shuffle using local memory 
  localShuffle(data, sMemx, sMemy, tid, (((tid & 15) * 65) + (tid >> 4))); 
  fftRadix16Pass(data);               // in-place radix-16 pass 
  twiddleFactorMul(data, tid, 64, 4); // twiddle factor multiplication 
  localShuffle(data, sMemx, sMemy, tid, (((tid >> 4) * 64) + (tid & 15))); 
  // four radix-4 function calls 
  fftRadix4Pass(data); fftRadix4Pass(data + 4); 
  fftRadix4Pass(data + 8); fftRadix4Pass(data + 12); 
  // coalesced global writes 
  globalStores(data, out, 64); 
}

Véase también

• GPGPU
• CUDA
• Close to Metal
• BrookGPU
• Lib Sh
• Larrabee (GPU)

Enlaces externos

• Sitio web oficial de OpenCL
• OpenCL: What you need to know - artículo publicado en Macworld, agosto del 2008