CUDA program giving garbage value

I really do not understand why 开发者_运维百科the output for the below code is not a and b.

#include<cutil.h>
#include<iostream>
__global__ void p(unsigned char **a){


unsigned char temp[2];
temp[0] = 'a';
temp[1] = 'b';
a[0] = temp;


}

void main(){

    unsigned char **a ;
    cudaMalloc((void**)&a, sizeof(unsigned char*));
    p<<<1,1>>>(a);
    unsigned char **c;
    unsigned char b[2];
    cudaMemcpy(c, a, sizeof(unsigned char *), cudaMemcpyDeviceToHost);
    cudaMemcpy(b, c[0], 2*sizeof(unsigned char), cudaMemcpyDeviceToHost);
    for( int i=0 ; i < 2; i++){
        printf("%c\n", b[i]);
    }


    getchar();


}

what is wrong with my logic?

Let's leave out CUDA for now. Let's just make a function that writes data to a user-provided array. The user passes the array via a pointer:

void fill_me_up(int * dst)
{
  // We sure hope that `dst` points to a large enough area of memory!

  dst[0] = 28;
  dst[1] = 75;
}

Now, what you're doing with the local variable doesn't make sense, because you want to use the address of a local variable, which becomes invalid after you leave the function scope. The next best thing you could do is memcpy(), or some equivalent C++ algorithm:

void fill_me_up_again(int * dst)
{
  int temp[] = { 28, 75 };
  memcpy((void *)dst, (const void *)temp, sizeof(temp));
}

OK, now on to calling that function: We first must provide the target memory, and then pass a pointer:

int main()
{
  int my_memory[2]; // here's our memory -- automatic local storage

  fill_me_up(my_memory);     // OK, array decays to pointer-to-beginning
  fill_me_up(&my_memory[0]); // A bit more explicit

  int * your_memory = malloc(sizeof(int) * 2); // more memory, this time dynamic
  fill_me_up_again(your_memory);
  /* ... */
  free(your_memory);
}

(In C++ you would probably have uses new int[2] and delete your_memory instead, but by using C malloc() the connection to CUDA hopefully becomes clear.)

When you're moving fill_me_up to the CUDA device, you have to give it a device pointer rather than a host pointer, so you have to set that one up first and afterwards copy the results back out, but that's about the only change.