From 894d1438ac8fb642bcf0a57d1d6289611f386238 Mon Sep 17 00:00:00 2001 From: bsegovia Date: Thu, 4 Aug 2011 04:16:20 +0000 Subject: [PATCH] Made test_imm_paramters / test_2d_copy pass on IVB --- kernels/test_imm_parameters_kernels.cl | 13 +++++++------ src/cl_kernel.c | 2 +- 2 files changed, 8 insertions(+), 7 deletions(-) diff --git a/kernels/test_imm_parameters_kernels.cl b/kernels/test_imm_parameters_kernels.cl index 433e8cc..dabcb69 100755 --- a/kernels/test_imm_parameters_kernels.cl +++ b/kernels/test_imm_parameters_kernels.cl @@ -8,11 +8,12 @@ __kernel void test_imm_parameters (__global int *dst, const int a5) { int tid = get_global_id(0); - dst[6*tid+0] = (int) a0 + src[tid]; - dst[6*tid+1] = (int) a1 + src[tid]; - dst[6*tid+2] = (int) a2 + src[tid]; - dst[6*tid+3] = (int) a3 + src[tid]; - dst[6*tid+4] = (int) a4 + src[tid]; - dst[6*tid+5] = (int) a5 + src[tid]; + const int from = src[tid]; + dst[6*tid+0] = (int) a0 + from; + dst[6*tid+1] = (int) a1 + from; + dst[6*tid+2] = (int) a2 + from; + dst[6*tid+3] = (int) a3 + from; + dst[6*tid+4] = (int) a4 + from; + dst[6*tid+5] = (int) a5 + from; } diff --git a/src/cl_kernel.c b/src/cl_kernel.c index aece675..1efb159 100644 --- a/src/cl_kernel.c +++ b/src/cl_kernel.c @@ -349,7 +349,7 @@ cl_kernel_setup_patch_list(cl_kernel k, const char *patch, size_t sz) ASSOC_ITEM (THREAD_PAYLOAD, thread_payload, thread_payload); case PATCH_TOKEN_DATA_PARAMETER_STREAM: - info->curbe.sz = *(uint32_t *) patch; + info->curbe.sz = *(uint32_t *) (patch + sizeof(cl_patch_item_header_t)); info->curbe.offset = 0; break; case PATCH_TOKEN_CONSTANT_MEMORY_KERNEL_ARGUMENT: -- 2.7.4