On 02/07/18 21:55, Ruiling Song wrote:
> Signed-off-by: Ruiling Song <ruiling.s...@intel.com>
> ---
> libavfilter/opencl.h | 4 ++--
> 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, 38 insertions(+), 139 deletions(-)
This doesn't apply - it seems like it's on top of another patch?
Code looks good to me.
Thanks,
- Mark
> diff --git a/libavfilter/opencl.h b/libavfilter/opencl.h
> index fd76f72..0ed360b 100644
> --- a/libavfilter/opencl.h
> +++ b/libavfilter/opencl.h
> @@ -116,9 +116,9 @@ int ff_opencl_filter_work_size_from_image(AVFilterContext
> *avctx,
> * A helper macro to handle OpenCL error. It will assign errcode to
> * variable err, log error msg, and jump to fail label on error.
> */
> -#define OCL_FAIL_ON_ERR(logctx, cle, errcode, ...) do {\
> +#define CL_FAIL_ON_ERROR(errcode, ...) do {\
> if (cle != CL_SUCCESS) {\
> - av_log(logctx, AV_LOG_ERROR, __VA_ARGS__);\
> + av_log(avctx, AV_LOG_ERROR, __VA_ARGS__);\
> err = errcode;\
> goto fail;\
> }\
> diff --git a/libavfilter/vf_avgblur_opencl.c b/libavfilter/vf_avgblur_opencl.c
> index d1d3eb1..bc6bcab 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 556ce35..e9c8532 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 a002792..dfb2565 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 36c7fbe..241f95e 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 5b1eb59..d76d1b1 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)
>
_______________________________________________
ffmpeg-devel mailing list
ffmpeg-devel@ffmpeg.org
http://ffmpeg.org/mailman/listinfo/ffmpeg-devel
