aboutsummaryrefslogtreecommitdiff
diff options
context:
space:
mode:
authorThomas White <taw@physics.org>2010-02-19 15:57:13 +0100
committerThomas White <taw@physics.org>2010-02-19 15:57:13 +0100
commit632c9738cac050bb6802ab839b07f3a7d3685d11 (patch)
tree1a505749ee5d6d4112f44984eafbdcc56c7338f2
parentcf163cb27898705b4f14344ad0b9a8edc2181d35 (diff)
Synchronise properly
-rw-r--r--src/diffraction-gpu.c14
1 files changed, 9 insertions, 5 deletions
diff --git a/src/diffraction-gpu.c b/src/diffraction-gpu.c
index 8139f961..344da6c6 100644
--- a/src/diffraction-gpu.c
+++ b/src/diffraction-gpu.c
@@ -151,7 +151,7 @@ void get_diffraction_gpu(struct gpu_context *gctx, struct image *image,
double cx, cy, cz;
float kc;
size_t dims[2];
- cl_event event_d;
+ cl_event *event;
int p;
float *tt_ptr;
int x, y;
@@ -223,6 +223,7 @@ void get_diffraction_gpu(struct gpu_context *gctx, struct image *image,
}
/* Iterate over panels */
+ event = malloc(image->det.n_panels * sizeof(cl_event));
for ( p=0; p<image->det.n_panels; p++ ) {
/* In a future version of OpenCL, this could be done
@@ -269,7 +270,7 @@ void get_diffraction_gpu(struct gpu_context *gctx, struct image *image,
}
err = clEnqueueNDRangeKernel(gctx->cq, gctx->kern, 2, NULL,
- dims, NULL, 0, NULL, &event_d);
+ dims, NULL, 0, NULL, &event[p]);
if ( err != CL_SUCCESS ) {
ERROR("Couldn't enqueue diffraction kernel: %s\n",
clError(err));
@@ -278,19 +279,22 @@ void get_diffraction_gpu(struct gpu_context *gctx, struct image *image,
}
diff_ptr = clEnqueueMapBuffer(gctx->cq, gctx->diff, CL_TRUE,
- CL_MAP_READ, 0, gctx->diff_size, 1,
- &event_d, NULL, &err);
+ CL_MAP_READ, 0, gctx->diff_size,
+ image->det.n_panels, event, NULL, &err);
if ( err != CL_SUCCESS ) {
ERROR("Couldn't map diffraction buffer: %s\n", clError(err));
return;
}
tt_ptr = clEnqueueMapBuffer(gctx->cq, gctx->tt, CL_TRUE, CL_MAP_READ, 0,
- gctx->tt_size, 1, &event_d, NULL, &err);
+ gctx->tt_size, image->det.n_panels, event,
+ NULL, &err);
if ( err != CL_SUCCESS ) {
ERROR("Couldn't map tt buffer\n");
return;
}
+ free(event);
+
image->sfacs = calloc(image->width * image->height,
sizeof(double complex));
image->twotheta = calloc(image->width * image->height, sizeof(double));