如题，我在主机端和显存端都加了printf函数，根据打印结果，在kernel函数里计算的结果都正确，然后调用完kernel函数之后，用一个cudaMemcpy把结果传回来，再打印就全都是0了，但是我在顶端定义了一个宏那个CUDA_CALL的查错的宏，程序中所有cuda函数都用了这个宏，都没有报错，但是就是结果传回来就成了0，这到底是为什么呢？
这就奇怪了，您最后一次的cudaMemcpy也返回成功么？发一下代码？
之前在centos5.5上最后一次cudaMemcpy会报一个unspecified lanch failure的错，换了系统centos6.3以后，这个错没有了，但是结果还是不正确。我发一下代码帮我看看吧千军兄，我快愁死了
include <stdio.h>
include <math.h>
include <stdlib.h>
include <string.h>
include <time.h>
include <cuda.h>
include <cuda_runtime.h>
include <device_launch_parameters.h>
include <assert.h>
include <malloc.h>
include
include
include <curand_kernel.h>
define CUDA_CALL(x) do { if((x) != cudaSuccess) {
printf(“Error at %s:%d\n”,FILE,LINE);
return EXIT_FAILURE;}} while(0)
define PI 3.1415926535
using namespace std;
int m, n, T, xsup, xinf, N,x0xn;
double sumResult(int n, double *result){
double s = 0.0;
for (int i = 0; i < n; ++i) {
s += exp(result[i]);
printf(“s = %f\n”, s);
}
return s;
}
global void mykernel(int *de_x0xn, int *de_T, int *de_n, int *de_N, int *de_xinf, int *de_xsup, double *de_x, double *de_result, curandState *de_States);
int main(int argc, const char * argv)
{
clock_t start,end;
start = clock();
int dev = 0, deviceCount = 0;
printf("CUDA Device Query (Runtime API) version (CUDART static linking)\n\n");
cudaError_t error_id = cudaGetDeviceCount(&deviceCount);
if (error_id != cudaSuccess)
{
printf("cudaGetDeviceCount returned %d\n-> %s\n", (int)error_id, cudaGetErrorString(error_id));
exit(EXIT_FAILURE);
}
// This function call returns 0 if there are no CUDA capable devices.
if (deviceCount == 0)
{
printf("There are no available device(s) that support CUDA\n");
}
else
{
printf("Detected %d CUDA Capable device(s)\n", deviceCount);
}
for (dev = 0; dev < deviceCount; ++dev)
{
cudaSetDevice(dev);
cudaDeviceProp deviceProp;
cudaGetDeviceProperties(&deviceProp, dev);
printf("\nDevice %d: \"%s\"\n", dev, deviceProp.name);
}
cudaSetDevice(0);
printf("\n");
printf("[example]:1 8 4 5 -5 1000000 0\n");
printf("please input:\nm n T xsup xinf N x0xn\n");
scanf("%d %d %d %d %d %d %d", &m, &n, &T, &xsup, &xinf, &N, &x0xn);
double expectaton = 0.0;
double *de_x;
CUDA_CALL(cudaMalloc((void **)&de_x, sizeof(double) * (n+1)));
CUDA_CALL(cudaMemset(de_x, 0, sizeof(double) * (n+1)));
double *de_result;
CUDA_CALL(cudaMalloc((void **)&de_result, sizeof(double) * (N + 1)));
CUDA_CALL(cudaMemset(de_result, 0, sizeof(double) * (N + 1)));
double *result;
result = (double *)malloc(sizeof(double) * (N + 1));
memset(result, 0, sizeof(double) * (N + 1));
int *de_n, *de_xsup, *de_xinf, *de_N, *de_T, *de_x0xn;
CUDA_CALL(cudaMalloc((void **)&de_n, sizeof(int)));
CUDA_CALL(cudaMalloc((void **)&de_xsup, sizeof(int)));
CUDA_CALL(cudaMalloc((void **)&de_xinf, sizeof(int)));
CUDA_CALL(cudaMalloc((void **)&de_N, sizeof(int)));
CUDA_CALL(cudaMalloc((void **)&de_T, sizeof(int)));
CUDA_CALL(cudaMalloc((void **)&de_x0xn, sizeof(int)));
CUDA_CALL(cudaMemcpy(de_n, &n, sizeof(int), cudaMemcpyHostToDevice));
CUDA_CALL(cudaMemcpy(de_N, &N, sizeof(int), cudaMemcpyHostToDevice));
CUDA_CALL(cudaMemcpy(de_xinf, &xinf, sizeof(int), cudaMemcpyHostToDevice));
CUDA_CALL(cudaMemcpy(de_xsup, &xsup, sizeof(int), cudaMemcpyHostToDevice));
CUDA_CALL(cudaMemcpy(de_T, &T, sizeof(int), cudaMemcpyHostToDevice));
CUDA_CALL(cudaMemcpy(de_x0xn, &x0xn, sizeof(int), cudaMemcpyHostToDevice));
///////////////////////////////////////////////////////////////////////////////
curandState *de_States;
CUDA_CALL(cudaMalloc((void **)&de_States, N * sizeof(curandState)));
mykernel<<<128, 1024>>>(de_x0xn, de_T, de_n, de_N, de_xinf, de_xsup, de_x, de_result, de_States);
///////////////////////////////////////////////////////////////////////////////////////
CUDA_CALL(cudaMemcpy(result, de_result, sizeof(double) * (N + 1), cudaMemcpyDeviceToHost));
for(int i = 0; i <= N; i++)
{
printf("result[%d] = %f\n", i, result[i]);
}
///////////////////////////////////////////////////////////////////////////////////////
expectaton = sumResult(N, result) / (double)N;
cudaThreadSynchronize();
cout << "expectaton: " << expectaton << endl;
double Y = pow(m * n / (PI * 2 * T), n / 2) * expectaton;
cout << "Y: " << Y << endl;
CUDA_CALL(cudaFree(de_x));
CUDA_CALL(cudaFree(de_result));
CUDA_CALL(cudaFree(de_n));
CUDA_CALL(cudaFree(de_xsup));
CUDA_CALL(cudaFree(de_xinf));
CUDA_CALL(cudaFree(de_N));
CUDA_CALL(cudaFree(de_T));
CUDA_CALL(cudaFree(de_x0xn));
end = clock();
printf("Total time is %fs\n", (double)(end - start) / CLOCKS_PER_SEC);
cudaDeviceReset();
return 0;
}
global void mykernel(int *de_x0xn, int *de_T, int *de_n, int *de_N, int *de_xinf, int *de_xsup, double *de_x, double *de_result, curandState *state)
{
int tid = blockDim.x * blockIdx.x + threadIdx.x;
long id = tid;
double r;
curandState localState = state[id + 21];
while(id <= *de_N) {
for(int t = 0; t < *de_n; t++)
{
r = curand_uniform(&localState);
printf("in gpu r[%d] = %f --%d\n", t, r, id);//every element in each array is different.
de_x[t] = (double)(*de_xinf) + (double)(*de_xsup - *de_xinf) * r;
//printf("in gpu de_x[%d] = %f --%d\n", t, de_x[t], id);
}
de_x[0] = de_x[*de_n] = (double)(*de_x0xn);
double sum = 0.0;
for(int t = 0; t < *de_n; ++t){
sum = sum - (pow(de_x[t + 1] - de_x[t], 2) + 0.5 * (*de_T) * pow(de_x[t], 2) / *de_n);
//printf("in gpu 2 for de_x[%d] = %f --%d\n", t, de_x[t], id);
//printf("in gpu sum[%d] = %f --%d \n", t, sum, id);
}
de_result[id] = sum + (double)(*de_n - 1) * log((double)(*de_xsup) - (double)(*de_xinf));
//printf("in gpu de_result[%d] = %f \n", id, de_result[id]);
id += blockDim.x * gridDim.x;
}
}
嗯嗯。最后一次的cudaMemcpy会报告之前的错误，而unspecified launch failure(ULP)一般意味着你的kernel有非法访存。（例如下标越界)。请仔细检查您的代码。
以及，换成Centos 6.3问题消失不一定代表BUG被自动消灭了，可能只是没有触发而已。
请考虑段落1的建议。
奇怪了。看了看您的kernel。似乎无问题（除了一些参数没必要通过设备指针传递外，但这个不影响）。
我将继续看看。
好，谢谢千军兄，我也继续看着呢，嘿嘿
楼主看下这行是否越界了：
curandState localState = state[id + 21];
这个一共有N个。但是你这里却直接用了id + 21, 是否不妥？id首先可能>N的。其次，就算不大于id + 21也有可能大于。不需要if一下吗？
以及，一般的随机数库是需要初始化的（或者自动初始化），请问curand不需要吗？（我对这个库不熟悉），是否在使用之前有一些初始化的函数需要调用？
这个随机数代码我之前在论坛问过，也没有什么答案，我就照着给我的那个curand官方文档，写的例子，我知道在主机端定义的state数组是N个元素，这里用的是state[id + 21],这里应该有越界，但是我就算把21去掉，结果也还是0，而且，如果不加一个数字（不一定是21），curand_uniform这个函数产生的每组随机数都是一样的，加上数字以后就能产生每组数据的每个元素都不同的数组。我也不知道是什么道理。刚刚我已经试了把21去掉，但是结果还是0.
还有，千军兄，就是这个程序，我上周运行的时候能得到正确的结果，但是这周怎么运行都是0，我觉得特别奇怪，同一个程序，怎么过不久运行就会不一样了呢？
我没有更好的建议了，我重复一下之前的吧：
（１）建议参数直接传递即可，不用手工复制到device memory, 然后用device memory的指针这么折腾。
（２）curandState localState = state[id + 21];
建议该成：
if (id >= 总状态数) return;
curandState localState = state[id]; (注意你简单的去掉你的21是不行的, 依然可能越界, 需要if下)
（３）看看curand是否需要特别的初始化函数(这个我不懂）。
没有其他建议了。
以及，欢迎其他朋友参谋下。
对于（1），我已经按照你说的去改了，结果依然是0；
对于（2），我把curandState localState = state[id]这句放进了while循环里面，这样应该能够保证不越界，我看的例子上没有curand_init,我又把这句给加上了，也就是改成有初始化的了，结果还是0……
不知道原因了。建议其他朋友帮忙看看。我没有找到问题。抱歉。
还有，以前在centos5.5系统下编译这个程序时不会报错，报那个unspecified lanch failuer那个是执行的时候报的错。现在系统都是6.3了，没有5.5的系统了，所以现在到底哪里出了错，真是不知道该怎么找了。
千军兄，curand_uniform()函数产生的是平均分布的随机数吗
LZ您好，关于此函数的解释，官方手册内容如下：
device float
curand_uniform (curandState_t *state)
This function returns a sequence of pseudorandom oats uniformly distributed between 0.0
and 1.0. It may return from 0.0 to 1.0, where 1.0 is included and 0.0 is excluded.
Distribution functions may use any number of unsigned integer values from a basic
generator. The number of values consumed is not guaranteed to be fixed.
您可以在Toolkit自带的CURAND_Library.pdf的第16页找到上述内容。
以及您可以在本论坛的资源下载区找到在线版文档。
祝您好运~
嗯，这个我谷歌翻译了一下，是均匀分布的……