cudaMemCpy2D崩潰

Question

我想在cuda中實現Sauvola二值化。爲此，我已經讀取了主機中的二維數組中的圖像，並在使用pitch的設備中爲二維數組分配內存。分配內存後，我嘗試複製主機使用cudaMemcpy2D二維數組構成2D陣列，它編譯罰款，但它崩潰這裏runtime.I我無法理解我在哪裏丟失的，請提示我已經寫是something.The代碼如下：

#include "BinMain.h" 
#include "Binarization.h" 
#include <stdlib.h> 
#include <stdio.h> 
#include <conio.h> 
#include <cuda.h> 
#include <cuda_runtime.h> 

void printDevProp(cudaDeviceProp); 
void CUDA_SAFE_CALL(cudaError_t); 


int main() 
{ 
    //Read an IplImage in imgOriginal as grayscale 
    IplImage * imgOriginal = cvLoadImage("E:\\1.tiff",CV_LOAD_IMAGE_GRAYSCALE); 

    //Create a size variable of type CvSize for cvCreateImage Parameter 
    CvSize size = cvSize(imgOriginal->width,imgOriginal->height); 

    //create an image for storing the result image with same height and width as  imgOriginal 
    IplImage * imgResult = cvCreateImage(size,imgOriginal->depth,imgOriginal- >nChannels); 

    //Create a 2D array for storing the pixels value of each of the pixel of imgOriginal grayscale image 
    int ** arrOriginal = (int **)malloc(imgOriginal->height * sizeof(int *)); 
    for (int i = 0; i < imgOriginal->height; i++) 
{ 
    arrOriginal[i] = (int*)malloc(imgOriginal->width * sizeof(int)); 
} 

//Create a 2D array for storing the returned device array 
int ** arrReturn = (int **)malloc(imgOriginal->height * sizeof(int *)); 
for (int i = 0; i < imgOriginal->height; i++) 
{ 
    arrReturn[i] = (int*)malloc(imgOriginal->width * sizeof(int)); 
} 

//Create a CvScalar variable to copy pixel values in 2D array (arrOriginal) 
CvScalar s; 

//Copying the pixl values 
for(int j = 0;j<imgOriginal->height;j++) 
{ 
    for(int k =0;k<imgOriginal->width;k++) 
    { 
     s = cvGet2D(imgOriginal,j,k); 
     arrOriginal[j][k] = s.val[0]; 
    } 
} 

//Cuda Device Property 
int devCount; 
cudaGetDeviceCount(&devCount); 
printf("CUDA Device Query...\n"); 
printf("There are %d CUDA devices.\n", devCount); 

// Iterate through devices 
for (int i = 0; i < devCount; ++i) 
{ 
    // Get device properties 
    printf("\nCUDA Device #%d\n", i); 
    cudaDeviceProp devProp; 
    cudaGetDeviceProperties(&devProp, i); 
    printDevProp(devProp); 
} 

//Start the clock 
clock_t start = clock(); 

//Allocating Device memory for 2D array using pitch 
size_t host_orig_pitch = imgOriginal->width * sizeof(int)* imgOriginal->height; //host original array pitch in bytes 
size_t dev_pitch; //device array pitch in bytes which will be used in cudaMallocPitch 
size_t dev_pitchReturn; //device return array pitch in bytes 
size_t host_ret_pitch = imgOriginal->width * sizeof(int)* imgOriginal->height; //host return array pitch in bytes 

int * devArrOriginal; //device 2d array of original image 
int * result; //device 2d array for returned array 
int dynmicRange = 128; //Dynamic Range for calculating the threshold from sauvola's formula 

//Allocating memory by using cudaMallocPitch 
CUDA_SAFE_CALL(cudaMallocPitch((void**)&devArrOriginal,&dev_pitch,imgOriginal->width * sizeof(int),imgOriginal->height * sizeof(int))); 

//Allocating memory for returned array 
CUDA_SAFE_CALL(cudaMallocPitch((void**)&result,&dev_pitchReturn,imgOriginal->width * sizeof(int),imgOriginal->height * sizeof(int))); 

//Copying 2D array from host memory to device mempry by using cudaMemCpy2D 
CUDA_SAFE_CALL(cudaMemcpy2D((void*)devArrOriginal,dev_pitch,(void*)arrOriginal,host_orig_pitch,imgOriginal->width * sizeof(float),imgOriginal->height,cudaMemcpyHostToDevice)); 
    int windowSize = 19; //Size of the window for calculating mean and variance 
    //Launching the kernel by calling myKernelLauncher function. 
    myKernelLauncher(devArrOriginal,result,windowSize,imgOriginal->width,imgOriginal- >height,dev_pitch,dynmicRange); 
    //Calling the sauvola binarization function by passing the parameters as 
    //1.arrOriginal 2D array 2.Original image height 3.Original image width 
    //int ** result = AdaptiveBinarization(arrOriginal,imgOriginal->height,imgOriginal- >width);//binarization(arrOriginal,imgOriginal->width,imgOriginal->height); 
    // 
CUDA_SAFE_CALL(cudaMemcpy2D(arrReturn,host_ret_pitch,result,dev_pitchReturn,imgOriginal->width * sizeof(int),imgOriginal->height * sizeof(int),cudaMemcpyDeviceToHost)); 
//create a CvScalar variable to set the data in imgResult 
CvScalar ss; 

//Copy the pixel values from returned array to imgResult 
for(int i=0;i<imgOriginal->height;i++) 
{ 
    for(int j=0;j<imgOriginal->width;j++) 
    { 
     ss = cvScalar(arrReturn[i][j]*255); 
     cvSet2D(imgResult,i,j,ss); 
     //k++; //No need for k if returned array is 2D 
    } 
} 

printf("Done \n"); 
//calculate and print the time elapsed 
printf("Time elapsed: %f\n", ((double)clock() - start)/CLOCKS_PER_SEC); 

//Create a windoe and show the resule image 
cvNamedWindow("Result",CV_WINDOW_AUTOSIZE); 
cvShowImage("Result",imgResult); 
cvWaitKey(0); 
getch(); 

//Release the various resources 
cvReleaseImage(&imgResult); 
cvReleaseImage(&imgOriginal); 
cvDestroyWindow("Result"); 
for(int i = 0; i < imgOriginal->height; i++) 
    free(arrOriginal[i]); 

free(arrOriginal); 
free(result); 
cudaFree(&devArrOriginal); 
cudaFree(&result); 

} 

// Print device properties 
void printDevProp(cudaDeviceProp devProp) 
{ 
printf("Major revision number:   %d\n", devProp.major); 
printf("Minor revision number:   %d\n", devProp.minor); 
printf("Name:       %s\n", devProp.name); 
printf("Total global memory:   %u\n", devProp.totalGlobalMem); 
printf("Total shared memory per block: %u\n", devProp.sharedMemPerBlock); 
printf("Total registers per block:  %d\n", devProp.regsPerBlock); 
printf("Warp size:      %d\n", devProp.warpSize); 
printf("Maximum memory pitch:   %u\n", devProp.memPitch); 
printf("Maximum threads per block:  %d\n", devProp.maxThreadsPerBlock); 
for (int i = 0; i < 3; ++i) 
printf("Maximum dimension %d of block: %d\n", i, devProp.maxThreadsDim[i]); 
for (int i = 0; i < 3; ++i) 
printf("Maximum dimension %d of grid: %d\n", i, devProp.maxGridSize[i]); 
printf("Clock rate:     %d\n", devProp.clockRate); 
printf("Total constant memory:   %u\n", devProp.totalConstMem); 
printf("Texture alignment:    %u\n", devProp.textureAlignment); 
printf("Concurrent copy and execution: %s\n", (devProp.deviceOverlap ? "Yes" : "No")); 
printf("Number of multiprocessors:  %d\n", devProp.multiProcessorCount); 
printf("Kernel execution timeout:  %s\n", (devProp.kernelExecTimeoutEnabled ? "Yes" : "No")); 
return; 
} 

/* Utility Macro : CUDA SAFE CALL */ 
void CUDA_SAFE_CALL(cudaError_t call) 
{ 

cudaError_t ret = call; 
switch(ret) 
{ 
    case cudaSuccess: 
     break; 
    default : 
      { 
       printf(" ERROR at line :%i.%d' ' %s\n", 
       __LINE__,ret,cudaGetErrorString(ret)); 
       exit(-1); 
       break; 
      } 
} 
} 

代碼流程如下： 1.在圖像主機中創建一個二維數組，從內核返回另一個數組。 2.使用CudaMallocPitch爲設備中的二維數組分配內存 3.爲將由內核返回的二維數組分配內存。 4.使用cudaMemcpy2d將原始2d陣列從主機複製到設備陣列。 5.啓動內核。 6.使用cudaMemcpy2D將返回的設備陣列複製到主機陣列。

程序在達到第4點時發生崩潰。它是一個未處理的異常，聲明「SauvolaBinarization_CUDA_OpenCV.exe中的0x773415de處未處理的異常：0xC0000005：訪問衝突讀取地址0x01611778」。

我認爲這個問題一定是在分配內存的時候，但是我第一次使用這個函數並且不知道它是如何工作的，好心建議。

Answer 1

首先，您沒有正確調用「cudaMallocPitch」。在「高度」參數應該代表的行數，所以不是：

imgOriginal->height * sizeof(int) 

，你只需要使用：

imgOriginal->height 

這是很好的，因爲每行的字節數已經包含在「音高」屬性。但主要問題在於你爲主機鏡像分配內存的方式。當你寫：

//Create a 2D array for storing the pixels value of each of the pixel of imgOriginal grayscale image 
    int ** arrOriginal = (int **)malloc(imgOriginal->height * sizeof(int *)); 
    for (int i = 0; i < imgOriginal->height; i++) 
{ 
    arrOriginal[i] = (int*)malloc(imgOriginal->width * sizeof(int)); 
} 

你正在有效地創建一個指向數組的指針數組。 CUDA的API調用，您 「重新制作：

CUDA_SAFE_CALL(cudaMemcpy2D((void*)devArrOriginal,dev_pitch,(void*)arrOriginal,host_orig_pitch,imgOriginal->width * sizeof(float),imgOriginal->height,cudaMemcpyHostToDevice)); 

預計，輸入內存緩衝區是連續的。因此，這裏的會發生什麼：從輸入圖像的第一行（總計「imgOriginal->寬*的sizeof（浮動）」字節）將被讀取開始地址：

(void*)arrOriginal 

然而，有效的數據量你從那個地址開始只有「imgOriginal-> height * sizeof（int *）」字節。這兩個字節數很可能不同，這將導致崩潰，因爲您最終會從未知位置讀取數據。

爲了解決這個問題，考慮將 「arrOriginal」 作爲一個連續的塊，如：

int * arrOriginal = (int *)malloc(imgOriginal->height * imgOriginal->width * sizeof(int)); 

而且，在這種情況下，您的間距應該是：

"imgOriginal->width * sizeof(int)"