OpenVX 1.3 - laplacian fix (#657)

* initialize output image buffer with zero

* code cleanup

* laplacian fix

* add log entry

* fix log entry

Co-authored-by: LakshmiKumar23 <[email protected]>

Author: hansely123

amd_openvx/openvx/ago/ago_haf_gpu_special_filters.cpp

@@ -1437,7 +1437,6 @@ int HafGpu_ScaleGaussianHalf(AgoNode * node)
 		"  int gx = get_global_id(0);\n"
 		"  int gy = get_global_id(1);\n"
 		"  p0_buf += p0_offset + (gy * p0_stride) + (gx << 2);\n"
-        "  *(__global uint *)p0_buf = 0;\n"
 		"  int gstride = p1_stride;\n"
 		"  __global uchar * gbuf = p1_buf + p1_offset + (((gy - ly) << 1) + 1) * gstride + ((gx - lx) << 3);\n"
 		"  bool valid = ((gx < %d) && (gy < %d)) ? true : false;\n" // (width+3)/4, height

amd_openvx/openvx/ago/ago_util_opencl.cpp

@@ -381,6 +381,14 @@ int agoGpuOclAllocBuffer(AgoData * data)
                 agoAddLogEntry(&context->ref, VX_FAILURE, "ERROR: agoGpuOclCreateBuffer(%p,CL_MEM_READ_WRITE,%d,0,*) => %d\n", context->opencl_context, (int)dataMaster->size + dataMaster->gpu_buffer_offset, err);
                 return -1;
             }
+            else {
+                vx_uint32 zero = 0;
+                err = clEnqueueFillBuffer(context->opencl_cmdq, dataMaster->opencl_buffer, &zero, sizeof(zero), 0, dataMaster->gpu_buffer_offset + dataMaster->size, 0, NULL, NULL);
+                if (err) {
+                    agoAddLogEntry(&context->ref, VX_FAILURE, "ERROR: clEnqueueFillBuffer() => %d\n", err);
+                    return -1;
+                }
+            }
             if (dataMaster->u.img.isUniform) {
                 // make sure that CPU buffer is allocated
                 if (!dataMaster->buffer) {