شرح تعدادی از دستورات مهم مثال CUDA ارائه شده در پست شماره 5
سلام
امروز چندتا از دستورات کلیدی cuda رو شرح میدم.
1- برای اینکه بتونید از cuda در برنامه تون استفاده کنید باید چندتا فایل رو include کنید. (اول همه تنظیماتی که در پست های قبل گفتم رو باید انجام داده باشید)
البته برای کارای پیشرفته تر باید فایل های اضافه دیگه ای رو هم include کنید که بستگی به کاری که می خواید انجام بدین داره اما همین اول کاری که ما فقط داریم روی آرایه ها کار می کنیم include این دوتا فایل برای استفاده از cuda کافیه
#include <cuda.h>
#include<cuda_runtime.h>
برای مثال ما می خوایم برنامه ای رو که در پست های قبل گذاشتم رو ببینیم هر کدوم از دستوراتش به چه کاری میان
شما میخواید روی یک آرایه سه بعدی کار کنید و ببینید محاسباتتون در cpu سریعتر انجام می شه یا در GPU (همون پردازنده گرافیکی تون)
اول باید مشخص کنید که هر کدوم از ابعاد آرایه تون چه اندازه ای داره؟ مثلا یک آرایه 10*10*10.
آرایه تون رو باید بصورت اشاره گر تعریف کنید نه به صورت استاتیک(درحالت معمول). بصورت زیر تعریف کنید:
اول مشخص کنید چه عددی رو برای هر بعد درنظر گرفتید
#define DATAXSIZE 10
#define DATAYSIZE 10
#define DATAZSIZE 10
حالا من برای راحتی کارم یه نوع داده ای جدید به اسم matrix ایجاد کردم که به وسیله اون می تونم به راحتی هرچندتا آرایه که خواستم تعریف کنم.
typedef float MATRIX[DATAYSIZE][DATAXSIZE];
این matrix ما فعلا فقط ابعاد x و y ش مشخصه. (چون داریم از اشاره گر ها استفاده می کنیم بعد z رو بعدا مشخص می کنیم حالا بعدا میگم چطوری)
یه اشاره گر به ماتریس ایجاد می کنیم بر روی RAM که CPU مون باهاش کار می کنه (از نوع داده ای جدید به اسم matrix)
MATRIX *vv; // storage for result stored on host
یه اشاره گر ماتریس دیگه ایجاد می کنیم که روی حافظه پردازنده گرافیکی هست و GPU مون باهاش کار می کنه. اونم از نوع matrix هست. (البته بگ اینجا مشخص نمی شه که شما هر کدوم از این ماتریس ها رو برای کدوم حافظه استفاده می کنید. بعدا براتون میگم از کجا مشخص کنید.
MATRIX *d_vv; // storage for result computed on device
این آرایه هم از نوع اشاره گر هست.
خب حالا باید مشخص کنیم که بلوک هی نخی که قراره روی پردازنده گرافیکی اجرا بشه باید چه اندازه باشه.
#define BLKXSIZE 32
#define BLKYSIZE 4
#define BLKZSIZE 4
این قسمت مشخص میکنه که بلوک شما سه بعدی هست و ابعادش چقدره
const dim3 blockSize(BLKXSIZE, BLKYSIZE, BLKZSIZE);
در قسمت بعد مشخص می کنه با توجه به اندازه آرایه تون gridی که رو پردازنده گرافیکی نخ های شما رو مدیریت می کنه به چه اندازه باید باشه و از چندتا بلوک نخ تشکیل شده
const dim3 gridSize(((DATAXSIZE+BLKXSIZE-1)/BLKXSIZE), ((DATAYSIZE+BLKYSIZE-1)/BLKYSIZE), ((DATAZSIZE+BLKZSIZE-1)/BLKZSIZE));
حالا این قسمت های سبز رو بپذیرید تا در یک پست جداگانه مفصل تر توضیح بدم که چرا این ابعاد رو استفاده کردیم و حکمتش چیه.
خب حالا بریم توی بدنه اصلی برنامه main()
اول از همه تعریف آرایه مون رو در RAM کامپیوتر تکمیل می کنیم. این قسمت مشخص می کنه که یه ماتریس vv رو که قبلا از نوع matrix تعریف کردیم دقیقا چه ابعادی داره. دستور malloc مقدار nx*ny*nz خونه از حافظه رو به اندازه نوع داده ای float به ماتریس vv میده. دستور if هم مشخص می کنه که اگر این تخصیص درست انجام نشد (یعنی تابع malloc مقدار 0 برگردوند) پیغام بده که خطا اتفاق افتاده.
// allocate storage for data set
if ((vv = (MATRIX *)malloc((nx*ny*nz)*sizeof(float))) == 0) {fprintf(stderr,"malloc1 Fail \n"); return 1;}
در این قسمت مشخص می کنیم که d_vv مربوط به حافظه پردازنده گرافیکی هست و اندهزه ش برابر nx*ny*nz هست و هر خونه هم اندازه ش به اندازه متغیر float هست.
cudaMalloc((void **) &d_vv, (nx*ny*nz)*sizeof(float));
اگر در تخصیص فضا به ماتریس در پردازنده گرافیکی به مشکلی برخورد با استفاده از این دستور رشته ای رو که قبل از تعریف تابع main() تعریف کردین نشون میده
و این مقدار رشته ای رو داخل رشته تعریف شده میریزه
cudaCheckErrors("Failed to allocate device buffer");
اگر در تخصیص فضا مشکلی نبود میاد محتوای ماتریس vv رو که داخل RAM قرار داره درون ماتریس d_vv که درون حافظه پردازنده گرافیکی قرار داره کپی می کنه(مخصوصا برای وقتایی که مقدار اولیه تون خارج از کرنل تولید میشه لازمه. به هرحال باید این دو فضا رو بهم ربط بدین چون در آخر مقادیر تولید شده توسط GPU باید به حافظه رم منتقل بشه.
cudaMemcpy(d_vv, vv, ((nx*ny*nz)*sizeof(float)), cudaMemcpyHostToDevice);
حالا این دستور مشخص می کنه که آیا در این عملیات کپی خطایی صورت گرفته یا خیر
cudaCheckErrors("CUDA memcpy failure");
حالا از این به بعد باید وارد کرنل GPU مون بشیم و بگیم که محاسبات از این به بعد درون GPU انجام بشه . این دستور یه کرنل ایجاد می کنه به نام set که مشخص می کنه از چندتا بلوک نخ استفاده شده(توسط gridSize) و هر بلوک نخ چندتا نخ می تونه داشته باشه(blockSize) و در قسمت پرانتز هم d_vv به کرنل فراخوانی می شه.
set<<<gridSize,blockSize>>>(d_vv);
این دستور مشخص میکنه که اگر در اجرای کرنل مشکلی ایجاد شد پیام خطا میده
cudaCheckErrors("Kernel launch failure");
خب بعد از اینکه محاسبات در هسته GPU اجرا شد لازمه که این نتیجه محاسبات دوباره به ماتریس داخل RAM کپی بشه.
cudaMemcpy(vv, d_vv, ((nx*ny*nz)*sizeof(float)), cudaMemcpyDeviceToHost);
حالا اگر خطایی در این عمل کپی کردن وجود داشته باشه دستور زیر اونو به ما اطلاع میده
cudaCheckErrors("CUDA memcpy failure");
کار با GPU تموم شد و بقیه دستورا مربوط به CPU هست.
نکته مهمی که وجود داره اینه که بخش زمان بر کار مربوط به کپی داده ها از RAM به حافظه GPU و برعکس هستش.
اجرای هسته GPU هم معمولا کمتر از یک میلی ثانیه زمان می بره. من که هرچه زمان اجرای هسته رو اندازه گرفتم نتیجه صفر میلی ثانیه بوده(لابد در حد هزارم ثانیه)
اما عملیات کپی داده ها از RAM به حافظه GPU و برعکس متناسب با قدرت GPU و پهنای باندش ممکنه چند ده ثانیه زمان بگیره (شایدم بیش از صد ثانیه) که این زمان کاملا به حجم ماتریس شما بستگی داره.
اگر زمان انجام محاسبات شما روی CPU مثلا ده هزار ثانیه زمان می بره خیلی معقوله شما صد ثانیه برای کپی داده ها به GPU و برعکس مصرف کنید اما محاسباتتون کمتر از یک میلی ثانیه اجرا بشه. اون وقت می تونید با خوشحالی بگید کل زمان اجرا مثلا 100 ثانیه بوده ( نه کمتر از یک میلی ثانیه)
اما اگر زمان محاسبات شما 5 ثانیه هست اون وقت این عملیات کپی شاید نیم ثانیه طول بکشه.
اون وقت نمی ارزه که شما یک میلیون و نیم بابت خرید یه کارت گرافیک قابل برنامه نویسی خوب هزینه کنید
توضیح دستورای داخل هسته GPU که تابعش قبل از تابع main() نوشته شده بماند برای بعد
امیدوارم مفید بوده باشه.
اگر شما هم توضیحی در این مورد دارید بفرمایید ما هم استفاده کنیم