內(nèi)存聲明：
設(shè)備和常量內(nèi)存：
使用__device__或__constant__內(nèi)存空間說明符在文件范圍內(nèi)聲明的內(nèi)存在使用設(shè)備運行時時具有相同的行為。 所有內(nèi)核都可以讀取或?qū)懭朐O(shè)備變量，無論內(nèi)核是由主機還是設(shè)備運行時初始啟動。 等同地，所有內(nèi)核將具有與在模塊范圍內(nèi)聲明的__constant__相同的視圖。
紋理和表面內(nèi)存：
CUDA支持動態(tài)創(chuàng)建的紋理和表面對象1，其中可以在主機上創(chuàng)建紋理參考，傳遞給內(nèi)核，由內(nèi)核使用，然后從主機中銷毀。 設(shè)備運行時不允許從設(shè)備代碼中創(chuàng)建或銷毀紋理或表面對象，但可以在設(shè)備上自由使用并自由傳送由主機創(chuàng)建的紋理和表面對象。 無論它們在哪里創(chuàng)建，動態(tài)創(chuàng)建的紋理對象總是有效的，并且可以從父級傳遞給子級內(nèi)核。
設(shè)備運行時不支持從設(shè)備啟動的內(nèi)核中的傳統(tǒng)模塊范圍（即費米架構(gòu)）紋理和表面。 模塊范圍（遺留）紋理可以從主機創(chuàng)建并在設(shè)備代碼中用于任何內(nèi)核，但只能由頂級內(nèi)核（即從主機啟動的）使用。

1：動態(tài)創(chuàng)建的紋理和曲面對象是CUDA 5.0引入的CUDA內(nèi)存模型的補充。 有關(guān)詳細信息，請參閱CUDA編程指南。
共享內(nèi)存變量聲明：
在CUDA中，C / C ++共享內(nèi)存可以聲明為靜態(tài)大小的文件范圍變量或函數(shù)范圍變量，也可以聲明為外部變量，其大小由內(nèi)核的調(diào)用者在運行時通過啟動配置參數(shù)確定。 這兩種類型的聲明在設(shè)備運行時間下均有效。

__global__ void permute(int n, int *data) {extern __shared__ int smem[];if (n <= 1)return;smem[threadIdx.x] = data[threadIdx.x];__syncthreads();permute_data(smem, n);__syncthreads();// Write back to GMEM since we can't pass SMEM to children.data[threadIdx.x] = smem[threadIdx.x];__syncthreads();if (threadIdx.x == 0) {permute << < 1, 256, n / 2 * sizeof(int) >> >(n / 2, data);permute << < 1, 256, n / 2 * sizeof(int) >> >(n / 2, data + n / 2);} } void host_launch(int *data) {permute << < 1, 256, 256 * sizeof(int) >> >(256, data); }

符號地址：
由于所有全局范圍的設(shè)備變量都在內(nèi)核的可見地址空間中，所以設(shè)備端符號（即標記為__ device_的那些符號）可以通過簡單的＆運算符從內(nèi)核中引用。 這也適用于__constant__符號，但在這種情況下，指針將引用只讀數(shù)據(jù)。
考慮到設(shè)備端符號可以直接引用，引用符號的CUDA運行時API（例如cudaMemcpyToSymbol（）或cudaGetSymbolAddress（））是多余的，因此設(shè)備運行時不支持。 注意這意味著即使在子內(nèi)核啟動之前，也不能在運行的內(nèi)核中更改常量數(shù)據(jù)，因為對__constant__空間的引用是只讀的。

總結(jié)

以上是生活随笔為你收集整理的CUDA学习（六十三）的全部內(nèi)容，希望文章能夠幫你解決所遇到的問題。

如果覺得生活随笔網(wǎng)站內(nèi)容還不錯，歡迎將生活随笔推薦給好友。