如何解决CUDA程序导致nvidia驱动程序崩溃
| 我的蒙特卡洛pi计算CUDA程序在我超过500次试用和256个完整块时导致nvidia驱动程序崩溃。它似乎发生在monteCarlo内核函数中,感谢您的帮助。#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>
#include <curand.h>
#include <curand_kernel.h>
#define NUM_THREAD 256
#define NUM_BLOCK 256
///////////////////////////////////////////////////////////////////////////////////////////
///////////////////////////////////////////////////////////////////////////////////////////
// Function to sum an array
__global__ void reduce0(float *g_odata) {
extern __shared__ int sdata[];
// each thread loads one element from global to shared mem
unsigned int tid = threadIdx.x;
unsigned int i = blockIdx.x*blockDim.x + threadIdx.x;
sdata[tid] = g_odata[i];
__syncthreads();
// do reduction in shared mem
for (unsigned int s=1; s < blockDim.x; s *= 2) { // step = s x 2
if (tid % (2*s) == 0) { // only threadIDs divisible by the step participate
sdata[tid] += sdata[tid + s];
}
__syncthreads();
}
// write result for this block to global mem
if (tid == 0) g_odata[blockIdx.x] = sdata[0];
}
///////////////////////////////////////////////////////////////////////////////////////////
///////////////////////////////////////////////////////////////////////////////////////////
__global__ void monteCarlo(float *g_odata,int trials,curandState *states){
// unsigned int tid = threadIdx.x;
unsigned int i = blockIdx.x*blockDim.x + threadIdx.x;
unsigned int incircle,k;
float x,y,z;
incircle = 0;
curand_init(1234,i,&states[i]);
for(k = 0; k < trials; k++){
x = curand_uniform(&states[i]);
y = curand_uniform(&states[i]);
z =(x*x + y*y);
if (z <= 1.0f) incircle++;
}
__syncthreads();
g_odata[i] = incircle;
}
///////////////////////////////////////////////////////////////////////////////////////////
///////////////////////////////////////////////////////////////////////////////////////////
int main() {
float* solution = (float*)calloc(100,sizeof(float));
float *sumDev,*sumHost,total;
const char *error;
int trials;
curandState *devStates;
trials = 500;
total = trials*NUM_THREAD*NUM_BLOCK;
dim3 dimGrid(NUM_BLOCK,1,1); // Grid dimensions
dim3 dimBlock(NUM_THREAD,1); // Block dimensions
size_t size = NUM_BLOCK*NUM_THREAD*sizeof(float); //Array memory size
sumHost = (float*)calloc(NUM_BLOCK*NUM_THREAD,sizeof(float));
cudaMalloc((void **) &sumDev,size); // Allocate array on device
error = cudaGetErrorString(cudaGetLastError());
printf(\"%s\\n\",error);
cudaMalloc((void **) &devStates,(NUM_THREAD*NUM_BLOCK)*sizeof(curandState));
error = cudaGetErrorString(cudaGetLastError());
printf(\"%s\\n\",error);
// Do calculation on device by calling CUDA kernel
monteCarlo <<<dimGrid,dimBlock>>> (sumDev,trials,devStates);
error = cudaGetErrorString(cudaGetLastError());
printf(\"%s\\n\",error);
// call reduction function to sum
reduce0 <<<dimGrid,dimBlock,(NUM_THREAD*sizeof(float))>>> (sumDev);
error = cudaGetErrorString(cudaGetLastError());
printf(\"%s\\n\",error);
dim3 dimGrid1(1,1);
dim3 dimBlock1(256,1);
reduce0 <<<dimGrid1,dimBlock1,error);
// Retrieve result from device and store it in host array
cudaMemcpy(sumHost,sumDev,sizeof(float),cudaMemcpyDeviceToHost);
error = cudaGetErrorString(cudaGetLastError());
printf(\"%s\\n\",error);
*solution = 4*(sumHost[0]/total);
printf(\"%.*f\\n\",1000,*solution);
free (solution);
free(sumHost);
cudaFree(sumDev);
cudaFree(devStates);
//*solution = NULL;
return 0;
}
解决方法
如果较小数量的试验可以正常工作,并且如果您在不带NVIDIA Tesla Compute Cluster(TCC)驱动程序和/或所使用GPU的MS Windows上运行显示器,则可能是操作系统超出了限制。的“看门狗”超时。如果内核占用显示设备(或Windows中没有TCC的Windows上的任何GPU)的时间过长,则OS将终止内核,因此系统不会变得非交互式。
该解决方案将在非显示器连接的GPU上运行,如果您使用的是Windows,请使用TCC驱动程序。否则,您将需要减少内核中的试用次数,并多次运行内核以计算所需的试用次数。
编辑:根据CUDA 4.0 curand文档(第15页,“性能注释”),您可以通过将生成器的状态复制到内核内部的本地存储,然后将状态存储回(如果需要再次使用它,则可以提高性能) )完成时:
curandState state = states[i];
for(k = 0; k < trials; k++){
x = curand_uniform(&state);
y = curand_uniform(&state);
z =(x*x + y*y);
if (z <= 1.0f) incircle++;
}
接下来,它提到设置非常昂贵,并建议您将curand_init移到单独的内核中。这可能有助于降低MC内核的成本,因此您不会遇到看门狗。
我建议阅读文档的这一部分,其中有一些有用的准则。
, 对于拥有不支持TCC驱动程序的geforce GPU的用户,还有另一种基于以下解决方案的解决方案:
http://msdn.microsoft.com/zh-CN/library/windows/hardware/ff569918(v=vs.85).aspx
开始注册,
导航到HKEY_LOCAL_MACHINE \\ System \\ CurrentControlSet \\ Control \\ GraphicsDrivers
创建一个名为TdrLevel的新DWORD键,将其值设置为0,
重新启动PC。
现在,不应将长时间运行的内核终止。该答案基于:
修改注册表以增加GPU超时,Windows 7
我只是认为在此处提供解决方案也可能有用。
版权声明:本文内容由互联网用户自发贡献,该文观点与技术仅代表作者本人。本站仅提供信息存储空间服务,不拥有所有权,不承担相关法律责任。如发现本站有涉嫌侵权/违法违规的内容, 请发送邮件至 dio@foxmail.com 举报,一经查实,本站将立刻删除。