unknown error

运行到一半,程序突然退出,报错cuda error in file"C:/…"in line 252:unknown error.

代码部分如下:
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result1, width, height, dLeftGradient);
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result2, width, height, dRightGradient);
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result3, width, height, dUpGradient);
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result4, width, height, dDownGradient);

float *hresult1;

hresult1 = new float[blockCount];
float *hresult2;
hresult2 = new float[blockCount];
float *hresult3;
hresult3 = new float[blockCount];
float *hresult4;
hresult4 = new float[blockCount];
for(int i =0;i<blockCount;i++)
{
hresult1[i] = 0;
hresult2[i] = 0;
hresult3[i] = 0;
hresult4[i] = 0;
}
CUDA_SAFE_CALL( cudaThreadSynchronize() );

运行到CUDA_SAFE_CALL( cudaThreadSynchronize() )就突然退出报错。有高手知道原因吗?

这个错误一般在两种情况下会出现。
(1)你人为的让你的device code挂掉,比如你嵌入了不正确的ptx汇编,或者干脆你故意,如asm(“trap;”);
(2)你的device code里面使用了非法指针。比如你的__global__里面尝试访问无效位置int *p=0; *p=1234;就会立刻出现unknown error而挂掉。

最快的解决方案:立刻检查你的device code, 仔细观察指针(包括数组名),找出使用不当的地方。

另:你这是在做什么?
for(int i =0;i<blockCount;i++)
{
hresult1 = 0;
hresult2 = 0;
hresult3 = 0;
hresult4 = 0;
}

谢谢啊~~代码calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result1, width, height, dLeftGradient)是计算数组dLeftGradient中的最大值返回给result1,result1也是一个数组,元素个数为block数(类似树状加法优化),每个元素为一个block中数字的最大值,然后这样计算出来四个result数组,再求四个result数组的最大值,所有需要再进行4个result数组数字的比较。
我想把result1到result4传到CPU做这个工作,但是执行到上面的地方就出错。那个循环语句我贴错了,
for(int i =0;i<blockCount;i++)
{
hresult1[i] = 0;
hresult2[i] = 0;
hresult3[i] = 0;
hresult4[i] = 0;
}
应该是这样,在CPU上实现数组初始化。我的device code是这样的:
printf(“hello cuda!”);
CUDA_SAFE_CALL( cudaThreadSynchronize() );
CUDA_SAFE_CALL(cudaMemcpy(hresult1,result1,sizeof(float)* blockCount, cudaMemcpyDeviceToHost));
CUDA_SAFE_CALL(cudaMemcpy(hresult2,result2,sizeof(float)* blockCount, cudaMemcpyDeviceToHost));
CUDA_SAFE_CALL(cudaMemcpy(hresult3,result3,sizeof(float)* blockCount, cudaMemcpyDeviceToHost));
CUDA_SAFE_CALL(cudaMemcpy(hresult4,result4,sizeof(float)* blockCount, cudaMemcpyDeviceToHost));

都是自带函数啊,怎么会出错呢。。(打印hello cuda后就退出)

汗,,贴过来不支持。。。。hresult1【i】这样的

汗,过一会跑又变结果了,原代码没动,还是到那个位置,同步和mencpy的地方然后显示器黑屏两秒然后报错the launch timed out and was treminated.还有显卡驱动程序停止工作并已经恢复。。。。

没有看懂你在说什么。2楼的回复的帖子。你照着检查了吗?
在一个context里面, 如果你一个调用出错,那么你会连续出错,无论下面是什么。也就是说,比如:你的一个launch失败了,那么它之后的所有的调用(比如memcpy, sync, 等等),都会返回这个错误的。
现在很明显你的设备代码里面有问题,如同2楼所说,你必须去检查一下。

你好~我把之前device里面的的 extern shared int 改为 shared float shared[1024],然后没有报错unknown error了,但是在CUDA_SAFE_CALL(cudaMemcpy(hresult1,result1,sizeof(float)* blockCount, cudaMemcpyDeviceToHost))的时候报错the launch timed out and was treminated,然后注释掉这些Memcpy就能继续往下运行。。。不知道这又怎么回事啊。。。

[
每个调用都可能返回上个调用的错误的。你自己检查你的kernel是不是过长时间执行导致。这个错误和返回它的memcpy无关。

谢谢哥们。我也看有说kernel运行时间过长。。。但是这之前我就八个kernel:
calcLeftGradientD<<<blockCount,BLOCK_SIZE>>>(dLeftData, width, height, dLeftGradient);
calcRightGradientD<<<blockCount,BLOCK_SIZE>>>(dLeftData, width, height, dRightGradient);
calcUpGradientD<<<blockCount,BLOCK_SIZE>>>(dLeftData, width, height, dUpGradient);
calcDownGradientD<<<blockCount,BLOCK_SIZE>>>(dLeftData, width, height, dDownGradient);

calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result1, width, height, dLeftGradient);
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result2, width, height, dRightGradient);
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result3, width, height, dUpGradient);
calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result4, width, height, dDownGradient);

其中calcMaxGradientD<<<blockCount,BLOCK_SIZE>>>(result4, width, height, dDownGradient)是我根据深入浅出里面的例程改的比较计算数组的最大元素,具体代码如下:
global static void calcMaxGradientD(float *result,int width,int height,unsigned char *dGradient)
{
shared int shared[1024];
int i;

shared[threadIdx.x] = 0;
for(i = blockIdx.x * blockDim.x + threadIdx.x;i < width*height;i += blockIdx.x * blockDim.x)
{
shared[threadIdx.x] = dGradient[blockIdx.x * blockDim.x + threadIdx.x];
}
__syncthreads();

if(threadIdx.x < 512) 
{
	if(shared[threadIdx.x] < shared[threadIdx.x + 512])
		shared[threadIdx.x] = shared[threadIdx.x + 512];
}

__syncthreads();
if(threadIdx.x < 256)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 256])
shared[threadIdx.x] = shared[threadIdx.x + 256];
}
__syncthreads();
if(threadIdx.x < 128)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 128])
shared[threadIdx.x] = shared[threadIdx.x+128];
}
__syncthreads();
if(threadIdx.x < 64)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 64])
shared[threadIdx.x] = shared[threadIdx.x+64];
}
__syncthreads();
if(threadIdx.x < 32)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 32])
shared[threadIdx.x] = shared[threadIdx.x+32];
}
__syncthreads();
if(threadIdx.x < 16)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 16])
shared[threadIdx.x] = shared[threadIdx.x+16];
}
__syncthreads();
if(threadIdx.x < 8)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 8])
shared[threadIdx.x] = shared[threadIdx.x+8];
}
__syncthreads();
if(threadIdx.x < 4)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 4])
shared[threadIdx.x] = shared[threadIdx.x + 4];
}
__syncthreads();
if(threadIdx.x < 2)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 2])
shared[threadIdx.x] = shared[threadIdx.x+2];
}
__syncthreads();
if(threadIdx.x < 1)
{
if(shared[threadIdx.x] < shared[threadIdx.x + 1])
shared[threadIdx.x] = shared[threadIdx.x + 1];
}
__syncthreads();

if(threadIdx.x == 0)
{
result[blockIdx.x] = shared[0];
}
}
这个kernel我按照例程改写的,运算时间也应该不多啊,数组有480*272个数。请指教。

看写字板太费力了。。sigh。。
貌似数组不是很大,用gpu不一定有cpu的release优化过的快。。

[ 本帖最后由 wscuiqiu 于 2010-11-2 17:42 编辑 ]