से कॉपी करें मैं क्यूडा और सी ++ का उपयोग करके जीपीयू पर दो कार्यों (2 कर्नेल में अलग) करने की कोशिश कर रहा हूं। इनपुट के रूप में मैं एक एनएक्सएम मैट्रिक्स लेता हूं (मेजबान पर एक फ्लोट सरणी के रूप में स्मृति में संग्रहीत)। मैं फिर एक कर्नेल का उपयोग करूंगा जो इस मैट्रिक्स पर कुछ संचालन करता है ताकि इसे एक NxMxD मैट्रिक्स बनाया जा सके। मेरे पास एक दूसरा कर्नेल है जो इस 3 डी मैट्रिक्स पर कुछ संचालन करता है (मैं केवल मानों को पढ़ता हूं, मुझे मूल्य लिखना नहीं है)।कूडा - डिवाइस ग्लोबल मेमोरी से बनावट मेमोरी
बनावट मेमोरी में ऑपरेटिंग मेरे काम के लिए बहुत तेज़ प्रतीत होता है, इसलिए मेरा सवाल यह है कि कर्नेल 1 के बाद डिवाइस पर ग्लोबल मेमोरी से अपने डेटा को प्रतिलिपि बनाना संभव है और इसे सीधे बिना कर्नेल 2 के बनावट मेमोरी में स्थानांतरित करना संभव है मेजबान वापस?
अद्यतन
मैं मेरी समस्या बेहतर वर्णन करने के लिए कुछ कोड जोड़ दिया है।
यहां दो कर्नल हैं। पहला अभी सिर्फ एक प्लेसहोल्डर है और 2 डी मैट्रिक्स को 3 डी में दोहराता है।
__global__ void computeFeatureVector(float* imData3D_dev, int imX, int imY, int imZ) {
//calculate each thread global index
int xindex=blockIdx.x*blockDim.x+threadIdx.x;
int yindex=blockIdx.y*blockDim.y+threadIdx.y;
#pragma unroll
for (int z=0; z<imZ; z++) {
imData3D_dev[xindex+yindex*imX + z*imX*imY] = tex2D(texImIp,xindex,yindex);
}
}
दूसरा यह 3 डी मैट्रिक्स लेगा, जिसे अब बनावट के रूप में दर्शाया गया है और कुछ संचालन करता है। अभी के लिए खाली
__global__ void kernel2(float* resData_dev, int imX) {
//calculate each thread global index
int xindex=blockIdx.x*blockDim.x+threadIdx.x;
int yindex=blockIdx.y*blockDim.y+threadIdx.y;
resData_dev[xindex+yindex*imX] = tex3D(texImIp3D,xindex,yindex, 0);
return;
}
तो कोड के मुख्य शरीर इस प्रकार है:
// declare textures
texture<float,2,cudaReadModeElementType> texImIp;
texture<float,3,cudaReadModeElementType> texImIp3D;
void main_fun() {
// constants
int imX = 1024;
int imY = 768;
int imZ = 16;
// input data
float* imData2D = new float[sizeof(float)*imX*imY];
for(int x=0; x<imX*imY; x++)
imData2D[x] = (float) rand()/RAND_MAX;
//create channel to describe data type
cudaArray* carrayImIp;
cudaChannelFormatDesc channel;
channel=cudaCreateChannelDesc<float>();
//allocate device memory for cuda array
cudaMallocArray(&carrayImIp,&channel,imX,imY);
//copy matrix from host to device memory
cudaMemcpyToArray(carrayImIp,0,0,imData2D,sizeof(float)*imX*imY,cudaMemcpyHostToDevice);
// Set texture properties
texImIp.filterMode=cudaFilterModePoint;
texImIp.addressMode[0]=cudaAddressModeClamp;
texImIp.addressMode[1]=cudaAddressModeClamp;
// bind texture reference with cuda array
cudaBindTextureToArray(texImIp,carrayImIp);
// kernel params
dim3 blocknum;
dim3 blocksize;
blocksize.x=16; blocksize.y=16; blocksize.z=1;
blocknum.x=(int)ceil((float)imX/16);
blocknum.y=(int)ceil((float)imY/16);
// store output here
float* imData3D_dev;
cudaMalloc((void**)&imData3D_dev,sizeof(float)*imX*imY*imZ);
// execute kernel
computeFeatureVector<<<blocknum,blocksize>>>(imData3D_dev, imX, imY, imZ);
//unbind texture reference to free resource
cudaUnbindTexture(texImIp);
// check copied ok
float* imData3D = new float[sizeof(float)*imX*imY*imZ];
cudaMemcpy(imData3D,imData3D_dev,sizeof(float)*imX*imY*imZ,cudaMemcpyDeviceToHost);
cout << " kernel 1" << endl;
for (int x=0; x<10;x++)
cout << imData3D[x] << " ";
cout << endl;
delete [] imData3D;
//
// kernel 2
//
// copy data on device to 3d array
cudaArray* carrayImIp3D;
cudaExtent volumesize;
volumesize = make_cudaExtent(imX, imY, imZ);
cudaMalloc3DArray(&carrayImIp3D,&channel,volumesize);
cudaMemcpyToArray(carrayImIp3D,0,0,imData3D_dev,sizeof(float)*imX*imY*imZ,cudaMemcpyDeviceToDevice);
// texture params and bind
texImIp3D.filterMode=cudaFilterModePoint;
texImIp3D.addressMode[0]=cudaAddressModeClamp;
texImIp3D.addressMode[1]=cudaAddressModeClamp;
texImIp3D.addressMode[2]=cudaAddressModeClamp;
cudaBindTextureToArray(texImIp3D,carrayImIp3D,channel);
// store output here
float* resData_dev;
cudaMalloc((void**)&resData_dev,sizeof(float)*imX*imY);
// kernel 2
kernel2<<<blocknum,blocksize>>>(resData_dev, imX);
cudaUnbindTexture(texImIp3D);
//copy result matrix from device to host memory
float* resData = new float[sizeof(float)*imX*imY];
cudaMemcpy(resData,resData_dev,sizeof(float)*imX*imY,cudaMemcpyDeviceToHost);
// check copied ok
cout << " kernel 2" << endl;
for (int x=0; x<10;x++)
cout << resData[x] << " ";
cout << endl;
delete [] imData2D;
delete [] resData;
cudaFree(imData3D_dev);
cudaFree(resData_dev);
cudaFreeArray(carrayImIp);
cudaFreeArray(carrayImIp3D);
}
इम खुशी है कि पहले गिरी सही ढंग से काम कर रहा है लेकिन 3 डी मैट्रिक्स imData3D_dev सही ढंग से बनावट texImIp3D लिए बाध्य होने के लिए प्रतीत नहीं होता ।
उत्तर
मैं cudaMemcpy3D का उपयोग कर मेरी समस्या हल। मुख्य समारोह के दूसरे भाग के लिए संशोधित कोड यहां दिया गया है। imData3D_dev में पहले कर्नेल से वैश्विक मेमोरी में 3 डी मैट्रिक्स शामिल है।
cudaArray* carrayImIp3D;
cudaExtent volumesize;
volumesize = make_cudaExtent(imX, imY, imZ);
cudaMalloc3DArray(&carrayImIp3D,&channel,volumesize);
cudaMemcpy3DParms copyparms={0};
copyparms.extent = volumesize;
copyparms.dstArray = carrayImIp3D;
copyparms.kind = cudaMemcpyDeviceToDevice;
copyparms.srcPtr = make_cudaPitchedPtr((void*)imData3D_dev, sizeof(float)*imX,imX,imY);
cudaMemcpy3D(©parms);
// texture params and bind
texImIp3D.filterMode=cudaFilterModePoint;
texImIp3D.addressMode[0]=cudaAddressModeClamp;
texImIp3D.addressMode[1]=cudaAddressModeClamp;
texImIp3D.addressMode[2]=cudaAddressModeClamp;
cudaBindTextureToArray(texImIp3D,carrayImIp3D,channel);
// store output here
float* resData_dev;
cudaMalloc((void**)&resData_dev,sizeof(float)*imX*imY);
kernel2<<<blocknum,blocksize>>>(resData_dev, imX);
// ... clean up
जो मेरी समस्या का समाधान करता है। – themush