मुझे पता है कि "प्रत्येक वार में निरंतर धागे होते हैं, थ्रेड 0 युक्त पहले वार्प के साथ थ्रेड आईडी बढ़ते हैं" तो पहले 32 थ्रेड पहले वार में होना चाहिए। साथ ही मुझे पता है कि एक वार में सभी धागे किसी भी उपलब्ध स्ट्रीमिंग मल्टीप्रोसेसर पर एक साथ निष्पादित किए जाते हैं।CUDA। पहले 32 धागे को अनलोल कैसे करें ताकि उन्हें समानांतर में निष्पादित किया जा सके?
जैसा कि मैंने समझा, क्योंकि थ्रेड सिंकिंग में कोई आवश्यकता नहीं है, अगर केवल एक वार्प निष्पादित किया जा रहा है। लेकिन यदि नीचे if ब्लॉक में __syncthreads() में से कोई भी हटा देता है तो नीचे दिया गया कोड गलत जवाब उत्पन्न करता है। मैंने कारण खोजने की कोशिश की लेकिन कुछ भी खत्म नहीं हुआ। मैं वास्तव में आपकी मदद के लिए आशा करता हूं, तो आप मुझे बता सकते हैं कि इस कोड के साथ क्या गलत है? मैं केवल अंतिम __syncthreads() क्यों नहीं छोड़ सकता और सही उत्तर प्राप्त कर सकता हूं?
#define BLOCK_SIZE 128
__global__ void reduce (int * inData, int * outData)
{
__shared__ int data [BLOCK_SIZE];
int tid = threadIdx.x;
int i = blockIdx.x * blockDim.x + threadIdx.x;
data [tid] = inData [i] + inData [i + blockDim.x/2 ];
__syncthreads();
for (int s = blockDim.x/4; s > 32; s >>= 1)
{
if (tid < s)
data [tid] += data [tid + s];
__syncthreads();
}
if (tid < 32)
{
data [tid] += data [tid + 32];
__syncthreads();
data [tid] += data [tid + 16];
__syncthreads();
data [tid] += data [tid + 8];
__syncthreads();
data [tid] += data [tid + 4];
__syncthreads();
data [tid] += data [tid + 2];
__syncthreads();
data [tid] += data [tid + 1];
__syncthreads();
}
if (tid == 0)
outData [blockIdx.x] = data [0];
}
void main()
{
...
reduce<<<dim3(128), dim3(128)>>>(dev_data, dev_res);
...
}
पीएस मैं GT560Ti