From 48a1abed133d005aa4d7f49e4fdd5bef9aa263f5 Mon Sep 17 00:00:00 2001 From: Ruiling Song Date: Wed, 4 Jul 2018 02:16:24 +0800 Subject: lavfi/opencl: add macro for opencl error handling. Signed-off-by: Ruiling Song --- libavfilter/opencl.h | 12 ++++++++++ libavfilter/vf_avgblur_opencl.c | 45 +++++++++-------------------------- libavfilter/vf_overlay_opencl.c | 29 +++++------------------ libavfilter/vf_program_opencl.c | 14 ++--------- libavfilter/vf_tonemap_opencl.c | 33 +++++--------------------- libavfilter/vf_unsharp_opencl.c | 52 +++++++++-------------------------------- 6 files changed, 48 insertions(+), 137 deletions(-) (limited to 'libavfilter') diff --git a/libavfilter/opencl.h b/libavfilter/opencl.h index 7441b11614..1b7f117865 100644 --- a/libavfilter/opencl.h +++ b/libavfilter/opencl.h @@ -61,6 +61,18 @@ typedef struct OpenCLFilterContext { goto fail; \ } +/** + * A helper macro to handle OpenCL errors. It will assign errcode to + * variable err, log error msg, and jump to fail label on error. + */ +#define CL_FAIL_ON_ERROR(errcode, ...) do { \ + if (cle != CL_SUCCESS) { \ + av_log(avctx, AV_LOG_ERROR, __VA_ARGS__); \ + err = errcode; \ + goto fail; \ + } \ + } while(0) + /** * Return that all inputs and outputs support only AV_PIX_FMT_OPENCL. */ diff --git a/libavfilter/vf_avgblur_opencl.c b/libavfilter/vf_avgblur_opencl.c index d1d3eb1983..bc6bcab682 100644 --- a/libavfilter/vf_avgblur_opencl.c +++ b/libavfilter/vf_avgblur_opencl.c @@ -64,26 +64,16 @@ static int avgblur_opencl_init(AVFilterContext *avctx) ctx->command_queue = clCreateCommandQueue(ctx->ocf.hwctx->context, ctx->ocf.hwctx->device_id, 0, &cle); - if (!ctx->command_queue) { - av_log(avctx, AV_LOG_ERROR, "Failed to create OpenCL " - "command queue: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create OpenCL " + "command queue %d.\n", cle); ctx->kernel_horiz = clCreateKernel(ctx->ocf.program,"avgblur_horiz", &cle); - if (!ctx->kernel_horiz) { - av_log(avctx, AV_LOG_ERROR, "Failed to create kernel: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create horizontal " + "kernel %d.\n", cle); ctx->kernel_vert = clCreateKernel(ctx->ocf.program,"avgblur_vert", &cle); - if (!ctx->kernel_vert) { - av_log(avctx, AV_LOG_ERROR, "Failed to create kernel: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create vertical " + "kernel %d.\n", cle); ctx->initialised = 1; return 0; @@ -236,12 +226,8 @@ static int avgblur_opencl_filter_frame(AVFilterLink *inlink, AVFrame *input) cle = clEnqueueNDRangeKernel(ctx->command_queue, ctx->kernel_horiz, 2, NULL, global_work, NULL, 0, NULL, NULL); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to enqueue kernel: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to enqueue horizontal " + "kernel: %d.\n", cle); cle = clFinish(ctx->command_queue); err = ff_opencl_filter_work_size_from_image(avctx, global_work, @@ -259,22 +245,13 @@ static int avgblur_opencl_filter_frame(AVFilterLink *inlink, AVFrame *input) cle = clEnqueueNDRangeKernel(ctx->command_queue, ctx->kernel_vert, 2, NULL, global_work, NULL, 0, NULL, NULL); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to enqueue kernel: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to enqueue vertical " + "kernel: %d.\n", cle); } } cle = clFinish(ctx->command_queue); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to finish command queue: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to finish command queue: %d.\n", cle); err = av_frame_copy_props(output, input); if (err < 0) diff --git a/libavfilter/vf_overlay_opencl.c b/libavfilter/vf_overlay_opencl.c index 556ce35070..e9c853203b 100644 --- a/libavfilter/vf_overlay_opencl.c +++ b/libavfilter/vf_overlay_opencl.c @@ -100,19 +100,11 @@ static int overlay_opencl_load(AVFilterContext *avctx, ctx->command_queue = clCreateCommandQueue(ctx->ocf.hwctx->context, ctx->ocf.hwctx->device_id, 0, &cle); - if (!ctx->command_queue) { - av_log(avctx, AV_LOG_ERROR, "Failed to create OpenCL " - "command queue: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create OpenCL " + "command queue %d.\n", cle); ctx->kernel = clCreateKernel(ctx->ocf.program, kernel, &cle); - if (!ctx->kernel) { - av_log(avctx, AV_LOG_ERROR, "Failed to create kernel: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create kernel %d.\n", cle); ctx->initialised = 1; return 0; @@ -209,21 +201,12 @@ static int overlay_opencl_blend(FFFrameSync *fs) cle = clEnqueueNDRangeKernel(ctx->command_queue, ctx->kernel, 2, NULL, global_work, NULL, 0, NULL, NULL); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to enqueue " - "overlay kernel for plane %d: %d.\n", cle, plane); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to enqueue overlay kernel " + "for plane %d: %d.\n", plane, cle); } cle = clFinish(ctx->command_queue); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to finish " - "command queue: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to finish command queue: %d.\n", cle); err = av_frame_copy_props(output, input_main); diff --git a/libavfilter/vf_program_opencl.c b/libavfilter/vf_program_opencl.c index a0027923fb..dfb25652bc 100644 --- a/libavfilter/vf_program_opencl.c +++ b/libavfilter/vf_program_opencl.c @@ -148,21 +148,11 @@ static int program_opencl_run(AVFilterContext *avctx) cle = clEnqueueNDRangeKernel(ctx->command_queue, ctx->kernel, 2, NULL, global_work, NULL, 0, NULL, NULL); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to enqueue kernel: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to enqueue kernel: %d.\n", cle); } cle = clFinish(ctx->command_queue); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to finish command queue: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to finish command queue: %d.\n", cle); if (ctx->nb_inputs > 0) { err = av_frame_copy_props(output, ctx->frames[0]); diff --git a/libavfilter/vf_tonemap_opencl.c b/libavfilter/vf_tonemap_opencl.c index 36c7fbe774..241f95e6c3 100644 --- a/libavfilter/vf_tonemap_opencl.c +++ b/libavfilter/vf_tonemap_opencl.c @@ -262,29 +262,17 @@ static int tonemap_opencl_init(AVFilterContext *avctx) ctx->command_queue = clCreateCommandQueue(ctx->ocf.hwctx->context, ctx->ocf.hwctx->device_id, 0, &cle); - if (!ctx->command_queue) { - av_log(avctx, AV_LOG_ERROR, "Failed to create OpenCL " - "command queue: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create OpenCL " + "command queue %d.\n", cle); ctx->kernel = clCreateKernel(ctx->ocf.program, "tonemap", &cle); - if (!ctx->kernel) { - av_log(avctx, AV_LOG_ERROR, "Failed to create kernel: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create kernel %d.\n", cle); ctx->util_mem = clCreateBuffer(ctx->ocf.hwctx->context, 0, (2 * DETECTION_FRAMES + 7) * sizeof(unsigned), NULL, &cle); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to create util buffer: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create util buffer: %d.\n", cle); ctx->initialised = 1; return 0; @@ -349,11 +337,7 @@ static int launch_kernel(AVFilterContext *avctx, cl_kernel kernel, cle = clEnqueueNDRangeKernel(ctx->command_queue, kernel, 2, NULL, global_work, local_work, 0, NULL, NULL); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to enqueue kernel: %d.\n", - cle); - return AVERROR(EIO); - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to enqueue kernel: %d.\n", cle); return 0; fail: return err; @@ -482,12 +466,7 @@ static int tonemap_opencl_filter_frame(AVFilterLink *inlink, AVFrame *input) } cle = clFinish(ctx->command_queue); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to finish command queue: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to finish command queue: %d.\n", cle); av_frame_free(&input); diff --git a/libavfilter/vf_unsharp_opencl.c b/libavfilter/vf_unsharp_opencl.c index 5b1eb592bf..d76d1b1733 100644 --- a/libavfilter/vf_unsharp_opencl.c +++ b/libavfilter/vf_unsharp_opencl.c @@ -76,12 +76,8 @@ static int unsharp_opencl_init(AVFilterContext *avctx) ctx->command_queue = clCreateCommandQueue(ctx->ocf.hwctx->context, ctx->ocf.hwctx->device_id, 0, &cle); - if (!ctx->command_queue) { - av_log(avctx, AV_LOG_ERROR, "Failed to create OpenCL " - "command queue: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create OpenCL " + "command queue %d.\n", cle); // Use global kernel if mask size will be too big for the local store.. ctx->global = (ctx->luma_size_x > 17.0f || @@ -92,11 +88,7 @@ static int unsharp_opencl_init(AVFilterContext *avctx) ctx->kernel = clCreateKernel(ctx->ocf.program, ctx->global ? "unsharp_global" : "unsharp_local", &cle); - if (!ctx->kernel) { - av_log(avctx, AV_LOG_ERROR, "Failed to create kernel: %d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create kernel %d.\n", cle); ctx->initialised = 1; return 0; @@ -176,12 +168,8 @@ static int unsharp_opencl_make_filter_params(AVFilterContext *avctx) CL_MEM_COPY_HOST_PTR | CL_MEM_HOST_NO_ACCESS, matrix_bytes, matrix, &cle); - if (!buffer) { - av_log(avctx, AV_LOG_ERROR, "Failed to create matrix buffer: " - "%d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create matrix buffer: " + "%d.\n", cle); ctx->plane[p].matrix = buffer; } else { buffer = clCreateBuffer(ctx->ocf.hwctx->context, @@ -190,12 +178,8 @@ static int unsharp_opencl_make_filter_params(AVFilterContext *avctx) CL_MEM_HOST_NO_ACCESS, sizeof(ctx->plane[p].blur_x), ctx->plane[p].blur_x, &cle); - if (!buffer) { - av_log(avctx, AV_LOG_ERROR, "Failed to create x-coef buffer: " - "%d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create x-coef buffer: " + "%d.\n", cle); ctx->plane[p].coef_x = buffer; buffer = clCreateBuffer(ctx->ocf.hwctx->context, @@ -204,12 +188,8 @@ static int unsharp_opencl_make_filter_params(AVFilterContext *avctx) CL_MEM_HOST_NO_ACCESS, sizeof(ctx->plane[p].blur_y), ctx->plane[p].blur_y, &cle); - if (!buffer) { - av_log(avctx, AV_LOG_ERROR, "Failed to create y-coef buffer: " - "%d.\n", cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to create y-coef buffer: " + "%d.\n", cle); ctx->plane[p].coef_y = buffer; } @@ -296,21 +276,11 @@ static int unsharp_opencl_filter_frame(AVFilterLink *inlink, AVFrame *input) cle = clEnqueueNDRangeKernel(ctx->command_queue, ctx->kernel, 2, NULL, global_work, ctx->global ? NULL : local_work, 0, NULL, NULL); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to enqueue kernel: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to enqueue kernel: %d.\n", cle); } cle = clFinish(ctx->command_queue); - if (cle != CL_SUCCESS) { - av_log(avctx, AV_LOG_ERROR, "Failed to finish command queue: %d.\n", - cle); - err = AVERROR(EIO); - goto fail; - } + CL_FAIL_ON_ERROR(AVERROR(EIO), "Failed to finish command queue: %d.\n", cle); err = av_frame_copy_props(output, input); if (err < 0) -- cgit v1.2.3