Мне нужна версия устройства следующего кода хоста:

double (**func)(double x);

double func1(double x)
{
 return x+1.;
}

double func2(double x)
{
 return x+2.;
}

double func3(double x)
{
 return x+3.;
}

void test(void)
{
 double x;

 for(int i=0;i<3;++i){
  x=func[i](2.0);
  printf("%g\n",x);
 }

}

int main(void)
{
 func=(double (**)(double))malloc(10*sizeof(double (*)(double)));

 test();

 return 0;
}

где func1, func2, func3 должны быть функциями __device__, а «test» должен быть (соответствующим образом измененным) ядром __global__.

У меня NVIDIA GeForce GTS 450 (вычислительные возможности 2.1) Заранее спасибо, Мишель

================================================== ======

Рабочий раствор

#define REAL double

typedef REAL (*func)(REAL x);

__host__ __device__ REAL func1(REAL x)
{
    return x+1.0f;
}

__host__ __device__ REAL func2(REAL x)
{
    return x+2.0f;
}

__host__ __device__ REAL func3(REAL x)
{
    return x+3.0f;
}

__device__ func func_list_d[3];
func func_list_h[3];

__global__ void assign_kernel(void)
{
    func_list_d[0]=func1;
    func_list_d[1]=func2;
    func_list_d[2]=func3;
}

void assign(void)
{
    func_list_h[0]=func1;
    func_list_h[1]=func2;
    func_list_h[2]=func3;
}


__global__ void test_kernel(void)
{
    REAL x;
    for(int i=0;i<3;++i){
        x=func_list_d[i](2.0);
        printf("%g\n",x);
  }
}

void test(void)
{
    REAL x;
    printf("=============\n");
    for(int i=0;i<3;++i){
        x=func_list_h[i](2.0);
        printf("%g\n",x);
  }
}

int main(void)
{
    assign_kernel<<<1,1>>>();
    test_kernel<<<1,1>>>();
    cudaThreadSynchronize();

    assign();
    test();

    return 0;
}