OpenCL
OpenCL (engl. Open Computing Language) on Khronos Groupin kehittämä rinnakkaisohjelmointiin tarkoitettu ohjelmistokehys, jonka tarkoitus on mahdollistaa ohjelmistojen hajauttaminen erilaisille laskenta-alustoille verrattain helposti.
OpenCL | |
---|---|
Kehittäjä | Khronos Group |
Kehityshistoria | |
Vakaa versio | 3.0 / 30. syyskuuta 2020[1] |
Tiedot | |
Lisenssi | Avoimen lähdekoodin lisenssi |
Aiheesta muualla | |
Verkkosivusto |
OpenCL:ää voi käyttää muun muassa moniytimisillä suorittimilla, grafiikkaprosessoreilla ja digitaalisilla signaaliprosessoreilla heterogeenisessä ympäristössä ja sitä voidaan käyttää korkean tason ohjelmointikielistä sekä rajapinnoista.[2]
OpenCL koostuu kohdelaitteilla suoritettavista ydinfunktioista (kernel) sekä isäntälaitteella suoritettavista ohjauskäskyistä. Isäntälaitteet ovat tyypillisesti tavanomaisia CPU-laitteita, joissa suoritettava koodi voidaan kirjoittaa eri ohjelmointikielillä, kunhan ne tukevat OpenCL APIa (esimerkiksi C/C++, C#, Java, Fortran, Python, ...). OpenCL on pyritty suunnittelemaan siten, että kohdealustana voivat toimia hyvinkin erilaiset suoritusympäristöt, kuten CPU, GPU, DSP,... Rinnakkaistaminen on mahdollista toteuttaa myös useammalla alustalla yhtä aikaa (esim NxCPU + MxGPU). OpenCL skaalautuu aina sulautetuista alustoista supertietokoneisiin. Suurimpia laitevalmistajia jotka tukevat OpenCL-rajapintaa ovat esimerkiksi NVIDIA, AMD, Apple ja Intel.
Kernel-sovellukset kehitetään OpenCL C -kielellä, joka perustuu C99-kieleen ja sisältää laajennuksia, jotka helpottavat rinnakkaissuorituksen hallittavuutta. Tällaisia laajennuksia ovat esimerkiksi osoiteavaruuden indeksointi globaalisti sekä lokaalisti rinnakkaistetussa tehtävälohkossa.
OpenCL-alustoista huomattava osa koostuu erilaisista grafiikkasuorittimista ja siksi on kieleen lisätty myös liittymärajapinnat sekä OpenGL- että DirectX-muistiobjekteihin. Käytännössä tämä näkyy siten, että esimerkiksi OpenGL-tekstuureita voidaan käsitellä suoraan OpenCL-kerneleillä.
OpenCL mahdollistaa sekä dataperusteisen rinnakkaistamisen että tehtävälähtöisen rinnakkaistamisen. Datalähtöisessä ohjelmointimallissa kernel-funktiot suorittavat samaa käskysarjaa eri tietolohkoilla yhtä aikaa (SIMD). Tehtävälähtöisessä tapauksessa työ voidaan jakaa eri kerneleihin, joiden työryhmiä suoritetaan yhtä aikaa (task parallel). Kernel-työryhmien suoritusta voidaan synkronoida tehtäväjonoilla ja näihin syötettävillä synkronointikäskyillä.
SPIR ja SPIR-V
muokkaaKohdelaitteiden suuren määrän ja vaihtelevuuden vuoksi OpenCL -ohjelmakoodi käännetään SPIR-välikielelle (engl. Standard Portable Intermediate Representation). SPIR käännetään kohdelaitteella alustakohtaiselle käskykannalle ja arkkitehtuurille. SPIR voidaan verrata myös HSAIL-käännökseen.[3] SPIR-V on tarkoitettu Vulkan-rajapinnan kanssa käytettäväksi ja pääasiassa tuotetaan korkeamman tason kielistä kuten GLSL ja HLSL.[4]
SPIR-V on avoin alustariippumaton standardi.[5]
Standardiversiot
muokkaaEnsimmäinen versio OpenCL 1.0 julkaistiin joulukuussa 2008.
Kesäkuussa 2010 esiteltiin laajennus 1.1, jota useat laitevalmistajat tukevat.
Spesifikaatiot laajennukseen OpenCL 1.2 esiteltiin marraskuussa 2011.
Tammikuussa 2017 on julkaistu OpenCL 2.1.[6]
Syyskuussa 2020 on julkaistu versio 3.0.[1] Versio 3.0 palasi aiempaan siten, että ydin osuus koostuu OpenCL 1.2:sta ja tähän on lisätty mahdollisuus kysyä ominaisuuden tukea.[7] Ydinosuuden päälle tulevat OpenCL 2.0:n ja 3.0:n uudet ominaisuudet, jotka ovat valinnaisia ja on mahdollista valikoida mitä ominaisuuksia tuetaan.[7]
Esimerkki
muokkaaTämä esimerkki toteuttaa Fourier’n muunnoksen (FFT).
// 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);
Todellinen laskelma:
// 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);
}
Katso myös
muokkaaLähteet
muokkaa- Aaftab Munshi et al. OpenCL Programming Guide, Addison-Wesley, 2012. ISBN 0-321-74964-2.
Viitteet
muokkaa- ↑ a b OpenCL 3.0 Specification Finalized and Initial Khronos Open Source OpenCL SDK Released khronos.org. 30.9.2020. Viitattu 12.10.2020. (englanniksi)
- ↑ https://www.khronos.org/opencl/
- ↑ Hindriksen, Vincent: OpenCL SPIR by example streamcomputing.eu. Viitattu 16.2.2017.
- ↑ HLSL in Vulkan There and Back Again (PDF) Khronos Group. Viitattu 12.10.2020. (englanniksi)
- ↑ The first open standard intermediate language for parallel compute and graphics Khronos Group. Viitattu 7.3.2017.
- ↑ The open standard for parallel programming of heterogeneous systems Khronos Group. Viitattu 21.1.2017.
- ↑ a b Ryan Smith: Khronos Announces OpenCL 3.0: Hitting the Reset Button on Compute Frameworks anandtech.com. 27.4.2020. Viitattu 9.2.2022. (englanniksi)