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
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;
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.
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!
CL_INVALID_ARG_SIZE. Try sizeof(G). You can refer to AMD SDK sample MonteCarloAsianMultiGPU.