原文鏈接

偶有興趣測試了一下題目中提到的這兩個函數，為了滿足對齊訪問數據，咱們平時可能會用到cudamallocPitch，以為它會帶來更高的效率。呵呵，這里給出一段測試程序，大家可以在自己的機器上跑跑，你會發現這兩個函數在某些情況下是一樣的。

[cpp]?view plaincopy

#include?<stdio.h>??

#include?<stdlib.h>??

#include?<cuda_runtime_api.h>??

??

int?main(int?argc,?char?**argv)??

{??

????//?device?pointers.??

????float?*d_pitch;??

????float?*d_normal;??

??

????//?matrix?size.??

????size_t?cols?=?63;??

????size_t?rows?=?16;??

??????

????size_t?pitch?=?0;??

??????

????//?alloc?the?data?form?gpu?memory.??

????cudaMallocPitch((void**)&d_pitch,?&pitch,?cols*sizeof(float),?rows);??

????cudaMalloc((void**)(&d_normal),?rows*cols*sizeof(float));??

??????

????//?test?the?data?address.??

????fprintf(stdout,?"row?size(in?bytes)?=?%.2f*128.\n",?pitch/128.0f);??

????fprintf(stdout,?"the?head?address?of?d_pitch??mod?128?=?%x.\n",?((unsigned?int)d_pitch)%128);??

????fprintf(stdout,?"the?head?address?of?d_normal?mod?128?=?%x.\n",?((unsigned?int)d_normal)%128);??

??????

????cudaFree(d_pitch);??

????cudaFree(d_normal);??

??

????getchar();??

????return?0;??

}??

?

上面這段程序的運行結果如下：

[cpp]?view plaincopy

row?size(in?bytes)?=?28.00*128.??

the?head?address?of?d_pitch?mod?128?=?0.??

the?head?address?of?d_normal?mod?128?=?0.???

?

我多次做過實驗，我覺得從以上實驗結果可以知道，無論如何改變實驗的參數，兩個顯存申請函數返回的數據首地址都是128，256的整數倍，我猜想GPU上的每個計算單元的數據在全局中加載的時候一次可以連續加載2的冪次個數據，并且這些數據的加載其實地址一定也是2的冪次，所以warp使用全局內存中的數據的時候應該盡量按照對齊的原則加載數據，這樣就可以獲得更高的效率了。至于對齊原則可以在CUDA的編程手冊中找到。

轉載于:https://www.cnblogs.com/liangliangdetianxia/p/4165946.html

總結

以上是生活随笔為你收集整理的cudaMalloc和cudaMallocPitch的全部內容，希望文章能夠幫你解決所遇到的問題。

如果覺得生活随笔網站內容還不錯，歡迎將生活随笔推薦給好友。