CODEWITHSUNDEEP
c++cuda
Alexey Korovin
Alexey Korovin

Reputation: 352

Local pointer arrays in CUDA global function

i'm new to CUDA. Can somebody explain me please why this code is invalid? I'm trying to run it on GT240 videocard and memory checker shows me an access violation error on line with buf[0][0].

Here's the code:

__global__ void addKernel(char *c)
{
    int i = threadIdx.x;

    if(i < 1) {
        char* buf[2];
        char some[3] = "ab";
        char another[3] = "cd";

        buf[0] = some;
        buf[1] = another;

        c[i] = buf[0][0];
    }
}

Thanks.

UPDATE: Possible solution to compute something in one function and pass results to other function is to store data in global memory (1.x computation), like that:

__device__ char* buf[2];
__device__ char some[3];
__device__ char another[3];

__global__ void addKernel(int *c, const int *a, const int *b)
{
    int i = threadIdx.x;

    if(i < 1) {
        some[0] = 'a';
        some[1] = 'b';

        another[0] = 'c';
        another[1] = 'd';

        buf[0] = some;
        buf[1] = another;

        buf[0][0] = 'b';
        c[i] = 1;
        }
}

Upvotes: 0

Views: 554

Answers (2)

talonmies
talonmies

Reputation: 72349

This is the classic dangling pointer problem, made worse by the fact that on your GT240, buff will be stored in registers or local memory, and c in global memory. Pointers are not portable in compute 1.x devices.

What you are trying to do is illegal in at least 2 different ways in the programming model/hardware you have, and can never be made to work.

Upvotes: 1

Soren
Soren

Reputation: 14688

Looks OK to me, and works fine when I run it on a Linux Desktop in a debugger.

Suggestions are;

  1. char *c points to an invalid address, or
  2. some weird compiler bug in the compiler you are using is unable to initialized the char some[3] -- try to break it up in simpler assignment statements, or
  3. The possibility of threadIdx.x being negative? i.e. if threadIdx.x is -1 then c[i] is c[-1] which may not be valid...

Upvotes: 1

Related Questions