Rumus Indeks Klasik

Inti dari kernel CUDA adalah setiap utas menangani satu bagian data. Untuk melakukannya, setiap utas memerlukan indeks global yang unik.

Di dalam sebuah blok, threadIdx.x hanya menghitung dari 0 hingga blockDim dikurangi satu. Banyak blok menggunakan kembali angka kecil yang sama, sehingga angka tersebut tidak dapat menjadi indeks akhir Anda.

Bayangkan kisi sebagai blok-blok yang disusun ujung ke ujung. Setiap blok memiliki bagian larik yang berurutan, dan blockIdx.x memberi tahu Anda bagian mana yang sedang digunakan.

blockDim.x adalah jumlah utas per blok. Nilai ini merupakan lebar setiap bagian, sehingga mengalikan pergeseran blok Anda ke posisi yang tepat.

Gabungkan ketiganya: lewati blok-blok sebelumnya, lalu tambahkan posisi Anda di dalam blok ini. Satu baris tersebut memberikan indeks unik kepada setiap utas.

Dengan 4 utas per blok, blok 0 mencakup 0 hingga 3, blok 1 mencakup 4 hingga 7, dan blok 2 mencakup 8 hingga 11. Pergeseran tersebut tidak pernah tumpang tindih.

Utas 2 di blok 3 dengan blockDim 256 berada pada 3 dikali 256 ditambah 2, yaitu 770. Itulah posisi global utas tersebut dalam data.

Setelah memiliki i, perlakukan i sebagai subskrip larik. Setiap utas hanya membaca dan menulis elemennya sendiri, tanpa tumpang tindih.

Jalankan n utas, dan rumus tersebut menghasilkan setiap nilai dari 0 hingga n dikurangi 1 tepat satu kali. Itulah cakupan larik yang benar-benar satu ke satu.

Selalu lakukan perkalian sebelum penjumlahan. blockIdx.x * blockDim.x adalah awal bagian Anda, sedangkan threadIdx.x adalah langkah di dalamnya.

Gagasan yang sama berlaku untuk 2D dan 3D dengan menggunakan anggota .y dan .z, tetapi untuk larik satu dimensi, rumus .x saja sudah cukup. 🚀

Anda telah mempelajari rumus yang digunakan setiap kernel: blockIdx.x * blockDim.x + threadIdx.x. Rumus tersebut memberikan satu posisi unik dalam larik kepada setiap utas. 🎉