- Open CL
-
OpenCL
OpenCL (Open Computing Language) est la combinaison d'une API et d'un langage de programmation dérivé du C, proposé comme un standard ouvert par le Khronos Group. L'objectif d'OpenCL est de faciliter la programmation de l'intersection naissante entre le monde des CPU et des GPU, les premiers étant de plus en plus parallèles, les seconds étant de plus en plus programmables. Dans un premier temps, OpenCL permettra aux applications nécessitant des calculs lourds de tirer parti de la puissance des circuits graphiques. Mais OpenCL vise aussi à tirer partie de CPU multi-coeurs, des multi-processeurs tels le CELL d'IBM, qui équipe notamment la Playstation 3 de Sony, ou d'autres systèmes de calcul intensifs. Par ses objectifs techniques, OpenCL se rapproche de C pour CUDA, modèle de programmation propriétaire de la société NVIDIA.
Sommaire
Historique
OpenCL a été initialement conçu par Apple[1] (qui en possède les droits d'auteur), et affiné dans le cadre d'une collaboration avec AMD, Intel et NVIDIA. Apple soumit d'abord sa proposition initiale au Khronos Group. Le 16 juin 2008, le Khronos Compute Working Group fut formé, comprenant des représentants des fabricants de matériel informatique et de logiciels. Il travailla cinq mois durant à boucler les détails technique de la spécification OpenCL 1.0. La spécification fut révisée par les membres de Khronos et approuvée pour une version d'essai le 8 décembre.
OpenCL est intégré dans Mac OS X 10.6[1] (Snow Leopard).
AMD a décidé de supporter OpenCL et DirectX 11 plutôt que Close to Metal dans son framework Stream SDK. RapidMind a annoncé l'adoption de OpenCL sous sa plate-forme de développement, afin de supporter les processeurs graphiques de plusieurs fabricants avec une interface unique. NVIDIA a également confirmé le 9 décembre 2008 le support complet de la spécification 1.0 dans son GPU Computing Toolkit[2].
Exemple
Cet exemple calcule une Transformée de Fourier rapide
// create a compute context with GPU device context = clCreateContextFromType(0, CL_DEVICE_TYPE_GPU, NULL, NULL, NULL); // 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);
Le calcul : (basé sur Fitting FFT onto the G80 Architecture)[3]
// 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); }
Références
- ↑ a et b Apple Previews Mac OS X Snow Leopard to Developers, 2008-06-09, Apple
- ↑ Communiqué de presse de NVIDIA
- ↑ Fitting FFT onto G80 Architecture, May 2008, Vasily Volkov and Brian Kazian, UC Berkeley CS258 project report
Voir aussi
- Portail de l’informatique
Catégorie : Langage de programmation
Wikimedia Foundation. 2010.