diff options
author | Thomas White <taw@physics.org> | 2010-02-18 12:10:15 +0100 |
---|---|---|
committer | Thomas White <taw@physics.org> | 2010-02-18 12:10:15 +0100 |
commit | 1cbd38bd817f4b73403e348fa715c6c5356be98a (patch) | |
tree | f5c5b7a3602eedbf7c76ae7efdaa50e7595e01eb /src | |
parent | b04ea39a8bf04b12765b73da3ff9eda7f2143bde (diff) |
CL fixes
Diffstat (limited to 'src')
-rw-r--r-- | src/diffraction-gpu.c | 53 |
1 files changed, 36 insertions, 17 deletions
diff --git a/src/diffraction-gpu.c b/src/diffraction-gpu.c index a6db1980..89655601 100644 --- a/src/diffraction-gpu.c +++ b/src/diffraction-gpu.c @@ -29,6 +29,33 @@ #define BANDWIDTH (1.0 / 100.0) +static const char *clstrerr(cl_int err) +{ + switch ( err ) { + case CL_SUCCESS : return "no error"; + case CL_INVALID_PLATFORM : return "invalid platform"; + case CL_INVALID_KERNEL : return "invalid kernel"; + case CL_INVALID_ARG_INDEX : return "invalid argument index"; + case CL_INVALID_ARG_VALUE : return "invalid argument value"; + case CL_INVALID_MEM_OBJECT : return "invalid memory object"; + case CL_INVALID_SAMPLER : return "invalid sampler"; + case CL_INVALID_ARG_SIZE : return "invalid argument size"; + case CL_INVALID_COMMAND_QUEUE : return "invalid command queue"; + case CL_INVALID_CONTEXT : return "invalid context"; + case CL_INVALID_VALUE : return "invalid value"; + case CL_INVALID_EVENT_WAIT_LIST : return "invalid wait list"; + case CL_MAP_FAILURE : return "map failure"; + case CL_MEM_OBJECT_ALLOCATION_FAILURE : return "object allocation failure"; + case CL_OUT_OF_HOST_MEMORY : return "out of host memory"; + case CL_OUT_OF_RESOURCES : return "out of resources"; + case CL_INVALID_KERNEL_NAME : return "invalid kernel name"; + default : + ERROR("Error code: %i\n", err); + return "unknown error"; + } +} + + static cl_device_id get_first_dev(cl_context ctx) { cl_device_id dev; @@ -86,7 +113,7 @@ static cl_program load_program(const char *filename, cl_context ctx, r = clBuildProgram(prog, 0, NULL, "-Werror", NULL, NULL); if ( r != CL_SUCCESS ) { - ERROR("Couldn't build program\n"); + ERROR("Couldn't build program '%s'\n", filename); show_build_log(prog, dev); *err = r; return 0; @@ -168,14 +195,6 @@ void get_diffraction_gpu(struct image *image, int na, int nb, int nc, ctx = clCreateContextFromType(prop, CL_DEVICE_TYPE_GPU, NULL, NULL, &err); if ( err != CL_SUCCESS ) { ERROR("Couldn't create OpenCL context: %i\n", err); - switch ( err ) { - case CL_INVALID_PLATFORM : - ERROR("Invalid platform\n"); - break; - case CL_OUT_OF_HOST_MEMORY : - ERROR("Out of memory\n"); - break; - } return; } @@ -189,7 +208,7 @@ void get_diffraction_gpu(struct image *image, int na, int nb, int nc, /* Create buffer for the picture */ diff_size = image->width*image->height*sizeof(cl_float)*2; /* complex */ - diff = clCreateBuffer(ctx, CL_MEM_READ_WRITE, diff_size, NULL, &err); + diff = clCreateBuffer(ctx, CL_MEM_WRITE_ONLY, diff_size, NULL, &err); if ( err != CL_SUCCESS ) { ERROR("Couldn't allocate diffraction memory\n"); return; @@ -202,7 +221,7 @@ void get_diffraction_gpu(struct image *image, int na, int nb, int nc, sfac_ptr[2*i+0] = creal(image->molecule->reflections[i]); sfac_ptr[2*i+1] = cimag(image->molecule->reflections[i]); } - sfacs = clCreateBuffer(ctx, CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR, + sfacs = clCreateBuffer(ctx, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sfac_size, sfac_ptr, &err); if ( err != CL_SUCCESS ) { ERROR("Couldn't allocate sfac memory\n"); @@ -210,7 +229,7 @@ void get_diffraction_gpu(struct image *image, int na, int nb, int nc, } tt_size = image->width*image->height*sizeof(cl_float); - tt = clCreateBuffer(ctx, CL_MEM_READ_WRITE, tt_size, NULL, &err); + tt = clCreateBuffer(ctx, CL_MEM_WRITE_ONLY, tt_size, NULL, &err); if ( err != CL_SUCCESS ) { ERROR("Couldn't allocate twotheta memory\n"); return; @@ -249,20 +268,20 @@ void get_diffraction_gpu(struct image *image, int na, int nb, int nc, clSetKernelArg(kern, 10, sizeof(cl_float4), &orientation); err = clEnqueueNDRangeKernel(cq, kern, 2, NULL, dims, NULL, - 0, NULL, NULL); + 1, &event_q, &event_d); if ( err != CL_SUCCESS ) { - ERROR("Couldn't enqueue kernel\n"); + ERROR("Couldn't enqueue diffraction kernel\n"); return; } diff_ptr = clEnqueueMapBuffer(cq, diff, CL_TRUE, CL_MAP_READ, 0, - diff_size, 0, NULL, NULL, &err); + diff_size, 1, &event_d, NULL, &err); if ( err != CL_SUCCESS ) { - ERROR("Couldn't map sfac buffer\n"); + ERROR("Couldn't map diffraction buffer: %s\n", clstrerr(err)); return; } tt_ptr = clEnqueueMapBuffer(cq, tt, CL_TRUE, CL_MAP_READ, 0, - tt_size, 0, NULL, NULL, &err); + tt_size, 1, &event_d, NULL, &err); if ( err != CL_SUCCESS ) { ERROR("Couldn't map tt buffer\n"); return; |