c++ - __device__ 클래스 멤버 함수가 장치 변수의 값을 변경 한 후 장치에서 호스트로 복사 할 때 cudaMemcpy 오류

내가 작성한 CUDA 코드의 동작에 대해서는 혼란 스럽다. 나는 __device__ 에 대한 테스트를 작성하는 중입니다   DimmedGridGPU 라는 클래스의 함수 . 이 클래스는 int DIM 에서 템플릿입니다  내가 문제가되는 함수는 입력 값에 가장 가까운 지점에서 그리드의 값을 반환하는 것입니다. x . 단위 테스트 목적 으로이 커널 네임 스페이스를 사용하여 각 __device__ 를 호출합니다.  독립된 기능.

이 코드의 바람직한 동작은 3.0 값을 반환하는 것입니다.   do_get_value(x, grid_) 에서  전화 및 d_target[0] 설정  이 값으로 설정 한 다음 단위 테스트 어설 션을 위해 호스트 측으로 다시 전송하십시오. 커널 전체가 제대로 작동하는 것처럼 보이지만 최종 호스트로 다시 전송하면 cudaErrorInvalidValue 가 수신됩니다.  오류가 발생하여 이유를 이해할 수 없습니다.

다음은 클래스의 구조와 기능을 유지하면서 최소한의 코드 예제입니다 :

#include <cuda_runtime.h>
#include <fstream>
#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
   if (code != cudaSuccess) 
   {
     fprintf(stderr,"GPUassert: \"%s\": %s %s %d\n", cudaGetErrorName(code), cudaGetErrorString(code), file, line);
      if (abort) exit(code);
   }
}

template <int DIM>
class DimmedGridGPU{
public:
  size_t grid_size_;//total size of grid
  int b_derivatives_;//if derivatives are going to be used
  int b_interpolate_;//if interpolation should be used on the grid
  double* grid_;//the grid values
  double* grid_deriv_;//derivatives    
  double dx_[DIM];//grid spacing
  double min_[DIM];//grid minimum
  double max_[DIM];//maximum
  int grid_number_[DIM];//number of points on grid
  int b_periodic_[DIM];//if a dimension is periodic
  int* d_b_interpolate_;
  int* d_b_derivatives_;

  DimmedGridGPU(const double* min, 
        const double* max, 
        const double* bin_spacing, 
        const int* b_periodic, 
        int b_derivatives, 
        int b_interpolate) :   b_derivatives_(b_derivatives), b_interpolate_(b_interpolate), grid_(NULL), grid_deriv_(NULL){
    size_t i;
    for(i = 0; i < DIM; i++) {
      min_[i] = min[i];
      max_[i] = max[i];
      b_periodic_[i] = b_periodic[i];
      grid_number_[i] = (int) ceil((max_[i] - min_[i]) / bin_spacing[i]);
      dx_[i] = (max_[i] - min_[i]) / grid_number_[i];
      //add one to grid points if 
      grid_number_[i] = b_periodic_[i] ? grid_number_[i] : grid_number_[i] + 1;
      //increment dx to compensate
      if(!b_periodic_[i])
    max_[i] += dx_[i];
    }
    grid_size_ = 1;
    for(i = 0; i < DIM; i++)
      grid_size_ *= grid_number_[i];
    gpuErrchk(cudaMallocManaged(&grid_, grid_size_ * sizeof(double)));
    if(b_derivatives_) {
      gpuErrchk(cudaMallocManaged(&grid_deriv_, DIM * grid_size_ * sizeof(double)));
      if(!grid_deriv_) {
    printf("Out of memory!! gpugrid.cuh:initialize");   
      }
    }
    gpuErrchk(cudaMalloc((void**)&d_b_interpolate_, sizeof(int)));
    gpuErrchk(cudaMemcpy(d_b_interpolate_, &b_interpolate, sizeof(int), cudaMemcpyHostToDevice));
    gpuErrchk(cudaMalloc((void**)&d_b_derivatives_, sizeof(int)));
    gpuErrchk(cudaMemcpy(d_b_derivatives_, &b_derivatives, sizeof(int), cudaMemcpyHostToDevice));
  }
  ~DimmedGridGPU(){
    gpuErrchk(cudaDeviceSynchronize());
    if(grid_ != NULL){
      gpuErrchk(cudaFree(grid_));
      grid_ = NULL;//need to do this so DimmedGrid's destructor functions properly
    }
    if(grid_deriv_ != NULL){
      gpuErrchk(cudaFree(grid_deriv_));
      grid_deriv_ = NULL;
    }
    gpuErrchk(cudaDeviceReset());
  }
//gets the value of the grid closest to x
  __host__ __device__ double do_get_value( double* x, double* grid_) {
    size_t index[DIM];
    get_index(x, index);
    printf("do_get_value was called on the GPU!, and index[0] is now %d\n", index[0]);
    printf("but multi2one(index) gives us %d\n", multi2one(index));
    double value = grid_[multi2one(index)];
    printf("and value to be returned is %f\n", value);
    return value;
  }
//gets grid's 1D index from an array of coordinates
   __host__ __device__ void get_index(const double* x, size_t result[DIM]) const {
    size_t i;
    double xi;
    printf("get_index was called on the GPU in %i dimension(s)\n", DIM);
    for(i = 0; i < DIM; i++) {
      xi = x[i];
      printf("xi is now %f, min_[i] is %f and dx_[i] is %f\n",xi, min_[i], dx_[i]);
      if(b_periodic_[i]){
    xi -= (max_[i] - min_[i]) * gpu_int_floor((xi - min_[i]) / (max_[i] - min_[i]));
      }
      result[i] = (size_t) floor((xi - min_[i]) / dx_[i]);
    }
  }
//takes a multidimensional index to a 1D index
  __host__ __device__ size_t multi2one(const size_t index[DIM]) const {
    size_t result = index[DIM-1];
    size_t i;    
    for(i = DIM - 1; i > 0; i--) {
      result = result * grid_number_[i-1] + index[i-1];
    }
    return result;
  }
};
__host__ __device__ int gpu_int_floor(double number) {
  return (int) number < 0.0 ? -ceil(fabs(number)) : floor(number);
}

namespace kernels{
  template <int DIM>
  __global__ void get_value_kernel(double* x, double* target_arr, double* grid_, DimmedGridGPU<DIM>  g){
    target_arr[0] = g.do_get_value(x, grid_);
    printf("get_value_kernel has set target[0] to be %f\n", target_arr[0]);//check if the value is set correctly
    return;
  }
}

int main(){
  using namespace kernels;
  double min[] = {0};
  double max[] = {10};
  double bin_spacing[] = {1};
  int periodic[] = {0};
  DimmedGridGPU<1> g (min, max, bin_spacing, periodic, 0, 0);
  for(int i = 0; i < 11; i++){
    g.grid_[i] = i;
    printf("g.grid_[%d] is now %f\n", i, g.grid_[i]);
  }
  gpuErrchk(cudaDeviceSynchronize());
  double x[] = {3.5};
  double* d_x;
  gpuErrchk(cudaMalloc(&d_x, sizeof(double)));
  gpuErrchk(cudaMemcpy(d_x, x, sizeof(double), cudaMemcpyHostToDevice));
  double target[] = {5.0};
  double* d_target;
  gpuErrchk(cudaMalloc((void**)&d_target, sizeof(double)));
  gpuErrchk(cudaMemcpy(d_target, target, sizeof(double), cudaMemcpyHostToDevice));
  gpuErrchk(cudaDeviceSynchronize());
  get_value_kernel<1><<<1,1>>>(d_x, d_target, g.grid_, g);
  gpuErrchk(cudaDeviceSynchronize());
  gpuErrchk(cudaMemcpy(target, d_target, sizeof(double), cudaMemcpyDeviceToHost));
  printf("and after GPU stuff, target[0] is now %f\n", target[0]);
  return(0);
}

그래서이 줄은 왜 마지막 줄인가? ) 오류 ' cudaMemcpy ", 포함 된 인쇄 설명을 보면 장치에서 올바른 값을 사용하고 있으며 CudaErrorInvalidValue 에서 반환 한 값을 명확하게 보여줍니다.  전화가 맞습니까?

이미 do_get_value(x, grid_) 를 사용해 보았습니다. , 어쩌면 과제가 어쨌든 값을 전달하고 변경하는 대신 기호를 생성한다고 생각했지만 cudaMemcpyFromSymbol 와 같은 경우는 아닙니다.  유효한 기호가 아닙니다.

내 코드의 샘플 출력은 다음과 같습니다.

와이즈 비즈 와이즈 비즈

     

d_target

     

     

g.grid_[0] is now 0.000000

     

g.grid_[1] is now 1.000000

     

g.grid_[2] is now 2.000000

     

g.grid_[3] is now 3.000000

     

g.grid_[4] is now 4.000000

     

g.grid_[5] is now 5.000000

     

g.grid_[6] is now 6.000000

     

g.grid_[7] is now 7.000000

     

g.grid_[8] is now 8.000000

     

g.grid_[9] is now 9.000000

     

g.grid_[10] is now 10.000000

     

get_index was called on the GPU in 1 dimension(s)

     

xi is now 3.500000, min_[i] is 0.000000 and dx_[i] is 1.000000

     

do_get_value was called on the GPU!, and index[0] is now 3

     

but multi2one(index) gives us 3

and value to be returned is 3.000000