Exemplo n.º 1
0
Arquivo: sin.c Projeto: arfoll/occcl
int init_sin ()
{
  cl_int error;

  if (sin_init)
    return 1;

  context = get_cl_context();
  device = get_cl_device();

  const char *src=sin_cl;
  size_t srcsize=strlen(sin_cl);
  const char *srcptr[]={src};

  // build CL program
  error = buildcl (srcptr, &srcsize, &prog, "", NUM_GPUS);
  // create kernel
  k_sin = clCreateKernel(prog, "sin_cl", &error);
  // get the shared CQ
  cq = get_command_queue();

  if (!error)
    sin_init = 1;

  return error;
}
Exemplo n.º 2
0
Arquivo: rot13.c Projeto: arfoll/occcl
int init_rot13 ()
{
  cl_int error;

  if (rot13_init)
    return 1;

  context = get_cl_context();
  device = get_cl_device();

  const char *src=rot13_cl;
  size_t srcsize=strlen(rot13_cl);
  const char *srcptr[]={src};

  // build CL program
  error = buildcl (srcptr, &srcsize, &prog, "", NUM_GPUS);
  // create kernel
  k_rot13 = clCreateKernel(prog, "rot13", &error);
  // get the shared CQ
  cq = get_command_queue();

  // we are initialised
  if (!error)
    rot13_init = 1;

  return error;
}
Exemplo n.º 3
0
XCamReturn
CL3aImageProcessor::create_handlers ()
{
    SmartPtr<CLImageHandler> image_handler;
    SmartPtr<CLContext> context = get_cl_context ();

    XCAM_ASSERT (context.ptr ());

    /* black leve as first */
    image_handler = create_cl_blc_image_handler (context);
    _black_level = image_handler;
    XCAM_FAIL_RETURN (
        WARNING,
        image_handler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create blc handler failed");
    add_handler (image_handler);

    /* hdr */
    if (_enable_hdr) {
        image_handler = create_cl_hdr_image_handler (context, CL_HDR_TYPE_RGB);
        _hdr = image_handler;
        XCAM_FAIL_RETURN (
            WARNING,
            _hdr.ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create hdr handler failed");
        add_handler (image_handler);
    }

    /* demosaic */
    image_handler = create_cl_demosaic_image_handler (context);
    _demosaic = image_handler.dynamic_cast_ptr<CLBayer2RGBImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _demosaic.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create demosaic handler failed");
    add_handler (image_handler);

    /* color space conversion */
    if (_out_smaple_type == OutSampleYuv) {
        image_handler = create_cl_csc_image_handler (context, CL_CSC_TYPE_RGBATONV12);
        _csc = image_handler.dynamic_cast_ptr<CLCscImageHandler> ();
        XCAM_FAIL_RETURN (
            WARNING,
            _csc .ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create csc handler failed");
        add_handler (image_handler);
    } else if (_out_smaple_type == OutSampleRGB) {
        _demosaic->set_output_format (_output_fourcc);
    }

    return XCAM_RETURN_NO_ERROR;
}
Exemplo n.º 4
0
int init_mandelbrot ()
{
  int i;
  cl_int error;

  if (mandelbrot_init)
    return 1;

  context = get_cl_context();
  device = get_cl_device();

  FILE *fp;
  fp = fopen(CLKERNELDEFS, "r");
  if (!fp) {
    fprintf(stderr, "Failed to load kernel.\n");
    return(1);
  }
  char *src = (char*) malloc (MAX_SOURCE_SIZE);
  size_t srcsize = fread (src, 1, MAX_SOURCE_SIZE, fp);
  fclose (fp);
  const char *srcptr[]={src};

  // get the number of GPUS/DEVICES available
  numdevices = getNumDevices();

#if CAN_USE_DOUBLE
  // build CL program with a USE_DOUBLE define if we found the correct extension
  if (getCorrectDevice("cl_khr_fp64") == CL_SUCCESS) {
    error = buildcl (srcptr, &srcsize, &prog[0], "-D USE_DOUBLE -cl-fast-relaxed-math -cl-mad-enable", numdevices);
  }
  else {
#endif
    mandelbrot_cl_float = 1;
    error = buildcl (srcptr, &srcsize, &prog[0], "-D USE_FLOAT -cl-fast-relaxed-math -cl-mad-enable", numdevices);
#if CAN_USE_DOUBLE
  }
#endif
  // create kernel
  for (i=0; i<numdevices; i++) {
    k_mandelbrot[i] = clCreateKernel(prog[i], "mandelbrot", &error);
  }
  // get the shared CQ
  cq = get_command_queue();

  if (!error)
    mandelbrot_init = 1;

  return error;
}
Exemplo n.º 5
0
int mandelbrotvis (cl_int *data, cl_fract *job)
{  
  cl_int error;
  int i;

#if MULTI_GPUS 
  int currentdevice = nextDevice();
#else
  int currentdevice = 0;
#endif
  cl_command_queue *cqm = get_command_queue();
  cl_context *cm = get_cl_context();

  // Allocate memory for the kernel to work with
  cl_mem mem1, mem2;
  mem1 = clCreateBuffer(*cm, CL_MEM_WRITE_ONLY, sizeof(cl_int)*(visheight*viswidth*framesperworker), 0, &error);

  if (mandelbrot_cl_float) {
    cl_float jobfloat[framesperworker*JOBS_PER_FRAME];
    for (i=0; i<framesperworker*JOBS_PER_FRAME; i++)
      jobfloat[i] = (cl_float) job[i];

    mem2 = clCreateBuffer(*cm, CL_MEM_COPY_HOST_PTR, sizeof(cl_float)*framesperworker*JOBS_PER_FRAME, jobfloat, &error);
  } else {
    mem2 = clCreateBuffer(*cm, CL_MEM_COPY_HOST_PTR, sizeof(cl_fract)*framesperworker*JOBS_PER_FRAME, job, &error);
  }
  
  // get a handle and map parameters for the kernel
  error = clSetKernelArg(k_mandelbrotvis[currentdevice], 0, sizeof(mem1), &mem1);
  error = clSetKernelArg(k_mandelbrotvis[currentdevice], 1, sizeof(mem2), &mem2);

  size_t worksize[3] = {visheight, viswidth, framesperworker};
  error = clEnqueueNDRangeKernel(*cqm, k_mandelbrotvis[currentdevice], 3, NULL, &worksize[0], 0, 0, 0, 0);
  // Read the result back into data
  error = clEnqueueReadBuffer(*cqm, mem1, CL_TRUE, 0, sizeof(cl_int)*(visheight*viswidth*framesperworker), data, 0, 0, 0);

  // cleanup - don't perform a flush as the queue is now shared between all executions. The
  // blocking clEnqueueReadBuffer should be enough
  clReleaseMemObject(mem1);
  clReleaseMemObject(mem2);

  if (error) {
    fprintf (stderr, "ERROR! : %s\n", errorMessageCL(error));
    exit(10);
  }

  return error;
}
XCamReturn
CLCscImageProcessor::create_handlers ()
{
    SmartPtr<CLImageHandler> image_handler;
    SmartPtr<CLContext> context = get_cl_context ();

    XCAM_ASSERT (context.ptr ());

    /* color space conversion */
    image_handler = create_cl_csc_image_handler (context, CL_CSC_TYPE_YUYVTORGBA);
    _csc = image_handler.dynamic_cast_ptr<CLCscImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _csc .ptr (),
        XCAM_RETURN_ERROR_CL,
        "CLCscImageProcessor create csc handler failed");

    image_handler->set_pool_type (CLImageHandler::CLVideoPoolType);
    add_handler (image_handler);

    return XCAM_RETURN_NO_ERROR;
}
Exemplo n.º 7
0
XCamReturn
CL3aImageProcessor::create_handlers ()
{
    SmartPtr<CLImageHandler> image_handler;
    SmartPtr<CLContext> context = get_cl_context ();

    XCAM_ASSERT (context.ptr ());

#if 1
    /* bayer pipeline */
    image_handler = create_cl_bayer_pipe_image_handler (context);
    _bayer_pipe = image_handler.dynamic_cast_ptr<CLBayerPipeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        image_handler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create bayer pipe handler failed");
    _bayer_pipe->set_stats_callback (_stats_callback);
#if 0
    if (get_profile () >= AdvancedPipelineProfile) {
        _bayer_pipe->set_output_format (V4L2_PIX_FMT_ABGR32);
    }
#endif
    _bayer_pipe->enable_denoise (XCAM_DENOISE_TYPE_BNR & _snr_mode);
    _bayer_pipe->enable_gamma (_enable_gamma);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE * 2);
    add_handler (image_handler);
    if(_capture_stage == BasicbayerStage)
        return XCAM_RETURN_NO_ERROR;

#else
    /* black leve as first */
    image_handler = create_cl_blc_image_handler (context);
    _black_level = image_handler.dynamic_cast_ptr<CLBlcImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        image_handler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create blc handler failed");
    add_handler (image_handler);

    image_handler = create_cl_dpc_image_handler (context);
    _dpc = image_handler.dynamic_cast_ptr<CLDpcImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        image_handler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create dpc handler failed");
    _dpc->set_kernels_enable(_enable_dpc);
    add_handler (image_handler);

    image_handler = create_cl_bnr_image_handler (context);
    _bnr = image_handler.dynamic_cast_ptr<CLBnrImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _bnr.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create bnr handler failed");
    _bnr->set_kernels_enable (XCAM_DENOISE_TYPE_BNR & _snr_mode);
    add_handler (image_handler);

    image_handler = create_cl_3a_stats_image_handler (context);
    _x3a_stats_calculator = image_handler.dynamic_cast_ptr<CL3AStatsCalculator> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _x3a_stats_calculator.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create 3a stats calculator failed");
    _x3a_stats_calculator->set_stats_callback (_stats_callback);
    add_handler (image_handler);

    image_handler = create_cl_wb_image_handler (context);
    _wb = image_handler.dynamic_cast_ptr<CLWbImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _wb.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create whitebalance handler failed");
    add_handler (image_handler);

    /* gamma */
    image_handler = create_cl_gamma_image_handler (context);
    _gamma = image_handler.dynamic_cast_ptr<CLGammaImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _gamma.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create gamma handler failed");
    _gamma->set_kernels_enable (_enable_gamma);
    add_handler (image_handler);

    /* hdr */
    image_handler = create_cl_hdr_image_handler (context, CL_HDR_DISABLE);
    _hdr = image_handler.dynamic_cast_ptr<CLHdrImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _hdr.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create hdr handler failed");
    if(_hdr_mode == CL_HDR_TYPE_RGB)
        _hdr->set_mode (_hdr_mode);
    add_handler (image_handler);

    /* demosaic */
    image_handler = create_cl_demosaic_image_handler (context);
    _demosaic = image_handler.dynamic_cast_ptr<CLBayer2RGBImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _demosaic.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create demosaic handler failed");
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);
#endif

    /* hdr-lab*/
    image_handler = create_cl_hdr_image_handler (context, CL_HDR_DISABLE);
    _hdr = image_handler.dynamic_cast_ptr<CLHdrImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _hdr.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create hdr handler failed");
    if(_hdr_mode == CL_HDR_TYPE_LAB)
        _hdr->set_mode (_hdr_mode);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

    /* bilateral noise reduction */
    image_handler = create_cl_denoise_image_handler (context);
    _binr = image_handler.dynamic_cast_ptr<CLDenoiseImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _binr.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create denoise handler failed");
    _binr->set_kernels_enable (XCAM_DENOISE_TYPE_BILATERAL & _snr_mode);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

#if 0
    image_handler = create_cl_rgb_pipe_image_handler (context);
    _rgb_pipe = image_handler.dynamic_cast_ptr<CLRgbPipeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _rgb_pipe.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create rgb pipe handler failed");
    _rgb_pipe->set_kernels_enable (get_profile () >= AdvancedPipelineProfile);
    add_handler (image_handler);
    /* Temporal Noise Reduction (RGB domain) */
    image_handler = create_cl_tnr_image_handler(context, CL_TNR_TYPE_RGB);
    _tnr_rgb = image_handler.dynamic_cast_ptr<CLTnrImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _tnr_rgb.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create tnr handler failed");
    _tnr_rgb->set_mode (CL_TNR_TYPE_RGB & _tnr_mode);
    add_handler (image_handler);
#else

    /* simple noise reduction */
    image_handler = create_cl_snr_image_handler (context);
    _snr = image_handler.dynamic_cast_ptr<CLSnrImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _snr.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create snr handler failed");
    _snr->set_kernels_enable (XCAM_DENOISE_TYPE_SIMPLE & _snr_mode);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);
#endif

    /* tone mapping*/
    image_handler = create_cl_tonemapping_image_handler (context);
    _tonemapping = image_handler.dynamic_cast_ptr<CLTonemappingImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _tonemapping.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create tonemapping handler failed");
    _tonemapping->set_kernels_enable (_enable_tonemapping);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

#if 1
    image_handler = create_cl_yuv_pipe_image_handler (context);
    _yuv_pipe = image_handler.dynamic_cast_ptr<CLYuvPipeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _yuv_pipe.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create macc handler failed");
    _yuv_pipe->set_tnr_enable (_tnr_mode & CL_TNR_TYPE_RGB, _tnr_mode & CL_TNR_TYPE_YUV);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);
#else
    /* macc */
    image_handler = create_cl_macc_image_handler (context);
    _macc = image_handler.dynamic_cast_ptr<CLMaccImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _macc.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create macc handler failed");
    _macc->set_kernels_enable (_enable_macc);
    add_handler (image_handler);

    /* color space conversion */
    image_handler = create_cl_csc_image_handler (context, CL_CSC_TYPE_RGBATONV12);
    _csc = image_handler.dynamic_cast_ptr<CLCscImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _csc .ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create csc handler failed");
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    add_handler (image_handler);


    /* Temporal Noise Reduction (YUV domain) */
    image_handler = create_cl_tnr_image_handler(context, CL_TNR_TYPE_YUV);
    _tnr_yuv = image_handler.dynamic_cast_ptr<CLTnrImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _tnr_yuv.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create tnr handler failed");
    _tnr_yuv->set_mode (CL_TNR_TYPE_YUV & _tnr_mode);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    add_handler (image_handler);
#endif

    /* ee */
    image_handler = create_cl_ee_image_handler (context);
    _ee = image_handler.dynamic_cast_ptr<CLEeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _ee.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create ee handler failed");
    _ee->set_kernels_enable (XCAM_DENOISE_TYPE_EE & _snr_mode);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);


    /* biyuv */
    image_handler = create_cl_biyuv_image_handler (context);
    _biyuv = image_handler.dynamic_cast_ptr<CLBiyuvImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _biyuv.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create biyuv handler failed");
    _biyuv->set_kernels_enable (XCAM_DENOISE_TYPE_BIYUV & _snr_mode);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

    /* image scaler */
    image_handler = create_cl_image_scaler_handler (context, V4L2_PIX_FMT_NV12);
    _scaler = image_handler.dynamic_cast_ptr<CLImageScaler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _scaler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create scaler handler failed");
    _scaler->set_scaler_factor (XCAM_CL_3A_IMAGE_SCALER_FACTOR);
    _scaler->set_buffer_callback (_stats_callback);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_kernels_enable (false);
    add_handler (image_handler);

    if (_out_smaple_type == OutSampleRGB) {
        image_handler = create_cl_csc_image_handler (context, CL_CSC_TYPE_NV12TORGBA);
        _csc = image_handler.dynamic_cast_ptr<CLCscImageHandler> ();
        XCAM_FAIL_RETURN (
            WARNING,
            _csc .ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create csc handler failed");
        image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
        image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
        add_handler (image_handler);
    }

    return XCAM_RETURN_NO_ERROR;
}
Exemplo n.º 8
0
int init_mandelbrotvis ()
{
  int i;
  cl_int error;

  if (mandelbrotvis_init)
    return 1;

  context = get_cl_context();
  device = get_cl_device();

  FILE *fp;
  fp = fopen(CLVISKERNELDEFS, "r");
  if (!fp) {
    fprintf(stderr, "Failed to load kernel.\n");
    return(1);
  }
  char *src = (char*) malloc (MAX_SOURCE_SIZE);
  size_t srcsize = fread (src, 1, MAX_SOURCE_SIZE, fp);
  fclose (fp);
  const char *srcptr[]={src};

  // get the number of GPUS/DEVICES available
  numdevices = getNumDevices();

  // build CL program with a USE_DOUBLE define if we found the correct extension
  char *precision = "             ";
#if CAN_USE_DOUBLE
  if (getCorrectDevice("cl_khr_fp64") == CL_SUCCESS) {
    precision = "-D USE_DOUBLE";
  }
  else {
#endif
    mandelbrot_cl_float = 1;
    precision = "-D USE_FLOAT";
#if CAN_USE_DOUBLE
  }
#endif

  char options[MAX_BUILD_LINE_LENGTH];
  // following options seem to speed things up a little
  char *compile_opt = "-cl-fast-relaxed-math -cl-mad-enable";
  snprintf(options, MAX_BUILD_LINE_LENGTH,
           "%s -D IMAGEWIDTHVIS=%d -D IMAGEHEIGHTVIS=%d %s", precision, viswidth, visheight, compile_opt);
  error = buildcl (srcptr, &srcsize, &progvis[0], options, numdevices);
//  printf("%s\n\n\n", options);

  // create kernel
  for (i=0; i<numdevices; i++) {
    k_mandelbrotvis[i] = clCreateKernel(progvis[i], "mandelbrot_vis", &error);
  }
  // get the shared CQ
  cq = get_command_queue();

  // initialise the jobs array
  initialiseJobs();

  if (!error)
    mandelbrotvis_init = 1;

  return error;
}
Exemplo n.º 9
0
int mandelbrot (cl_char (*data)[200], cl_fract *job)
{  
  cl_int error;
  int i;

#if ERROR_CHECK
  if (prog == NULL) {
    init_mandelbrot();
  }
#endif

#if MULTI_GPUS 
  // move to new context/cq
  int currentdevice = nextDevice();
  cl_command_queue *cq = get_command_queue();
  cl_context *context = get_cl_context();
#else
  int currentdevice = 0;
#endif

  // Allocate memory for the kernel to work with
  cl_mem mem1, mem2;
  mem1 = clCreateBuffer(*context, CL_MEM_WRITE_ONLY, sizeof(cl_char)*(IMAGEHEIGHT*IMAGEWIDTH*2), 0, &error);

  if (mandelbrot_cl_float) {
    cl_float jobfloat[4];
    for (i=0; i<4; i++)
      jobfloat[i] = (cl_float) job[i];

    mem2 = clCreateBuffer(*context, CL_MEM_COPY_HOST_PTR, sizeof(cl_float)*4, jobfloat, &error);
  } else {
    mem2 = clCreateBuffer(*context, CL_MEM_COPY_HOST_PTR, sizeof(cl_fract)*4, job, &error);
  }
  
  // get a handle and map parameters for the kernel
  error = clSetKernelArg(k_mandelbrot[currentdevice], 0, sizeof(cl_mem), &mem1);
  error = clSetKernelArg(k_mandelbrot[currentdevice], 1, sizeof(cl_mem), &mem2);

  // Perform the operation (width is 100 in this example)
  size_t worksize[3] = {IMAGEHEIGHT, IMAGEWIDTH, 0};
  error = clEnqueueNDRangeKernel(*cq, k_mandelbrot[currentdevice], 2, NULL, &worksize[0], 0, 0, 0, 0);
  // Read the result back into data
  error = clEnqueueReadBuffer(*cq, mem1, CL_TRUE, 0, sizeof(cl_char)*(IMAGEHEIGHT*IMAGEWIDTH*2), data, 0, 0, 0);

  // cleanup - don't perform a flush as the queue is now shared between all executions. The
  // blocking clEnqueueReadBuffer should be enough
  clReleaseMemObject(mem1);
  clReleaseMemObject(mem2);

  if (error) {
    fprintf (stderr, "ERROR! : %s\n", errorMessageCL(error));
    exit(10);
  }

#if C_PRINT
  // this will print a frame coming out of the CL kernel in a dirty but functional manner
  int j;
  int colour = -1;
  for (i=0; i < IMAGEHEIGHT; i++) {
    for (j=0; j < IMAGEWIDTH*2; j++) {
      if (colour != data[i][j]) {
        colour = data[i][j];
        textcolour(colour);
      }
      j++;
      fprintf (stdout, "%c", data[i][j]);
    }
    fprintf(stdout, "\n");
  }
#endif

  return error;
}
Exemplo n.º 10
0
XCamReturn
CL3aImageProcessor::create_handlers ()
{
    SmartPtr<CLImageHandler> image_handler;
    SmartPtr<CLContext> context = get_cl_context ();

    XCAM_ASSERT (context.ptr ());

    /* bayer pipeline */
    image_handler = create_cl_bayer_basic_image_handler (context, _enable_gamma, _3a_stats_bits);
    _bayer_basic_pipe = image_handler.dynamic_cast_ptr<CLBayerBasicImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _bayer_basic_pipe.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create bayer basic pipe handler failed");
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    _bayer_basic_pipe->set_stats_callback (_stats_callback);
    add_handler (image_handler);

    /* tone mapping */
    switch(_wdr_mode) {
    case Gaussian: {
        image_handler = create_cl_tonemapping_image_handler (context);
        _tonemapping = image_handler.dynamic_cast_ptr<CLTonemappingImageHandler> ();
        XCAM_FAIL_RETURN (
            WARNING,
            _tonemapping.ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create tonemapping handler failed");
        _tonemapping->set_kernels_enable (true);
        image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
        add_handler (image_handler);
        break;
    }
    case Haleq: {
        image_handler = create_cl_newtonemapping_image_handler (context);
        _newtonemapping = image_handler.dynamic_cast_ptr<CLNewTonemappingImageHandler> ();
        XCAM_FAIL_RETURN (
            WARNING,
            _newtonemapping.ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create tonemapping handler failed");
        _newtonemapping->set_kernels_enable (true);
        image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
        add_handler (image_handler);
        break;
    }
    default:
        XCAM_LOG_DEBUG ("WDR disabled");
        break;
    }

    /* bayer pipe */
    image_handler = create_cl_bayer_pipe_image_handler (context);
    _bayer_pipe = image_handler.dynamic_cast_ptr<CLBayerPipeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        image_handler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create bayer pipe handler failed");

    _bayer_pipe->enable_denoise (XCAM_DENOISE_TYPE_BNR & _snr_mode);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE * 2);
    //image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    add_handler (image_handler);
    if(_capture_stage == BasicbayerStage)
        return XCAM_RETURN_NO_ERROR;

    image_handler = create_cl_yuv_pipe_image_handler (context);
    _yuv_pipe = image_handler.dynamic_cast_ptr<CLYuvPipeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _yuv_pipe.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create yuv pipe handler failed");
    _yuv_pipe->set_tnr_enable (_tnr_mode & CL_TNR_TYPE_YUV);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE * 2);
    add_handler (image_handler);

#if ENABLE_YEENR_HANDLER
    /* ee */
    image_handler = create_cl_ee_image_handler (context);
    _ee = image_handler.dynamic_cast_ptr<CLEeImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _ee.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create ee handler failed");
    _ee->set_kernels_enable (XCAM_DENOISE_TYPE_EE & _snr_mode);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);
#endif

    /* wavelet denoise */
    switch (_wavelet_basis) {
    case CL_WAVELET_HAT: {
        image_handler = create_cl_wavelet_denoise_image_handler (context, _wavelet_channel);
        _wavelet = image_handler.dynamic_cast_ptr<CLWaveletDenoiseImageHandler> ();
        XCAM_FAIL_RETURN (
            WARNING,
            _wavelet.ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create wavelet denoise handler failed");
        _wavelet->set_kernels_enable (true);
        image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
        image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
        add_handler (image_handler);
        break;
    }
    case CL_WAVELET_HAAR: {
        image_handler = create_cl_newwavelet_denoise_image_handler (context, _wavelet_channel, _wavelet_bayes_shrink);
        _newwavelet = image_handler.dynamic_cast_ptr<CLNewWaveletDenoiseImageHandler> ();
        XCAM_FAIL_RETURN (
            WARNING,
            _newwavelet.ptr (),
            XCAM_RETURN_ERROR_CL,
            "CL3aImageProcessor create new wavelet denoise handler failed");
        _newwavelet->set_kernels_enable (true);
        image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
        image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
        add_handler (image_handler);
        break;
    }
    case CL_WAVELET_DISABLED:
    default :
        XCAM_LOG_DEBUG ("unknown or disable wavelet (%d)", _wavelet_basis);
        break;
    }

    /* image scaler */
    image_handler = create_cl_image_scaler_handler (context, V4L2_PIX_FMT_NV12);
    _scaler = image_handler.dynamic_cast_ptr<CLImageScaler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _scaler.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create scaler handler failed");
    _scaler->set_scaler_factor (_scaler_factor);
    _scaler->set_buffer_callback (_stats_callback);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_kernels_enable (_enable_scaler);
    add_handler (image_handler);

    /* wire frame */
    image_handler = create_cl_wire_frame_image_handler (context);
    _wire_frame = image_handler.dynamic_cast_ptr<CLWireFrameImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _wire_frame.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor create wire frame handler failed");
    _wire_frame->set_kernels_enable (_enable_wireframe);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_3A_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

    XCAM_FAIL_RETURN (
        WARNING,
        post_config (),
        XCAM_RETURN_ERROR_CL,
        "CL3aImageProcessor post_config failed");

    return XCAM_RETURN_NO_ERROR;
}
Exemplo n.º 11
0
int main(int argc, char **argv)
{
    cl_int           err     = 0;
    cl_context       context = 0;
    cl_device_id *   devices = NULL;
    cl_command_queue queue   = 0;
    cl_program       program = 0;
    cl_mem           cl_a = 0, cl_b = 0, cl_res = 0;
    cl_kernel        adder = 0;
    cl_event         event;
    // The iteration variable
    int i;
    // Define our data set
    cl_float a[DATA_SIZE], b[DATA_SIZE], res[DATA_SIZE];

    // Initialize array
    srand(time(0));
    for (i = 0; i < DATA_SIZE; i++) {
        a[i]   = (rand() % 100) / 100.0;
        b[i]   = (rand() % 100) / 100.0;
        res[i] = 0;
    }

    check_release(get_cl_context(&context, &devices, 0) == false,
                  "Fail to create context");

    // Specify the queue to be profile-able
    queue = clCreateCommandQueue(context, devices[0], CL_QUEUE_PROFILING_ENABLE, 0);
    check_release(queue == NULL, "Can't create command queue");

    program = load_program(context, devices[0], "shader.cl");
    check_release(program == NULL, "Fail to build program");

    cl_a =
      clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(cl_float) * DATA_SIZE, NULL, NULL);
    cl_b =
      clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(cl_float) * DATA_SIZE, NULL, NULL);
    cl_res = clCreateBuffer(
      context, CL_MEM_WRITE_ONLY, sizeof(cl_float) * DATA_SIZE, NULL, NULL);
    if (cl_a == 0 || cl_b == 0 || cl_res == 0) {
        printf("Can't create OpenCL buffer\n");
        goto release;
    }

    check_release(clEnqueueWriteBuffer(
                    queue, cl_a, CL_TRUE, 0, sizeof(cl_float) * DATA_SIZE, a, 0, 0, 0),
                  "Write Buffer 1");
    check_release(clEnqueueWriteBuffer(
                    queue, cl_b, CL_TRUE, 0, sizeof(cl_float) * DATA_SIZE, b, 0, 0, 0),
                  "Write Buffer 2");

    adder = clCreateKernel(program, "adder", &err);
    if (err == CL_INVALID_KERNEL_NAME) printf("CL_INVALID_KERNEL_NAME\n");
    check_release(adder == NULL, "Can't load kernel");

    clSetKernelArg(adder, 0, sizeof(cl_mem), &cl_a);
    clSetKernelArg(adder, 1, sizeof(cl_mem), &cl_b);
    clSetKernelArg(adder, 2, sizeof(cl_mem), &cl_res);

    size_t work_size = DATA_SIZE;

    check_release(clEnqueueNDRangeKernel(queue, adder, 1, 0, &work_size, 0, 0, 0, &event),
                  "Can't enqueue kernel");
    check_release(
      clEnqueueReadBuffer(
        queue, cl_res, CL_TRUE, 0, sizeof(cl_float) * DATA_SIZE, res, 0, 0, 0),
      "Can't enqueue read buffer");

    clWaitForEvents(1, &event);
    printf("Execution Time: %.04lf ms\n\n", get_event_exec_time(event));

    // Make sure everything is done before we do anything
    clFinish(queue);
    err = 0;
    for (i = 0; i < DATA_SIZE; i++) {
        if (res[i] != a[i] + b[i]) {
            printf("%f + %f = %f(answer %f)\n", a[i], b[i], res[i], a[i] + b[i]);
            err++;
        }
    }
    if (err == 0)
        printf("Validation passed\n");
    else
        printf("Validation failed\n");
    printf("------\n");

    //--------------------------------
    // Second test
    for (i = 0; i < DATA_SIZE; i++) {
        a[i]   = i;
        b[i]   = i;
        res[i] = 0;
    }

    check_err(clEnqueueWriteBuffer(
                queue, cl_a, CL_TRUE, 0, sizeof(cl_float) * DATA_SIZE, a, 0, 0, 0),
              "Write Buffer 1");
    check_err(clEnqueueWriteBuffer(
                queue, cl_b, CL_TRUE, 0, sizeof(cl_float) * DATA_SIZE, b, 0, 0, 0),
              "Write Buffer 2");

    check_err(clEnqueueNDRangeKernel(queue, adder, 1, 0, &work_size, 0, 0, 0, &event),
              "Can't enqueue kernel");
    check_err(clEnqueueReadBuffer(
                queue, cl_res, CL_TRUE, 0, sizeof(cl_float) * DATA_SIZE, res, 0, 0, 0),
              "Can't enqueue read buffer");

    clWaitForEvents(1, &event);
    printf("Execution Time: %.04lf ms\n\n", get_event_exec_time(event));

    // Make sure everything is done before we do anything
    clFinish(queue);
    err = 0;
    for (i = 0; i < DATA_SIZE; i++) {
        if (res[i] != a[i] + b[i]) {
            printf("%f + %f = %f(answer %f)\n", a[i], b[i], res[i], a[i] + b[i]);
            err++;
        }
    }
    if (err == 0)
        printf("Validation passed\n");
    else
        printf("Validation failed\n");

release:
    clReleaseKernel(adder);
    clReleaseProgram(program);
    clReleaseMemObject(cl_a);
    clReleaseMemObject(cl_b);
    clReleaseMemObject(cl_res);
    clReleaseCommandQueue(queue);
    clReleaseContext(context);

    return 0;
}
Exemplo n.º 12
0
XCamReturn
CLPostImageProcessor::create_handlers ()
{
    SmartPtr<CLImageHandler> image_handler;
    SmartPtr<CLContext> context = get_cl_context ();

    XCAM_ASSERT (context.ptr ());

    /* defog: retinex */
    image_handler = create_cl_retinex_image_handler (context);
    _retinex = image_handler.dynamic_cast_ptr<CLRetinexImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _retinex.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CLPostImageProcessor create retinex handler failed");
    _retinex->set_kernels_enable (_defog_mode == CLPostImageProcessor::DefogRetinex);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_POST_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

    /* defog: dark channel prior */
    image_handler = create_cl_defog_dcp_image_handler (context);
    _defog_dcp = image_handler.dynamic_cast_ptr<CLDefogDcpImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _retinex.ptr (),
        XCAM_RETURN_ERROR_CL,
        "CLPostImageProcessor create retinex handler failed");
    _defog_dcp->set_kernels_enable (_defog_mode == CLPostImageProcessor::DefogDarkChannelPrior);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_POST_IMAGE_MAX_POOL_SIZE);
    add_handler (image_handler);

    /* Temporal Noise Reduction */
    if (_defog_mode != CLPostImageProcessor::DefogDisabled) {
        switch (_tnr_mode) {
        case TnrYuv: {
            image_handler = create_cl_tnr_image_handler (context, CL_TNR_TYPE_YUV);
            _tnr = image_handler.dynamic_cast_ptr<CLTnrImageHandler> ();
            XCAM_FAIL_RETURN (
                WARNING,
                _tnr.ptr (),
                XCAM_RETURN_ERROR_CL,
                "CLPostImageProcessor create tnr handler failed");
            _tnr->set_mode (CL_TNR_TYPE_YUV);
            image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
            image_handler->set_pool_size (XCAM_CL_POST_IMAGE_DEFAULT_POOL_SIZE);
            add_handler (image_handler);
            break;
        }
        case TnrDisable:
            XCAM_LOG_DEBUG ("CLPostImageProcessor disable tnr");
            break;
        default:
            XCAM_LOG_WARNING ("CLPostImageProcessor unknown tnr mode (%d)", _tnr_mode);
            break;
        }
    }

    /* csc (nv12torgba) */
    image_handler = create_cl_csc_image_handler (context, CL_CSC_TYPE_NV12TORGBA);
    _csc = image_handler.dynamic_cast_ptr<CLCscImageHandler> ();
    XCAM_FAIL_RETURN (
        WARNING,
        _csc .ptr (),
        XCAM_RETURN_ERROR_CL,
        "CLPostImageProcessor create csc handler failed");
    _csc->set_kernels_enable (_out_sample_type == OutSampleRGB);
    _csc->set_output_format (_output_fourcc);
    image_handler->set_pool_type (CLImageHandler::DrmBoPoolType);
    image_handler->set_pool_size (XCAM_CL_POST_IMAGE_DEFAULT_POOL_SIZE);
    add_handler (image_handler);

    return XCAM_RETURN_NO_ERROR;
}