เริ่มต้น CUDA Programming
วันนี้เราจะมาเรียนรู้เรื่องการประมวลผลแบบขนาน ด้วย CUDA ในภาษา Python โดยเฉพาะอย่างยิ่งในงานการประมวลผลภาพ ที่เราจะได้เห็นการแปลงภาพ RGB เป็นภาพเฉดสีเทา และการคูณเมทริกซ์ ซึ่งเป็นพื้นฐานสำคัญในงานประมวลผลภาพ และความเข้าใจในการทำงานของ CUDA
CUDA คืออะไร
CUDA เป็นเทคโนโลยีที่เปิดประตูสู่การใช้งาน GPU สำหรับการคำนวณความเร็วสูง โดยเฉพาะอย่างยิ่งในการประมวลผลภาพ การเริ่มต้นทำงานกับ CUDA ใน Python นั้นไม่ยากเลยครับ เราเริ่มจากการโหลดภาพจาก URL และทำการปรับขนาดภาพ ก่อนจะแปลงภาพจากระบบสี RGB เป็นเฉดสีเทา โดยใช้ CUDA kernels ดังตัวอย่างใน Jupyter Notebook ด้านล่าง
ตัวอย่างการแปลงภาพสี RGB เป็นภาพขาวดำ Greyscale
การแปลงภาพสี RGB เป็น ภาพขาวดำ Greyscale เป็นหนึ่งในงานที่เหมาะสมที่สุดสำหรับการแสดงศักยภาพของ CUDA เนื่องจากการประมวลผลแต่ละพิกเซลสามารถทำได้แบบอิสระ การใช้ CUDA ช่วยเร่งกระบวนการนี้ได้อย่างมาก ทำให้เห็นความแตกต่างชัดเจนเมื่อเทียบกับการใช้ Python อย่างเดียว
Python Code
def blk_kernel(f, blocks, threads, *args):
for i in range(blocks):
for j in range(threads): f(i, j, threads, *args)
def rgb2grey_bk(blockidx, threadidx, blockdim, x, out, n):
i = blockidx*blockdim + threadidx
if i<n: out[i] = 0.2989*x[i] + 0.5870*x[i+n] + 0.1140*x[i+2*n]
def rgb2grey_pybk(x):
c,h,w = x.shape
n = h*w
x = x.flatten()
res = torch.empty(n, dtype=x.dtype, device=x.device)
threads = 256
blocks = int(math.ceil(h*w/threads))
blk_kernel(rgb2grey_bk, blocks, threads, x, res, n)
return res.view(h,w)CPP Code
cuda_src = cuda_begin + r'''
__global__ void rgb_to_grayscale_kernel(unsigned char* x, unsigned char* out, int n) {
int i = blockIdx.x*blockDim.x + threadIdx.x;
if (i<n) out[i] = 0.2989*x[i] + 0.5870*x[i+n] + 0.1140*x[i+2*n];
}
torch::Tensor rgb_to_grayscale(torch::Tensor input) {
CHECK_INPUT(input);
int h = input.size(1);
int w = input.size(2);
printf("h*w: %d*%d\n", h, w);
auto output = torch::empty({h,w}, input.options());
int threads = 256;
rgb_to_grayscale_kernel<<<cdiv(w*h,threads), threads>>>(
input.data_ptr<unsigned char>(), output.data_ptr<unsigned char>(), w*h);
C10_CUDA_KERNEL_LAUNCH_CHECK();
return output;
}'''ตัวอย่างการคูณเมทริกซ์
การคูณเมทริกซ์คืออีกหนึ่งงานที่สามารถประโยชน์จาก CUDA ได้มาก การใช้ CUDA kernels ในการคูณเมทริกซ์ไม่เพียงแต่เร่งความเร็วในการคำนวณ แต่ยังช่วยให้เราสามารถจัดการกับข้อมูลขนาดใหญ่ได้มากขึ้น โดยไม่ทำให้เกิดปัญหาด้านประสิทธิภาพ
Python Code
def blk_kernel2d(f, blocks, threads, *args):
for i0 in range(blocks.y):
for i1 in range(blocks.x):
for j0 in range(threads.y):
for j1 in range(threads.x): f(ns(x=i0,y=i1), ns(x=j0,y=j1), threads, *args)
def matmul_bk(blockidx, threadidx, blockdim, m, n, out, h, w, k):
r = blockidx.y*blockdim.y + threadidx.y
c = blockidx.x*blockdim.x + threadidx.x
if (r>=h or c>=w): return
o = 0.
for i in range(k): o += m[r*k+i] * n[i*w+c]
out[r*w+c] = o
def matmul_2d(m, n):
h,k = m.shape
k2,w = n.shape
assert k==k2, "Size mismatch!"
output = torch.zeros(h, w, dtype=m.dtype)
tpb = ns(x=16,y=16)
blocks = ns(x=math.ceil(w/tpb.x), y=math.ceil(h/tpb.y))
blk_kernel2d(matmul_bk, blocks, tpb,
m.flatten(), n.flatten(), output.flatten(), h, w, k)
return outputCPP Code
cuda_src = cuda_begin + r'''
__global__ void matmul_k(float* m, float* n, float* out, int h, int w, int k) {
int r = blockIdx.y*blockDim.y + threadIdx.y;
int c = blockIdx.x*blockDim.x + threadIdx.x;
if (r>=h || c>=w) return;
float o = 0;
for (int i = 0; i<k; ++i) o += m[r*k+i] * n[i*w+c];
out[r*w+c] = o;
}
torch::Tensor matmul(torch::Tensor m, torch::Tensor n) {
CHECK_INPUT(m); CHECK_INPUT(n);
int h = m.size(0);
int w = n.size(1);
int k = m.size(1);
TORCH_CHECK(k==n.size(0), "Size mismatch!");
auto output = torch::zeros({h, w}, m.options());
dim3 tpb(16,16);
dim3 blocks(cdiv(w, tpb.x), cdiv(h, tpb.y));
matmul_k<<<blocks, tpb>>>(
m.data_ptr<float>(), n.data_ptr<float>(), output.data_ptr<float>(), h, w, k);
C10_CUDA_KERNEL_LAUNCH_CHECK();
return output;
}
'''ตัวอย่างการประยุกต์ใช้ในงาน AI
ในการนำเสนอนี้ เรายังได้เห็นตัวอย่างการใช้งาน CUDA กับชุดข้อมูล MNIST ซึ่งเป็นชุดข้อมูลที่มีความสำคัญในโลกของ Machine Learning และ Computer Vision การใช้ CUDA ในการคูณเมทริกซ์บนข้อมูลภาพเหล่านี้ช่วยเปิดโอกาสให้เราสามารถประมวลผลข้อมูลขนาดใหญ่ได้เร็วขึ้นอย่างมาก
สรุป
การศึกษา CUDA Programming ในการประมวลผลภาพด้วย Python นั้นเป็นทักษะที่มีค่ามากในยุคปัจจุบัน ไม่ว่าจะเป็นตัวอย่างการแปลงภาพสี RGB เป็นภาพขาวดำ Greyscale หรือการคูณเมทริกซ์ การใช้ CUDA ไม่เพียงแต่ช่วยเพิ่มความเร็วในการประมวลผลเท่านั้น แต่ยังช่วยให้การพัฒนาเป็นไปได้ด้วยความยืดหยุ่นและประสิทธิภาพสูงสุด
เรามาเริ่มกันเลยดีกว่า
แชร์ให้เพื่อน:
