cancel
Showing results for 
Search instead for 
Did you mean: 

Archives Discussions

t-man
Adept II

error while building a program due to kernel structure use

So this is my kernel:

const char *KernelSource =

" typedef struct\n" \

"{ \n"\

" int nv;            // number of vertices\n"\

" int ne;            // number of edges \n"\

" int *nbr;          // array of neighbors of all vertices \n" \

" int *firstnbr;     // index in nbr[] of first neighbor of each vtx\n" \

"} graph; \n" \

" \n" \

"__kernel void square(                                                       \n" \

"   __global int* input,                                              \n" \

"   __global int* output,                                             \n" \

"   const unsigned int count,                                           \n" \

"   __global graph* graf) \n" \

"{                                                                      \n" \

"   int i = get_global_id(0);                                           \n" \

"   if(i < count)                                                       \n" \

"       output = input*input;                                \n" \

"}                                                                      \n" \

"\n";

This is the host code:

program = clCreateProgramWithSource(context, 1, (const char **) & KernelSource, NULL, &err);

    if (!program)

    {

        printf("Error: Failed to create compute program!\n");

        return EXIT_FAILURE;

    }

    // Build the program executable

    //

    err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);

    if (err != CL_SUCCESS)

    {

        size_t len;

        printf("%s", KernelSource);

        char buffer[2048];

        printf("Error: Failed to build program executable!\n");

        clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len);

        printf("%s\n", buffer);

        exit(1);

    }

And the error I get is :

Error: Failed to build program executable!

"/tmp/OCLuQgTS9.cl", line 13: error: kernel pointer arguments must point to

          addrSpace global, local, or constant

     __global graph* graf)

                     ^

1 error detected in the compilation of "/tmp/OCLuQgTS9.cl".

Internal error: clc compiler invocation failed.

Any suggestions? Thank you in advance

0 Likes
4 Replies
t-man
Adept II

ok so apparently I had to put in the kernel function __global const struct graph* graf, but if I try to do graf->ne in the kernel function I Get :

Failed to build program executable!

"/tmp/OCLRrsniW.cl", line 13: warning: declaration is not visible outside of

          function

     __global struct graph* graf)

                     ^

"/tmp/OCLRrsniW.cl", line 17: error: pointer to incomplete class type is not

          allowed

         output = graf->ne;           

0 Likes

Hi t-man,

I think it's the struct caused the error. In your struct, there are two pointers, and they point to int, but the address is host memory or device memory, it's not defined. So you can try: 1.Change your struct. 2. the parameter list like this

"__kernel void square(                                                       \n" \

"   __global int* input,                                              \n" \

"   __global int* output,                                             \n" \

"   const unsigned int count,                                           \n" \

"   graph graf) \n" \

Have a try.

0 Likes

Hey Wenju,

Thank you for your reply,

This is how I declared the buffer containing the graf on host memory

graf = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR | CL_MEM_READ_ONLY, sizeof(graph), G, NULL);

err = clEnqueueWriteBuffer( commands, graf, CL_TRUE, 0, sizeof(graph),(void *) &G, 0, NULL, NULL);

where G is G = (graph *) malloc(sizeof(graph));

int nv,ne,nrows=5,ncols=6,k,r,c,vtx;

  nv = nrows*ncols;

  ne = 4*nv - 2*(nrows+ncols); 

  G->nv = nv;

  G->ne = ne;                   

  G->nbr = (int *) malloc(ne*sizeof(int));

  G->firstnbr = (int *) malloc((nv+1)*sizeof(int));

  k = 0;

  for (r = 0; r < nrows; r++) {

    for (c = 0; c < ncols; c++) {

      vtx = r*ncols+c;

      G->firstnbr[vtx] = k;

      if (r!=0)       G->nbr[k++] = vtx-ncols; // all but first grid row

      if (c!=0)       G->nbr[k++] = vtx-1;     // all but first grid col

      if (c!=ncols-1) G->nbr[k++] = vtx+1;     // all but last grid col

      if (r!=nrows-1) G->nbr[k++] = vtx+ncols; // all but last grid row

    }

  }

I did what you told me and now I get error -51 when setting up the kernel arguments. Thank you again for your help!

0 Likes

CL_INVALID_ARG_SIZE. Try sizeof(G). You can refer to AMD SDK sample MonteCarloAsianMultiGPU.

0 Likes