2
$\begingroup$

Following is the CUDA kernel:

odCOde="
__global__ void oneDimensionalIntegral(Real_t* randomInput, mint \
listSize, Real_t* total)
{
    mint index = threadIdx.x + blockIdx.x * blockDim.x;
    mint tid = threadIdx.x;
    extern __shared__ float sdata[];
    if( index<listSize)
    {
        float x;
        x = randomInput[index];
        sdata[tid]=x;
    }
    __syncthreads();
}";

The kernel actually did nothing but just assigning values to sdata[tid](my real aim is to do parallel reduction).

To execute the code, the following is needed:

len = 10;
inputList = RandomReal[{0, \[Pi]/2}, len];
listSize = len;
(*load the function*)
fun = CUDAFunctionLoad[odCOde, 
   "oneDimensionalIntegral", {{_Real, _, 
     "Input"}, _Integer, {_Real, _, "InputOutput"}}, 16];
fun[inputList, listSize, ConstantArray[0., len]]

My problem is, if I delete the expression sdata[tid]=x; in odCOde, the code worked fine. But, with it, mathematica printed:CUDALink experienced an internal error. It seems that the problem is due to this assigning.

$\endgroup$
2
$\begingroup$

It seems that extern is not supported in Mathematica CUDALink. So I just need to delete extern. Also the shared array should be defined with constant size. Therefore, the code is revised as following:

odCOde="
__global__ void oneDimensionalIntegral(Real_t* randomInput, mint \
listSize, Real_t* total)
{
    mint index = threadIdx.x + blockIdx.x * blockDim.x;
    mint tid = threadIdx.x;
    __shared__ float sdata[10];
    if( index<listSize)
    {
        float x;
        x = randomInput[index];
        sdata[tid]=x;
    }
    __syncthreads();
}";
$\endgroup$

Your Answer

By clicking “Post Your Answer”, you agree to our terms of service, privacy policy and cookie policy

Not the answer you're looking for? Browse other questions tagged or ask your own question.