Geekly articles weekly

Contoh Pemrograman Akselerator FPGA



Belum lama ini kami berbicara tentang layanan Selectel baru - komputasi awan kinerja tinggi pada akselerator FPGA . Dalam artikel baru tentang hal ini, kami mempertimbangkan contoh pemrograman FPGA untuk membangun set Mandelbrot, sebuah algoritma matematika yang terkenal untuk memvisualisasikan gambar fraktal. Artikel tersebut menggunakan bahan dari situs Euler Project .



Alih-alih kata pengantar


Pertama, beberapa istilah. Sistem komputer dengan akselerator FPGA - sebagai aturan, ini adalah adaptor PCIe dengan chip FPGA sebagai bagian dari server x64. Akselerator mengambil tugas intensif sumber daya yang terpisah di mana komputasi paralel dapat dilibatkan dan melakukan banyak perintah lebih cepat dari prosesor x64, menurunkannya dan meningkatkan kinerja seluruh sistem komputasi. Misalnya, siklus perhitungan dengan 100 ribu repetisi dapat dilakukan pada FPGA hanya dalam satu pass, alih-alih mengeksekusi 100 ribu kali secara berurutan pada prosesor x64 klasik. Elemen logis, sumber daya perangkat keras, tautan komunikasi, chip FPGA diprogram oleh pengguna secara langsung untuk tugas itu sendiri, yang memungkinkan Anda untuk mengimplementasikan tugas sebagai implementasi algoritma dalam silikon - Algoritma dalam Silikon dan dengan demikian mencapai kinerja tinggi, dan dengan konsumsi daya yang sangat sederhana.

Saat ini, ambang untuk memasukkan teknologi FPGA cukup mudah diakses bahkan untuk startup - server dengan akselerator FPGA dan semua perangkat lunak yang diperlukan (SDK) dapat disewa di cloud Selectel dengan harga yang wajar (apa yang disebut "cloud FPGA"), dan dukungan untuk standar Open CL di FPGA mengarah ke bahwa seorang programmer yang tahu bagaimana bekerja dengan C dapat mempersiapkan dan menjalankan program pada FPGA.

Ke depan: coba FPGA di tempat kerja


Contoh pemrograman yang dijelaskan di bawah ini untuk membangun set Mandelbrot telah diterapkan pada server uji di Selectel Lab , di mana siapa pun dapat mengevaluasi kinerjanya (pendaftaran akan diperlukan).


Proyek disediakan dalam kode dan disiapkan untuk kompilasi. Selectel menawarkan akses jarak jauh ke server dengan akselerator Intel Arria 10 FPGA. Di sisi server, alat SDK dan BSP untuk pengembangan, debugging dan kompilasi OpenCL, kode Visual Studio untuk mempersiapkan aplikasi host (aplikasi kontrol untuk prosesor pusat server) dikerahkan.
Perhatikan bahwa contoh itu sendiri tidak memiliki nilai yang diterapkan, itu dipilih karena alasan menunjukkan metode percepatan menggunakan prinsip-prinsip paralelisme. Dengan contoh ini, pembaca berkenalan dengan rute merancang aplikasi dalam sistem komputasi heterogen dengan FPGA - nantinya rute ini dapat digunakan untuk mengembangkan aplikasi Anda sendiri dengan komputasi paralel.
UPDATE : Pada musim semi 2018, Intel memperkenalkan prosesor hybrid kinerja tinggi Xeon Gold 6138P dengan chip Arria 10 FPGA terintegrasi. Pada akhir 2018, prosesor serial jenis ini diharapkan tersedia untuk pelanggan melalui mitra Intel. Kami di Selectel menantikan chip ini, dan berharap untuk menjadi yang pertama di Rusia untuk memberikan pelanggan kami kesempatan untuk menguji produk baru yang unik ini.

Tentang Standar OpenCL untuk Pemrograman FPGA


Standar OpenCL dikembangkan oleh Khronos Group, produsen chip dan perangkat lunak terkemuka di dunia yang terdiri dari Intel, AMD, Apple, ARM, Nvidia, Sony Computer Entertainment, dan lainnya, dirancang untuk menulis aplikasi yang menggunakan komputasi paralel pada berbagai jenis prosesor, termasuk FPGA. Standar OpenCL mencakup bahasa pemrograman C yang didasarkan pada versi bahasa C99 (versi terbaru C99 adalah ISO / IEC 9899: 1999 / Cor 3: 2007 dari 2007-11-15) dan lingkungan pemrograman aplikasi.

Popularitas menggunakan OpenCL untuk komputasi berkinerja tinggi didasarkan pada kenyataan bahwa itu adalah standar terbuka, dan penggunaannya tidak memerlukan lisensi. Selain itu, OpenCL tidak membatasi jangkauan perangkat yang didukung untuk merek tertentu, memungkinkan penggunaan perangkat keras dari berbagai produsen pada platform perangkat lunak yang sama.

Selain itu tentang OpenCL: Pengantar OpenCL tentang Habr .


Sedikit sejarah - rute desain FPGA yang ada sebelum standar OpenCL sangat spesifik dan memakan waktu, sementara dalam hal kompleksitas bahkan lebih unggul daripada desain chip kustom (ASIC, sirkuit terintegrasi khusus aplikasi, "sirkuit terintegrasi tujuan khusus"). Pemahaman yang mendalam tentang struktur perangkat keras FPGA diperlukan, konfigurasi yang harus dilakukan dalam bahasa deskripsi perangkat keras tingkat rendah (HDL). Kepemilikan desain dan rute verifikasi ini telah dan masih menjadi seni yang, karena kompleksitasnya yang ekstrem, tersedia untuk kalangan pengembang terbatas.

Munculnya toolkit dukungan OpenCL Intel untuk FPGA telah sebagian membahas masalah aksesibilitas pemrograman FPGA untuk pengembang perangkat lunak. Programmer secara independen memilih bagian dari algoritma yang cocok untuk pemrosesan paralel dan menjelaskannya dalam C, kemudian kompiler Intel OpenCL untuk FPGA membuat file konfigurasi biner untuk menjalankan fragmen algoritma ini pada akselerator.
Menggunakan lingkungan Visual Studio biasa atau kompiler gcc standar, aplikasi host disiapkan (aplikasi tipe .exe, dieksekusi pada prosesor x64 utama), sementara semua pustaka dukungan yang diperlukan dimasukkan dalam SDK. Ketika aplikasi host diluncurkan, firmware FPGA dimuat, data akan dimuat ke dalam inti chip dan pemrosesan akan dimulai sesuai dengan algoritma yang dikandung.

FPGA (FPGA) adalah struktur perangkat keras paralel masif yang dapat diprogram ulang pengguna dengan jutaan elemen logika, ribuan unit sinyal DSP, dan puluhan megabita cache untuk perhitungan di papan, tanpa mengakses modul memori utama server. Antarmuka I / O yang cepat (10GE, 40GE, 100GE, PCIe Gen 3, dll.) Memungkinkan Anda untuk bertukar data secara efektif dengan prosesor utama server.


Standar OpenCL adalah lingkungan untuk menjalankan perangkat lunak heterogen. Lingkungan terdiri dari dua bagian yang terpisah:

  1. Perangkat lunak host - aplikasi yang berjalan pada prosesor pusat utama server, ditulis dalam C / C ++ dan menggunakan serangkaian fungsi OpenCL API. Server host mengatur seluruh proses komputasi, memasok sumber dan menerima data keluaran, dan berinteraksi dengan semua sistem server dengan akselerator FPGA.
  2. Perangkat lunak akselerator - program yang ditulis dalam bahasa OpenCL C (bahasa C dengan sejumlah batasan), dikompilasi untuk dijalankan pada chip FPGA.

Server khas untuk komputasi paralel adalah komputer berbasis x64 (untuk menjalankan aplikasi host), yang mencakup akselerator FPGA perangkat keras, paling sering terhubung melalui bus PCI-Express. Omong-omong, sistem seperti itu disajikan di Selectel Lab.

Urutan pemrograman dan kompilasi untuk akselerator FPGA terdiri dari dua tahap. Kode aplikasi host dikompilasi oleh kompiler standar (Visual C ++, GCC) untuk mendapatkan file yang dapat dieksekusi di sistem operasi server (misalnya, * .exe). Kode sumber akselerator FPGA (kernel, kernel) disiapkan oleh kompiler AOC sebagai bagian dari SDK, dengan penerimaan file biner (* .aocx). File ini hanya untuk pemrograman akselerator.

Arsitektur lingkungan kompilasi perangkat lunak OpenCL
Fig. Arsitektur lingkungan kompilasi perangkat lunak OpenCL

Pertimbangkan beberapa contoh kode untuk menghitung vektor besar dengan dua cara
( PS Jangan tembak pianis - selanjutnya kode dari situs Euler Project digunakan ):

void inc (float *a, float c, int N) { for (int i = 0; i<N; i++) a[i] = a[i] + c; } void main() { ... inc(a,c,N); ... } 

 _kernel void inc (_global float *a, float c) { int i = get_global_id(0); a[i] = a[i] + c; } void main() { ... clEnqueueNDRangeKernel(...,&N,...) ... }

Kode di awal adalah contoh bagaimana implementasi single-threaded di C mungkin terlihat seperti menggunakan metode perhitungan sekuensial elemen skalar.

Versi kedua dari kode adalah implementasi yang mungkin dari algoritma pada OpenCL dalam bentuk fungsi yang dihitung pada akselerator FPGA. Tidak ada loop, dan perhitungan berlangsung dalam satu iterasi dari loop. Perhitungan array vektor terjadi sebagai eksekusi N salinan fungsi ini. Setiap salinan memiliki indeks sendiri, diganti ke dalam iterator dalam satu lingkaran, dan jumlah percobaan ulang diatur dari host ketika kode dieksekusi. Tindakan iterator disediakan oleh fungsi get_global_id (), yang berfungsi dengan indeks dalam 0 ≤ indeks <N.

Dapatkan ke titik: membangun fraktal


Set Mandelbrot adalah array titik "c" pada bidang kompleks yang relasinya berulang Zn + 1 = Zn² + c untuk Z0 = 0 mendefinisikan urutan yang dibatasi.

Kami mendefinisikan Zn = Zn + IYn, dan juga c = p + iq.
Untuk setiap titik, urutan berikut dihitung:

Xn + 1 = Xn² + Yn² + p
Yn + 1 = 2XnYn + q


Perhitungan milik titik ke set di setiap iterasi dilakukan sebagai persamaan
Xn² + Yn² <4.

Untuk menampilkan set Mandelbrot di layar, kami menetapkan aturan:

  1. Jika ketidaksetaraan berlaku pada setiap iterasi, maka titik memasuki set dan akan ditampilkan dalam warna hitam.
  2. Jika ketidaksetaraan tidak berlaku, dimulai dengan nilai iterasi tertentu n = N, maka warna ditentukan oleh jumlah iterasi N.

Proses perhitungan pada host adalah sebagai berikut:

  • Penghitungan jumlah iterasi untuk setiap titik di dalam jendela piksel ditetapkan ke fungsi mandel_pixel ().
  • Pencacahan berurutan poin gambar akan disediakan oleh fungsi softwareCalculateFrame (). Parameter menentukan interval nyata dari titik yang dihitung, langkah nyata dari algoritma, dan penunjuk ke buffer warna ukuran gambar (Lebar * theHeight).
  • Warna titik disesuaikan dengan theSoftColorTable.

Mari kita beralih ke kode:

 inline unsigned int mandel_pixel( double x0, double y0, unsigned int maxIterations ) { // variables for the calculation double x = 0.0; double y = 0.0; double xSqr = 0.0; double ySqr = 0.0; unsigned int iterations = 0; // perform up to the maximum number of iterations to solve // the current point in the image while ( xSqr + ySqr < 4.0 &&iterations < maxIterations ) { // perform the current iteration xSqr = x*x; ySqr = y*y; y = 2*x*y + y0; x = xSqr - ySqr + x0; // increment iteration count iterations++; } // return the iteration count return iterations; } 

 int softwareCalculateFrame( double aStartX, double aStartY, double aScale, unsigned int* aFrameBuffer ) { // temporary pointer and index variables unsigned int * fb_ptr = aFrameBuffer; unsigned int j, k, pixel; // window position variables double x = aStartX; double y = aStartY; double cur_x, cur_y; double cur_step_size = aScale; // for each pixel in the y dimension window for ( j = 0, cur_y = y; j < theHeight; j++, cur_y -= cur_step_size ) { // for each pixel in the x dimension of the window for ( cur_x = x, k = 0; k< theWidth; k++, cur_x += cur_step_size ) { // set the value of the pixel in the window pixel = mandel_pixel(cur_x, cur_y, theSoftColorTableSize); if ( pixel == theSoftColorTableSize ) *fb_ptr++ = 0x0; else *fb_ptr++ = theSoftColorTable[pixel]; } } return 0; }

Setiap piksel dihitung secara terpisah dari piksel lainnya, dan karenanya proses ini dapat diparalelkan. Saat menerapkan algoritma untuk akselerator FPGA, instruksi SIMD dibuat untuk menghitung angka untuk setiap piksel iterasi (menentukan kode warna dari palet). Implementasi dua loop bersarang pada buffer gambar dibingkai melalui OpenCL dengan menjalankan operasi (theWidth * theHeight).

Instance kernel dalam daftar di bawah ini disebut work-item, dan himpunan semua instance disebut ruang indeks. Fitur-fitur fungsi perangkat keras meliputi:

  • Deklarasi fungsi dimulai dengan kata kunci __kernel.
  • Jenis fungsi perangkat keras - jenis kembali selalu batal.
  • Nilai pengembalian dilakukan melalui buffer yang diteruskan sebagai parameter.
    • Tiga parameter pertama menentukan kisi material, node yang sesuai dengan piksel dari gambar output.
    • Parameter keempat membatasi jumlah iterasi, mencegah perulangan untuk titik-titik yang dimiliki set Mandelbrot.
    • Parameter kelima adalah pointer ke buffer warna output.
    • Kata kunci __global menunjukkan jenis memori yang akan digunakan buffernya: ini adalah memori DDR (QDR) umum pada akselerator itu sendiri.
    • Batasan kata kunci memberi optimizer larangan menggunakan referensi buffer tidak langsung.
    • Pada parameter ke-6, sebuah pointer ke palet dilewatkan.
    • Kata kunci __constant mengoptimalkan akses buffer dengan menghasilkan cache dengan atribut read-only.

    Deskripsi fungsi dalam daftar dekat dengan implementasi untuk prosesor x64. Di sini, definisi instance kernel saat ini dilakukan melalui fungsi get_global_id, di mana nomor dimensi (0, 1) diteruskan sebagai parameter.

    Untuk optimalisasi yang lebih baik, sebuah indikasi eksplisit dari awal siklus telah diperkenalkan. Dengan tidak adanya informasi tentang jumlah iterasi pada saat kompilasi, jumlah langkah loop ditunjukkan secara eksplisit, karena blok perangkat keras mereka sendiri akan dibuat untuk mereka. Dengan pengkodean semacam ini, seseorang harus “melihat ke belakang” pada kapasitas chip tertentu yang dipasang pada akselerator, karena konsumsi sumber daya FPGA untuk jumlah siklus yang lebih besar.

     //////////////////////////////////////////////////////////////////// // mandelbrot_kernel.cl : Hardware implementation of the mandelbrot algorithm //////////////////////////////////////////////////////////////////// // Amount of loop unrolling. #ifndef UNROLL #define UNROLL 20 #endif // Define the color black as 0 #define BLACK 0x00000000 __kernel void hw_mandelbrot_frame ( const double x0, const double y0, const double stepSize, const unsigned int maxIterations, __global unsigned int *restrict framebuffer, __constant const unsigned int *restrict colorLUT, const unsigned int windowWidth) { // Work-item position const size_t windowPosX = get_global_id(0); const size_t windowPosY = get_global_id(1); const double stepPosX = x0 + (windowPosX * stepSize); const double stepPosY = y0 - (windowPosY * stepSize); // Variables for the calculation double x = 0.0; double y = 0.0; double xSqr = 0.0; double ySqr = 0.0;</code> <code>unsigned #pragma while { int iterations = 0; // Perform up to the maximum number of iterations to solve // the current work-item's position in the image // The loop unrolling factor can be adjusted based on the amount of FPGA // resources available. unroll UNROLL xSqr + ySqr < 4.0 && iterations < maxIterations ) // Perform the current iteration xSqr = x*x; ySqr = y*y; y = 2*x*y + stepPosY; x = xSqr - ySqr + stepPosX; // Increment iteration count iterations++; } // Output black if we never finished, and a color from the look up table otherwise framebuffer[windowWidth * windowPosY + windowPosX] = (iterations == maxIterations) ? BLACK : colorLUT[iterations]; }

    Paket utilitas Intel FPGA SDK untuk OpenCL perlu diinstal pada host sebelum mengkompilasi implementasi algoritma dari perangkat keras. Di antara perangkat lunak yang telah diinstal sebelumnya, Anda harus menyertakan BSP (Board Support Package) dari produsen papan akselerator tertentu. Dalam contoh tersebut, Intel Quartus Prime Pro 16.1 diinstal dengan dukungan untuk OpenCL dan BSP dari akselerator Euler Thread (Intel Arria 10).

    Berikut ini adalah konfigurasi lintasan dan variabel lingkungan. Variabel ALTERAOCLSDKROOT berisi lintasan ke Intel FPGA SDK, variabel AOCL_BOARD_PACKAGE_ROOT berisi lintasan ke BSP akselerator.

     set ALTERAOCLSDKROOT=C:\intelFPGA_pro\16.1\hld set AOCL_BOARD_PACKAGE_ROOT=C:\intelFPGA_pro\16.1\hld\board\euler_thread set path=%path%;C:\intelFPGA_pro\16.1\hld\bin set path=%path%;C:\intelFPGA_pro\16.1\quartus\bin64 set path=%path%;C:\intelFPGA_pro\16.1\hld\board\a10_ref\windows64\bin set path=%path%;C:\intelFPGA_pro\16.1\hld\host\windows64\bin set path=%path%;C:\intelFPGA_pro\16.1\qsys\bin set path=%path%;C:\Program Files (x86)\GnuWin32\bin\

    Untuk kompilasi, aoc compiler dari SDK digunakan.

     aoc mandelbrot_kernel.cl -o mandelbrot_kernel.aocx --board thread -v -v --report

    Kami mendekripsi: mandelbrot_kernel.cl - file dengan teks sumber, mandelbrot_kernel.aocx - file objek output untuk pemrograman FPGA, utas - nama akselerator dari paket BSP. Switch --report menampilkan laporan penggunaan sumber daya FPGA. Sakelar –v menampilkan informasi diagnostik selama kompilasi. Laporan konsumsi sumber daya untuk kernel adalah sebagai berikut:

    + ------------------------------------------------- ------------------- +
    ; Ringkasan Penggunaan Sumber Daya yang Diperkirakan;
    + ---------------------------------------- + -------- ------------------- +
    ; Sumber Daya + Penggunaan;
    + ---------------------------------------- + -------- ------------------- +
    ; Pemanfaatan logika; 49%
    ; ALUT; 26%;
    ; Register logika khusus; 25%;
    ; Blok memori; 21%
    ; Blok DSP; 16%;
    + ---------------------------------------- + -------- -------------------;

    Untuk mengkompilasi aplikasi host, contoh menggunakan paket Microsoft Visual Studio 2010 Express dengan Microsoft SDK 7.1 diinstal. Dalam pengaturan proyek, konfigurasi untuk x64 dipilih. Selanjutnya, hubungkan folder untuk file header eksternal dan tentukan path ke pustaka Intel FPGA SDK tambahan di pengaturan linker.
    Direktori tambahan untuk menyertakan file = $ (ALTERAOCLSDKROOT) \ host \ include;
    Direktori perpustakaan tambahan = $ (AOCL_BOARD_PACKAGE_ROOT) \ windows64 \ lib;
     $(ALTERAOCLSDKROOT)\host\windows64\lib;

    Rencana umum tindakan untuk meluncurkan kernel pada akselerator adalah sebagai berikut:

    1. Dapatkan daftar platform
    2. Dapatkan daftar perangkat
    3. buat konteks;
    4. memuat kernel ke dalam perangkat;
    5. mengirim buffer input ke perangkat;
    6. jalankan kernel untuk dieksekusi;
    7. baca buffer output dari perangkat;
    8. konteks bebas.

    Pertimbangkan beberapa poin yang terkait langsung dengan peluncuran kernel. Jadi, satu inti dirancang untuk memproses satu piksel gambar. Jadi, Anda perlu menjalankan instance kernel N, di mana N adalah jumlah total piksel dalam gambar.

    Di bawah ini, kami mencatat kasus ketika ada beberapa papan akselerator di server, maka tugas dapat didistribusikan di antara mereka. Di setiap akselerator, Anda perlu memuat kernel (file mandelbrot_kernel.aocx). Misalkan jumlah akselerator adalah numDevices, dan garis gambar dibagi antara semua akselerator:

     #define MAXDEV 10 static cl_context theContext; static cl_program theProgram; static cl_kernel theKernels[MAXDEV]; //.. // Create the program object theProgram = createProgramFromBinary( theContext, "mandelbrot_kernel.aocx", theDevices, numDevices); // Create the kernels for ( unsigned i = 0; i < numDevices; ++i ) theKernels[i] = clCreateKernel( theProgram, "hw_mandelbrot_frame", &theStatus ); // Create output pixel buffers for every kernel for( unsigned i = 0; i < numDevices; ++i ) thePixelData[i] = clCreateBuffer(theContext, CL_MEM_WRITE_ONLY, thePixelDataWidth*rowsPerDevice[i]*sizeof(unsigned int), NULL, &theStatus); // Preparing and writing palette buffer to every device theHardColorTable = clCreateBuffer(theContext, CL_MEM_READ_ONLY, aColorTableSize*sizeof(unsigned int), NULL, &theStatus); for( unsigned i = 0; i < numDevices; i++ ) theStatus = clEnqueueWriteBuffer(theQueues[i], theHardColorTable, CL_TRUE, 0, aColorTableSize*sizeof(unsigned int), aColorTable, 0, NULL, NULL); // Preparing kernels and run unsigned rowOffset = 0; for ( unsigned i = 0; i < numDevices; rowOffset += rowsPerDevice[i++] ) { // Create ND range size size_t globalSize[2] = { thePixelDataWidth, rowsPerDevice[i] }; // Set the arguments unsigned argi = 0; theStatus = clSetKernelArg (theKernels[i], argi++, sizeof(cl_double), (void*) &aStartX ); const double offsetedStartY = aStartY - rowOffset * aScale; theStatus = clSetKernelArg(theKernels[i], argi++, sizeof(cl_double), (void*)&offsetedStartY); theStatus = clSetKernelArg(theKernels[i], argi++, sizeof(cl_double), (void*)&aScale); theStatus = clSetKernelArg(theKernels[i], argi++, sizeof(cl_uint), (void*)&theHardColorTableSize); theStatus = clSetKernelArg(theKernels[i], argi++, sizeof(cl_mem), (void*)&thePixelData[i]); theStatus = clSetKernelArg(theKernels[i], argi++, sizeof(cl_mem), (void*)&theHardColorTable); theStatus = clSetKernelArg(theKernels[i], argi++, sizeof(cl_uint), (void*)&theWidth); // Launch kernel theStatus = clEnqueueNDRangeKernel(theQueues[i], theKernels[i], 2, NULL, globalSize, NULL, 0, NULL, NULL); } rowOffset = 0; for( unsigned i = 0; i < numDevices; rowOffset += rowsPerDevice[i++] ) { // Read the output theStatus = clEnqueueReadBuffer(theQueues[i], thePixelData[i], CL_TRUE, 0, thePixelDataWidth*rowsPerDevice[i]*sizeof(unsigned int), &aFrameBuffer[rowOffset * theWidth], 0, NULL, NULL); } / / . .

    • Fungsi createProgramFromBinary membuat objek program OpenCL dari file objek.
    • Selanjutnya, untuk setiap perangkat, kernel dibuat berdasarkan objek program.
    • Buffer ThePixelData dibuat untuk menerima output dari setiap inti.
    • Buffer dibuat untuk menyimpan palet warna dan dimuat ke masing-masing akselerator.
    • Selanjutnya, untuk setiap perangkat, pengikatan parameter aplikasi lokal dan parameter kernel diatur menggunakan fungsi clSetKernelArg.
    • Parameter ditentukan oleh nomor seri dalam deklarasi fungsi kernel, mulai dari nol.

    Poin penting berikutnya adalah menentukan ukuran tugas berdasarkan ruang indeks menurut larik globalSize. Array ini bisa satu, dua atau tiga dimensi. Untuk setiap dimensi, dimensi diberikan sebagai bilangan bulat. Dimensi ruang akan menentukan urutan indeks item-pekerjaan di kernel.

    Dalam contoh, untuk setiap inti, ruang dua dimensi ditentukan, di mana salah satu sumbu adalah elemen baris piksel, yang kedua adalah serangkaian garis gambar yang diproses pada perangkat ini. Dalam kode kernel, nomor piksel dalam baris diperoleh dengan memanggil get_global_id (0), nomor barisnya adalah get_global_id (1). Variabel globalSize diteruskan ke fungsi clEnqueueNDRangeKernel untuk memulai jumlah instance kernel yang diperlukan untuk dieksekusi.

    Setelah menyelesaikan eksekusi core, buffer pixel dibaca dari perangkat ke array lokal. Mari kita evaluasi kinerja dengan jumlah frame per detik - hasilnya terlihat pada demonstrasi yang dilakukan di konferensi SelectelTechDay ( lihat bagian awal artikel ).

    Kesimpulan


    Memprogram akselerator FPGA dalam bahasa tingkat tinggi tidak diragukan lagi menurunkan ambang akses ke teknologi ini untuk pengembang dengan urutan besarnya. Misalnya, bagi mereka yang baru menguasai perangkat ini, bahkan ada implementasi FPGA dari contoh "Hello World" yang terkenal.

    Tapi tidak sesederhana itu. Menulis - dan terutama - men-debug algoritma yang bekerja dengan jelas dari masalah nyata masih membutuhkan profesionalisme tinggi. Keterbatasan lain adalah bahwa setiap chip FPGA hanya dapat melakukan satu tugas komputasi dalam aplikasi. Untuk tugas lain, harus diprogram ulang lagi.
    Omong-omong, model penggunaan platform memungkinkan Anda untuk memiliki lebih dari satu akselerator FPGA pada host, meskipun ini adalah solusi yang agak mahal.
    Host (aplikasi host) mengelola proses pembuatan konteks (struktur data untuk akselerator) dan antrian perintah. Yaitu Aplikasi host tunggal, di mana terdapat berbagai subtugas untuk komputasi paralel pada FPGA, dapat memuatnya pada akselerator yang berbeda:
    KERNEL1 => ACCELERATOR A
    KERNEL2 => ACCELERATOR B

    Namun demikian, upaya untuk menguasai akselerator FPGA tidak sia-sia - dalam banyak bidang aplikasi teknologi ini menjadi sangat diperlukan: telekomunikasi, bioteknologi, pemrosesan data besar, pengenalan pola, pemrosesan sinyal dan gambar, dalam matematika komputasi dan pemodelan bidang fisik.

    Informasi tambahan untuk artikel:
    www.altera.com adalah sumber daya inti Intel FPGA.
    www.eulerproject.com adalah situs web resmi Proyek Euler.
    Altera + OpenCL: kami memprogram di bawah FPGA tanpa sepengetahuan VHDL / Verilog - sebuah artikel tentang Habr.

Source: https://habr.com/ru/post/id418403/

More articles:


All Articles