💡 💊 ⚱️ Menulis di Rust + CUDA C 👴🏿 🚺 💬

Halo semuanya!

Dalam panduan ini saya ingin memberi tahu cara berteman CUDA C / C ++ dan Rust. Dan sebagai contoh, kita akan menulis sebuah program kecil di Rust untuk menghitung produk skalar vektor, perhitungan produk skalar akan dilakukan pada GPU menggunakan CUDA C.

Siapa yang peduli dengan kucing!

CUDA C

Langkah pertama adalah menginstal kompiler CUDA - nvcc. Apa itu CUDA dan mengapa saya perlu menggambarkannya, saya dapat membacanya di sini, misalnya. Saya hanya dapat mengatakan bahwa dengan bantuannya Anda dapat menulis kode yang akan berjalan pada kartu video NVIDIA (selanjutnya disebut sebagai GPU) dan menggunakan semua kekuatan mereka untuk komputasi paralel dan pemrosesan grafis. Sekali lagi, tutorial ini bukan tentang bagaimana menulis kode dalam CUDA, tetapi tentang bagaimana menggunakan kelebihannya dari kode Rust dan menulis komputasi paralel pada GPU.

Jadi instal nvcc dan CUDA Toolkit. Dengan kerumitan ini, instruksi rinci tidak boleh muncul: di situs off .

RUST + CUDA C

Dalam tutorial ini, seperti yang disebutkan sebelumnya, kita akan menulis sebuah program di Rust untuk menemukan produk skalar dari dua vektor, proses perhitungan itu sendiri akan terjadi pada GPU.

Produk skalar dari dua vektor.

Misalkan kita memiliki dua vektor:

a = [a_{1}, a_{2}, . . . a_{n}]

$a = [a_1, a_2, ... a_n]$ dan

b = [b_{1}, b_{2}, . . ., b_{n}]

$b = [b_1, b_2, ..., b_n]$ , produk skalar dari vektor-vektor ini:

a c d o t b = s u m_{i = 1}^{n} a_{i} b_{i}

$a \ cdot b = \ sum_ {i = 1} ^ {n} {a_ib_i}$

Mari mulai membuat program kami. Selanjutnya, saya berasumsi bahwa nvcc berhasil diinstal, rustc dan kargo juga berarti mengkompilasi kode karat.

Pertama, buat folder proyek. Di folder proyek, buat file Cargo.toml, yang berisi instruksi untuk pengumpul kargo. File terlihat seperti ini:

[package] name = "rust_cuda" #   version = "0.1.0" #   authors = ["MoonL1ght <ixav1@icloud.com>"] #    build = "build.rs" #    rust links = "cudart" #  cuda,    [dependencies] libc = "0.2" #  rust     rand = "0.5.5" #  rust      [build-dependencies] cc = "1.0" # rust

Juga di folder root proyek, buat file build.rs yang akan berisi instruksi untuk membangun program karat dan mengkompilasi kode CUDA C.

Tambahkan folder src ke root proyek di mana kita akan menempatkan file kode sumber. Dalam folder src, buat empat file: main.rs - kode program utama, dot.cpp - C ++ binding (pembungkus untuk CUDA C), dot_gpu.h, dot_gpu.cu - file yang berisi kode yang dieksekusi pada GPU.

Total kami memiliki struktur proyek seperti itu:

 rust-cuda/ src/ main.rs dot.cpp dot_gpu.h dot_gpu.cu Cargo.toml build.rs

Dalam file build.rs, yang paling penting adalah menulis ini:

 println!("cargo:rustc-link-search=native=/Developer/NVIDIA/CUDA-10.1/lib"); println!("cargo:rustc-link-search=/Developer/NVIDIA/CUDA-10.1/lib"); println!("cargo:rustc-env=LD_LIBRARY_PATH=/Developer/NVIDIA/CUDA-10.1/lib"); println!("cargo:rustc-link-lib=dylib=cudart");

di mana /Developer/NVIDIA/CUDA-10.1/lib adalah path ke file executable CUDA, dalam sistem seperti unix, path ini dapat ditemukan, misalnya, dengan perintah:

 which nvcc

Selain itu, dalam file build.rs, Anda perlu menentukan path ke file dot.cpp dan dot_gpu.cpp:

 .files(&["./src/dot.cpp", "./src/dot_gpu.cu"])

Semua kode build.rs

 extern crate cc; fn main() { cc::Build::new() .cuda(true) .cpp(true) .flag("-cudart=shared") .files(&["./src/dot.cpp", "./src/dot_gpu.cu"]) .compile("dot.a"); println!("cargo:rustc-link-search=native=/Developer/NVIDIA/CUDA-10.1/lib"); println!("cargo:rustc-link-search=/Developer/NVIDIA/CUDA-10.1/lib"); println!("cargo:rustc-env=LD_LIBRARY_PATH=/Developer/NVIDIA/CUDA-10.1/lib"); println!("cargo:rustc-link-lib=dylib=cudart"); }

Sekarang Anda dapat mulai menulis kode program utama. Dalam file main.rs, Anda perlu membuat antarmuka untuk fungsi C / C ++ untuk menelepon langsung dari kode Rust. Anda dapat membaca lebih lanjut tentang ini di dokumentasi resmi di bagian FFI .

 extern "C" { //  C        fn dot(v1: *mut c_float, v2: *mut c_float, N: size_t) -> c_float; }

Untuk menyebutnya, Anda perlu menggunakan blok kode yang tidak aman, sebagai argumen kami meneruskan pointer yang bisa diubah ke tipe Vec:

 unsafe { gpu_res = dot(v1.as_mut_ptr(), v2.as_mut_ptr(), VEC_SIZE); }

Kode file main.rs penuh

 extern crate libc; extern crate rand; use libc::{c_float, size_t}; use rand::Rng; const VEC_SIZE: usize = 10; const MAX: f32 = 10.; const MIN: f32 = 0.; extern "C" { fn dot(v1: *mut c_float, v2: *mut c_float, N: size_t) -> c_float; } fn cpu_dot(v1: Vec<f32>, v2: Vec<f32>) -> f32 { let mut res: f32 = 0.; for i in 0..v1.len() { res += v1[i] * v2[i]; } return res; } fn main() { let mut v1: Vec<f32> = Vec::new(); let mut v2: Vec<f32> = Vec::new(); let mut gpu_res: c_float; let mut cpu_res: f32 = 0.; let mut rng = rand::thread_rng(); for _ in 0..VEC_SIZE { v1.push(rng.gen_range(MIN, MAX)); v2.push(rng.gen_range(MIN, MAX)); } println!("{:?}", v1); println!("{:?}", v2); println!("GPU computing started"); unsafe { gpu_res = dot(v1.as_mut_ptr(), v2.as_mut_ptr(), VEC_SIZE); } println!("GPU computing finished"); println!("GPU dot product result: {}", gpu_res); cpu_res = cpu_dot(v1, v2); println!("CPU dot product result: {}", cpu_res); }

Sekarang kita mulai menulis binding dalam C ++, serta kode untuk menghitung produk skalar vektor di CUDA C.

Dalam file dot.cpp, kami menulis fungsi penjilidan, sebenarnya kami memanggil fungsi ini dari kode Rust:

 extern "C" { float dot(float *v1, float *v2, size_t N) { float *gpu_res; float res = 0.0; gpu_res = gpu_dot(v1, v2, N); //   GPU for (int i = 0; i < blocksPerGrid; i++) { res += gpu_res[i]; } free(gpu_res); return res; } }

Kode file dot.cpp penuh

 #include <iostream> #include "dot_gpu.h" using namespace std; void display_vector(float *v, size_t N) { cout << "["; for (size_t i = 0; i < N; i++) { cout << v[i]; if (i != N - 1) { cout << ", "; } } cout << "]" << endl; } extern "C" { float dot(float *v1, float *v2, size_t N) { cout << "Calling gpu dot product" << endl; cout << "Got two vectors from rust:" << endl; display_vector(v1, N); display_vector(v2, N); float *gpu_res; float res = 0.0; gpu_res = gpu_dot(v1, v2, N); for (int i = 0; i < blocksPerGrid; i++) { res += gpu_res[i]; } free(gpu_res); return res; } }

Berikut ini adalah kode dari file dot_gpu.cu di mana perhitungan utama dilakukan, saya tidak akan menjelaskan kode itu sendiri dalam tutorial ini, karena tidak didedikasikan untuk pemrograman CUDA.

dot_gpu.cu

 #include "dot_gpu.h" __global__ void dot__(float *v1, float *v2, float *res, int N) { __shared__ float cache [threadsPerBlock]; int tid = threadIdx.x + blockIdx.x * blockDim.x; int cacheIndex = threadIdx.x; float temp = 0.0; while (tid < N) { temp += v1[tid] * v2[tid]; tid += blockDim.x * gridDim.x; } cache[cacheIndex] = temp; __syncthreads(); int i = blockDim.x / 2; while (i != 0) { if (cacheIndex < i) { cache[cacheIndex] += cache[cacheIndex + i]; } __syncthreads(); i /= 2; } if (cacheIndex == 0) { res[blockIdx.x] = cache[0]; } } float * gpu_dot (float *v1, float *v2, size_t N) { float *dev_v1, *dev_v2, *dev_res, *res; res = new float[blocksPerGrid]; cudaMalloc((void**)&dev_v1, N * sizeof(float)); cudaMalloc((void**)&dev_v2, N * sizeof(float)); cudaMalloc((void**)&dev_res, blocksPerGrid * sizeof(float)); cudaMemcpy(dev_v1, v1, N * sizeof(float), cudaMemcpyHostToDevice); cudaMemcpy(dev_v2, v2, N * sizeof(float), cudaMemcpyHostToDevice); dot__<<<blocksPerGrid, threadsPerBlock>>>(dev_v1, dev_v2, dev_res, (int)N); cudaMemcpy(res, dev_res, blocksPerGrid * sizeof(float), cudaMemcpyDeviceToHost); cudaFree(dev_v1); cudaFree(dev_v2); cudaFree(dev_res); return res; }

Semua program kecil kami ditulis dan siap dibangun. Untuk memasangnya di konsol, panggil perintah:

 cargo build

Untuk menjalankan:

 cargo run

Setelah membangun program, folder target akan muncul di direktori utama proyek. File yang dapat dieksekusi dari program kami akan berada di folder: ./target/debug/

Selain itu, jika kita hanya menjalankan file yang dapat dieksekusi, kita mendapatkan kesalahan: pustaka dyld tidak dimuat. Artinya, dia tidak dapat menemukan jalur ke perpustakaan dinamis cuda. Untuk mengatasi masalah ini, Anda dapat mendaftarkan variabel lingkungan LD_LIBRARY_PATH = path_to_CUDA_lib_directory / atau menjalankan tautan simbolik di folder rust toolchain untuk CUDA sebelum memulai file yang dapat dieksekusi di konsol:

 ln -s /Developer/NVIDIA/CUDA-10.1/lib/* /Users/Alexander/.rustup/toolchains/nightly-x86_64-apple-darwin/lib

di mana / Pengguna /Alexander/.rustup/toolchains/nightly-x86_64-apple-darwin/lib adalah jalur saya ke rantai alat karat yang diinstal, ini mungkin sedikit berbeda untuk Anda.

Saat memulai program melalui proses kargo, kesalahan seperti itu tidak terjadi, karena kami mendaftarkan variabel lingkungan LD_LIBRARY_PATH dalam file build.rs.

Pada akhirnya

Kami memiliki kemampuan untuk menjalankan kode CUDA C langsung dari kode Rust. Untuk memverifikasi ini, kami membuat program kecil, ia bekerja dengan vektor dan melakukan semua perhitungan pada GPU. Kode lengkap juga dapat dilihat di github .

Menulis di Rust + CUDA C

CUDA C

RUST + CUDA C

Pada akhirnya

More articles: