My .cl is somehow like this:
void evaluation(float *L)
{...}
__kernel void memdata(__global float *d_v, __global float LC, const int num)
{ const int tid = get_global_id(0);
__global float *L_loc = LC + num*tid;
evaluation(L_loc);
}
This always brings me a programbuild error. The reason I have to add "__global" in front of L_loc is because LC is in global memory, otherwise L_loc can't point to LC at all. However, "evaluation" is a non-kernel function, and it seems it can't reach global memory. Is that right?
This compiles fine after changing the argument pointer of the function to point to global memory.
void evaluation(__global float *L)
{}
__kernel void memdata(__global float *d_v, __global float *LC, const int num)
{ const int tid = get_global_id(0);
__global float *L_loc = LC + num*tid;
evaluation(L_loc);
}
Yes, I tried that solution, too. It compiles fine, but fails when I try to run it.
Fails as in correctness failure? Can you post the complete code?
BTW The following code works just fine for me -
unsigned int func(__global unsigned int *ptr, unsigned int mul) { return ptr[get_global_id(0)] * mul; } __kernel void templateKernel(__global unsigned int * output, __global unsigned int * input, const unsigned int multiplier) { uint tid = get_global_id(0); output[tid] = func(input,multiplier); }
Originally posted by: n0thing Fails as in correctness failure? Can you post the complete code?
BTW The following code works just fine for me -
Thanks again. Let me try your code and get back to you.
Originally posted by: n0thing This compiles fine after changing the argument pointer of the function to point to global memory.
void evaluation(__global float *L)
Seems like in a non-kernel function, it's not allow to add __global in front a parameter. You can try to add any irrelevant function with that kind of parameter in you own code and see. Thanks,
I can confirm I'm using __global parameters in non-kernel functions successfully. So maybe try out changing your code to something following:
void evaluation(__global float *L, const int startIndex) { //acces the array as L[startIndex] } __kernel void memdata(__global float *d_v, __global float *LC, const int num) { const int tid = get_global_id(0); evaluation(LC, num*tid); }
Thanks, n0thing and karbous. Yes, n0thing's code works fine. I will figure out what's wrong with mine. I couldn't post the whole thing before i get the permission from the project manager.
Originally posted by: karbous I can confirm I'm using __global parameters in non-kernel functions successfully. So maybe try out changing your code to something following:
I think I figured out what my problem is. Please look at the code:
float valuation(const int t, float *L, const int numFactors_, const int numLibors_, const int numTimeSteps_, const InstrumentType instrumentType_)
{float *npv;
float npv_;
npv_ = 0.0f;
npv = &npv_;
if (instrumentType_ == SWAP)
{swap_valuation(t, L, npv);}
else if (instrumentType_ == CAP)
{cap_valuation(t, L, npv);}
else if (instrumentType_ == SWAPTION)
{swaption_valuation(t, L, npv, numFactors_, numLibors_, numTimeSteps_);
}
return *npv;
}
__kernel void fun1(__global float *d_v,...)
{...
evaluation(...);
}
__kernel void fun2(__global float *d_v, __global float *L,...)
{...
evaluation(...);
}
As in above, I have two kernel functions, and they both call the "evalution" function. However, in kernel one, I dont have a global array argument named "L" passed from my host code, and the "L" array required by function "evaluation" will be generated locally. In the second kernel function, I do have an global argument named "L" passed from the host code. If I add "__global" to the parameter "L" in function "evaluation", the first kernel function wont work any more. Further more, all the functions inside "evaluation" would have problems, too, since they dont have a "__global" sign, either.
OpenCL should really allows overriding.
You can create a__global array inside the kernel 1 instead of __local and then there would be no pointer conversion issues.
Or have two differrent functions.
Thanks.