* Foundation, Inc., 51 Franklin Street, Fifth Floor, Boston, MA 02110-1301 USA
*/
-#include "libavutil/avassert.h"
-#include "libavutil/buffer.h"
-#include "libavutil/common.h"
-#include "libavutil/hwcontext.h"
-#include "libavutil/hwcontext_opencl.h"
#include "libavutil/log.h"
-#include "libavutil/mathematics.h"
#include "libavutil/mem.h"
-#include "libavutil/pixdesc.h"
#include "libavutil/opt.h"
+#include "libavutil/pixdesc.h"
#include "avfilter.h"
#include "framesync.h"
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;
kernel_arg = 0;
mem = (cl_mem)output->data[plane];
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_mem), &mem);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_mem, &mem);
+ kernel_arg++;
mem = (cl_mem)input_main->data[plane];
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_mem), &mem);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_mem, &mem);
+ kernel_arg++;
mem = (cl_mem)input_overlay->data[plane];
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_mem), &mem);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_mem, &mem);
+ kernel_arg++;
if (ctx->alpha_separate) {
mem = (cl_mem)input_overlay->data[ctx->nb_planes];
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_mem), &mem);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_mem, &mem);
+ kernel_arg++;
}
x = ctx->x_position / (plane == 0 ? 1 : ctx->x_subsample);
y = ctx->y_position / (plane == 0 ? 1 : ctx->y_subsample);
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_int), &x);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_int), &y);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_int, &x);
+ kernel_arg++;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_int, &y);
+ kernel_arg++;
if (ctx->alpha_separate) {
cl_int alpha_adj_x = plane == 0 ? 1 : ctx->x_subsample;
cl_int alpha_adj_y = plane == 0 ? 1 : ctx->y_subsample;
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_int), &alpha_adj_x);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
- cle = clSetKernelArg(ctx->kernel, kernel_arg++, sizeof(cl_int), &alpha_adj_y);
- if (cle != CL_SUCCESS)
- goto fail_kernel_arg;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_int, &alpha_adj_x);
+ kernel_arg++;
+ CL_SET_KERNEL_ARG(ctx->kernel, kernel_arg, cl_int, &alpha_adj_y);
+ kernel_arg++;
}
- global_work[0] = output->width;
- global_work[1] = output->height;
+ err = ff_opencl_filter_work_size_from_image(avctx, global_work,
+ output, plane, 0);
+ if (err < 0)
+ goto fail;
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);
return ff_filter_frame(outlink, output);
-fail_kernel_arg:
- av_log(avctx, AV_LOG_ERROR, "Failed to set kernel arg %d: %d.\n",
- kernel_arg, cle);
- err = AVERROR(EIO);
fail:
av_frame_free(&output);
return err;