kartu nvidia cuda. CUDA Kami Menggulung: Teknologi NVIDIA CUDA. Konversi video, penyaringan audio

Mari kita kembali ke sejarah - kembali ke tahun 2003, ketika Intel dan AMD berada dalam perlombaan bersama untuk prosesor yang paling kuat. Hanya dalam beberapa tahun, kecepatan clock telah meningkat secara signifikan sebagai hasil dari perlombaan ini, terutama setelah rilis Intel Pentium 4.

Tapi balapan dengan cepat mendekati batas. Setelah gelombang peningkatan besar dalam kecepatan clock (antara 2001 dan 2003, kecepatan clock Pentium 4 berlipat ganda dari 1,5 menjadi 3 GHz), pengguna harus puas dengan sepersepuluh gigahertz yang dapat diperas oleh produsen (dari 2003 hingga 2005 , kecepatan clock meningkat dari hanya 3 menjadi 3,8 GHz).

Bahkan arsitektur yang dioptimalkan untuk kecepatan clock tinggi, seperti Prescott, mulai mengalami kesulitan, dan kali ini tidak hanya yang produksi. Pembuat chip baru saja melanggar hukum fisika. Beberapa analis bahkan meramalkan bahwa hukum Moore akan berhenti beroperasi. Tapi itu tidak terjadi. Arti asli dari hukum sering disalahartikan, tetapi mengacu pada jumlah transistor pada permukaan inti silikon. Untuk waktu yang lama, peningkatan jumlah transistor dalam CPU disertai dengan peningkatan kinerja yang sesuai - yang menyebabkan distorsi makna. Tapi kemudian situasinya menjadi lebih rumit. Perancang arsitektur CPU mendekati hukum pengurangan gain: jumlah transistor yang perlu ditambahkan untuk peningkatan kinerja yang diinginkan menjadi semakin banyak, yang mengarah ke jalan buntu.

Sementara pembuat CPU telah berusaha keras mencari solusi untuk masalah mereka, pembuat GPU terus mendapat manfaat luar biasa dari manfaat Hukum Moore.

Mengapa mereka tidak berakhir di jalan buntu yang sama dengan para perancang arsitektur CPU? Alasannya sangat sederhana: CPU dirancang untuk mendapatkan kinerja terbaik pada aliran instruksi yang memproses data yang berbeda (baik bilangan bulat dan angka floating point), melakukan akses memori acak, dan sebagainya. Sampai sekarang, pengembang telah mencoba untuk menyediakan paralelisme instruksi yang lebih besar - yaitu, untuk mengeksekusi instruksi sebanyak mungkin secara paralel. Jadi, misalnya, eksekusi superskalar muncul dengan Pentium, ketika dalam kondisi tertentu dimungkinkan untuk mengeksekusi dua instruksi per jam. Pentium Pro menerima eksekusi instruksi yang tidak sesuai pesanan, yang memungkinkan untuk mengoptimalkan kinerja unit komputasi. Masalahnya adalah bahwa eksekusi paralel dari aliran instruksi berurutan memiliki batasan yang jelas, sehingga meningkatkan jumlah unit komputasi secara membabi buta tidak memberikan keuntungan, karena sebagian besar waktu mereka masih akan menganggur.

Sebaliknya, pekerjaan GPU relatif sederhana. Ini terdiri dari mengambil sekelompok poligon di satu sisi dan menghasilkan sekelompok piksel di sisi lain. Poligon dan piksel tidak tergantung satu sama lain, sehingga dapat diproses secara paralel. Jadi, di GPU, dimungkinkan untuk mengalokasikan sebagian besar kristal untuk unit komputasi, yang, tidak seperti CPU, akan benar-benar digunakan.

Klik pada gambar untuk memperbesar.

GPU berbeda dari CPU tidak hanya dalam hal ini. Akses memori di GPU sangat digabungkan - jika texel dibaca, maka setelah beberapa siklus, texel tetangga akan dibaca; ketika sebuah piksel ditulis, piksel tetangga akan ditulis setelah beberapa siklus. Dengan mengatur memori secara cerdas, Anda bisa mendapatkan kinerja yang mendekati bandwidth teoritis. Ini berarti bahwa GPU, tidak seperti CPU, tidak memerlukan cache yang besar, karena perannya adalah untuk mempercepat operasi texturing. Yang diperlukan hanyalah beberapa kilobyte yang berisi beberapa texel yang digunakan dalam filter bilinear dan trilinear.

Klik pada gambar untuk memperbesar.

Hidup GeForce FX!

Kedua dunia tetap terpisah untuk waktu yang lama. Kami menggunakan CPU (atau bahkan beberapa CPU) untuk tugas kantor dan aplikasi Internet, dan GPU hanya cocok untuk mempercepat rendering. Tetapi satu fitur mengubah segalanya: yaitu, munculnya GPU yang dapat diprogram. Pada awalnya, CPU tidak perlu takut. Yang pertama disebut GPU yang dapat diprogram (NV20 dan R200) ​​hampir tidak menjadi ancaman. Jumlah instruksi dalam program tetap terbatas pada sekitar 10, mereka bekerja pada tipe data yang sangat eksotis, seperti angka titik tetap 9 atau 12-bit.

Klik pada gambar untuk memperbesar.

Namun hukum Moore kembali menunjukkan sisi terbaiknya. Peningkatan jumlah transistor tidak hanya meningkatkan jumlah unit komputasi, tetapi juga meningkatkan fleksibilitasnya. Munculnya NV30 dapat dianggap sebagai langkah maju yang signifikan karena beberapa alasan. Tentu saja, para gamer tidak terlalu menyukai kartu NV30, tetapi GPU baru mulai mengandalkan dua fitur yang dirancang untuk mengubah persepsi GPU sebagai lebih dari sekadar akselerator grafis.

• Dukungan untuk perhitungan floating-point presisi tunggal (meskipun tidak sesuai dengan standar IEEE754);
• dukungan untuk lebih dari seribu instruksi.

Jadi kami mendapatkan semua kondisi yang dapat menarik para peneliti perintis yang selalu mencari daya komputasi tambahan.

Gagasan menggunakan akselerator grafis untuk perhitungan matematis bukanlah hal baru. Upaya pertama dilakukan pada tahun 90-an abad terakhir. Tentu saja, mereka sangat primitif - terbatas, sebagian besar, untuk penggunaan beberapa fitur berbasis perangkat keras, seperti rasterisasi dan buffer-Z untuk mempercepat tugas seperti pencarian rute atau keluaran diagram Voronoi .

Klik pada gambar untuk memperbesar.

Pada tahun 2003, dengan munculnya shader yang berevolusi, bilah baru tercapai - kali ini melakukan perhitungan matriks. Ini adalah tahun di mana seluruh bagian SIGGRAPH ("Komputasi pada GPU") didedikasikan untuk area baru TI. Inisiatif awal ini disebut GPGPU (General-Purpose Computing pada GPU). Dan munculnya .

Untuk memahami peran BrookGPU, Anda perlu memahami bagaimana segala sesuatu terjadi sebelum kemunculannya. Satu-satunya cara untuk mendapatkan sumber daya GPU pada tahun 2003 adalah dengan menggunakan salah satu dari dua API grafis - Direct3D atau OpenGL. Akibatnya, pengembang yang ingin mendapatkan kekuatan GPU untuk komputasi mereka harus bergantung pada dua API yang disebutkan. Masalahnya adalah mereka tidak selalu ahli dalam pemrograman kartu grafis, dan ini membuat akses ke teknologi menjadi sangat sulit. Jika pemrogram 3D beroperasi dengan shader, tekstur, dan fragmen, maka spesialis di bidang pemrograman paralel mengandalkan utas, inti, penghambur, dll. Oleh karena itu, pada awalnya perlu untuk menarik analogi antara dua dunia.

• sungai kecil adalah aliran elemen dari jenis yang sama, di GPU dapat diwakili oleh tekstur. Pada prinsipnya, dalam pemrograman klasik ada analog seperti array.
• Inti- fungsi yang akan diterapkan secara independen ke setiap elemen aliran; adalah setara dengan pixel shader. Dalam pemrograman klasik, Anda dapat memberikan analogi untuk sebuah siklus - ini diterapkan pada sejumlah besar elemen.
• Untuk membaca hasil penerapan kernel ke aliran, tekstur harus dibuat. Tidak ada yang setara pada CPU karena ada akses memori penuh.
• Lokasi di memori yang akan ditulisi (dalam operasi scatter/scatter) dikendalikan melalui vertex shader, karena pixel shader tidak dapat mengubah koordinat piksel yang diproses.

Seperti yang Anda lihat, bahkan dengan mempertimbangkan analogi di atas, tugasnya tidak terlihat sederhana. Dan Brook datang untuk menyelamatkan. Nama ini mengacu pada ekstensi ke bahasa C ("C dengan aliran", "C dengan aliran"), sebagaimana disebut oleh para pengembang di Stanford. Pada intinya, tugas Brook adalah menyembunyikan semua komponen API 3D dari programmer, yang memungkinkan untuk menghadirkan GPU sebagai koprosesor untuk komputasi paralel. Untuk melakukan ini, kompiler Brook memproses file .br dengan kode dan ekstensi C++, lalu menghasilkan kode C++ yang ditautkan ke pustaka dengan dukungan untuk keluaran yang berbeda (DirectX, OpenGL ARB, OpenGL NV3x, x86).

Klik pada gambar untuk memperbesar.

Brook memiliki beberapa penghargaan, yang pertama adalah membawa GPGPU keluar dari bayang-bayang sehingga teknologinya dapat dilihat oleh masyarakat umum. Meski, setelah pengumuman proyek tersebut, sejumlah situs IT terlalu optimis bahwa rilis Brook meragukan keberadaan CPU, yang akan segera digantikan oleh GPU yang lebih kuat. Tapi, seperti yang kita lihat, bahkan setelah lima tahun ini tidak terjadi. Sejujurnya, kami tidak berpikir itu akan pernah terjadi sama sekali. Di sisi lain, melihat keberhasilan evolusi CPU, yang semakin berorientasi paralelisme (lebih banyak core, teknologi multithreading SMT, perluasan blok SIMD), serta GPU, yang, sebaliknya, menjadi lebih universal (mendukung untuk perhitungan floating point). presisi tunggal, perhitungan bilangan bulat, dukungan untuk perhitungan presisi ganda), sepertinya GPU dan CPU akan segera bergabung. Apa yang akan terjadi kemudian? Apakah GPU akan ditelan oleh CPU, seperti yang terjadi pada koprosesor matematika? Cukup mungkin. Intel dan AMD sedang mengerjakan proyek serupa hari ini. Tapi masih banyak yang bisa berubah.

Tapi kembali ke topik kita. Keuntungan Brook adalah mempopulerkan konsep GPGPU, ini sangat menyederhanakan akses ke sumber daya GPU, yang memungkinkan semakin banyak pengguna untuk menguasai model pemrograman baru. Di sisi lain, terlepas dari semua kualitas Brook, masih ada jalan panjang sebelum sumber daya GPU dapat digunakan untuk komputasi.

Salah satu masalah terkait dengan tingkat abstraksi yang berbeda, dan juga, khususnya, beban tambahan berlebihan yang dibuat oleh API 3D, yang bisa sangat terlihat. Tetapi yang lebih serius dapat dianggap sebagai masalah kompatibilitas, di mana pengembang Brook tidak dapat melakukan apa pun. Ada persaingan sengit antara produsen GPU, sehingga mereka sering mengoptimalkan driver mereka. Jika pengoptimalan seperti itu, sebagian besar, bagus untuk para gamer, mereka dapat menghilangkan kompatibilitas Brook dalam sekejap. Oleh karena itu, sulit membayangkan penggunaan API ini dalam kode industri yang akan berfungsi di suatu tempat. Dan untuk waktu yang lama, Brook tetap menjadi peneliti dan pemrogram amatir.

Namun, keberhasilan Brook cukup menarik perhatian ATI dan Nvidia, dan mereka tertarik dengan inisiatif seperti itu, karena dapat memperluas pasar, membuka sektor penting baru bagi perusahaan.

Para peneliti yang awalnya terlibat dalam proyek Brook dengan cepat bergabung dengan tim pengembangan di Santa Clara untuk mempresentasikan strategi global untuk mengembangkan pasar baru. Idenya adalah untuk membuat kombinasi perangkat keras dan perangkat lunak yang cocok untuk tugas GPGPU. Karena pengembang nVidia mengetahui semua rahasia GPU mereka, dimungkinkan untuk tidak mengandalkan API grafis, tetapi untuk berkomunikasi dengan prosesor grafis melalui driver. Meskipun, tentu saja, ini datang dengan masalahnya sendiri. Jadi, tim pengembangan CUDA (Compute Unified Device Architecture) telah membuat satu set lapisan perangkat lunak untuk bekerja dengan GPU.

Klik pada gambar untuk memperbesar.

Seperti yang Anda lihat dalam diagram, CUDA menyediakan dua API.

• API tingkat tinggi: CUDA Runtime API;
• API tingkat rendah: CUDA Driver API.

Karena API tingkat tinggi diimplementasikan di atas API tingkat rendah, setiap panggilan fungsi runtime dipecah menjadi instruksi sederhana yang diproses oleh Driver API. Perhatikan bahwa kedua API tersebut saling eksklusif: seorang programmer dapat menggunakan satu atau API lainnya, tetapi tidak mungkin untuk mencampur pemanggilan fungsi dari kedua API tersebut. Secara umum, istilah "API tingkat tinggi" bersifat relatif. Bahkan API Runtime sedemikian rupa sehingga banyak orang akan menganggapnya tingkat rendah; namun, ia masih menyediakan fungsi yang sangat nyaman untuk inisialisasi atau manajemen konteks. Tapi jangan berharap tingkat abstraksi yang sangat tinggi - Anda masih harus memiliki pengetahuan yang baik tentang GPU nVidia dan cara kerjanya.

Driver API bahkan lebih sulit untuk digunakan; Anda membutuhkan lebih banyak upaya untuk menjalankan pemrosesan GPU. Di sisi lain, API tingkat rendah lebih fleksibel, memberi programmer lebih banyak kontrol jika diperlukan. Dua API mampu bekerja dengan sumber daya OpenGL atau Direct3D (hanya versi kesembilan pada hari ini). Manfaat fitur ini jelas - CUDA dapat digunakan untuk membuat sumber daya (geometri, tekstur prosedural, dll.) yang dapat diteruskan ke API grafis atau, sebaliknya, API 3D dapat dibuat untuk mengirim hasil rendering ke CUDA program, yang, pada gilirannya, akan melakukan pasca-pemrosesan. Ada banyak contoh interaksi seperti itu, dan keuntungannya adalah bahwa sumber daya terus disimpan dalam memori GPU, mereka tidak perlu ditransfer melalui bus PCI Express, yang masih menjadi hambatan.

Namun, perlu dicatat bahwa berbagi sumber daya dalam memori video tidak selalu ideal dan dapat menyebabkan beberapa "sakit kepala". Misalnya, saat mengubah resolusi atau kedalaman warna, data grafis diutamakan. Oleh karena itu, jika diperlukan untuk meningkatkan sumber daya dalam buffer bingkai, maka driver akan dengan mudah melakukan ini dengan mengorbankan sumber daya aplikasi CUDA, yang hanya akan macet dengan kesalahan. Tentu saja, tidak terlalu elegan, tetapi situasi ini seharusnya tidak terlalu sering terjadi. Dan karena kami mulai berbicara tentang kerugiannya: jika Anda ingin menggunakan beberapa GPU untuk aplikasi CUDA, maka Anda harus terlebih dahulu menonaktifkan mode SLI, jika tidak, aplikasi CUDA hanya akan dapat "melihat" satu GPU.

Akhirnya, tingkat perangkat lunak ketiga diberikan ke perpustakaan - dua, tepatnya.

• CUBLAS, yang berisi blok yang diperlukan untuk menghitung aljabar linier pada GPU;
• Cufft, yang mendukung perhitungan transformasi Fourier, adalah algoritma yang banyak digunakan di bidang pemrosesan sinyal.

Sebelum kita masuk ke CUDA, mari kita definisikan beberapa istilah yang tersebar di seluruh dokumentasi nVidia. Perusahaan telah memilih terminologi yang sangat spesifik yang sulit untuk digunakan. Pertama-tama, kami perhatikan bahwa benang di CUDA tidak memiliki arti yang sama dengan utas CPU, juga tidak setara dengan utas di artikel GPU kami. Utas GPU dalam hal ini adalah kumpulan data dasar yang perlu diproses. Tidak seperti utas CPU, utas CUDA sangat "ringan", artinya sakelar konteks di antara dua utas sama sekali bukan operasi intensif sumber daya.

Istilah kedua yang sering ditemukan dalam dokumentasi CUDA adalah melengkung. Tidak ada kebingungan di sini, karena tidak ada analog dalam bahasa Rusia (kecuali jika Anda adalah penggemar game Start Trek atau Warhammer). Padahal, istilah ini diambil dari industri tekstil, di mana benang pakan ditarik melalui benang lusi, yang direntangkan pada alat tenun. Warp di CUDA adalah sekelompok 32 utas dan merupakan jumlah minimum data yang diproses dengan cara SIMD di multiprosesor CUDA.

Tetapi "kekasaran" seperti itu tidak selalu nyaman bagi programmer. Oleh karena itu, di CUDA, alih-alih bekerja dengan warps secara langsung, Anda dapat bekerja dengan memblokir, berisi 64 hingga 512 utas.

Akhirnya, balok-balok ini disatukan dalam kisi-kisi. Keuntungan dari pengelompokan ini adalah jumlah blok yang diproses oleh GPU pada saat yang sama berkaitan erat dengan sumber daya perangkat keras, seperti yang akan kita lihat di bawah. Mengelompokkan blok ke dalam kisi memungkinkan Anda untuk sepenuhnya abstrak dari batasan ini dan menerapkan kernel / kernel ke lebih banyak utas dalam satu panggilan, tanpa memikirkan sumber daya tetap. Perpustakaan CUDA bertanggung jawab untuk semua ini. Selain itu, model ini berskala dengan baik. Jika GPU kekurangan sumber daya, maka GPU akan mengeksekusi blok secara berurutan. Jika jumlah prosesor komputasi besar, maka blok dapat dieksekusi secara paralel. Artinya, kode yang sama dapat berjalan pada GPU entry-level dan model top-end dan bahkan di masa mendatang.

Ada beberapa istilah lagi di CUDA API yang merupakan singkatan dari CPU ( tuan rumah / tuan rumah) dan GPU ( perangkat/perangkat). Jika perkenalan kecil ini tidak membuat Anda takut, maka inilah saatnya untuk mengenal CUDA lebih baik.

Jika Anda sering membaca Tom's Hardware Guide, maka arsitektur GPU terbaru dari nVidia sudah tidak asing lagi bagi Anda. Jika belum, kami sarankan Anda membaca artikel " nVidia GeForce GTX 260 dan 280: kartu grafis generasi baru". Berkenaan dengan CUDA, nVidia menyajikan arsitektur yang sedikit berbeda, menunjukkan beberapa detail yang sebelumnya tersembunyi.

Seperti yang Anda lihat dari ilustrasi di atas, inti shader nVidia terdiri dari beberapa kelompok prosesor tekstur. (Kluster Prosesor Tekstur, TPC). 8800 GTX, misalnya, menggunakan delapan cluster, 8800 GTS menggunakan enam, dan seterusnya. Setiap cluster pada dasarnya terdiri dari unit tekstur dan dua streaming multiprosesor. Yang terakhir termasuk awal pipa (ujung depan), yang membaca dan mendekode instruksi, serta mengirimkannya untuk dieksekusi, dan akhir pipa (ujung belakang), yang terdiri dari delapan perangkat komputasi dan dua perangkat superfungsional. SFU (Unit Fungsi Super), di mana instruksi dieksekusi sesuai dengan prinsip SIMD, yaitu, satu instruksi diterapkan ke semua utas di warp. nVidia menyebut cara melakukannya SIMT(satu instruksi banyak utas, satu instruksi, banyak utas). Penting untuk dicatat bahwa ujung pipa beroperasi pada dua kali frekuensi awalnya. Dalam praktiknya, ini berarti bahwa bagian tersebut terlihat dua kali lebih "lebar" daripada yang sebenarnya (yaitu seperti blok SIMD 16 saluran, bukan yang delapan saluran). Multiprosesor streaming bekerja seperti ini: setiap siklus, awal pipeline memilih warp yang siap dieksekusi dan mulai mengeksekusi instruksi. Agar instruksi diterapkan ke semua 32 utas di warp, akhir pipa akan membutuhkan empat siklus clock, tetapi karena berjalan pada dua kali frekuensi awal, itu hanya akan membutuhkan dua siklus clock (dalam hal awal pipa). Oleh karena itu, agar awal pipa tidak menganggur selama satu siklus, dan perangkat keras dimuat secara maksimal, dalam kasus yang ideal, Anda dapat mengganti instruksi setiap siklus - instruksi klasik dalam satu siklus dan instruksi untuk SFU - di siklus lain .

Setiap multiprosesor memiliki seperangkat sumber daya tertentu yang layak untuk dipahami. Ada area kecil di memori yang disebut "Berbagi memori", 16 KB per multiprosesor. Ini sama sekali bukan memori cache: programmer dapat menggunakannya sesuai kebijaksanaannya. Artinya, kami memiliki sesuatu yang dekat dengan Toko Lokal SPU pada prosesor Sel. Detail ini cukup menarik karena menekankan bahwa CUDA merupakan kombinasi dari teknologi perangkat lunak dan perangkat keras. Area memori ini tidak digunakan untuk pixel shader, karena Nvidia dengan cerdik menunjukkan "kami tidak suka ketika piksel berbicara satu sama lain".

Area memori ini membuka kemungkinan pertukaran informasi antar utas. dalam satu blok. Penting untuk menekankan batasan ini: semua utas dalam satu blok dijamin akan dieksekusi oleh multiprosesor tunggal. Sebaliknya, pengikatan blok ke multiprosesor yang berbeda tidak ditetapkan sama sekali, dan dua utas dari blok yang berbeda tidak dapat bertukar informasi satu sama lain selama eksekusi. Artinya, menggunakan memori bersama tidak sesederhana itu. Namun, memori bersama masih dibenarkan, kecuali ketika beberapa utas mencoba mengakses bank memori yang sama, menyebabkan konflik. Dalam situasi lain, akses memori bersama secepat akses register.

Memori bersama bukan satu-satunya yang dapat diakses oleh multiprosesor. Mereka dapat menggunakan memori video, tetapi dengan bandwidth yang lebih rendah dan latensi yang lebih tinggi. Oleh karena itu, untuk mengurangi frekuensi mengakses memori ini, nVidia melengkapi multiprosesor dengan cache (sekitar 8 KB per multiprosesor) yang menyimpan konstanta dan tekstur.

Multiprosesor memiliki 8.192 register yang umum untuk semua utas dari semua blok yang aktif pada multiprosesor. Jumlah blok aktif per multiprosesor tidak boleh melebihi delapan, dan jumlah warps aktif dibatasi hingga 24 (768 utas). Oleh karena itu, 8800 GTX dapat memproses hingga 12.288 utas sekaligus. Semua batasan ini layak disebutkan karena memungkinkan algoritme dioptimalkan berdasarkan sumber daya yang tersedia.

Optimalisasi program CUDA, oleh karena itu, terdiri dari memperoleh keseimbangan optimal antara jumlah blok dan ukurannya. Lebih banyak utas per blok akan membantu dalam mengurangi latensi memori, tetapi jumlah register yang tersedia per utas juga akan berkurang. Selain itu, blok 512 utas tidak akan efisien, karena hanya satu blok yang dapat aktif pada multiprosesor, yang mengakibatkan hilangnya 256 utas. Oleh karena itu, nVidia merekomendasikan penggunaan blok 128 atau 256 utas, yang memberikan kompromi terbaik antara latensi yang lebih rendah dan jumlah register untuk sebagian besar inti/kernel.

Dari sudut pandang program, CUDA terdiri dari satu set ekstensi ke bahasa C, yang menyerupai BrookGPU, serta beberapa panggilan API tertentu. Di antara ekstensi adalah penentu tipe yang terkait dengan fungsi dan variabel. Penting untuk mengingat kata kunci __global__, yang, ketika diberikan di depan fungsi, menunjukkan bahwa yang terakhir mengacu pada kernel / kernel - fungsi ini akan dipanggil oleh CPU, dan akan dijalankan pada GPU. Awalan __perangkat__ menentukan bahwa fungsi tersebut akan dijalankan pada GPU (yang CUDA menyebutnya sebagai "perangkat/perangkat") tetapi hanya dapat dipanggil dari GPU (dengan kata lain, dari fungsi __device__ lain atau dari fungsi __global__). Akhirnya, awalan __tuan rumah__ opsional, ini menunjukkan fungsi yang dipanggil oleh CPU dan dijalankan oleh CPU - dengan kata lain, fungsi biasa.

Ada sejumlah batasan yang terkait dengan fungsi __device__ dan __global__: mereka tidak dapat rekursif (yaitu, menyebut diri mereka sendiri), dan mereka tidak dapat memiliki sejumlah argumen yang bervariasi. Terakhir, karena fungsi __device__ berada di ruang memori GPU, masuk akal jika alamatnya tidak dapat diambil. Variabel juga memiliki sejumlah kualifikasi yang menunjukkan lokasi memori di mana mereka akan disimpan. Variabel dengan awalan __bersama__ berarti akan disimpan dalam memori bersama dari multiprosesor streaming. Panggilan ke fungsi __global__ sedikit berbeda. Masalahnya, saat menelepon, Anda perlu mengatur konfigurasi eksekusi - lebih khusus lagi, ukuran grid / grid yang akan diterapkan kernel / kernel, serta ukuran setiap blok. Ambil, misalnya, kernel dengan tanda tangan berikut.

__global__ void Func(parameter float*);

Ini akan disebut sebagai

fungsi<<< Dg, Db >>> (parameter);

di mana Dg adalah ukuran grid dan Db adalah ukuran blok. Kedua variabel ini mengacu pada tipe vektor baru yang diperkenalkan dengan CUDA.

CUDA API berisi fungsi untuk bekerja dengan memori di VRAM: cudaMalloc untuk mengalokasikan memori, cudaFree untuk membebaskan, dan cudaMemcpy untuk menyalin memori antara RAM dan VRAM dan sebaliknya.

Kami akan mengakhiri ulasan ini dengan cara yang sangat menarik di mana program CUDA dikompilasi: kompilasi dilakukan dalam beberapa langkah. Pertama, kode khusus CPU diekstraksi dan diteruskan ke kompiler standar. Kode yang ditujukan untuk GPU pertama kali dikonversi ke bahasa perantara PTX. Ini mirip dengan bahasa rakitan dan memungkinkan Anda mempelajari kode untuk mencari bagian yang tidak efisien. Akhirnya, fase terakhir adalah menerjemahkan bahasa perantara ke dalam instruksi khusus GPU dan membuat biner.

Melihat melalui dokumentasi nVidia membuat saya ingin mencoba CUDA minggu ini. Memang, apa yang bisa lebih baik daripada mengevaluasi API dengan membuat program Anda sendiri? Saat itulah sebagian besar masalah harus muncul ke permukaan, bahkan jika semuanya terlihat sempurna di atas kertas. Selain itu, praktik terbaik akan menunjukkan seberapa baik Anda memahami semua prinsip yang diuraikan dalam dokumentasi CUDA.

Sangat mudah untuk terjun ke proyek semacam itu. Saat ini, sejumlah besar alat gratis, tetapi berkualitas tinggi tersedia untuk diunduh. Untuk pengujian kami, kami menggunakan Visual C++ Express 2005, yang memiliki semua yang Anda butuhkan. Bagian tersulit adalah menemukan program yang tidak membutuhkan waktu berminggu-minggu untuk diport ke GPU, tetapi cukup menarik sehingga upaya kami tidak sia-sia. Pada akhirnya, kami memilih sepotong kode yang mengambil peta ketinggian dan menghitung peta normal yang sesuai. Kami tidak akan menyelidiki fungsi ini secara rinci, karena hampir tidak menarik dalam artikel ini. Singkatnya, program berurusan dengan kelengkungan area: untuk setiap piksel gambar awal, kami menerapkan matriks yang menentukan warna piksel yang dihasilkan dalam gambar yang dihasilkan dari piksel yang berdekatan, menggunakan rumus yang kurang lebih kompleks. Keuntungan dari fungsi ini adalah sangat mudah untuk diparalelkan, sehingga tes ini menunjukkan kemampuan CUDA dengan sempurna.

Keuntungan lain adalah bahwa kami telah memiliki implementasi pada CPU, sehingga kami dapat membandingkan hasilnya dengan versi CUDA - dan tidak menemukan kembali kemudi.

Kami ulangi sekali lagi bahwa tujuan pengujian adalah untuk berkenalan dengan utilitas CUDA SDK, dan bukan untuk membandingkan versi untuk CPU dan GPU. Karena ini adalah upaya pertama kami untuk membuat program CUDA, kami tidak terlalu mengharapkan kinerja yang tinggi. Karena bagian kode ini tidak kritis, versi CPU tidak dioptimalkan, jadi perbandingan langsung dari hasilnya hampir tidak menarik.

Pertunjukan

Namun, kami mengukur waktu eksekusi untuk melihat apakah ada keuntungan menggunakan CUDA bahkan dengan implementasi yang paling kasar, atau jika kami membutuhkan latihan yang lama dan melelahkan untuk mendapatkan keuntungan apa pun saat menggunakan GPU. Mesin uji diambil dari lab pengembangan kami - laptop dengan prosesor Core 2 Duo T5450 dan kartu grafis GeForce 8600M GT yang menjalankan Vista. Ini jauh dari superkomputer, tetapi hasilnya sangat menarik, karena pengujiannya tidak "dipertajam" untuk GPU. Itu selalu menyenangkan untuk melihat nVidia menunjukkan keuntungan besar pada sistem dengan GPU mengerikan dan banyak bandwidth, tetapi dalam prakteknya, banyak dari 70 juta CUDA-enabled GPU di pasar PC saat ini sama sekali tidak kuat, sehingga pengujian kami dibenarkan.

Untuk gambar 2048 x 2048 piksel, kami mendapatkan hasil sebagai berikut.

• CPU 1 utas: 1419ms;
• CPU 2 utas: 749ms;
• CPU 4 utas: 593ms
• GPU (8600M GT) blok 256 utas: 109ms;
• GPU (8600M GT) blok 128 thread: 94ms;
• GPU (8800 GTX) blok 128 utas / 256 utas: 31ms.

Beberapa kesimpulan dapat ditarik dari hasil tersebut. Mari kita mulai dengan fakta bahwa, terlepas dari pembicaraan tentang kemalasan yang jelas dari programmer, kami memodifikasi versi awal CPU untuk beberapa utas. Seperti yang telah kami sebutkan, kode ini ideal untuk situasi ini - yang diperlukan hanyalah membagi gambar awal menjadi zona sebanyak aliran yang ada. Perhatikan bahwa dari beralih dari satu utas ke dua utas pada CPU dual-core kami, akselerasinya hampir linier, yang juga menunjukkan sifat paralel dari program pengujian. Secara tidak terduga, versi dengan empat utas juga lebih cepat, meskipun ini sangat aneh pada prosesor kami - sebaliknya, orang dapat mengharapkan penurunan efisiensi karena biaya pengelolaan utas tambahan. Bagaimana hasil seperti itu dapat dijelaskan? Sulit untuk mengatakannya, tetapi penjadwal utas Windows mungkin adalah pelakunya; dalam hal apapun, hasilnya dapat diulang. Dengan tekstur yang lebih kecil (512x512), keuntungan dari pemisahan tidak begitu terasa (sekitar 35% berbanding 100%), dan perilaku versi empat utas lebih logis, tanpa keuntungan dibandingkan dengan versi dua utas. GPU masih lebih cepat, tetapi tidak begitu menonjol (8600M GT tiga kali lebih cepat daripada versi ulir ganda).

Klik pada gambar untuk memperbesar.

Pengamatan signifikan kedua adalah bahwa implementasi GPU yang paling lambat pun ternyata hampir enam kali lebih cepat daripada versi CPU dengan kinerja tertinggi. Untuk program pertama dan versi algoritma yang tidak dioptimalkan, hasilnya sangat menggembirakan. Harap dicatat bahwa kami mendapat hasil yang jauh lebih baik pada blok kecil, meskipun intuisi mungkin menyarankan sebaliknya. Penjelasannya sederhana - program kami menggunakan 14 register per utas, dan dengan 256 blok utas, diperlukan 3.584 register per blok, dan 768 utas diperlukan untuk beban prosesor penuh, seperti yang kami tunjukkan. Dalam kasus kami, ini adalah tiga blok atau 10.572 register. Tetapi multiprosesor hanya memiliki 8.192 register, sehingga hanya dapat membuat dua blok aktif. Sebaliknya, dengan blok 128 utas, kita membutuhkan 1.792 register per blok; jika 8.192 dibagi 1.792 dan dibulatkan ke bilangan bulat terdekat, kita mendapatkan empat balok. Dalam praktiknya, jumlah utas akan sama (512 per multiprosesor, meskipun 768 secara teoritis diperlukan untuk beban penuh), tetapi peningkatan jumlah blok memberi GPU keuntungan fleksibilitas dalam akses memori - saat operasi sedang berlangsung dengan penundaan yang besar, ia dapat mulai mengeksekusi instruksi dari blok lain, menunggu hasil yang diterima. Empat blok jelas mengurangi latensi, terutama karena program kami menggunakan banyak akses memori.

Analisis

Akhirnya, terlepas dari apa yang kami katakan di atas, kami tidak dapat menahan godaan dan menjalankan program pada 8800 GTX, yang tiga kali lebih cepat daripada 8600, terlepas dari ukuran bloknya. Anda mungkin berpikir bahwa dalam praktiknya, pada arsitektur yang sesuai, hasilnya akan empat kali atau lebih tinggi: 128 ALU / prosesor shader versus 32 dan kecepatan clock yang lebih tinggi (1,35 GHz versus 950 MHz), tetapi ini tidak terjadi. Kemungkinan besar, faktor pembatasnya adalah akses memori. Untuk lebih tepatnya, gambar awal diakses sebagai array CUDA multi-dimensi - istilah yang agak rumit untuk sesuatu yang tidak lebih dari tekstur. Tapi makan beberapa manfaat.

• mengakses manfaat dari cache tekstur;
• kami menggunakan mode pembungkus, yang tidak perlu menangani batas gambar, tidak seperti versi CPU.

Selain itu, kita dapat memperoleh manfaat dari pemfilteran "gratis" dengan pengalamatan yang dinormalisasi antara alih-alih dan , tetapi ini hampir tidak berguna dalam kasus kita. Seperti yang Anda ketahui, 8600 memiliki 16 unit tekstur, dibandingkan dengan 32 untuk 8800 GTX. Oleh karena itu, rasio antara dua arsitektur hanya dua banding satu. Tambahkan ke perbedaan frekuensi dan kami mendapatkan rasio (32 x 0,575) / (16 x 0,475) = 2,4 - mendekati "tiga banding satu" yang sebenarnya kami dapatkan. Teori ini juga menjelaskan mengapa ukuran blok tidak banyak berubah pada G80, karena ALU masih bertumpu pada unit tekstur.

Klik pada gambar untuk memperbesar.

Selain hasil yang menjanjikan, pertemuan pertama kami dengan CUDA berjalan sangat baik, mengingat kondisi yang dipilih tidak terlalu menguntungkan. Mengembangkan pada laptop Vista berarti menggunakan CUDA SDK 2.0, yang masih dalam versi beta, dengan driver 174.55, yang juga dalam versi beta. Meskipun demikian, kami tidak dapat melaporkan kejutan yang tidak menyenangkan - hanya kesalahan awal selama debugging pertama, ketika program kami, masih sangat "berjalan", mencoba mengatasi memori di luar ruang yang dialokasikan.

Monitor mulai berkedip-kedip liar, kemudian layar menjadi hitam...sampai Vista menjalankan layanan perbaikan driver dan semuanya baik-baik saja. Tetapi masih agak mengejutkan untuk melihat apakah Anda terbiasa melihat Segmentation Fault yang khas pada program standar seperti milik kami. Akhirnya, kritik kecil terhadap nVidia: di semua dokumentasi yang tersedia untuk CUDA, tidak ada panduan kecil yang akan memandu Anda langkah demi langkah tentang cara menyiapkan lingkungan pengembangan untuk Visual Studio. Sebenarnya, masalahnya tidak besar, karena SDK memiliki serangkaian contoh lengkap yang dapat dipelajari untuk memahami kerangka kerja untuk aplikasi CUDA, tetapi panduan pemula akan lebih baik.

Klik pada gambar untuk memperbesar.

Nvidia memperkenalkan CUDA dengan merilis GeForce 8800. Dan pada saat itu janji-janji itu tampak sangat menggiurkan, tetapi kami tetap antusias untuk ujian yang sebenarnya. Memang, pada saat itu sepertinya lebih seperti menandai wilayah untuk tetap berada di gelombang GPGPU. Tanpa SDK yang tersedia, sulit untuk mengatakan bahwa kami tidak menghadapi dummy pemasaran lain yang tidak akan berfungsi. Ini bukan pertama kalinya inisiatif yang baik diumumkan terlalu dini dan pada saat itu tidak terungkap karena kurangnya dukungan - terutama di sektor yang begitu kompetitif. Sekarang, satu setengah tahun setelah pengumuman, kita dapat dengan aman mengatakan bahwa nVidia telah menepati janjinya.

SDK masuk ke beta cukup cepat pada awal 2007, dan telah diperbarui dengan cepat sejak saat itu, membuktikan pentingnya proyek ini untuk nVidia. Saat ini, CUDA berkembang dengan sangat baik: SDK sudah tersedia dalam versi beta 2.0 untuk sistem operasi utama (Windows XP dan Vista, Linux, serta 1.1 untuk Mac OS X), dan nVidia telah mendedikasikan seluruh bagian situs untuk pengembang.

Di tingkat yang lebih profesional, kesan langkah awal bersama CUDA ternyata sangat positif. Bahkan jika Anda terbiasa dengan arsitektur GPU, Anda dapat dengan mudah mengetahuinya. Ketika API terlihat jelas pada pandangan pertama, Anda segera mulai percaya bahwa Anda akan mendapatkan hasil yang meyakinkan. Tetapi bukankah waktu komputasi akan terbuang sia-sia dari beberapa transfer dari CPU ke GPU? Dan bagaimana cara menggunakan ribuan utas ini dengan hampir tidak ada sinkronisasi primitif? Kami memulai eksperimen kami dengan semua ketakutan ini dalam pikiran. Tetapi mereka dengan cepat menghilang ketika versi pertama dari algoritme kami, meskipun sangat sepele, ternyata jauh lebih cepat daripada di CPU.

Jadi CUDA bukanlah penyelamat bagi para peneliti yang ingin meyakinkan pejabat universitas untuk membelikan mereka GeForce. CUDA sudah merupakan teknologi yang tersedia sepenuhnya yang dapat digunakan oleh programmer C mana pun jika mereka bersedia meluangkan waktu dan upaya untuk membiasakan diri dengan paradigma pemrograman baru. Upaya ini tidak akan sia-sia jika algoritme Anda diparalelkan dengan baik. Kami juga ingin berterima kasih kepada nVidia karena telah menyediakan dokumentasi yang lengkap dan berkualitas tinggi, di mana para programmer CUDA pemula akan menemukan jawabannya.

Apa yang diperlukan CUDA untuk menjadi API yang dapat dikenali? Dalam satu kata: portabilitas. Kita tahu bahwa masa depan TI terletak pada komputasi paralel - hari ini semua orang sudah bersiap untuk perubahan tersebut, dan semua inisiatif, baik perangkat lunak maupun perangkat keras, diarahkan ke arah ini. Namun, saat ini, jika Anda melihat perkembangan paradigma, kami masih pada tahap awal: kami membuat utas secara manual dan mencoba menjadwalkan akses ke sumber daya bersama; semua ini entah bagaimana bisa diatasi jika jumlah core dapat dihitung dengan jari satu tangan. Namun dalam beberapa tahun, ketika jumlah prosesor akan mencapai ratusan, kemungkinan ini tidak akan ada lagi. Dengan dirilisnya CUDA, nVidia mengambil langkah pertama dalam memecahkan masalah ini - tetapi, tentu saja, solusi ini hanya cocok untuk GPU dari perusahaan ini, dan itupun tidak untuk semua orang. Hanya GF8 dan 9 (dan turunan Quadro/Tesla mereka) yang dapat menjalankan program CUDA hari ini. Dan baris baru 260/280, tentu saja.

Klik pada gambar untuk memperbesar.

Nvidia mungkin membanggakan bahwa ia telah menjual 70 juta GPU yang kompatibel dengan CUDA di seluruh dunia, tetapi itu masih belum cukup untuk menjadi standar de facto. Mempertimbangkan fakta bahwa pesaing tidak tinggal diam. AMD menawarkan SDK (Stream Computing) sendiri, dan Intel telah mengumumkan solusi (Ct), meskipun belum tersedia. Perang standar akan datang, dan jelas tidak akan ada ruang di pasar untuk tiga pesaing sampai pemain lain seperti Microsoft keluar dengan proposal API umum, yang tentu saja akan membuat hidup lebih mudah bagi pengembang.

Oleh karena itu, nVidia memiliki banyak kesulitan dalam hal persetujuan CUDA. Meskipun secara teknologi kami memiliki solusi yang sukses di hadapan kami, tanpa diragukan lagi, tetap meyakinkan para pengembang tentang prospeknya - dan ini tidak akan mudah. Namun, dilihat dari banyaknya pengumuman dan berita API baru-baru ini, masa depan tidak terlihat suram sama sekali.

Perangkat untuk mengubah komputer pribadi menjadi superkomputer kecil telah dikenal sejak lama. Kembali di tahun 80-an abad terakhir, yang disebut transputer ditawarkan di pasar, yang dimasukkan ke dalam slot ekspansi ISA yang umum. Pada awalnya, kinerja mereka dalam tugas yang sesuai sangat mengesankan, tetapi kemudian pertumbuhan kinerja prosesor universal dipercepat, mereka memperkuat posisi mereka dalam komputasi paralel, dan tidak ada gunanya transputer. Meskipun perangkat tersebut masih ada, mereka adalah berbagai akselerator khusus. Namun seringkali cakupan aplikasinya sempit dan akselerator semacam itu tidak banyak digunakan.

Namun baru-baru ini, tongkat estafet komputasi paralel telah pindah ke pasar massal, dengan satu atau lain cara terhubung dengan game tiga dimensi. Perangkat serba guna dengan prosesor multi-inti untuk komputasi vektor paralel yang digunakan dalam grafik 3D mencapai kinerja puncak tinggi yang tidak dapat dilakukan oleh prosesor serba guna. Tentu saja, kecepatan maksimum hanya dicapai dalam sejumlah tugas yang nyaman dan memiliki beberapa batasan, tetapi perangkat semacam itu sudah mulai digunakan secara luas di area yang awalnya tidak dimaksudkan. Contoh luar biasa dari prosesor paralel semacam itu adalah prosesor Sel yang dikembangkan oleh aliansi Sony-Toshiba-IBM dan digunakan di konsol game Sony PlayStation 3, serta semua kartu video modern dari para pemimpin pasar - Nvidia dan AMD.

Kami tidak akan menyentuh Cell hari ini, meskipun itu muncul sebelumnya dan merupakan prosesor universal dengan kemampuan vektor tambahan, kami tidak membicarakannya hari ini. Untuk akselerator video 3D, teknologi komputasi non-grafis serba guna pertama GPGPU (Komputasi Tujuan Umum pada GPU) muncul beberapa tahun lalu. Bagaimanapun, chip video modern berisi ratusan unit eksekusi matematis, dan kekuatan ini dapat digunakan untuk mempercepat banyak aplikasi intensif komputasi secara signifikan. Dan generasi GPU saat ini memiliki arsitektur yang cukup fleksibel yang, bersama dengan bahasa pemrograman tingkat tinggi dan arsitektur perangkat keras-perangkat lunak seperti yang dibahas dalam artikel ini, membuka kemungkinan ini dan membuatnya lebih mudah diakses.

Penciptaan GPCPU didorong oleh munculnya program shader yang cukup cepat dan fleksibel yang mampu mengeksekusi chip video modern. Pengembang memutuskan untuk membuat GPU menghitung tidak hanya gambar dalam aplikasi 3D, tetapi juga digunakan dalam perhitungan paralel lainnya. GPGPU menggunakan API grafis untuk ini: OpenGL dan Direct3D, ketika data ditransfer ke chip video dalam bentuk tekstur, dan program kalkulasi dimuat dalam bentuk shader. Kekurangan dari metode ini adalah kompleksitas pemrograman yang relatif tinggi, kecepatan pertukaran data yang rendah antara CPU dan GPU, dan keterbatasan lainnya, yang akan kita bahas nanti.

Komputasi GPU telah berkembang dan berkembang sangat pesat. Dan selanjutnya, dua produsen chip video besar, Nvidia dan AMD, mengembangkan dan mengumumkan masing-masing platform yang disebut CUDA (Compute Unified Device Architecture) dan CTM (Close To Metal atau AMD Stream Computing). Tidak seperti model pemrograman GPU sebelumnya, ini dilakukan dengan akses langsung ke kemampuan perangkat keras kartu grafis. Platform tidak kompatibel satu sama lain, CUDA adalah perpanjangan dari bahasa pemrograman C, dan CTM adalah mesin virtual yang mengeksekusi kode perakitan. Namun kedua platform telah menghilangkan beberapa batasan penting dari model GPGPU sebelumnya yang menggunakan jalur grafis tradisional dan antarmuka Direct3D atau OpenGL yang sesuai.

Tentu saja, standar terbuka yang menggunakan OpenGL tampaknya paling portabel dan universal, mereka memungkinkan Anda untuk menggunakan kode yang sama untuk chip video dari produsen yang berbeda. Tetapi metode seperti itu memiliki banyak kelemahan, mereka kurang fleksibel dan tidak nyaman untuk digunakan. Selain itu, mereka mencegah penggunaan fitur khusus dari kartu video tertentu, seperti memori bersama (bersama) yang cepat yang ada dalam prosesor komputasi modern.

Itulah sebabnya Nvidia merilis platform CUDA, bahasa pemrograman mirip-C dengan kompiler dan pustakanya sendiri untuk komputasi GPU. Tentu saja, menulis kode optimal untuk chip video sama sekali tidak mudah dan tugas ini membutuhkan pekerjaan manual yang lama, tetapi CUDA hanya mengungkapkan semua kemungkinan dan memberi programmer lebih banyak kendali atas kemampuan perangkat keras GPU. Penting agar dukungan Nvidia CUDA tersedia untuk chip G8x, G9x dan GT2xx yang digunakan dalam kartu video seri Geforce 8, 9 dan 200, yang tersebar luas. Versi final CUDA 2.0 kini telah dirilis, yang memiliki beberapa fitur baru, seperti dukungan untuk perhitungan presisi ganda. CUDA tersedia di sistem operasi Linux, Windows, dan MacOS X 32-bit dan 64-bit.

Perbedaan antara CPU dan GPU dalam Komputasi Paralel

Pertumbuhan frekuensi prosesor universal mengalami keterbatasan fisik dan konsumsi daya yang tinggi, dan kinerjanya semakin meningkat karena penempatan beberapa inti dalam satu chip. Prosesor yang dijual sekarang hanya berisi hingga empat inti (pertumbuhan lebih lanjut tidak akan cepat) dan ditujukan untuk aplikasi umum, gunakan MIMD - beberapa instruksi dan aliran data. Setiap inti beroperasi secara terpisah dari yang lain, mengeksekusi instruksi yang berbeda untuk proses yang berbeda.

Kemampuan vektor khusus (SSE2 dan SSE3) untuk vektor 4 komponen (titik-mengambang presisi tunggal) dan dua komponen (presisi ganda) telah muncul di prosesor tujuan umum karena meningkatnya permintaan aplikasi grafis. Itulah sebabnya untuk tugas-tugas tertentu penggunaan GPU lebih menguntungkan, karena pada awalnya dibuat untuk mereka.

Misalnya, dalam chip video Nvidia, unit utamanya adalah multiprosesor dengan delapan hingga sepuluh core dan total ratusan ALU, beberapa ribu register, dan sejumlah kecil shared memory. Selain itu, kartu video berisi memori global cepat yang dapat diakses oleh semua multiprosesor, memori lokal di setiap multiprosesor, dan memori khusus untuk konstanta.

Yang terpenting, beberapa inti multiprosesor di GPU ini adalah inti SIMD (aliran instruksi tunggal, aliran data ganda). Dan inti ini menjalankan instruksi yang sama pada saat yang sama, gaya pemrograman yang umum dalam algoritme grafik dan banyak tugas ilmiah, tetapi memerlukan pemrograman khusus. Tetapi pendekatan ini memungkinkan Anda untuk meningkatkan jumlah unit eksekusi karena penyederhanaannya.

Jadi, mari daftar perbedaan utama antara arsitektur CPU dan GPU. Core CPU dirancang untuk mengeksekusi satu aliran instruksi berurutan dengan kinerja maksimum, sementara GPU dirancang untuk mengeksekusi sejumlah besar aliran instruksi paralel dengan cepat. Prosesor tujuan umum dioptimalkan untuk mencapai kinerja tinggi pada aliran instruksi tunggal yang memproses bilangan bulat dan angka floating point. Akses memori acak.

Perancang CPU mencoba untuk mendapatkan instruksi sebanyak mungkin untuk dieksekusi secara paralel untuk meningkatkan kinerja. Untuk melakukan ini, dimulai dengan prosesor Intel Pentium, eksekusi superskalar muncul, menyediakan eksekusi dua instruksi per jam, dan Pentium Pro membedakan dirinya dengan eksekusi instruksi yang tidak berurutan. Tetapi eksekusi paralel dari aliran instruksi sekuensial memiliki batasan dasar tertentu, dan dengan meningkatkan jumlah unit eksekusi, peningkatan kecepatan berganda tidak dapat dicapai.

Chip video memiliki operasi yang sederhana dan paralel sejak awal. Chip video mengambil sekelompok poligon pada input, melakukan semua operasi yang diperlukan, dan menghasilkan piksel pada output. Pemrosesan poligon dan piksel bersifat independen, mereka dapat diproses secara paralel, terpisah satu sama lain. Oleh karena itu, karena organisasi kerja paralel yang inheren di GPU, sejumlah besar unit eksekusi digunakan, yang mudah dimuat, berbeda dengan aliran instruksi berurutan untuk CPU. Selain itu, GPU modern juga dapat mengeksekusi lebih dari satu instruksi per jam (masalah ganda). Dengan demikian, arsitektur Tesla, dalam kondisi tertentu, meluncurkan operasi MAD+MUL atau MAD+SFU secara bersamaan.

GPU berbeda dari CPU juga dalam hal prinsip akses memori. Di GPU, ini terhubung dan mudah diprediksi - jika texel tekstur dibaca dari memori, maka setelah beberapa saat waktunya akan tiba untuk texel tetangga. Ya, dan saat merekam yang sama - piksel ditulis ke framebuffer, dan setelah beberapa siklus, yang terletak di sebelahnya akan direkam. Oleh karena itu, organisasi memori berbeda dari yang digunakan di CPU. Dan chip video, tidak seperti prosesor universal, sama sekali tidak memerlukan memori cache yang besar, dan tekstur hanya membutuhkan beberapa kilobyte (hingga 128-256 dalam GPU saat ini).

Dan itu sendiri, bekerja dengan memori untuk GPU dan CPU agak berbeda. Jadi, tidak semua CPU memiliki pengontrol memori internal, dan semua GPU biasanya memiliki beberapa pengontrol, hingga delapan saluran 64-bit dalam chip Nvidia GT200. Selain itu, kartu video menggunakan memori yang lebih cepat, dan sebagai hasilnya, chip video memiliki lebih banyak bandwidth memori yang tersedia, yang juga sangat penting untuk perhitungan paralel yang beroperasi dengan aliran data yang besar.

Dalam prosesor tujuan umum, sejumlah besar transistor dan area chip masuk ke buffer instruksi, prediksi cabang perangkat keras, dan sejumlah besar memori cache on-chip. Semua blok perangkat keras ini diperlukan untuk mempercepat eksekusi beberapa aliran instruksi. Chip video menghabiskan transistor pada array unit eksekusi, unit kontrol aliran, memori bersama kecil, dan pengontrol memori multi-saluran. Hal di atas tidak mempercepat eksekusi masing-masing utas, ini memungkinkan chip memproses beberapa ribu utas yang secara bersamaan dieksekusi pada chip dan membutuhkan bandwidth memori yang tinggi.

Tentang perbedaan dalam caching. CPU tujuan umum menggunakan cache untuk meningkatkan kinerja dengan mengurangi latensi akses memori, sementara GPU menggunakan cache atau memori bersama untuk meningkatkan bandwidth. CPU mengurangi latensi akses memori dengan cache besar dan prediksi cabang kode. Potongan perangkat keras ini mengambil sebagian besar area chip dan menghabiskan banyak daya. Chip video mengatasi masalah penundaan akses memori dengan mengeksekusi ribuan utas secara bersamaan - sementara salah satu utas menunggu data dari memori, chip video dapat melakukan perhitungan utas lain tanpa menunggu dan menunda.

Ada banyak perbedaan dalam dukungan multithreading juga. CPU mengeksekusi 1-2 utas komputasi per inti prosesor, dan chip video dapat mendukung hingga 1024 utas per multiprosesor, di mana ada beberapa di dalam chip. Dan jika beralih dari satu utas ke utas lainnya untuk CPU menghabiskan ratusan siklus, maka GPU mengalihkan beberapa utas dalam satu siklus.

Selain itu, CPU menggunakan blok SIMD (instruksi tunggal, banyak data) untuk komputasi vektor, dan GPU menggunakan SIMT (instruksi tunggal, beberapa utas) untuk pemrosesan utas skalar. SIMT tidak mengharuskan pengembang untuk mengubah data menjadi vektor dan memungkinkan percabangan sewenang-wenang dalam aliran.

Singkatnya, kita dapat mengatakan bahwa, tidak seperti CPU universal modern, chip video dirancang untuk komputasi paralel dengan sejumlah besar operasi aritmatika. Dan jumlah transistor GPU yang jauh lebih besar bekerja untuk tujuan yang dimaksudkan - pemrosesan array data, dan tidak mengontrol eksekusi (kontrol aliran) dari beberapa utas komputasi berurutan. Ini adalah diagram berapa banyak ruang di CPU dan GPU membutuhkan berbagai logika:

Akibatnya, dasar untuk penggunaan kekuatan GPU yang efektif dalam perhitungan ilmiah dan non-grafis lainnya adalah paralelisasi algoritme menjadi ratusan unit eksekusi yang tersedia dalam chip video. Misalnya, banyak aplikasi pemodelan molekul sangat cocok untuk perhitungan pada chip video, mereka membutuhkan daya komputasi yang besar dan oleh karena itu nyaman untuk komputasi paralel. Dan penggunaan beberapa GPU memberikan lebih banyak daya komputasi untuk memecahkan masalah tersebut.

Melakukan perhitungan pada GPU menunjukkan hasil yang sangat baik dalam algoritme yang menggunakan pemrosesan data paralel. Yaitu, ketika urutan operasi matematika yang sama diterapkan pada sejumlah besar data. Dalam hal ini, hasil terbaik dicapai jika rasio jumlah instruksi aritmatika dengan jumlah akses memori cukup besar. Ini menempatkan lebih sedikit tuntutan pada kontrol aliran, dan kepadatan matematika yang tinggi dan sejumlah besar data menghilangkan kebutuhan akan cache yang besar, seperti pada CPU.

Sebagai hasil dari semua perbedaan yang dijelaskan di atas, kinerja teoritis chip video secara signifikan melebihi kinerja CPU. Nvidia memberikan grafik pertumbuhan kinerja CPU dan GPU berikut selama beberapa tahun terakhir:

Secara alami, data ini bukan tanpa bagian dari kelicikan. Memang, pada CPU jauh lebih mudah untuk mencapai angka teoretis dalam praktik, dan angka diberikan untuk presisi tunggal dalam kasus GPU, dan untuk presisi ganda dalam kasus CPU. Bagaimanapun, presisi tunggal sudah cukup untuk beberapa tugas paralel, dan perbedaan kecepatan antara prosesor universal dan grafis sangat besar, dan oleh karena itu gim ini layak untuk dinyalakan.

Upaya pertama untuk menerapkan perhitungan pada GPU

Chip video telah digunakan dalam perhitungan matematika paralel untuk waktu yang lama. Upaya pertama pada aplikasi semacam itu sangat primitif dan terbatas pada penggunaan beberapa fitur perangkat keras, seperti rasterisasi dan buffering Z. Tetapi di abad ini, dengan munculnya shader, mereka mulai mempercepat perhitungan matriks. Pada tahun 2003, di SIGGRAPH, bagian terpisah dialokasikan untuk komputasi GPU, dan itu disebut GPGPU (Komputasi Tujuan Umum pada GPU) - komputasi GPU universal).

BrookGPU yang paling terkenal adalah kompiler bahasa pemrograman aliran Brook, yang dirancang untuk melakukan perhitungan non-grafis pada GPU. Sebelum kemunculannya, pengembang yang menggunakan kemampuan chip video untuk perhitungan memilih salah satu dari dua API umum: Direct3D atau OpenGL. Ini sangat membatasi penggunaan GPU, karena grafik 3D menggunakan shader dan tekstur yang tidak perlu diketahui oleh programmer paralel, mereka menggunakan utas dan inti. Brook dapat membantu membuat tugas mereka lebih mudah. Ekstensi streaming ke bahasa C ini, yang dikembangkan di Universitas Stanford, menyembunyikan API 3D dari pemrogram dan menampilkan chip video sebagai koprosesor paralel. Kompilator mem-parsing file .br dengan kode dan ekstensi C++, menghasilkan kode yang ditautkan ke pustaka yang mendukung DirectX, OpenGL, atau x86.

Tentu saja, Brook memiliki banyak kekurangan, yang kami permasalahkan dan yang akan kami bahas lebih detail nanti. Tetapi bahkan penampilannya saja telah menyebabkan gelombang perhatian yang signifikan dari Nvidia dan ATI yang sama ke inisiatif komputasi GPU, karena pengembangan kemampuan ini secara serius mengubah pasar di masa depan, membuka sektor yang sama sekali baru - komputasi paralel berdasarkan chip video.

Selanjutnya, beberapa peneliti dari proyek Brook bergabung dengan tim pengembangan Nvidia untuk memperkenalkan strategi komputasi paralel perangkat keras-perangkat lunak, membuka pangsa pasar baru. Dan keuntungan utama dari inisiatif Nvidia ini adalah pengembang mengetahui dengan sempurna semua kemampuan GPU mereka hingga detail terkecil, dan tidak perlu menggunakan API grafis, dan Anda dapat bekerja dengan perangkat keras secara langsung menggunakan driver. Hasil dari upaya tim ini adalah Nvidia CUDA (Compute Unified Device Architecture), arsitektur perangkat keras dan perangkat lunak baru untuk komputasi paralel pada GPU Nvidia, yang menjadi pokok bahasan artikel ini.

Area penerapan komputasi paralel pada GPU

Untuk memahami keuntungan apa yang dibawa oleh transfer perhitungan ke chip video, kami akan menyajikan angka rata-rata yang diperoleh oleh para peneliti di seluruh dunia. Rata-rata, saat mentransfer perhitungan ke GPU, dalam banyak tugas, akselerasi dicapai 5-30 kali lipat dibandingkan dengan prosesor universal yang cepat. Angka terbesar (dari urutan kecepatan 100x dan bahkan lebih!) dicapai pada kode yang tidak terlalu cocok untuk perhitungan menggunakan blok SSE, tetapi cukup nyaman untuk GPU.

Ini hanya beberapa contoh percepatan kode sintetis pada GPU versus kode vektor SSE pada CPU (menurut Nvidia):

• Mikroskop fluoresensi: 12x;
• Dinamika molekul (kalkulasi gaya tak terikat): 8-16x;
• Elektrostatika (penjumlahan Coulomb langsung dan multi-level): 40-120x dan 7x.

Dan ini adalah piring yang sangat disukai Nvidia, menunjukkannya di semua presentasi, yang akan kita bahas secara lebih rinci di bagian kedua artikel, yang dikhususkan untuk contoh spesifik aplikasi praktis komputasi CUDA:

Seperti yang Anda lihat, jumlahnya sangat menarik, terutama keuntungan 100-150 kali lipat yang mengesankan. Dalam artikel CUDA berikutnya, kita akan melihat lebih dekat beberapa angka ini. Dan sekarang kami membuat daftar aplikasi utama di mana komputasi GPU sekarang digunakan: analisis dan pemrosesan gambar dan sinyal, simulasi fisika, matematika komputasi, biologi komputasi, perhitungan keuangan, database, dinamika gas dan cairan, kriptografi, terapi radiasi adaptif, astronomi , pemrosesan suara, bioinformatika, simulasi biologi, visi komputer, penambangan data, sinema digital dan televisi, simulasi elektromagnetik, sistem informasi geografis, aplikasi militer, perencanaan penambangan, dinamika molekul, pencitraan resonansi magnetik (MRI), jaringan saraf, penelitian oseanografi, partikel fisika, simulasi pelipatan protein, kimia kuantum, ray tracing, pencitraan, radar, simulasi reservoir, kecerdasan buatan, analisis data satelit, eksplorasi seismik, pembedahan, ultrasound, konferensi video.

Detail dari banyak aplikasi dapat ditemukan di situs web Nvidia di bagian . Seperti yang Anda lihat, daftarnya cukup besar, tapi bukan itu saja! Ini dapat dilanjutkan, dan kita pasti dapat berasumsi bahwa di masa depan area lain dari penerapan perhitungan paralel pada chip video akan ditemukan, yang masih belum kita ketahui.

Kemampuan Nvidia CUDA

Teknologi CUDA adalah arsitektur komputasi perangkat lunak dan perangkat keras Nvidia berdasarkan ekstensi bahasa C, yang memungkinkan untuk mengakses set instruksi akselerator grafis dan mengelola memorinya dalam komputasi paralel. CUDA membantu mengimplementasikan algoritme yang dapat diimplementasikan pada prosesor grafis akselerator video Geforce generasi kedelapan dan yang lebih lama (seri Geforce 8, Geforce 9, Geforce 200), serta Quadro dan Tesla.

Meskipun kompleksitas pemrograman GPU dengan CUDA cukup tinggi, namun lebih rendah dibandingkan dengan solusi GPGPU awal. Program tersebut memerlukan partisi aplikasi di beberapa multiprosesor yang mirip dengan pemrograman MPI, tetapi tanpa berbagi data yang disimpan dalam memori video bersama. Dan karena pemrograman CUDA untuk setiap multiprosesor mirip dengan pemrograman OpenMP, maka diperlukan pemahaman yang baik tentang organisasi memori. Namun, tentu saja, kerumitan pengembangan dan porting ke CUDA sangat bergantung pada aplikasinya.

Kit pengembang berisi banyak contoh kode dan didokumentasikan dengan baik. Proses pembelajaran akan memakan waktu sekitar dua hingga empat minggu bagi mereka yang sudah terbiasa dengan OpenMP dan MPI. API didasarkan pada bahasa C yang diperluas, dan untuk menerjemahkan kode dari bahasa ini, CUDA SDK menyertakan kompiler baris perintah nvcc, berdasarkan kompiler Open64 terbuka.

Kami mencantumkan karakteristik utama CUDA:

• solusi perangkat lunak dan perangkat keras terpadu untuk komputasi paralel pada chip video Nvidia;
• berbagai solusi yang didukung, dari seluler hingga multi-chip
• bahasa pemrograman C standar;
• pustaka standar untuk analisis numerik FFT (Fast Fourier Transform) dan BLAS (Aljabar Linier);
• pertukaran data yang dioptimalkan antara CPU dan GPU;
• interaksi dengan API grafis OpenGL dan DirectX;
• dukungan untuk sistem operasi 32- dan 64-bit: Windows XP, Windows Vista, Linux dan MacOS X;
• kemampuan untuk berkembang pada tingkat yang rendah.

Mengenai dukungan sistem operasi, harus ditambahkan bahwa semua distribusi Linux utama secara resmi didukung (Red Hat Enterprise Linux 3.x/4.x/5.x, SUSE Linux 10.x), tetapi, menurut para penggemar, CUDA berfungsi dengan baik di build lain: Fedora Core, Ubuntu, Gentoo, dll.

Lingkungan Pengembangan CUDA (CUDA Toolkit) meliputi:

• kompiler nvcc;
• perpustakaan FFT dan BLAS;
• pembuat profil;
• debugger gdb untuk GPU;
• Driver runtime CUDA disertakan dengan driver Nvidia standar
• panduan pemrograman;
• CUDA Developer SDK (kode sumber, utilitas, dan dokumentasi).

Dalam contoh kode sumber: penyortiran bitonic paralel (pengurutan bitonic), transposisi matriks, penjumlahan awalan paralel dari array besar, konvolusi gambar, transformasi wavelet diskrit, contoh interaksi dengan OpenGL dan Direct3D, penggunaan pustaka CUBLAS dan Cufft, perhitungan harga opsi ( Black-Scholes, model binomial, metode Monte Carlo), generator bilangan acak paralel Mersenne Twister, perhitungan histogram array besar, pengurangan noise, filter Sobel (deteksi terikat).

Manfaat dan Keterbatasan CUDA

Dari sudut pandang programmer, pipa grafis adalah serangkaian tahapan pemrosesan. Blok geometri menghasilkan segitiga, dan blok rasterisasi menghasilkan piksel yang ditampilkan di monitor. Model pemrograman GPGPU tradisional adalah sebagai berikut:

Untuk mentransfer komputasi ke GPU dalam kerangka model seperti itu, diperlukan pendekatan khusus. Bahkan penambahan elemen demi elemen dari dua vektor akan membutuhkan menggambar bentuk ke layar atau ke buffer di luar layar. Angka tersebut dirasterisasi, warna setiap piksel dihitung sesuai dengan program yang diberikan (pixel shader). Program membaca data input dari tekstur untuk setiap piksel, menambahkannya, dan menulisnya ke buffer output. Dan semua banyak operasi ini diperlukan untuk apa yang ditulis dalam satu operator dalam bahasa pemrograman konvensional!

Oleh karena itu, penggunaan GPGPU untuk komputasi tujuan umum memiliki keterbatasan berupa terlalu banyak kerumitan untuk dipelajari oleh pengembang. Dan ada batasan lain yang cukup, karena pixel shader hanyalah formula untuk ketergantungan warna akhir piksel pada koordinatnya, dan bahasa pixel shader adalah bahasa untuk menulis formula ini dengan sintaks seperti C. Metode GPGPU awal adalah trik cerdas untuk memanfaatkan kekuatan GPU, tetapi tanpa kenyamanan apa pun. Data yang ada direpresentasikan dengan gambar (tekstur), dan algoritma direpresentasikan dengan proses rasterisasi. Perlu dicatat dan model yang sangat spesifik dari memori dan eksekusi.

Arsitektur perangkat keras dan perangkat lunak Nvidia untuk komputasi pada GPU dari Nvidia berbeda dari model GPGPU sebelumnya karena memungkinkan penulisan program untuk GPU dalam C nyata dengan sintaks standar, pointer, dan kebutuhan akan ekstensi minimum untuk mengakses sumber daya komputasi chip video. CUDA tidak bergantung pada API grafis, dan memiliki beberapa fitur yang dirancang khusus untuk komputasi tujuan umum.

Keuntungan CUDA dibandingkan pendekatan tradisional untuk komputasi GPGPU:

• antarmuka pemrograman aplikasi CUDA didasarkan pada bahasa pemrograman C standar dengan ekstensi, yang menyederhanakan proses pembelajaran dan penerapan arsitektur CUDA;
• CUDA menyediakan akses ke 16 KB memori bersama per multiprosesor, yang dapat digunakan untuk mengatur cache dengan bandwidth yang lebih tinggi daripada pengambilan tekstur;
• transfer data yang lebih efisien antara sistem dan memori video
• tidak perlu API grafis dengan redundansi dan overhead;
• pengalamatan memori linier, dan mengumpulkan dan menyebarkan, kemampuan untuk menulis ke alamat sewenang-wenang;
• dukungan perangkat keras untuk operasi integer dan bit.

Keterbatasan utama CUDA:

• kurangnya dukungan rekursi untuk fungsi yang dapat dieksekusi;
• lebar blok minimum adalah 32 utas;
• arsitektur CUDA eksklusif yang dimiliki oleh Nvidia.

Kelemahan pemrograman dengan metode GPGPU sebelumnya adalah metode ini tidak menggunakan unit eksekusi vertex shader pada arsitektur non-unified sebelumnya, data disimpan dalam tekstur dan output ke buffer off-screen, dan algoritma multi-pass menggunakan unit pixel shader. Keterbatasan GPGPU meliputi: penggunaan kemampuan perangkat keras yang tidak cukup efisien, keterbatasan bandwidth memori, tidak ada operasi pencar (hanya pengumpulan), penggunaan wajib API grafis.

Keuntungan utama CUDA dibandingkan metode GPGPU sebelumnya berasal dari fakta bahwa arsitektur ini dirancang untuk secara efisien menggunakan komputasi non-grafis pada GPU dan menggunakan bahasa pemrograman C tanpa memerlukan algoritme untuk porting ke bentuk yang sesuai untuk konsep grafis. pipa. CUDA menawarkan jalur komputasi GPU baru yang tidak menggunakan API grafis, menawarkan akses memori acak (menyebarkan atau mengumpulkan). Arsitektur seperti itu bebas dari kelemahan GPGPU dan menggunakan semua unit eksekusi, dan juga memperluas kemampuan melalui matematika bilangan bulat dan operasi pergeseran bit.

Selain itu, CUDA membuka beberapa fitur perangkat keras yang tidak tersedia dari API grafis, seperti memori bersama. Ini adalah sejumlah kecil memori (16 kilobyte per multiprosesor) yang dapat diakses oleh blok utas. Ini memungkinkan Anda untuk men-cache data yang paling sering diakses dan dapat memberikan kinerja yang lebih cepat daripada menggunakan pengambilan tekstur untuk tugas ini. Ini, pada gilirannya, mengurangi sensitivitas throughput dari algoritma paralel di banyak aplikasi. Misalnya, berguna untuk aljabar linier, transformasi Fourier cepat, dan filter pemrosesan gambar.

Lebih nyaman di CUDA dan akses memori. Kode dalam API grafis mengeluarkan data sebagai 32 nilai floating-point presisi tunggal (nilai RGBA secara bersamaan ke delapan target render) di area yang telah ditentukan, dan CUDA mendukung perekaman sebar - jumlah catatan yang tidak terbatas di alamat mana pun. Keunggulan tersebut memungkinkan untuk mengeksekusi beberapa algoritma pada GPU yang tidak dapat diimplementasikan secara efisien menggunakan metode GPGPU berdasarkan API grafis.

Juga, API grafis harus menyimpan data dalam tekstur, yang membutuhkan pengemasan sebelumnya dari array besar ke dalam tekstur, yang memperumit algoritme dan memaksa penggunaan pengalamatan khusus. Dan CUDA memungkinkan Anda membaca data di alamat mana pun. Keuntungan lain dari CUDA adalah komunikasi yang dioptimalkan antara CPU dan GPU. Dan untuk pengembang yang ingin mengakses level rendah (misalnya, saat menulis bahasa pemrograman lain), CUDA menawarkan kemungkinan pemrograman bahasa assembly level rendah.

Sejarah perkembangan CUDA

Pengembangan CUDA diumumkan dengan chip G80 pada November 2006, dan versi beta publik dari CUDA SDK dirilis pada Februari 2007. Versi 1.0 dirilis pada Juni 2007 untuk meluncurkan solusi Tesla berdasarkan chip G80 untuk pasar komputasi kinerja tinggi. Kemudian, pada akhir tahun, CUDA 1.1 beta dirilis, yang meskipun sedikit meningkat dalam jumlah versi, memperkenalkan cukup banyak hal baru.

Dari apa yang muncul di CUDA 1.1, kami dapat mencatat penyertaan fungsionalitas CUDA di driver video Nvidia biasa. Ini berarti bahwa dalam persyaratan untuk program CUDA apa pun, cukup untuk menentukan kartu video seri Geforce 8 dan lebih tinggi, serta versi driver minimum 169.xx. Ini sangat penting bagi pengembang, jika kondisi ini terpenuhi, program CUDA akan berfungsi untuk pengguna mana pun. Selain itu, eksekusi asinkron ditambahkan bersama dengan penyalinan data (hanya untuk G84, G86, G92 dan chip yang lebih tinggi), transfer data asinkron ke memori video, operasi akses memori atom, dukungan untuk Windows versi 64-bit, dan kemungkinan multi -operasi CUDA chip dalam mode SLI.

Saat ini, versi untuk solusi berbasis GT200 adalah CUDA 2.0, yang dirilis bersama dengan lini Geforce GTX 200. Versi beta dirilis pada musim semi 2008. Versi kedua memiliki: dukungan untuk perhitungan presisi ganda (hanya dukungan perangkat keras untuk GT200), Windows Vista (versi 32 dan 64-bit) dan Mac OS X akhirnya didukung, alat debugging dan profil telah ditambahkan, tekstur 3D didukung, dioptimalkan transfer data.

Sedangkan untuk perhitungan dengan presisi ganda, kecepatannya pada generasi perangkat keras saat ini beberapa kali lebih rendah daripada presisi tunggal. Alasan dibahas di kami. Implementasi dukungan ini di GT200 terletak pada kenyataan bahwa blok FP32 tidak digunakan untuk mendapatkan hasil dengan kecepatan empat kali lebih lambat, untuk mendukung perhitungan FP64, Nvidia memutuskan untuk membuat blok komputasi khusus. Dan di GT200 jumlahnya sepuluh kali lebih sedikit daripada blok FP32 (satu blok presisi ganda untuk setiap multiprosesor).

Kenyataannya, kinerjanya bisa lebih rendah lagi, karena arsitekturnya dioptimalkan untuk pembacaan 32-bit dari memori dan register, selain itu, presisi ganda tidak diperlukan dalam aplikasi grafis, dan di GT200 itu dibuat lebih mungkin. Ya, dan prosesor quad-core modern menunjukkan kinerja yang tidak kalah nyata. Tetapi bahkan 10 kali lebih lambat dari presisi tunggal, dukungan ini berguna untuk sirkuit presisi campuran. Salah satu teknik umum adalah untuk mendapatkan hasil perkiraan awalnya dalam presisi tunggal, dan kemudian memperbaikinya dalam presisi ganda. Sekarang ini dapat dilakukan langsung pada kartu video, tanpa mengirimkan data perantara ke CPU.

Fitur lain yang berguna dari CUDA 2.0 tidak ada hubungannya dengan GPU, anehnya. Sekarang mungkin untuk mengkompilasi kode CUDA menjadi kode SSE multi-utas yang sangat efisien untuk eksekusi cepat pada CPU. Artinya, sekarang fitur ini cocok tidak hanya untuk debugging, tetapi juga untuk penggunaan nyata pada sistem tanpa kartu video Nvidia. Bagaimanapun, penggunaan CUDA dalam kode normal dibatasi oleh fakta bahwa kartu video Nvidia, meskipun yang paling populer di antara solusi video khusus, tidak tersedia di semua sistem. Dan sebelum versi 2.0, dalam kasus seperti itu, dua kode berbeda harus ditulis: untuk CUDA dan secara terpisah untuk CPU. Dan sekarang Anda dapat menjalankan program CUDA apa pun pada CPU dengan efisiensi tinggi, meskipun pada kecepatan yang lebih rendah daripada pada chip video.

Solusi yang Didukung Nvidia CUDA

Semua kartu grafis berkemampuan CUDA dapat membantu mempercepat tugas yang paling menuntut, mulai dari pemrosesan audio dan video hingga penelitian medis dan ilmiah. Satu-satunya batasan nyata adalah bahwa banyak program CUDA memerlukan minimal 256 megabyte memori video, dan ini adalah salah satu spesifikasi paling penting untuk aplikasi CUDA.

Daftar terbaru produk yang mendukung CUDA dapat ditemukan di . Pada saat penulisan ini, perhitungan CUDA mendukung semua produk seri Geforce 200, Geforce 9 dan Geforce 8, termasuk produk seluler, dimulai dengan Geforce 8400M, serta chipset Geforce 8100, 8200, dan 8300. Quadro modern dan semua Tesla: S1070, C1060, C870, D870 dan S870.

Kami secara khusus mencatat bahwa bersama dengan kartu video Geforce GTX 260 dan 280 yang baru, solusi komputasi kinerja tinggi yang sesuai diumumkan: Tesla C1060 dan S1070 (ditunjukkan pada foto di atas), yang akan tersedia untuk pembelian musim gugur ini. GPU yang sama digunakan di dalamnya - GT200, di C1060 itu satu, di S1070 - empat. Namun, tidak seperti solusi game, mereka menggunakan empat gigabyte memori per chip. Dari kekurangannya, mungkin frekuensi memori dan bandwidth memori yang lebih rendah daripada kartu game, menyediakan 102 gigabyte / s per chip.

Komposisi Nvidia CUDA

CUDA mencakup dua API: tingkat tinggi (CUDA Runtime API) dan tingkat rendah (CUDA Driver API), meskipun tidak mungkin untuk menggunakan keduanya secara bersamaan dalam satu program, Anda harus menggunakan satu atau yang lain. Yang tingkat tinggi bekerja "di atas" dari yang tingkat rendah, semua panggilan runtime diterjemahkan ke dalam instruksi sederhana yang diproses oleh API Driver tingkat rendah. Tetapi bahkan API "tingkat tinggi" mengasumsikan pengetahuan tentang desain dan pengoperasian chip video Nvidia; tidak ada tingkat abstraksi yang terlalu tinggi di sana.

Ada level lain, bahkan lebih tinggi - dua perpustakaan:

CUBLAS- Versi CUDA dari BLAS (Subprogram Aljabar Linier Dasar), dirancang untuk menghitung masalah aljabar linier dan menggunakan akses langsung ke sumber daya GPU;

manset- Versi CUDA dari pustaka Fast Fourier Transform untuk menghitung Fast Fourier Transform, banyak digunakan dalam pemrosesan sinyal. Jenis transformasi berikut didukung: kompleks-kompleks (C2C), real-kompleks (R2C), dan kompleks-nyata (C2R).

Mari kita lihat lebih dekat perpustakaan ini. CUBLAS adalah algoritma aljabar linier standar yang diterjemahkan ke dalam bahasa CUDA, saat ini hanya beberapa fungsi dasar CUBLAS yang didukung. Pustaka ini sangat mudah digunakan: Anda perlu membuat objek matriks dan vektor dalam memori video, mengisinya dengan data, memanggil fungsi CUBLAS yang diperlukan, dan memuat kembali hasil dari memori video ke dalam memori sistem. CUBLAS berisi fungsi khusus untuk membuat dan menghancurkan objek di memori GPU, serta untuk membaca dan menulis data ke memori ini. Fungsi BLAS yang didukung: level 1, 2 dan 3 untuk bilangan real, level 1 CGEMM untuk kompleks. Level 1 adalah operasi vektor-vektor, level 2 adalah operasi matriks-vektor, level 3 adalah operasi matriks-matriks.

Cufft - varian CUDA dari Fast Fourier Transform - banyak digunakan dan sangat penting dalam analisis sinyal, penyaringan, dll. Cufft menyediakan antarmuka sederhana untuk komputasi FFT yang efisien pada GPU Nvidia tanpa perlu mengembangkan FFT khusus untuk GPU. Varian CUDA FFT mendukung transformasi 1D, 2D, dan 3D dari data yang kompleks dan nyata, eksekusi batch untuk beberapa transformasi 1D secara paralel, ukuran transformasi 2D dan 3D dapat berada di dalam , untuk 1D ukuran hingga 8 juta elemen didukung.

Dasar-dasar membuat program di CUDA

Untuk memahami teks di bawah ini, Anda harus memahami fitur arsitektur dasar chip video Nvidia. GPU terdiri dari beberapa cluster unit tekstur (Texture Processing Cluster). Setiap cluster terdiri dari blok pengambilan tekstur yang diperbesar dan dua atau tiga multiprosesor streaming, yang masing-masing terdiri dari delapan perangkat komputasi dan dua blok superfungsional. Semua instruksi dieksekusi sesuai dengan prinsip SIMD, ketika satu instruksi diterapkan ke semua utas dalam warp (istilah dari industri tekstil, di CUDA ini adalah sekelompok 32 utas - jumlah minimum data yang diproses oleh multiprosesor). Metode eksekusi ini disebut SIMT (satu instruksi banyak utas - satu instruksi dan banyak utas).

Setiap multiprosesor memiliki sumber daya tertentu. Jadi, ada memori bersama khusus dengan kapasitas 16 kilobyte per multiprosesor. Tetapi ini bukan cache, karena programmer dapat menggunakannya untuk kebutuhan apa pun, mirip dengan Penyimpanan Lokal di SPU prosesor Sel. Memori bersama ini memungkinkan pertukaran informasi antara utas dari blok yang sama. Adalah penting bahwa semua utas dari satu blok selalu dieksekusi oleh multiprosesor yang sama. Dan utas dari blok yang berbeda tidak dapat bertukar data, dan Anda harus mengingat batasan ini. Memori bersama seringkali berguna, kecuali ketika beberapa utas mengakses bank memori yang sama. Multiprosesor juga dapat mengakses memori video, tetapi dengan latensi yang lebih tinggi dan bandwidth yang lebih rendah. Untuk mempercepat akses dan mengurangi frekuensi mengakses memori video, multiprosesor memiliki 8 kilobyte cache untuk konstanta dan data tekstur.

Multiprosesor menggunakan 8192-16384 (untuk G8x/G9x dan GT2xx, masing-masing) register umum untuk semua utas dari semua blok yang mengeksekusinya. Jumlah maksimum blok per multiprosesor untuk G8x/G9x adalah delapan, dan jumlah warps adalah 24 (768 utas per multiprosesor). Secara total, kartu video teratas dari seri Geforce 8 dan 9 dapat memproses hingga 12288 utas sekaligus. GeForce GTX 280 berdasarkan GT200 menawarkan hingga 1024 thread per multiprosesor, ia memiliki 10 cluster dari tiga multiprosesor yang memproses hingga 30720 thread. Mengetahui batasan ini memungkinkan Anda mengoptimalkan algoritme untuk sumber daya yang tersedia.

Langkah pertama dalam mem-porting aplikasi yang ada ke CUDA adalah membuat profilnya dan mengidentifikasi area kode yang merupakan hambatan yang memperlambat kerja. Jika di antara bagian-bagian tersebut ada yang cocok untuk eksekusi paralel cepat, fungsi-fungsi ini ditransfer ke ekstensi C dan CUDA untuk dieksekusi pada GPU. Program ini dikompilasi menggunakan compiler yang disediakan Nvidia, yang menghasilkan kode untuk CPU dan GPU. Ketika sebuah program dijalankan, CPU mengeksekusi bagian kodenya, dan GPU mengeksekusi kode CUDA dengan komputasi paralel terberat. Bagian ini, dirancang untuk GPU, disebut kernel (kernel). Kernel mendefinisikan operasi yang akan dilakukan pada data.

Chip video menerima inti dan membuat salinan untuk setiap elemen data. Salinan ini disebut utas. Sebuah stream berisi counter, register, dan state. Untuk data dalam jumlah besar, seperti pemrosesan gambar, jutaan utas diluncurkan. Utas berjalan dalam kelompok 32 yang disebut warps. Warps ditugaskan untuk berjalan pada multiprosesor streaming tertentu. Setiap multiprosesor terdiri dari delapan inti - prosesor aliran yang mengeksekusi satu instruksi MAD per siklus clock. Untuk mengeksekusi satu warp 32-utas, diperlukan empat siklus multiprosesor (kita berbicara tentang frekuensi domain shader, yaitu 1,5 GHz dan lebih tinggi).

Multiprosesor bukanlah prosesor multi-core tradisional, ini sangat cocok untuk multi-threading, mendukung hingga 32 warps pada satu waktu.Setiap siklus clock, perangkat keras memilih warps mana yang akan dieksekusi, dan beralih dari satu ke yang lain tanpa kehilangan siklus. Jika kita menggambar analogi dengan prosesor pusat, ini seperti menjalankan 32 program pada saat yang sama dan beralih di antara mereka setiap siklus jam tanpa kehilangan sakelar konteks. Pada kenyataannya, inti CPU mendukung eksekusi simultan dari satu program dan beralih ke yang lain dengan penundaan ratusan siklus.

Model Pemrograman CUDA

Sekali lagi, CUDA menggunakan model komputasi paralel, di mana masing-masing prosesor SIMD mengeksekusi instruksi yang sama pada item data yang berbeda secara paralel. GPU adalah perangkat komputasi, coprocessor (perangkat) untuk prosesor pusat (host), yang memiliki memori sendiri dan memproses sejumlah besar utas secara paralel. Kernel (kernel) adalah fungsi untuk GPU, dijalankan oleh utas (analogi dari grafik 3D - shader).

Kami mengatakan di atas bahwa chip video berbeda dari CPU karena dapat memproses puluhan ribu utas secara bersamaan, yang biasanya untuk grafik yang diparalelkan dengan baik. Setiap aliran adalah skalar, tidak memerlukan data untuk dikemas ke dalam vektor 4 komponen, yang lebih nyaman untuk sebagian besar tugas. Jumlah utas logis dan blok utas melebihi jumlah unit eksekusi fisik, yang memberikan skalabilitas yang baik untuk seluruh rentang solusi perusahaan.

Model pemrograman di CUDA mengasumsikan pengelompokan thread. Utas digabungkan menjadi blok utas - kisi satu atau dua dimensi dari utas yang berinteraksi satu sama lain menggunakan memori bersama dan titik sinkronisasi. Program (kernel) dijalankan di atas kotak blok utas, lihat gambar di bawah. Satu grid dieksekusi pada waktu yang sama. Setiap blok dapat berbentuk satu, dua, atau tiga dimensi, dan dapat terdiri dari 512 utas pada perangkat keras saat ini.

Blok utas berjalan dalam kelompok kecil yang disebut warps, yang berukuran 32 utas. Ini adalah jumlah minimum data yang dapat diproses dalam multiprosesor. Dan karena ini tidak selalu nyaman, CUDA memungkinkan Anda untuk bekerja dengan blok yang berisi 64 hingga 512 utas.

Mengelompokkan blok ke dalam kisi memungkinkan Anda untuk keluar dari batasan dan menerapkan kernel ke sejumlah besar utas dalam satu panggilan. Ini juga membantu dengan penskalaan. Jika GPU tidak memiliki sumber daya yang cukup, ia akan mengeksekusi blok secara berurutan. Jika tidak, blok dapat dieksekusi secara paralel, yang penting untuk distribusi kerja yang optimal pada chip video dari berbagai level, mulai dari yang seluler dan terintegrasi.

Model memori CUDA

Model memori di CUDA dibedakan oleh kemungkinan pengalamatan byte, dukungan untuk pengumpulan dan pencar. Jumlah register yang cukup besar tersedia untuk setiap stream processor, hingga 1024 buah. Akses ke mereka sangat cepat, Anda dapat menyimpan bilangan bulat 32-bit atau angka floating point di dalamnya.

Setiap utas memiliki akses ke jenis memori berikut:

memori global- jumlah memori terbesar yang tersedia untuk semua multiprosesor pada chip video, ukurannya berkisar dari 256 megabyte hingga 1,5 gigabyte untuk solusi saat ini (dan hingga 4 GB untuk Tesla). Ini memiliki throughput yang tinggi, lebih dari 100 gigabyte / s untuk solusi Nvidia teratas, tetapi penundaan yang sangat besar hingga beberapa ratus siklus. Tidak dapat di-cache, mendukung instruksi pemuatan dan penyimpanan umum, dan pointer memori biasa.

memori lokal adalah sejumlah kecil memori yang hanya dapat diakses oleh satu prosesor aliran. Ini relatif lambat - sama dengan yang global.

Berbagi memori adalah blok memori 16-kilobyte (dalam chip video arsitektur saat ini) dengan akses bersama untuk semua prosesor aliran dalam multiprosesor. Memori ini sangat cepat, sama seperti register. Ini menyediakan interaksi utas, dikelola langsung oleh pengembang, dan memiliki latensi rendah. Keuntungan dari memori bersama: digunakan dalam bentuk cache tingkat pertama yang dikelola oleh programmer, mengurangi penundaan pengaksesan data oleh unit eksekusi (ALU), mengurangi jumlah akses memori global.

Memori konstan- area memori 64 kilobyte (sama untuk GPU saat ini), hanya-baca oleh semua multiprosesor. Itu di-cache pada 8 kilobyte per multiprosesor. Cukup lambat - penundaan beberapa ratus siklus tanpa adanya data yang diperlukan dalam cache.

memori tekstur- blok memori yang tersedia untuk dibaca oleh semua multiprosesor. Pengambilan sampel data dilakukan dengan menggunakan unit tekstur chip video, sehingga kemungkinan interpolasi data linier disediakan tanpa biaya tambahan. 8 kilobyte di-cache per multiprosesor. Lambat seperti global - ratusan siklus penundaan karena tidak adanya data dalam cache.

Secara alami, memori global, lokal, tekstur, dan konstan secara fisik adalah memori yang sama, yang dikenal sebagai memori video lokal kartu video. Perbedaannya terletak pada algoritma caching dan model akses yang berbeda. CPU hanya dapat memperbarui dan meminta memori eksternal: global, konstan, dan tekstur.

Dari apa yang telah ditulis di atas, jelas bahwa CUDA menyiratkan pendekatan khusus untuk pengembangan, tidak persis sama dengan yang diadopsi dalam program untuk CPU. Anda perlu ingat tentang berbagai jenis memori, bahwa memori lokal dan global tidak di-cache dan penundaan dalam mengaksesnya jauh lebih tinggi daripada memori terdaftar, karena secara fisik terletak di sirkuit mikro yang terpisah.

Pola pemecahan masalah yang khas, tetapi tidak wajib:

• tugas dibagi menjadi subtugas;
• data input dibagi menjadi blok-blok yang sesuai dengan memori bersama;
• setiap blok diproses oleh blok utas;
• subblok dimuat ke memori bersama dari yang global;
• perhitungan yang sesuai dilakukan pada data dalam memori bersama;
• hasilnya disalin dari memori bersama kembali ke global.

Lingkungan pemrograman

CUDA termasuk perpustakaan runtime:

• bagian umum yang menyediakan tipe vektor bawaan dan subset panggilan RTL yang didukung pada CPU dan GPU;
• Komponen CPU, untuk mengontrol satu atau lebih GPU;
• Komponen GPU yang menyediakan fungsionalitas khusus GPU.

Proses utama aplikasi CUDA berjalan pada prosesor generik (host), menjalankan beberapa salinan dari proses kernel pada kartu video. Kode untuk CPU melakukan hal berikut: menginisialisasi GPU, mengalokasikan memori pada kartu video dan sistem, menyalin konstanta ke memori kartu video, menjalankan beberapa salinan proses kernel pada kartu video, menyalin hasil dari memori video, membebaskan memori dan keluar.

Sebagai contoh untuk pemahaman, berikut adalah kode CPU untuk penjumlahan vektor yang disajikan dalam CUDA:

Fungsi yang dijalankan oleh chip video memiliki batasan sebagai berikut: tidak ada rekursi, tidak ada variabel statis di dalam fungsi, dan tidak ada jumlah variabel argumen. Dua jenis manajemen memori yang didukung: memori linier diakses oleh pointer 32-bit, dan array CUDA diakses hanya melalui fungsi tekstur fetching.

Program CUDA dapat berinteraksi dengan API grafis: untuk merender data yang dihasilkan dalam program, untuk membaca hasil rendering dan memprosesnya menggunakan alat CUDA (misalnya, saat menerapkan filter pascapemrosesan). Untuk melakukan ini, sumber daya API grafis dapat dipetakan (memperoleh alamat sumber daya) ke dalam ruang memori global CUDA. Jenis sumber daya API grafis berikut didukung: Objek Buffer (PBO / VBO) di OpenGL, Buffer Vertex dan tekstur (2D, 3D, dan peta kubus) Direct3D9.

Langkah-langkah kompilasi aplikasi CUDA:

File kode sumber CUDA C dikompilasi menggunakan program NVCC, yang membungkus alat lain dan memanggilnya: cudacc, g++, cl, dll. NVCC menghasilkan: kode CPU yang dikompilasi bersama dengan aplikasi lainnya yang ditulis dalam C murni, dan Kode objek PTX untuk chip video. File yang dapat dieksekusi dengan kode CUDA memerlukan keberadaan perpustakaan runtime CUDA (cudart) dan perpustakaan inti CUDA (cuda).

Optimalisasi program di CUDA

Secara alami, dalam kerangka artikel ulasan, tidak mungkin untuk mempertimbangkan masalah pengoptimalan yang serius dalam pemrograman CUDA. Oleh karena itu, kami hanya akan berbicara secara singkat tentang hal-hal dasar. Untuk menggunakan kemampuan CUDA secara efektif, Anda harus melupakan metode penulisan program yang biasa untuk CPU, dan menggunakan algoritme yang diparalelkan dengan baik untuk ribuan utas. Penting juga untuk menemukan tempat yang optimal untuk menyimpan data (register, memori bersama, dll.), meminimalkan transfer data antara CPU dan GPU, dan menggunakan buffering.

Secara umum, ketika mengoptimalkan program CUDA, seseorang harus mencoba mencapai keseimbangan optimal antara ukuran dan jumlah blok. Lebih banyak utas dalam satu blok akan mengurangi dampak latensi memori, tetapi juga akan mengurangi jumlah register yang tersedia. Selain itu, blok 512 utas tidak efisien, Nvidia sendiri merekomendasikan penggunaan blok 128 atau 256 utas sebagai nilai kompromi untuk mencapai latensi dan jumlah register yang optimal.

Di antara poin utama pengoptimalan program CUDA: penggunaan memori bersama seaktif mungkin, karena jauh lebih cepat daripada memori video global kartu video; membaca dan menulis dari memori global harus digabungkan bila memungkinkan. Untuk melakukan ini, Anda perlu menggunakan tipe data khusus untuk membaca dan menulis 32/64/128 bit data sekaligus dalam satu operasi. Jika operasi baca sulit untuk digabungkan, Anda dapat mencoba menggunakan pengambilan tekstur.

kesimpulan

Arsitektur perangkat keras dan perangkat lunak yang disajikan oleh Nvidia untuk perhitungan pada chip video CUDA sangat cocok untuk menyelesaikan berbagai tugas dengan paralelisme tinggi. CUDA berjalan pada sejumlah besar chip video Nvidia, dan meningkatkan model pemrograman GPU dengan sangat menyederhanakannya dan menambahkan banyak fitur seperti memori bersama, kemampuan untuk menyinkronkan utas, perhitungan presisi ganda, dan operasi bilangan bulat.

CUDA adalah teknologi yang tersedia untuk setiap pengembang perangkat lunak, dapat digunakan oleh programmer mana pun yang mengetahui bahasa C. Anda hanya perlu membiasakan diri dengan paradigma pemrograman berbeda yang melekat pada komputasi paralel. Tetapi jika algoritme, pada prinsipnya, diparalelkan dengan baik, maka studi dan waktu yang dihabiskan untuk pemrograman CUDA akan kembali dalam beberapa ukuran.

Kemungkinan karena meluasnya penggunaan kartu video di dunia, perkembangan komputasi paralel pada GPU akan sangat mempengaruhi industri komputasi kinerja tinggi. Kemungkinan-kemungkinan ini telah membangkitkan minat besar di kalangan ilmiah, dan tidak hanya di dalamnya. Lagi pula, potensi untuk mempercepat algoritme yang cocok untuk paralelisasi (pada perangkat keras yang terjangkau, yang tidak kalah pentingnya) sekaligus hingga puluhan kali tidak begitu umum.

Prosesor tujuan umum berkembang cukup lambat, mereka tidak memiliki lonjakan kinerja seperti itu. Bahkan, sekeras kedengarannya, semua orang yang membutuhkan komputasi cepat sekarang bisa mendapatkan superkomputer pribadi yang murah di meja mereka, kadang-kadang bahkan tanpa menginvestasikan uang ekstra, karena kartu video Nvidia tersedia secara luas. Belum lagi peningkatan efisiensi dalam hal GFLOPS/$ dan GFLOPS/W yang sangat disukai oleh produsen GPU.

Masa depan banyak komputasi jelas dalam algoritma paralel, hampir semua solusi dan inisiatif baru diarahkan ke arah ini. Sejauh ini, bagaimanapun, pengembangan paradigma baru adalah pada tahap awal, Anda harus secara manual membuat utas dan menjadwalkan akses memori, yang memperumit tugas dibandingkan dengan pemrograman konvensional. Tetapi teknologi CUDA telah mengambil langkah ke arah yang benar, dan itu jelas terlihat seperti solusi yang sukses, terutama jika Nvidia berhasil meyakinkan pengembang sebanyak mungkin tentang manfaat dan prospeknya.

Tapi, tentu saja, GPU tidak akan menggantikan CPU. Dalam bentuknya saat ini, mereka tidak dirancang untuk ini. Sekarang chip video secara bertahap bergerak menuju CPU, menjadi lebih dan lebih universal (perhitungan dengan floating point presisi tunggal dan ganda, perhitungan integer), sehingga CPU menjadi lebih dan lebih "paralel", memperoleh sejumlah besar core, multithreading teknologi, belum lagi munculnya blok SIMD dan proyek prosesor yang heterogen. Kemungkinan besar, GPU dan CPU hanya akan bergabung di masa depan. Diketahui bahwa banyak perusahaan, termasuk Intel dan AMD, sedang mengerjakan proyek serupa. Dan tidak masalah jika GPU dikonsumsi oleh CPU, atau sebaliknya.

Dalam artikel tersebut, kami terutama berbicara tentang manfaat CUDA. Tapi ada juga lalat di salep. Salah satu dari sedikit kelemahan CUDA adalah portabilitasnya yang buruk. Arsitektur ini hanya berfungsi pada chip video perusahaan ini, dan tidak pada semuanya, tetapi mulai dari seri Geforce 8 dan 9 serta Quadro dan Tesla yang sesuai. Ya, ada banyak solusi seperti itu di dunia, Nvidia memberikan angka 90 juta chip video yang kompatibel dengan CUDA. Ini bagus, tetapi pesaing menawarkan solusi mereka sendiri yang berbeda dari CUDA. Jadi, AMD memiliki Stream Computing, Intel akan memiliki Ct di masa depan.

Teknologi mana yang akan menang, tersebar luas, dan hidup lebih lama dari yang lain - hanya waktu yang akan menjawab. Tetapi CUDA memiliki peluang bagus, karena dibandingkan dengan Stream Computing, misalnya, ia menyediakan lingkungan pemrograman yang lebih berkembang dan mudah digunakan dalam bahasa C biasa. Mungkin pihak ketiga akan membantu dengan definisi dengan merilis semacam solusi umum. Misalnya, dalam pembaruan DirectX berikutnya di bawah versi 11, Microsoft menjanjikan shader komputasi, yang dapat menjadi semacam solusi rata-rata yang cocok untuk semua orang, atau hampir semua orang.

Dilihat dari data awal, shader jenis baru ini banyak meminjam dari model CUDA. Dan dengan memprogram di lingkungan ini sekarang, Anda dapat memperoleh manfaat langsung dan keterampilan yang diperlukan untuk masa depan. Dari sudut pandang komputasi kinerja tinggi, DirectX juga memiliki kelemahan yang berbeda dari portabilitas yang buruk, karena API terbatas pada platform Windows. Namun, standar lain sedang dikembangkan - inisiatif multi-platform terbuka OpenCL, yang didukung oleh sebagian besar perusahaan, termasuk Nvidia, AMD, Intel, IBM, dan banyak lainnya.

Jangan lupa bahwa dalam artikel CUDA berikutnya, Anda akan menjelajahi aplikasi praktis spesifik dari komputasi ilmiah dan komputasi non-grafis lainnya yang dilakukan oleh pengembang di seluruh dunia menggunakan Nvidia CUDA.

– satu set antarmuka pemrograman tingkat rendah ( API) untuk membuat game dan aplikasi multimedia berkinerja tinggi lainnya. Termasuk dukungan untuk kinerja tinggi 2D- Dan 3D-grafis, suara dan perangkat input.

Langsung3D (D3D) – antarmuka keluaran 3D primitif(benda geometris). Termasuk dalam .

OpenGL(dari bahasa Inggris. Buka Pustaka Grafik, secara harfiah - perpustakaan grafik terbuka) adalah spesifikasi yang mendefinisikan antarmuka pemrograman lintas platform independen bahasa pemrograman untuk menulis aplikasi yang menggunakan grafik komputer dua dimensi dan tiga dimensi. Termasuk lebih dari 250 fungsi untuk menggambar adegan 3D kompleks dari primitif sederhana. Ini digunakan dalam pembuatan video game, realitas virtual, visualisasi dalam penelitian ilmiah. Di platform jendela bersaing dengan .

OpenCL(dari bahasa Inggris. Buka Bahasa Komputasi, secara harfiah - bahasa komputasi terbuka) - kerangka(kerangka sistem perangkat lunak) untuk menulis program komputer yang terkait dengan komputasi paralel pada berbagai grafik ( GPU) Dan ( ). Untuk kerangka OpenCL termasuk bahasa pemrograman dan antarmuka pemrograman aplikasi ( API). OpenCL memberikan paralelisme pada tingkat instruksi dan pada tingkat data dan merupakan implementasi dari teknik GPGPU.

GPGPU(disingkat dari bahasa Inggris. Unit Pemrosesan Grafik Tujuan Umum, secara harfiah - GPU tujuan umum) - teknik untuk menggunakan prosesor grafis kartu video untuk perhitungan umum, yang biasanya dilakukan.

shader(Bahasa Inggris) shader) adalah program untuk membangun bayangan pada gambar yang disintesis, digunakan dalam grafik tiga dimensi untuk menentukan parameter akhir suatu objek atau gambar. Sebagai aturan, ini mencakup deskripsi penyerapan dan hamburan cahaya, pemetaan tekstur, refleksi dan refraksi, bayangan, perpindahan permukaan dan efek pasca-pemrosesan dari kompleksitas yang berubah-ubah. Permukaan kompleks dapat dirender menggunakan bentuk geometris sederhana.

rendering(Bahasa Inggris) rendering) - visualisasi, dalam grafik komputer, proses memperoleh gambar dari model menggunakan perangkat lunak.

SDK(disingkat dari bahasa Inggris. Kit Pengembangan Perangkat Lunak) adalah seperangkat alat pengembangan perangkat lunak.

CPU(disingkat dari bahasa Inggris. Unit pemrosesan utama, secara harfiah - perangkat komputasi pusat / utama / utama) - pusat (mikro); perangkat yang mengeksekusi instruksi mesin; bagian dari perangkat keras yang bertanggung jawab untuk melakukan operasi komputasi (diberikan oleh sistem operasi dan perangkat lunak aplikasi) dan mengoordinasikan pekerjaan semua perangkat.

GPU(disingkat dari bahasa Inggris. Unit Pemrosesan Grafis, secara harfiah - perangkat komputasi grafis) - prosesor grafis; perangkat atau konsol game terpisah yang melakukan rendering grafis (visualisasi). GPU modern sangat efisien dalam memproses dan merender grafik komputer secara realistis. Prosesor grafis dalam adaptor video modern digunakan sebagai akselerator grafis 3D, tetapi dalam beberapa kasus juga dapat digunakan untuk perhitungan ( GPGPU).

Masalah CPU

Untuk waktu yang lama, peningkatan kinerja yang tradisional terutama disebabkan oleh peningkatan berurutan dalam frekuensi clock (sekitar 80% dari kinerja ditentukan oleh frekuensi clock) dengan peningkatan simultan dalam jumlah transistor pada satu chip. Namun, peningkatan lebih lanjut dalam frekuensi clock (pada frekuensi clock lebih dari 3,8 GHz, chip terlalu panas!) Menolak sejumlah hambatan fisik mendasar (karena proses teknologi hampir mendekati ukuran atom: , dan ukuran atom silikon kira-kira 0,543 nm):

Pertama, dengan penurunan ukuran kristal dan dengan peningkatan frekuensi clock, arus bocor transistor meningkat. Hal ini menyebabkan peningkatan konsumsi daya dan peningkatan emisi panas;

Kedua, manfaat dari kecepatan clock yang lebih tinggi sebagian diimbangi oleh latensi akses memori, karena waktu akses memori tidak sejalan dengan peningkatan kecepatan clock;

Ketiga, untuk beberapa aplikasi, arsitektur serial tradisional menjadi tidak efisien karena kecepatan clock meningkat karena apa yang disebut “Von Neumann bottleneck,” kemacetan kinerja yang dihasilkan dari aliran komputasi yang berurutan. Pada saat yang sama, penundaan transmisi sinyal resistif-kapasitif meningkat, yang merupakan hambatan tambahan yang terkait dengan peningkatan frekuensi clock.

Perkembangan GPU

Sejalan dengan perkembangan GPU:

…

Nopember 2008 - Intel memperkenalkan garis 4-core Intel Core i7 berdasarkan mikroarsitektur generasi berikutnya Nehalem. Prosesor beroperasi pada frekuensi clock 2,6-3,2 GHz. Dibuat dalam teknologi proses 45nm.

Desember 2008 - Pengiriman quad core dimulai AMD Phenom II 940(nama kode - Deneb). Beroperasi pada frekuensi 3 GHz, diproduksi sesuai dengan teknologi proses 45-nm.

…

Mei 2009 - perusahaan AMD memperkenalkan versi GPU ATI Radeon HD 4890 dengan kecepatan core clock meningkat dari 850 MHz menjadi 1 GHz. Ini yang pertama grafis prosesor berjalan pada 1 GHz. Kekuatan pemrosesan chip, karena peningkatan frekuensi, telah berkembang dari 1,36 menjadi 1,6 teraflops. Prosesor berisi 800 (!) core, mendukung memori video GDDR5, DirectX 10.1, ATI CrossFireX dan semua teknologi lain yang melekat pada model kartu video modern. Chip ini dibuat berdasarkan teknologi 55-nm.

Perbedaan utama GPU

Fitur khas GPU(dibandingkan dengan ) adalah:

– arsitektur yang secara maksimal ditujukan untuk meningkatkan kecepatan penghitungan tekstur dan objek grafik yang kompleks;

adalah kekuatan puncak dari tipikal GPU jauh lebih tinggi dari ;

– berkat arsitektur pipa khusus, GPU jauh lebih efisien dalam memproses informasi grafis daripada .

"Krisis Genre"

"Krisis Genre" untuk matang pada tahun 2005 - saat itulah mereka muncul. Namun, terlepas dari perkembangan teknologi, pertumbuhan produktivitas konvensional menurun secara nyata. Pada saat yang sama kinerja GPU terus berkembang. Jadi, pada tahun 2003, ide revolusioner ini mengkristal - gunakan kekuatan komputasi grafik. GPU telah digunakan secara aktif untuk komputasi "non-grafis" (simulasi fisika, pemrosesan sinyal, matematika/geometri komputasi, operasi basis data, biologi komputasi, ekonomi komputasi, visi komputer, dll.).

Masalah utamanya adalah tidak ada antarmuka standar untuk pemrograman GPU. Pengembang menggunakan OpenGL atau Langsung3D tapi itu sangat nyaman. Perusahaan NVIDIA(salah satu produsen prosesor grafis, media dan komunikasi terbesar, serta prosesor media nirkabel; didirikan pada tahun 1993) terlibat dalam pengembangan semacam standar terpadu dan nyaman - dan memperkenalkan teknologi CUDA.

Bagaimana itu dimulai?

2006 - NVIDIA menunjukkan CUDA™; awal dari sebuah revolusi dalam komputasi GPU.

2007 - NVIDIA merilis arsitektur CUDA(versi asli SDK CUDA dipresentasikan 15 Februari 2007); nominasi "Kebaruan terbaik" dari majalah Ilmu pengetahuan populer dan "Pilihan Pembaca" dari publikasi HPCWire.

2008 - teknologi NVIDIA CUDA menang dalam nominasi "Keunggulan Teknis" dari Majalah PC.

Apa yang terjadi CUDA

CUDA(disingkat dari bahasa Inggris. Hitung Arsitektur Perangkat Terpadu, secara harfiah - arsitektur komputasi terpadu perangkat) - arsitektur (satu set perangkat lunak dan perangkat keras) yang memungkinkan Anda untuk memproduksi di GPU perhitungan tujuan umum GPU sebenarnya bertindak sebagai koprosesor yang kuat.

Teknologi NVIDIA CUDA™ adalah satu-satunya lingkungan pengembangan dalam bahasa pemrograman C, yang memungkinkan pengembang membuat perangkat lunak untuk memecahkan masalah komputasi yang kompleks dalam waktu yang lebih singkat, berkat kekuatan pemrosesan GPU. Jutaan sudah bekerja di dunia GPU dengan dukungan CUDA, dan ribuan programmer sudah menggunakan (gratis!) alat CUDA untuk mempercepat aplikasi dan menyelesaikan tugas intensif sumber daya yang paling kompleks - mulai dari encoding video dan audio hingga eksplorasi minyak dan gas, pemodelan produk, pencitraan medis, dan penelitian ilmiah.

CUDA memberi pengembang kesempatan, atas kebijaksanaannya sendiri, untuk mengatur akses ke set instruksi akselerator grafis dan mengelola memorinya, mengatur perhitungan paralel yang kompleks di atasnya. Akselerator grafis yang didukung CUDA menjadi arsitektur terbuka yang dapat diprogram yang kuat seperti sekarang ini. Semua ini memberi pengembang akses tingkat rendah, terdistribusi, dan berkecepatan tinggi ke peralatan, membuat CUDA dasar yang diperlukan untuk membangun alat tingkat tinggi yang serius seperti kompiler, debugger, perpustakaan matematika, platform perangkat lunak.

Uralsky, Spesialis Teknologi Utama NVIDIA, membandingkan GPU Dan , mengatakan seperti ini: - Ini SUV. Dia selalu bepergian dan kemana-mana, tapi tidak terlalu cepat. TETAPI GPU adalah mobil sport. Di jalan yang buruk, dia tidak akan pergi ke mana pun, tetapi memberikan perlindungan yang baik - dan dia akan menunjukkan semua kecepatannya, yang tidak pernah diimpikan oleh SUV! ..».

Kemampuan Teknologi CUDA

Dalam perkembangan prosesor modern, ada kecenderungan peningkatan bertahap dalam jumlah inti, yang meningkatkan kemampuan mereka dalam komputasi paralel. Namun, sudah lama ada GPU yang jauh lebih unggul daripada CPU dalam hal ini. Dan kemampuan GPU ini telah diperhitungkan oleh beberapa perusahaan. Upaya pertama untuk menggunakan akselerator grafis untuk perhitungan non-target telah dilakukan sejak akhir 90-an. Tetapi hanya penampilan shader yang menjadi pendorong pengembangan teknologi yang sama sekali baru, dan pada tahun 2003 konsep GPGPU (Unit pemrosesan grafis serba guna) muncul. Peran penting dalam pengembangan inisiatif ini dimainkan oleh BrookGPU, yang merupakan ekstensi khusus untuk bahasa C. Sebelum munculnya BrookGPU, programmer hanya dapat bekerja dengan GPU melalui Direct3D atau OpenGL API. Brook memungkinkan pengembang untuk bekerja dengan lingkungan yang akrab, dan kompiler itu sendiri, menggunakan perpustakaan khusus, mengimplementasikan interaksi dengan GPU pada tingkat rendah.

Kemajuan seperti itu pasti menarik perhatian para pemimpin industri ini - AMD dan NVIDIA, yang mulai mengembangkan platform perangkat lunak mereka sendiri untuk komputasi non-grafis pada kartu video mereka. Tidak ada yang lebih baik dari pengembang GPU yang mengetahui dengan sempurna semua nuansa dan fitur produk mereka, yang memungkinkan perusahaan-perusahaan ini untuk mengoptimalkan paket perangkat lunak untuk solusi perangkat keras tertentu seefisien mungkin. Saat ini NVIDIA sedang mengembangkan platform CUDA (Compute Unified Device Architecture), AMD menyebut teknologi ini CTM (Close To Metal) atau AMD Stream Computing. Kami akan melihat beberapa fitur CUDA dan mengevaluasi kemampuan komputasi chip grafis G92 dari kartu video GeForce 8800 GT dalam praktiknya.

Tapi pertama-tama, mari kita lihat beberapa nuansa melakukan perhitungan menggunakan GPU. Keuntungan utama mereka terletak pada kenyataan bahwa chip grafis pada awalnya dirancang untuk eksekusi banyak utas, dan setiap inti dari CPU konvensional mengeksekusi aliran instruksi sekuensial. Setiap GPU modern adalah multiprosesor yang terdiri dari beberapa cluster komputasi, dengan banyak ALU di masing-masingnya. Chip GT200 modern yang paling kuat terdiri dari 10 kluster, yang masing-masing memiliki 24 prosesor aliran. Kartu video yang diuji GeForce 8800 GT berdasarkan chip G92 memiliki tujuh unit komputasi besar dengan masing-masing 16 prosesor aliran. CPU menggunakan blok SIMD SSE untuk perhitungan vektor (instruksi tunggal beberapa data - satu instruksi dieksekusi pada banyak data), yang membutuhkan transformasi data menjadi 4x vektor. GPU memproses utas secara skalar, mis. satu instruksi diterapkan pada beberapa utas (SIMT - instruksi tunggal beberapa utas). Ini menyelamatkan pengembang dari mengubah data menjadi vektor, dan memungkinkan percabangan sewenang-wenang dalam aliran. Setiap unit komputasi GPU memiliki akses memori langsung. Dan bandwidth memori video lebih tinggi karena penggunaan beberapa pengontrol memori terpisah (di bagian atas G200 adalah 8 saluran masing-masing 64-bit) dan frekuensi operasi yang tinggi.

Secara umum, dalam tugas-tugas tertentu, saat bekerja dengan data dalam jumlah besar, GPU jauh lebih cepat daripada CPU. Di bawah ini Anda melihat ilustrasi pernyataan ini:

Grafik menunjukkan evolusi kinerja CPU dan GPU sejak tahun 2003. NVIDIA suka mengutip data ini sebagai iklan dalam dokumennya, tetapi itu hanya perhitungan teoretis dan sebenarnya kesenjangannya, tentu saja, bisa jauh lebih kecil.

Namun, bagaimanapun juga, ada potensi besar GPU yang dapat digunakan, dan yang memerlukan pendekatan khusus untuk pengembangan perangkat lunak. Semua ini diimplementasikan di lingkungan perangkat keras-perangkat lunak CUDA, yang terdiri dari beberapa lapisan perangkat lunak - API Runtime CUDA tingkat tinggi dan API Driver CUDA tingkat rendah.

CUDA menggunakan bahasa C standar untuk pemrograman, yang merupakan salah satu keuntungan utama bagi pengembang. Awalnya, CUDA menyertakan library BLAS (Basic Linear Algebra) dan FFT (Fourier Transform Calculation). CUDA juga memiliki kemampuan untuk berinteraksi dengan API grafis OpenGL atau DirectX, kemampuan untuk berkembang pada tingkat rendah, dan dicirikan oleh distribusi aliran data yang dioptimalkan antara CPU dan GPU. Perhitungan CUDA dilakukan secara bersamaan dengan yang grafis, berbeda dengan platform AMD serupa, di mana mesin virtual khusus umumnya diluncurkan untuk perhitungan GPU. Tetapi "kohabitasi" seperti itu juga penuh dengan kesalahan jika beban besar dibuat oleh API grafis saat CUDA berjalan pada saat yang sama - lagipula, operasi grafis masih memiliki prioritas yang lebih tinggi. Platform ini kompatibel dengan sistem operasi 32-bit dan 64-bit Windows XP, Windows Vista, MacOS X dan berbagai versi Linux. Platform terbuka dan di situs, selain driver khusus untuk kartu video, Anda dapat mengunduh CUDA Toolkit, paket perangkat lunak CUDA Developer SDK, yang mencakup kompiler, debugger, pustaka standar, dan dokumentasi.

Adapun penerapan praktis CUDA, untuk waktu yang lama teknologi ini hanya digunakan untuk perhitungan matematis yang sangat khusus di bidang fisika partikel dasar, astrofisika, kedokteran, atau peramalan perubahan di pasar keuangan, dll. Namun teknologi ini secara bertahap semakin dekat dengan pengguna biasa, khususnya, ada plug-in khusus untuk Photoshop yang dapat menggunakan daya komputasi GPU. Pada halaman khusus, Anda dapat menjelajahi seluruh daftar program yang menggunakan kemampuan NVIDIA CUDA.

Sebagai uji praktik teknologi baru pada kartu video MSI NX8800GT-T2D256E-OC, kami akan menggunakan program TMPGEnc. Produk ini komersial (versi lengkap berharga $100), tetapi untuk kartu video MSI, produk ini hadir sebagai bonus dalam versi uji coba selama 30 hari. Anda dapat mengunduh versi ini dari situs pengembang, tetapi untuk menginstal TMPGEnc 4.0 XPress MSI Edisi Khusus, Anda memerlukan disk asli dengan driver dari kartu MSI - program tidak dapat diinstal tanpanya.

Untuk menampilkan informasi paling lengkap tentang kemampuan komputasi di CUDA dan membandingkannya dengan adaptor video lainnya, Anda dapat menggunakan utilitas CUDA-Z khusus. Berikut adalah informasi yang diberikannya tentang kartu video GeForce 8800GT kami:

Sehubungan dengan model referensi, instans kami berjalan pada frekuensi yang lebih tinggi: domain raster 63 MHz lebih tinggi dari nominal, dan unit shader 174 MHz lebih cepat, memori 100 MHz lebih cepat.

Kami akan membandingkan kecepatan konversi dari video HD yang sama saat menghitung hanya dengan bantuan CPU dan dengan aktivasi tambahan CUDA dalam program TMPGEnc pada konfigurasi berikut:

• Prosesor: Pentium Dual-Core E5200 2.5 GHz;
• Papan Utama: Gigabyte P35-S3;
• Memori: 2x1GB GoodRam PC6400 (5-5-5-18-2T)
• Kartu video: MSI NX8800GT-T2D256E-OC;
• Harddisk: 320GB WD3200AAKS;
• Catu daya: CoolerMaster eXtreme Power 500-PCAP;
• Sistem operasi: Windows XP SP2;
• TMPGEnc 4.0 XPress 4.6.3.268;
• Driver kartu video: ForceWare 180.60.

Untuk pengujian, prosesor di-overclock hingga 3 GHz (dalam konfigurasi 11.5x261 MHz) dan hingga 4 GHz (11.5x348 MHz) dengan frekuensi RAM 835 MHz dalam kasus pertama dan kedua. Video dalam resolusi Full HD 1920x1080 selama satu menit dua puluh detik. Untuk membuat beban tambahan, filter pengurangan kebisingan dihidupkan, yang pengaturannya dibiarkan secara default.

Pengkodean dilakukan dengan menggunakan codec DivX 6.8.4. Dalam pengaturan kualitas codec ini, semua nilai dibiarkan secara default, multithreading diaktifkan.

Dukungan multithreading di TMPGEnc awalnya diaktifkan di tab pengaturan CPU/GPU. Di bagian yang sama, CUDA juga diaktifkan.

Seperti yang Anda lihat dari tangkapan layar di atas, pemrosesan filter CUDA diaktifkan, dan dekoder video perangkat keras tidak diaktifkan. Dokumentasi untuk program memperingatkan bahwa mengaktifkan opsi terakhir menyebabkan peningkatan waktu pemrosesan file.

Berdasarkan hasil pengujian diperoleh data sebagai berikut:

Pada 4 GHz dengan CUDA diaktifkan, kami hanya memperoleh beberapa detik (atau 2%), yang tidak terlalu mengesankan. Tetapi pada frekuensi yang lebih rendah, keuntungan dari mengaktifkan teknologi ini memungkinkan Anda menghemat sekitar 13% dari waktu, yang akan sangat terlihat saat memproses file besar. Tetap saja, hasilnya tidak semenarik yang diharapkan.

Program TMPGEnc memiliki indikator beban CPU dan CUDA; dalam konfigurasi pengujian ini, menunjukkan beban CPU sekitar 20%, dan inti grafis pada 80% sisanya. Hasilnya, kami memiliki 100% yang sama seperti saat mengonversi tanpa CUDA, dan mungkin tidak ada perbedaan waktu sama sekali (tetapi masih ada). Memori yang kecil sebesar 256 MB juga bukan merupakan faktor pembatas. Dilihat dari pembacaan RivaTuner, tidak lebih dari 154 MB memori video yang digunakan selama operasi.

kesimpulan

Program TMPGEnc adalah salah satu yang membawa teknologi CUDA ke masyarakat. Penggunaan GPU dalam program ini memungkinkan Anda untuk mempercepat proses pemrosesan video dan secara signifikan menurunkan beban CPU, yang akan memungkinkan pengguna untuk menangani tugas-tugas lain secara bersamaan dengan nyaman. Dalam contoh khusus kami, kartu video GeForce 8800GT 256MB sedikit meningkatkan kinerja waktu saat mengonversi video berdasarkan prosesor Pentium Dual-Core E5200 yang di-overclock. Tetapi terlihat jelas bahwa dengan penurunan frekuensi, keuntungan dari aktivasi CUDA meningkat; pada prosesor yang lemah, keuntungan dari penggunaannya akan jauh lebih besar. Dengan latar belakang ketergantungan seperti itu, cukup logis untuk mengasumsikan bahwa bahkan dengan peningkatan beban (misalnya, penggunaan sejumlah besar filter video tambahan), hasil sistem dengan CUDA akan dibedakan oleh delta yang lebih signifikan dari perbedaan waktu yang dihabiskan untuk proses pengkodean. Juga, jangan lupa bahwa G92 bukanlah chip yang paling kuat saat ini, dan kartu video yang lebih modern akan memberikan kinerja yang jauh lebih tinggi dalam aplikasi tersebut. Namun, selama pengoperasian aplikasi, GPU tidak terisi penuh dan, mungkin, distribusi beban tergantung pada setiap konfigurasi secara terpisah, yaitu pada kombinasi prosesor / kartu video, yang pada akhirnya dapat memberikan peningkatan yang lebih besar (atau lebih kecil). dalam persentase dari aktivasi CUDA. Bagaimanapun, bagi mereka yang bekerja dengan data video dalam jumlah besar, teknologi ini masih akan menghemat waktu mereka secara signifikan.

Benar, CUDA belum mendapatkan popularitas yang luas, kualitas perangkat lunak yang bekerja dengan teknologi ini perlu ditingkatkan. Dalam program TMPGEnc 4.0 XPress yang kami ulas, teknologi ini tidak selalu berfungsi. Video yang sama dapat di-encode ulang beberapa kali, dan kemudian tiba-tiba, pada awal berikutnya, beban CUDA sudah 0%. Dan fenomena ini benar-benar acak pada sistem operasi yang sama sekali berbeda. Selain itu, program yang ditinjau menolak untuk menggunakan CUDA saat menyandikan ke format XviD, tetapi tidak ada masalah dengan codec DivX yang populer.

Akibatnya, sementara teknologi CUDA dapat secara signifikan meningkatkan kinerja komputer pribadi hanya dalam tugas-tugas tertentu. Tetapi cakupan teknologi tersebut akan berkembang, dan proses peningkatan jumlah inti dalam prosesor konvensional menunjukkan meningkatnya permintaan untuk komputasi multi-utas paralel dalam aplikasi perangkat lunak modern. Bukan tanpa alasan bahwa semua pemimpin industri baru-baru ini membakar gagasan untuk menggabungkan CPU dan GPU dalam satu arsitektur terpadu (ingat setidaknya AMD Fusion yang diiklankan). Mungkin CUDA adalah salah satu tahapan dalam proses unifikasi ini.

Kami berterima kasih kepada perusahaan-perusahaan berikut untuk menyediakan peralatan uji:

Dan itu dirancang untuk menerjemahkan kode host (utama, kode kontrol) dan kode perangkat (kode perangkat keras) (file dengan ekstensi .cu) menjadi file objek yang cocok untuk membangun program akhir atau perpustakaan di lingkungan pemrograman apa pun, misalnya , di NetBeans .

Arsitektur CUDA menggunakan model memori grid, pemodelan thread berkerumun, dan instruksi SIMD. Berlaku tidak hanya untuk komputasi grafis berperforma tinggi, tetapi juga untuk berbagai komputasi ilmiah menggunakan kartu grafis nVidia. Para ilmuwan dan peneliti menggunakan CUDA secara ekstensif di berbagai bidang, termasuk astrofisika, biologi komputasi dan kimia, pemodelan dinamika fluida, interaksi elektromagnetik, computed tomography, analisis seismik, dan banyak lagi. CUDA memiliki kemampuan untuk terhubung ke aplikasi yang menggunakan OpenGL dan Direct3D. CUDA adalah perangkat lunak lintas platform untuk sistem operasi seperti Linux, Mac OS X dan Windows.

Pada 22 Maret 2010, nVidia merilis CUDA Toolkit 3.0 yang berisi dukungan untuk OpenCL.

Peralatan

Platform CUDA pertama kali muncul di pasar dengan merilis chip G80 generasi kedelapan NVIDIA dan telah hadir di semua seri chip grafis berikutnya yang digunakan dalam keluarga akselerator GeForce, Quadro, dan NVidia Tesla.

Seri perangkat keras pertama yang mendukung CUDA SDK, G8x, memiliki prosesor vektor presisi tunggal 32-bit menggunakan CUDA SDK sebagai API (CUDA mendukung tipe ganda C, tetapi sekarang presisinya diturunkan ke floating point 32-bit) . Prosesor GT200 yang lebih baru memiliki dukungan untuk presisi 64-bit (hanya untuk SFU), tetapi kinerjanya jauh lebih buruk daripada untuk presisi 32-bit (karena hanya dua SFU per streaming multiprosesor, dan delapan prosesor skalar). GPU mengatur multithreading perangkat keras, yang memungkinkan Anda menggunakan semua sumber daya GPU. Dengan demikian, prospek terbuka untuk mengalihkan fungsi akselerator fisik ke akselerator grafis (contoh implementasi - nVidia PhysX). Ini juga membuka peluang luas untuk menggunakan peralatan grafis komputer untuk melakukan perhitungan non-grafis yang kompleks: misalnya, dalam biologi komputasi dan cabang ilmu pengetahuan lainnya.

Keuntungan

Dibandingkan dengan pendekatan tradisional untuk mengatur komputasi tujuan umum melalui kemampuan API grafis, arsitektur CUDA memiliki keunggulan berikut di bidang ini:

Pembatasan

• Semua fungsi yang dapat dieksekusi perangkat tidak mendukung rekursi (dalam CUDA Toolkit 3.1 mendukung pointer dan rekursi) dan memiliki beberapa batasan lain

GPU dan GPU yang didukung

Daftar perangkat dari produsen perangkat keras Nvidia dengan dukungan penuh yang dinyatakan untuk teknologi CUDA diberikan di situs web resmi Nvidia: Produk GPU yang Diaktifkan CUDA (Bahasa Inggris).

Faktanya, periferal berikut saat ini mendukung teknologi CUDA di pasar perangkat keras PC:

Versi spesifikasi	GPU	Kartu video
1.0	G80, G92, G92b, G94, G94b	GeForce 8800GTX/Ultra, 9400GT, 9600GT, 9800GT, Tesla C/D/S870, FX4/5600, 360M, GT 420
1.1	G86, G84, G98, G96, G96b, G94, G94b, G92, G92b	GeForce 8400GS/GT, 8600GT/GTS, 8800GT/GTS, 9600 GSO, 9800GTX/GX2, GTS250 /370M, 3/5/770M, 16/17/27/28/36/37/3800M, NVS420/50
1.2	GT218, GT216, GT215	GeForce 210, GT 220/40, FX380 LP, 1800M, 370/380M, NVS 2/3100M
1.3	GT200, GT200b	GeForce GTX 260, GTX 275, GTX 280, GTX 285, GTX 295, Tesla C/M1060, S1070, Quadro CX, FX 3/4/5800
2.0	GF100, GF110	GeForce (GF100) GTX 465, GTX 470, GTX 480, Tesla C2050, C2070, S/M2050/70, Quadro Plex 7000, Quadro 4000, 5000, 6000, GeForce (110) GTX GF560 TI 448, GTX570, GTX580, GTX590
2.1	GF104, GF114, GF116, GF108, GF106	GeForce 610M, GT 430, GT 440, GTS 450, GTX 460, GTX 550 Ti, GTX 560, GTX 560 Ti, 500M, Quadro 600, 2000
3.0	GK104, GK106, GK107	GeForce GTX 690, GTX 680, GTX 670, GTX 660 Ti, GTX 660, GTX 650 Ti, GTX 650, GT 640, GeForce GTX 680MX, GeForce GTX 680M, GeForce GTX 675MX, GeForce GTX 670MX, GTX 660M, GeForce GT 650M, GeForce GT 645M, GeForce GT 640M
3.5	GK110

Nvidia GeForce Desktop GeForce GTX 590 GeForce GTX 580 GeForce GTX 570 GeForce GTX 560 Ti GeForce GTX 560 GeForce GTX 550 Ti GeForce GTX 520 GeForce GTX 480 GeForce GTX 470 GeForce GTX 465 GeForce GTX 460 GeForce GTS450 GeForce GTX 295 GeForce GTX 285 GeForce GTX 280 GeForce GTX 275 GeForce GTX 260 GeForce GTS 250 GeForce GT 240 GeForce GT 220 GeForce 210 GeForce GTS 150 GeForce GT 130 GeForce GT120 GeForce G100 GeForce 9800 GX2 GeForce 9800 GTX+ GeForce 9800 GTX GeForce 9800 GT GeForce 9600 GSO GeForce 9600 GT GeForce 9500 GT GeForce 9400 GT GeForce 9400mGPU GeForce 9300mGPU GeForce 8800 GTS 512 GeForce 8800 GT GeForce 8600 GTS GeForce 8600 GT GeForce 8500 GT GeForce 8400GS

Nvidia GeForce untuk Komputer Seluler GeForce GTX 580M GeForce GTX 570M GeForce GTX 560M GeForce GT 555M GeForce GT 540M GeForce GT 525M GeForce GT 520M GeForce GTX 485M GeForce GTX 480M GeForce GTX 470M GeForce GTX 460M GeForce GT 445M GeForce GT 435M GeForce GT 425M GeForce GT 420M GeForce GT 415M GeForce GTX 285M GeForce GTX 280M GeForce GTX 260M GeForce GTS 360M GeForce GTS 350M GeForce GTS 160M GeForce GTS 150M GeForce GT 335M GeForce GT 330M GeForce GT 325M GeForce GT 240M GeForce GT 130M GeForce G210M GeForce G110M GeForce G105M GeForce 310M GeForce 305M GeForce 9800M GTX GeForce 9800M GT GeForce 9800M GTS GeForce 9700M GTS GeForce 9700M GT GeForce 9650M GS GeForce 9600M GT GeForce 9600M GS GeForce 9500M GS GeForce 9500M G GeForce 9300M GS GeForce 9300MG GeForce 9200M GS GeForce 9100MG GeForce 8800M GTS GeForce 8700M GT GeForce 8600M GT GeForce 8600M GS GeForce 8400M GT GeForce 8400M GS

NvidiaTesla * Tesla C2050/C2070 Tesla M2050/M2070/M2090 Tesla S2050 Tesla S1070 Tesla M1060 Tesla C1060 Tesla C870 Tesla D870 Tesla S870

Desktop Nvidia Quadro Quadro 6000 Quadro 5000 Quadro 4000 Quadro 2000 Quadro 600 QuadroFX 5800 QuadroFX 5600 QuadroFX4800 Quadro FX 4700X2 QuadroFX4600 QuadroFX 3700 QuadroFX 1700 QuadroFX 570 QuadroFX470 Profil Rendah Quadro FX 380 QuadroFX 370 Profil Rendah Quadro FX 370 Quadro CX Quadro NVS450 Quadro NVS 420 Quadro NVS 290 Quadro Plex 2100 D4 Quadro Plex 2200 D2 Quadro Plex 2100 S4 Quadro Plex 1000 Model IV

Nvidia Quadro untuk Komputer Seluler Quadro 5010M Quadro 5000M Quadro 4000M Quadro 3000M Quadro 2000M Quadro 1000M QuadroFX 3800M QuadroFX 3700M QuadroFX 3600M QuadroFX 2800M QuadroFX 2700M QuadroFX 1800M QuadroFX 1700M QuadroFX 1600M QuadroFX 880M QuadroFX 770M QuadroFX 570M QuadroFX 380M QuadroFX 370M QuadroFX 360M Quadro NVS 5100M Quadro NVS 4200M Quadro NVS 3100M Quadro NVS 2100M Quadro NVS 320M Quadro NVS 160M Quadro NVS 150M Quadro NVS 140M Quadro NVS 135M Quadro NVS 130M

• Model Tesla C1060, Tesla S1070, Tesla C2050/C2070, Tesla M2050/M2070, Tesla S2050 memungkinkan perhitungan GPU presisi ganda.

Fitur dan spesifikasi versi yang berbeda

Dukungan fitur (fitur yang tidak terdaftar adalah didukung untuk semua kemampuan komputasi)	Kemampuan komputasi (versi)
	1.0	1.1	1.2	1.3	2.x
Kata 32-bit dalam memori global	Bukan	Ya
nilai floating point dalam memori global
Fungsi atom bilangan bulat beroperasi pada Kata 32-bit dalam memori bersama	Bukan		Ya
atomExch() beroperasi pada 32-bit nilai floating point dalam memori bersama
Fungsi atom bilangan bulat beroperasi pada Kata 64-bit dalam memori global
Fungsi suara warp
Operasi floating-point presisi ganda	Bukan			Ya
Fungsi atom yang beroperasi pada 64-bit nilai integer dalam memori bersama	Bukan				Ya
Penambahan atom titik-mengambang beroperasi pada Kata 32-bit dalam memori global dan bersama
_suara()
_threadfence_system()
_syncthreads_count(), _syncthreads_and(), _syncthreads_or()
Fungsi permukaan
Kotak 3D dari blok utas

spesifikasi teknis	Kemampuan komputasi (versi)
	1.0	1.1	1.2	1.3	2.x
Dimensi maksimum kisi-kisi blok ulir	2				3
Dimensi x-, y-, atau z maksimum dari kisi blok ulir	65535
Dimensi maksimum blok ulir	3
Dimensi x atau y maksimum dari sebuah balok	512				1024
Dimensi z maksimum dari sebuah blok	64
Jumlah maksimum utas per blok	512				1024
Ukuran lusi	32
Jumlah maksimum blok penduduk per multiprosesor	8
Jumlah maksimum warps penduduk per multiprosesor	24		32		48
Jumlah maksimum utas penduduk per multiprosesor	768		1024		1536
Jumlah register 32-bit per multiprosesor	8K		16K		32K
Jumlah maksimum memori bersama per multiprosesor	16KB				48KB
Jumlah bank memori bersama	16				32
Jumlah memori lokal per utas	16KB				512KB
Ukuran memori konstan	64KB
Set kerja cache per multiprosesor untuk memori konstan	8KB
Set kerja cache per multiprosesor untuk memori tekstur	Tergantung perangkat, antara 6 KB dan 8 KB
Lebar maksimum untuk tekstur 1D	8192				32768
Lebar maksimum untuk tekstur 1D referensi terikat ke memori linier	2 27
Lebar maksimum dan jumlah lapisan untuk referensi tekstur berlapis 1D	8192x512				16384x2048
Lebar dan tinggi maksimum untuk 2D referensi tekstur terikat pada memori linier atau array CUDA	65536x32768				65536 x 65535
Lebar, tinggi, dan jumlah maksimum lapisan untuk referensi tekstur berlapis 2D	8192 x 8192 x 512				16384x16384x2048
Lebar, tinggi, dan kedalaman maksimum untuk referensi tekstur 3D terikat linier memori atau array CUDA	2048x2048x2048
Jumlah maksimum tekstur yang dapat diikat ke kernel	128
Lebar maksimum untuk permukaan 1D referensi terikat ke array CUDA	Bukan didukung				8192
Lebar dan tinggi maksimum untuk 2D referensi permukaan terikat ke array CUDA					8192x8192
Jumlah maksimum permukaan yang dapat diikat ke kernel					8
Jumlah maksimum instruksi per inti	2 juta

Contoh

CudaArray* cu_array; tekstur< float , 2 >teks; // Alokasikan array cudaMalloc( & cu_array, cudaCreateChannelDesc< float>() , lebar tinggi ) ; // Salin data gambar ke array cudaMemcpy( cu_array, image, width* height, cudaMemcpyHostToDevice) ; // Ikat array ke tekstur cudaBindTexture( tex, cu_array) ; // Jalankan kernel dim3 blockDim(16 , 16 , 1 ); dim3 gridDim(width / blockDim.x , height / blockDim.y , 1 ); inti<<< gridDim, blockDim, 0 >>> (d_odata, lebar, tinggi) ; cudaUnbindTexture(tex) ; __global__ void kernel(float * odata, int height, int width) ( unsigned int x = blockIdx.x * blockDim.x + threadIdx.x ; unsigned int y = blockIdx.y * blockDim.y + threadIdx.y ; float c = texfetch(tex, x, y) ; odata[ y* lebar+ x] = c; )

impor pycuda.driver sebagai drv impor numpy drv.init() dev = drv.Device(0) ctx = dev.make_context() mod = drv.SourceModule( """ __global__ void multiply_them(float *dest, float *a, float *b) ( const int i = threadIdx.x; dest[i] = a[i] * b[i]; ) """) multiply_them = mod.get_function ("multiply_them" ) a = numpy.random .randn (400 ) .astype (numpy.float32 ) b = numpy.random .randn (400 ) .astype (numpy.float32 ) dest = numpy.zeros_like (a) multiply_them( drv.Out (dest) , drv.In (a) , drv.In (b) , block= (400 , 1 , 1 ) ) print dest-a*b

CUDA sebagai mata pelajaran di universitas

Pada Desember 2009, model pemrograman CUDA diajarkan di 269 universitas di seluruh dunia. Di Rusia, kursus pelatihan CUDA diajarkan di Universitas Politeknik St. Petersburg, Universitas Negeri Yaroslavl. PG Demidov, Moskow, Nizhny Novgorod, St. Petersburg, Tver, Kazan, Novosibirsk, Universitas Teknik Negeri Novosibirsk Universitas Negeri Omsk dan Perm, Universitas Internasional Sifat Masyarakat dan Manusia "Dubna", Universitas Teknik Tenaga Negeri Ivanovo, Universitas Negeri Belgorod , MSTU mereka. Bauman, RKhTU im. Mendeleev, Pusat Superkomputer Antar Wilayah Akademi Ilmu Pengetahuan Rusia, . Selain itu, pada bulan Desember 2009, pembukaan pusat ilmiah dan pendidikan pertama di Rusia "Komputasi Paralel", yang terletak di kota Dubna, diumumkan, yang tugasnya meliputi pelatihan dan konsultasi untuk memecahkan masalah komputasi yang kompleks pada GPU.

Di Ukraina, kursus tentang CUDA diajarkan di Institut Analisis Sistem Kiev.

Tautan

Sumber daya resmi

• CUDA Zone (Rusia) - situs web resmi CUDA
• CUDA GPU Computing (Bahasa Inggris) - forum web resmi yang didedikasikan untuk komputasi CUDA

Sumber Daya Tidak Resmi

Perangkat keras Tom

• Dmitry Chekanov. nVidia CUDA: komputasi kartu grafis atau CPU mati? . Perangkat Keras Tom (22 Juni 2008).
• Dmitry Chekanov. nVidia CUDA: Tolok Ukur Aplikasi GPU untuk Pasar Massal. Perangkat Keras Tom (19 Mei 2009). Diarsipkan dari versi asli tanggal 4 Maret 2012. Diakses tanggal 19 Mei 2009.

iXBT.com

• Alexey Berillo. NVIDIA CUDA - komputasi non-grafis pada GPU. Bagian 1 . iXBT.com (23 September 2008). Diarsipkan dari versi asli tanggal 4 Maret 2012. Diakses tanggal 20 Januari 2009.
• Alexey Berillo. NVIDIA CUDA - komputasi non-grafis pada GPU. Bagian 2 . iXBT.com (22 Oktober 2008). - Contoh implementasi NVIDIA CUDA. Diarsipkan dari versi asli tanggal 4 Maret 2012. Diakses tanggal 20 Januari 2009.

Sumber daya lainnya

• Boreskov Alexey Viktorovich. Dasar-dasar CUDA (20 Januari 2009). Diarsipkan dari versi asli tanggal 4 Maret 2012. Diakses tanggal 20 Januari 2009.
• Vladimir Frolov. Pengantar Teknologi CUDA. Jurnal Online “Grafik Komputer dan Multimedia” (19 Desember 2008). Diarsipkan dari versi asli tanggal 4 Maret 2012. Diakses tanggal 28 Oktober 2009.
• Igor Oskolkov. NVIDIA CUDA adalah tiket terjangkau untuk komputasi besar. Computerra (30 April 2009). Diakses pada 3 Mei 2009.
• Vladimir Frolov. Pengantar Teknologi CUDA (1 Agustus 2009). Diarsipkan dari versi asli tanggal 4 Maret 2012. Diakses tanggal 3 April 2010.
• GPGPU.ru. Menggunakan kartu grafis untuk komputasi
• . Pusat Komputasi Paralel

Catatan

Lihat juga

Nvidia
Grafis prosesor	Dini NV1 NV2 keluarga RIVA TNT TNT2 keluarga GeForce GeForce 256 GeForce 2 GeForce 3 GeForce 4 GeForce FX GeForce 6 GeForce 7 GeForce 8 GeForce 9 GeForce 100 GeForce 200 GeForce 400 GeForce 500 GeForce 600