AnsweredAssumed Answered

ulong16 not implemented fully in libamdocl64??

Question asked by kd2 on Apr 20, 2013
Latest reply on Apr 24, 2013 by himanshu.gautam

I've got the following code crashing within the clBuildProgram() function. Specifically gdb SIGSEGVs at

   SCRegSpill::CreateSplitReload(SCInst*, int, unsigned short, SCBlock*, bitset*, bitset*) () from /usr/lib64/libamdocl64.so


I'm using AMD-APP-SDK-v2.8-lnx64.tgz with linux kernel 3.4.4, Catalyst-13.1, on a CL_DEVICE_NAME=Tahiti card (if it matters, CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG=1).

 

I seem to be able to get along just fine coding with ulong8s. It's just when I try to jump to using ulong16s that I eventually run into trouble like this.

Should ulong16s just be avoided for the time being?


 

-----------------------------------------------------------------------------------------------------------------------

#include <CL/cl.h>

#include <stdio.h>

#include <string.h>

 

const char *src = "\n\

__kernel void kernel_0(__global ulong16 *in, __global int *out, uint N) \n\

{ \n\

    ulong16 A = in[0];          \n\

    ulong16 B0 = (ulong16)0;    \n\

    ulong16 B1 = (ulong16)0;    \n\

    ulong16 B2 = (ulong16)0;    \n\

    ulong16 B3 = (ulong16)0;    \n\

    for(uint i=0; i < N; i++) { \n\

        B0 += A << 0;           \n\

        B1 += A << 1;           \n\

        B2 += A << 2;           \n\

        B3 += A << 3;           \n\

    }                           \n\

    ulong16 C0 = B0 + B1;       \n\

    ulong16 C1 = B2 + B3;       \n\

    double16 D0 = as_double16(C0);  \n\

    double16 D1 = as_double16(C1);  \n\

    long16 D = isgreater(D0,D1);        \n\

    out[4] = any(D);            \n\

}\n";

 

int main(int argc, char *argv[])

{

    cl_platform_id platform;

    cl_device_id dev;

    cl_uint platforms, devs;

    clGetPlatformIDs(1,&platform,&platforms);

    clGetDeviceIDs(platform,CL_DEVICE_TYPE_ALL,1,&dev,&devs);

    cl_context_properties properties[] = {CL_CONTEXT_PLATFORM,(cl_context_properties)platform,0};

    cl_int err;

    cl_context ctx = clCreateContext(properties,1,&dev,NULL,NULL,&err);

    size_t src_sz(::strlen(src));

    cl_program prog = clCreateProgramWithSource(ctx,1,&src,&src_sz,&err);

    clBuildProgram(prog,0,NULL,"",NULL,NULL);

}

Outcomes