ELF@@8 @@@@ȄȔȔ@@@ppRtdPtd,,QtdStd 88 Android!GNUt:rB19AGNU3Y}G^)KrNa2@`gmv{>~"8 B X ` q x qD`D4LD0DFDO D L D |LLL D6 \LTH@q$LDD L <L DpD (D LL lL L  L` DTLqtLW HD D hD. Lg DDpL LYDPDL 8D!DD!L-L D$ LL L 0 0 07q  c  r 8 A@H @ " @@   A@X$@ H 0 8AELPSV[]_dgiHZcFHjQPLAV{> AX~8p#F@~ 6jE^0ӟ}-6AA#d+!( ~CI/M #% ,zFҳ3ج^E##:/ҐjT^qMUw|\/7?Y8#MqJRvQ__cxa_finalize_ZN12ChiNodeUtils8SNPrintFEPcmPKcz_ZN12ChiNodeUtils14LibMapFullNameEPKc_ZN12ChiNodeUtils10LibGetAddrEPvPKc__android_log_print_ZN6ChiLog9LogSystemEPKcS1_S1_S1_iS1_z_ZN8SocUtils10GetSocNameEb__cfi_slowpath_ZN4CamX5Mutex6CreateEPKc_ZN12ChiNodeUtils11FloatToHalfEf__ubsan_handle_add_overflow_minimal_abort_Z10CamxCallocmjiiPKcj_Z8CamxFreePviPKcj__ubsan_handle_mul_overflow_minimal_abort_ZN12ChiNodeUtils5FOpenEPKcS1__ZN12ChiNodeUtils6FWriteEPKvmmP7__sFILE_ZN12ChiNodeUtils6FCloseEP7__sFILE_ZN12ChiNodeUtils5FSeekEP7__sFILEli_ZN12ChiNodeUtils5FTellEP7__sFILE_ZN12ChiNodeUtils5FReadEPvmmmP7__sFILE_ZN4CamX5Mutex4LockEv_ZN4CamX5Mutex6UnlockEv__ubsan_handle_sub_overflow_minimal_abort_ZN4CamX5Mutex7DestroyEv_ZN12ChiNodeUtils8LibUnmapEPv_ZNKSt3__112basic_stringIcNS_11char_traitsIcEENS_9allocatorIcEEE7compareEPKc_ZnwmChiNodeEntry_ZN12ChiNodeUtils11GetMetaDataEmj15ChiMetadataTypeP16ChiNodeInterfacePvjproperty_getatoi_Znam_ZN12ChiNodeUtils12GetAlignmentEiij9ChiFormatPiS1__ZdaPv_ZdlPv_ZN12ChiNodeUtils9VSNPrintFEPcmPKcSt9__va_list__strrchr_chkstrrchr_ZN8SocUtils8GetSocIdEvaccessclose__open_2readatrace_begin_bodyatrace_get_enabled_tagsatrace_end_body_ZN12ChiNodeUtils16GetVendorTagBaseEPKcS1_P16ChiNodeInterface_ZNSt3__112basic_stringIcNS_11char_traitsIcEENS_9allocatorIcEEE6assignEPKc_ZNKSt3__121__basic_string_commonILb1EE20__throw_length_errorEvstrlen__cfi_checkvsnprintfg_enableChxLogsg_enableSystemLog_ZZN8SocUtils8GetSocIdEvE5socId__ubsan_handle_type_mismatch_minimal__ubsan_handle_type_mismatch_minimal_abort__ubsan_handle_alignment_assumption_minimal__ubsan_handle_alignment_assumption_minimal_abort__ubsan_handle_add_overflow_minimalwriteandroid_set_abort_messageabort__ubsan_handle_sub_overflow_minimal__ubsan_handle_mul_overflow_minimal__ubsan_handle_negate_overflow_minimal__ubsan_handle_negate_overflow_minimal_abort__ubsan_handle_divrem_overflow_minimal__ubsan_handle_divrem_overflow_minimal_abort__ubsan_handle_shift_out_of_bounds_minimal__ubsan_handle_shift_out_of_bounds_minimal_abort__ubsan_handle_out_of_bounds_minimal__ubsan_handle_out_of_bounds_minimal_abort__ubsan_handle_builtin_unreachable_minimal__ubsan_handle_missing_return_minimal__ubsan_handle_vla_bound_not_positive_minimal__ubsan_handle_vla_bound_not_positive_minimal_abort__ubsan_handle_float_cast_overflow_minimal__ubsan_handle_float_cast_overflow_minimal_abort__ubsan_handle_load_invalid_value_minimal__ubsan_handle_load_invalid_value_minimal_abort__ubsan_handle_invalid_builtin_minimal__ubsan_handle_invalid_builtin_minimal_abort__ubsan_handle_invalid_objc_cast_minimal__ubsan_handle_invalid_objc_cast_minimal_abort__ubsan_handle_function_type_mismatch_minimal__ubsan_handle_function_type_mismatch_minimal_abort__ubsan_handle_implicit_conversion_minimal__ubsan_handle_implicit_conversion_minimal_abort__ubsan_handle_nonnull_arg_minimal__ubsan_handle_nonnull_arg_minimal_abort__ubsan_handle_nonnull_return_minimal__ubsan_handle_nonnull_return_minimal_abort__ubsan_handle_nullability_arg_minimal__ubsan_handle_nullability_arg_minimal_abort__ubsan_handle_nullability_return_minimal__ubsan_handle_nullability_return_minimal_abort__ubsan_handle_pointer_overflow_minimal__ubsan_handle_pointer_overflow_minimal_abort__ubsan_handle_cfi_check_fail_minimal__ubsan_handle_cfi_check_fail_minimal_abortgetauxval__system_property_getstrncmp__stack_chk_failmemcpymemsetliblog.soLIBLOGlibc.soLIBClibdl.soLIBC_OMR1libc++.solibcamera_metadata.solibcamxcommonutils.solibchilog.solibcom.qti.chinodeutils.solibcutils.solibsync.solibz.solibm.socom.qti.node.gpu.soAPS2  ȔEDDL S(08@]H P X8` h p xMȃЃ؃ F !("0#8$@%H&P'X(`)h*p+x,/01234567Failed to create ds4 weights image with error %d[ERROR ] %s:%d %s() Error querying for number of devices. errNum = %d Error querying for program binary sizes. errNum = %drotateImagep010_to_pd10Soc: %s, kernel: %s, dimensiont: %d, buffer size: %lux%lu, localWGSize: NULL[ERROR ] %s:%d %s() Invalid pSetBufferInfo argument [ERROR ] %s:%d %s() Number of linked cameras %d, exceeds max supported linked cameras %d [ERROR ] %s:%d %s() Invalid physical camera id:%d After m_nodeFlags 0x%x m_bIsSkipProcessing %d[VERBOSE] %s:%d %s() numOutputPorts = %d, numInputPorts = %d Sensor width %d height %d aspect ratio %fnot skipping FD frameProcessRequest Entry nodeinstance %d m_nodeCaps %u Frame Num %lu Seq number %d numInputs: %uGPU node %d frameID: %lu seq %dkamortasm8350kodiakCould not obtain buffer from: %sSupportedFeatureubsan: add-overflowubsan: out-of-bounds ubsan: invalid-builtinubsan: implicit-conversion ubsan: nullability-return clCreateKernelclGetEventProfilingInfo[ERROR ] %s:%d %s() Error querying for program binaries. errNum = %d [ERROR ] %s:%d %s() GPUOpenCL::FlipImage failed to create clBuffer srcY: error: %d GpuNodeDestroyCHINODEQUERYBUFFERINFO is smaller than expected[ERROR ] %s:%d %s() Invalid argument: pInfo = %p [VERBOSE] %s:%d %s() StreamTypePresent [VERBOSE] %s:%d %s() OutputPortId: %u - (%u x %u) isSinkPortwithBuffer %d Process request %llu based on DS, FD vendor tag and skip property[ERROR ] %s:%d %s() Force DownscaleImage before flip GPUIntance:%d ApplyDownscale:inputPort %d->outputPort%d For Request: %lu Xvendor/qcom/proprietary/chi-cdk-legacy/rtwo/oem/qcom/node/gpu//camxchinodegpu.cppError creating the OpenCL Command Queue: %dclBuildProgram failed: error: %d, Log: %sFail to open file %s for device id = %dflipImage[ERROR ] %s:%d %s() clCreateKernel for the p010_to_tp10 Kernel failed: error: %d CreateSeparateYUVImagePlanesReleaseYUVImagePlanes[ERROR ] %s:%d %s() ExecuteBoxFilterSinglePlane failed for CLGetEventProfilingInfo err %d Invalide argument: pQueryVendorTag is NULL[ERROR ] %s:%d %s() Unable to get StreamTypePresent Vendor Tag Id %d Change Opt AspectRatio %f to Sensor AR: %f, Updated opt Height %d[E] - Preview Dimensions: (%u x %u)[VERBOSE] %s:%d %s() Flush is in Progress for the session, aborting [BYPASS] ProcessRequest GPU Rotation as rotation = %d[ERROR ] %s:%d %s() Null GpuNodePostPipelineCreate pNodeSession ubsan: mul-overflow[ERROR ][ERROR ] %s:%d %s() Error querying the maximum work group size: %d CreateProgramFromBinaryclCreateKernel for the BoxFilter2SinglePlane Kernel failed: error: %dMapped fd %d YUVImage=%p YImage=%p UVImage=%p pAddr=%pGPUOpenCL::DownscaleBy4Image failed: result: %d, error: %d[ERROR ] %s:%d %s() Invalide argument: pQueryVendorTag is NULL [ERROR ] %s:%d %s() CHINODESETBUFFERPROPERTIESINFO is smaller than expected pPhysicalDiffMap is NULLInput port id = %d, Format = %d m_nodeCaps: %u max %dx%d min %dx%dOutputPortId: %u - (%u x %u) isSinkPortwithBuffer %dProcessDownscaleRequest[VERBOSE] %s:%d %s() Skipping frame %llu from DS for FD [VERBOSE] %s:%d %s() Process request %llu based on DS, FD vendor tag and skip property ubsan: type-mismatchubsan: nullability-arg ubsan: too many errors clCreateImage2DError querying the maximum work group size: %dCreateProgram[WARN ] %s:%d %s() Failed to write program binary clCreateKernel for the p010_to_pd10 Kernel failed: error: %dExecuteRotateImage failed: error: %dExecuteConvertP010ImageToPD10 failed: error: %dGPUOpenCL::FlipImage failed to create clBuffer dstUV: error: %dGPUOpenCL::ConvertP010ImageToPD10 failed to create clBuffer dst: error: %d[VERBOSE] %s:%d %s() Soc: %s, kernel: %s, dimensiont: %d, buffer size: %lux%lu, localWGSize: NULL Invalid argument: pDestroyInfo->hNodeSession is NULL[ERROR ] %s:%d %s() Invalid argument: session handle = %p PhysicalDifferentDimensionMap[ERROR ] %s:%d %s() Invalid portid %d SetBufferInfo[WARN ] %s:%d %s() Unsupported operation (%d), just pass. Null GpuNodePostPipelineCreate pNodeSession[ERROR ] %s:%d %s() Could not create: %s ubsan: missing-return ubsan: nonnull-return Load program binary %s successful![ERROR ] %s:%d %s() clCreateProgramWithBinary failed: error: %d, status: %d Output format: %d, width: %d, height: %d, outStride: %d, OutScanline: %d, planeY: %p, planeUV: %pGPUOpenCL::FlipImage failed to create clBuffer dstY: error: %dExecuteBoxFilterSinglePlane Scales %fx%fpQueryVendorTag is smaller than expectedcamera index:%d, cameraid:%dBefore m_nodeFlags 0x%xm_hBufferManagerDS16[VERBOSE] %s:%d %s() Request: %lu Bypass: %u Output Port: %u Fmt: %u - (%u x %u) ProcessRequestProcessRequest failed for request = %luSeq number %dUpdateFlipLast Request Id flushed %lupreviewclGetProgramInfoclCreateProgramWithSource failed: error: %d[VERBOSE] %s:%d %s() Fail to open file %s for device id = %d clCreateKernel for the FlipImage Kernel failed: error: %dds4_single_planeboxfilter_single_plane_twoops[ERROR ] %s:%d %s() GPUOpenCL::DownscaleImage failed: result: %d, error: %d BoxFilter Waitforevents %dx%d[ERROR ] %s:%d %s() GPUOpenCL::ConvertP010ImageToPD10 failed to create clBuffer dst: error: %d [ERROR ] %s:%d %s() GPUOpenCL::ConvertP010ImageToUBWCTP10 failed: error: %d CopyUBWCTP10ImageToUBWCTP10GPUOpenCL::Uninitialize Failed to unmap lib: %dInvalid pProcessRequestInfo argumentPhysicalCameraConfigs[BYPASS] ProcessRequest GPU NOFLIP %d , rotation = %dsdm865taroubsan: nonnull-argubsan: nullability-argexynos9810clCreateContext[ERROR ] %s:%d %s() Error creating an OpenCL context: %d [ERROR ] %s:%d %s() Failed to create ds4 sampler with error %d WFailed to write program binary[ERROR ] %s:%d %s() clCreateKernel for the BoxFilter2SinglePlane Kernel failed: error: %d ExecuteRotateImageCopyImage[ERROR ] %s:%d %s() ExecuteDownscaleBy4SinglePlane failed setting arg %u with err %d [ERROR ] %s:%d %s() ExecuteBoxFilterSinglePlane failed setting arg %u with err %d ConvertP010ImageToPD10[ERROR ] %s:%d %s() GPUOpenCL::ConvertP010ImageToPD10 failed: error: %d ConvertP010ImageToUBWCTP10lanaiGetLocalWGSize[ERROR ] %s:%d %s() pQueryVendorTag is smaller than expected Invalid argument: session handle = %pCHINODEFLUSHREQUESTINFO is smaller than expected: pInfo->size = %d, sizeof(CHINODEFLUSHREQUESTINFO) = %d[VERBOSE] %s:%d %s() pInfo %p pInfo->hChiSession %p [VERBOSE] %s:%d %s() Number of linked cameras %d index:%d, camId:%d, isDiff:%dCHINODEPROCREQUESTINFO is smaller than expectedFrameDimensionubsan: vla-bound-not-positiveclReleaseCommandQueue[VERBOSE] %s:%d %s() Deleting fd %d YUVImage=%p YImage=%p UVImage=%p pAddr=%p [ERROR ] %s:%d %s() Unknown rotation angle %d [ERROR ] %s:%d %s() GPUOpenCL::Uninitialize Failed to unmap lib: %d pCameraConfig: NULL[VERBOSE] %s:%d %s() planeId %d input alignment %d %d Instanceid:%d Invalid input port map for this output port:%d(%d)One of the o/p ports are invalid so not possible to apply SimoUnable to read endorTag IsPreviewStreamPrasent request %llu[VERBOSE] %s:%d %s() [NoBYPASS] ProcessRequest GPU Rotation rotation = %d Unsupported downscale format from Input (Port: %u, Format: %u) to Output (Port: %u, Format: %u)ubsan: sub-overflow clGetProgramBuildInfoCHIGPUEnableBufferMappingforPort[ERROR ] %s:%d %s() ExecuteCopyImage failed: error: %d ExecuteCopyUBWCTP10ImageToUBWCTP10[ERROR ] %s:%d %s() ExecuteFlipImage failed: error: %d DownscaleImageUnable to downscale image with current formatglobal WGS %zu x %zu SCales %f and %f[ERROR ] %s:%d %s() GPUOpenCL::CopyUBWCTP10ImageToUBWCTP10 failed: error: %d [ERROR ] %s:%d %s() Invalid argument: pTagTypeInfo is NULL [VERBOSE] %s:%d %s() [%p] - InstanceId: %u ReqId: %u pConfig: %p - numConfigurations: %u InitializeDownscaleBypassMap[VERBOSE] %s:%d %s() Physical Camera: %u -> Node: %u:%u @ InputPort: %u | Type: %u OpenCL Initialization failed !Unable to get FD based process skip Tag Id[VERBOSE] %s:%d %s() cameraIdx:%d, bypassCameraId:%d [ERROR ] %s:%d %s() Instanceid:%d Invalid input port map for this output port:%d(%d) FD Frame skip so not possible to apply SIMOcamano[ERROR ] %s:%d %s() Invalid argument: pSetupRequestInfo: %p, pSetupRequestOutput: %p Invalid argument: pSetupRequestInfo: %p, pSetupRequestOutput: %pubsan: add-overflow libOpenCLstatic const int Y_COMPONENT = 0; static const int U_COMPONENT = 1; static const int V_COMPONENT = 2; __kernel void tp10_to_tp10(__read_only image2d_t src_tp10, __write_only image2d_t dest_tp10_y, __write_only image2d_t dest_tp10_uv, sampler_t sampler) { const int wid_x = get_global_id(0); const int wid_y = get_global_id(1); const int2 read_coord = (int2)(6 * wid_x, wid_y); const int2 y_write_coord = (int2)(6 * wid_x, wid_y); const int2 uv_write_coord = (int2)(3 * wid_x, wid_y / 2); const float4 pixels_in[] = { read_imagef(src_tp10, sampler, read_coord), read_imagef(src_tp10, sampler, read_coord + (int2)(1, 0)), read_imagef(src_tp10, sampler, read_coord + (int2)(2, 0)), read_imagef(src_tp10, sampler, read_coord + (int2)(3, 0)), read_imagef(src_tp10, sampler, read_coord + (int2)(4, 0)), read_imagef(src_tp10, sampler, read_coord + (int2)(5, 0)), }; float y_pixels_out[2][3] = { { pixels_in[0].s0, pixels_in[1].s0, pixels_in[2].s0 }, { pixels_in[3].s0, pixels_in[4].s0, pixels_in[5].s0 }, }; float2 uv_pixels_out[3] = { { pixels_in[0].s1, pixels_in[0].s2 }, { pixels_in[2].s1, pixels_in[2].s2 }, { pixels_in[4].s1, pixels_in[4].s2 }, }; qcom_write_imagefv_3x1_n10t00(dest_tp10_y, y_write_coord, y_pixels_out[0]); qcom_write_imagefv_3x1_n10t00(dest_tp10_y, y_write_coord + (int2)(3, 0), y_pixels_out[1]); if (wid_y % 2 == 0) qcom_write_imagefv_3x1_n10t01(dest_tp10_uv, uv_write_coord, uv_pixels_out); } __kernel void p010_to_tp10(__read_only image2d_t src_p010, __write_only image2d_t dest_tp10_y, __write_only image2d_t dest_tp10_uv, sampler_t sampler) { const int wid_x = get_global_id(0); const int wid_y = get_global_id(1); const int2 read_coord = (int2)(6 * wid_x, wid_y); const int2 y_write_coord = (int2)(6 * wid_x, wid_y); const int2 uv_write_coord = (int2)(3 * wid_x, wid_y / 2); const float4 pixels_in[] = { read_imagef(src_p010, sampler, read_coord ), read_imagef(src_p010, sampler, read_coord + (int2)(1, 0)), read_imagef(src_p010, sampler, read_coord + (int2)(2, 0)), read_imagef(src_p010, sampler, read_coord + (int2)(3, 0)), read_imagef(src_p010, sampler, read_coord + (int2)(4, 0)), read_imagef(src_p010, sampler, read_coord + (int2)(5, 0)), }; float y_pixels_out[2][3] = { {pixels_in[0].s0, pixels_in[1].s0, pixels_in[2].s0}, {pixels_in[3].s0, pixels_in[4].s0, pixels_in[5].s0}, }; float2 uv_pixels_out[3] = { {pixels_in[0].s1, pixels_in[0].s2}, {pixels_in[2].s1, pixels_in[2].s2}, {pixels_in[4].s1, pixels_in[4].s2}, }; qcom_write_imagefv_3x1_n10t00(dest_tp10_y, y_write_coord, y_pixels_out[0]); qcom_write_imagefv_3x1_n10t00(dest_tp10_y, y_write_coord + (int2)(3, 0), y_pixels_out[1]); if (wid_y % 2 == 0) qcom_write_imagefv_3x1_n10t01(dest_tp10_uv, uv_write_coord, uv_pixels_out); } __kernel void tp10_to_p010(__read_only image2d_t src_image, __write_only image2d_t dest_p010_y, __write_only image2d_t dest_p010_uv, sampler_t sampler) { const int wid_x = get_global_id(0); const int wid_y = get_global_id(1); const float2 read_coord = (float2)(4 * wid_x, 4 * wid_y) + 0.5; const int2 y_write_coord = (int2)(4 * wid_x, 4 * wid_y); const int2 uv_write_coord = (int2)(2 * wid_x, 2 * wid_y); const float4 y_pixels_in[] = { qcom_read_imagef_2x2(src_image, sampler, read_coord, Y_COMPONENT), qcom_read_imagef_2x2(src_image, sampler, read_coord + (float2)(2., 0.), Y_COMPONENT), qcom_read_imagef_2x2(src_image, sampler, read_coord + (float2)(0., 2.), Y_COMPONENT), qcom_read_imagef_2x2(src_image, sampler, read_coord + (float2)(2., 2.), Y_COMPONENT), }; float y_pixels_out[4][4] = { {y_pixels_in[0].s3, y_pixels_in[0].s2, y_pixels_in[1].s3, y_pixels_in[1].s2}, {y_pixels_in[0].s0, y_pixels_in[0].s1, y_pixels_in[1].s0, y_pixels_in[1].s1}, {y_pixels_in[2].s3, y_pixels_in[2].s2, y_pixels_in[3].s3, y_pixels_in[3].s2}, {y_pixels_in[2].s0, y_pixels_in[2].s1, y_pixels_in[3].s0, y_pixels_in[3].s1}, }; qcom_write_imagefv_4x1_n10p00(dest_p010_y, y_write_coord, y_pixels_out[0]); qcom_write_imagefv_4x1_n10p00(dest_p010_y, y_write_coord + (int2)(0, 1), y_pixels_out[1]); qcom_write_imagefv_4x1_n10p00(dest_p010_y, y_write_coord + (int2)(0, 2), y_pixels_out[2]); qcom_write_imagefv_4x1_n10p00(dest_p010_y, y_write_coord + (int2)(0, 3), y_pixels_out[3]); const float4 u_pixels_in = qcom_read_imagef_2x2(src_image, sampler, read_coord, U_COMPONENT); const float4 v_pixels_in = qcom_read_imagef_2x2(src_image, sampler, read_coord, V_COMPONENT); float2 uv_pixels_out[2][2] = { {{u_pixels_in.s3, v_pixels_in.s3}, {u_pixels_in.s2, v_pixels_in.s2}}, {{u_pixels_in.s0, v_pixels_in.s0}, {u_pixels_in.s1, v_pixels_in.s1}}, }; qcom_write_imagefv_2x1_n10p01(dest_p010_uv, uv_write_coord, uv_pixels_out[0]); qcom_write_imagefv_2x1_n10p01(dest_p010_uv, uv_write_coord + (int2)(0, 1), uv_pixels_out[1]); } __kernel void copyImage(__write_only image2d_t dst, __read_only image2d_t src, unsigned int isSwapNeeded) { int2 coord; float4 color; coord.x = get_global_id(0); coord.y = get_global_id(1); const sampler_t sampler1 = CLK_NORMALIZED_COORDS_TRUE | CLK_ADDRESS_CLAMP | CLK_FILTER_NEAREST; color = read_imagef(src, sampler1, coord); if (0 == isSwapNeeded) { write_imagef(dst, coord, color); } else { write_imagef(dst, coord, color.yxzw); } } __kernel void rotateImage(__write_only image2d_t dst, __read_only image2d_t src, int dstWidth, int dstHeight, int srcWidth, int srcHeight, float sinTheta, float cosTheta) { const int ix = get_global_id(0); const int iy = get_global_id(1); const int halfWidth = dstWidth / 2; const int halfHeight = dstHeight / 2; int2 dstCoord = (int2)(ix, iy); int2 srcCoord; int translatedX = ix - halfWidth; int translatedY = iy - halfHeight; float rotatedX = (((float) translatedX) * cosTheta + ((float) translatedY) * sinTheta); float rotatedY = (((float) translatedX) * -sinTheta + ((float) translatedY) * cosTheta); srcCoord.x = rotatedX + (srcWidth / 2); srcCoord.y = rotatedY + (srcHeight / 2); const sampler_t sampler1 = CLK_NORMALIZED_COORDS_TRUE | CLK_ADDRESS_REPEAT | CLK_FILTER_NEAREST; float4 color; color = read_imagef(src, sampler1, srcCoord); write_imagef(dst, dstCoord, color); } __kernel void flipImage(__global unsigned char* dst, __global unsigned char* src, int width, int height, unsigned int inStride, unsigned int outStride, unsigned int inUVOffset, unsigned int outUVOffset, unsigned int swapneeded, int direction) { int2 srcCoord; int2 dstCoord; float4 color; dstCoord.x = get_global_id(0); dstCoord.y = get_global_id(1); const sampler_t sampler1 = CLK_NORMALIZED_COORDS_TRUE | CLK_ADDRESS_CLAMP | CLK_FILTER_NEAREST; if (0 == direction) { srcCoord.x = dstCoord.x; srcCoord.y = dstCoord.y; } else if (1 == direction) { srcCoord.x = width - dstCoord.x - 1; srcCoord.y = dstCoord.y; } else if (2 == direction) { srcCoord.x = dstCoord.x; srcCoord.y = height - dstCoord.y - 1; } else { srcCoord.x = width - dstCoord.x - 1; srcCoord.y = height - dstCoord.y - 1; } if (0 == inUVOffset) { dst[(dstCoord.y * outStride) + dstCoord.x] = src[(srcCoord.y * inStride) + srcCoord.x]; } else { if (1 == swapneeded) { dst[(dstCoord.y * outStride) + (dstCoord.x * 2) + outUVOffset] = src[(srcCoord.y * inStride) + (srcCoord.x * 2) + 1 + inUVOffset]; dst[(dstCoord.y * outStride) + (dstCoord.x * 2) + 1 + outUVOffset] = src[(srcCoord.y * inStride) + (srcCoord.x * 2) + inUVOffset]; } else { dst[(dstCoord.y * outStride) + (dstCoord.x * 2) + outUVOffset] = src[(srcCoord.y * inStride) + (srcCoord.x * 2) + inUVOffset]; dst[(dstCoord.y * outStride) + (dstCoord.x * 2) + 1 + outUVOffset] = src[(srcCoord.y * inStride) + (srcCoord.x * 2) + 1 + inUVOffset]; } } } __kernel void ds4_single_plane ( __read_only image2d_t src_image, __write_only image2d_t downscaled_image, __read_only qcom_weight_image_t Kvalues, sampler_t sampler, float src_y_max ) { int wid_x = get_global_id(0); int wid_y = get_global_id(1); int2 dst_coord = (int2)(wid_x, wid_y); float4 result; float2 src_coord = convert_float2(dst_coord) * 4.0f + 0.5 + 0.5f; result = qcom_convolve_imagef(src_image, sampler, src_coord, Kvalues); src_coord.y = fmin(src_coord.y, src_y_max); write_imagef(downscaled_image, dst_coord, result); } __kernel void boxfilter_single_plane( __read_only image2d_t src_image, __write_only image2d_t downscaled_image, sampler_t sampler, qcom_box_size_t scale, float2 unpacked_scale, unsigned int isSwapNeeded, float src_y_max, int2 crop ) { int wid_x = get_global_id(0); int wid_y = get_global_id(1); float2 src_coord = (float2)((((float)wid_x+0.5f)* unpacked_scale.x + (float)crop.x), (((float)wid_y+0.5f) * unpacked_scale.y + (float)crop.y)); int2 downscaled_coord= (int2)(wid_x, wid_y); float4 downscaled_pixel= (float4)(0.0f, 0.0f, 0.0f, 0.0f); src_coord.y = fmin(src_coord.y, src_y_max); // Do down_scaling downscaled_pixel = qcom_box_filter_imagef(src_image, sampler, src_coord, scale); // Writeout the down_scaled pixel if (0 == isSwapNeeded) { write_imagef(downscaled_image, downscaled_coord, downscaled_pixel); } else { write_imagef(downscaled_image, downscaled_coord, downscaled_pixel.yxzw); } } __kernel void boxfilter_single_plane_twoops( __read_only image2d_t src_image, __write_only image2d_t downscaled_image, __write_only image2d_t downscaled_image2, sampler_t sampler, qcom_box_size_t scale, qcom_box_size_t scale2, float2 unpacked_scale, float2 unpacked_scale2, float src_y_max, int dstwidth, int dstheight, int dstwidth2, int dstheight2, unsigned int isSwapNeeded, unsigned int isSwapNeeded2, int2 crop, int2 crop2 ) { int wid_x = get_global_id(0); int wid_y = get_global_id(1); float2 src_coord = (float2)(((float)wid_x+0.5f)* unpacked_scale.x + (float)crop.x, ((float)wid_y+0.5f) * unpacked_scale.y + (float)crop.y); float2 src_coord2 = (float2)(((float)wid_x+0.5f)* unpacked_scale2.x + (float)crop2.x, ((float)wid_y+0.5f) * unpacked_scale2.y + (float)crop2.y); int2 downscaled_coord = (int2)(wid_x, wid_y); float4 downscaled_pixel = (float4)(0.0f, 0.0f, 0.0f, 0.0f); float4 downscaled_pixel2 = (float4)(0.0f, 0.0f, 0.0f, 0.0f); src_coord.y = fmin(src_coord.y, src_y_max); src_coord2.y = fmin(src_coord2.y, src_y_max); if ((wid_x <= dstwidth) || (wid_y <= dstheight)) { // Do down_scaling downscaled_pixel = qcom_box_filter_imagef(src_image, sampler, src_coord, scale); if (0 == isSwapNeeded) { write_imagef(downscaled_image, downscaled_coord, downscaled_pixel); } else { write_imagef(downscaled_image, downscaled_coord, downscaled_pixel.yxzw); } } if ((wid_x <= dstwidth2) || (wid_y <= dstheight2)) { // Downscale 2 downscaled_pixel2 = qcom_box_filter_imagef(src_image, sampler, src_coord2, scale2); if (0 == isSwapNeeded2) { write_imagef(downscaled_image2, downscaled_coord, downscaled_pixel2); } else { write_imagef(downscaled_image2, downscaled_coord, downscaled_pixel2.yxzw); } } } __kernel void p010_to_pd10( __global unsigned char* dst, __global unsigned char* src, uint in_stride_bytes, uint in_scanline_bytes, uint out_stride_bytes) { int wid_x = get_global_id(0); int wid_y = get_global_id(1); const ushort in_stride_pixels = (in_stride_bytes / 2); const ushort *src_luma16 = src; const ushort *src_chroma16 = src + (in_stride_bytes * in_scanline_bytes); const ushort *src_luma_row = src_luma16 + (in_stride_bytes * wid_y); const ushort *src_chroma_row = src_chroma16 + (in_stride_pixels * wid_y); uchar *dst_row = dst + (out_stride_bytes * wid_y); ushort src_x0 = (wid_x * 2); ushort src_x1 = (src_x0 + 1); ushort dst_idx = (wid_x * 8); ushort y0, y1, y2, y3, cb, cr; y0 = src_luma_row[src_x0]>>6; y1 = src_luma_row[src_x1]>>6; y2 = src_luma_row[in_stride_pixels+src_x0]>>6; y3 = src_luma_row[in_stride_pixels+src_x1]>>6; cb = src_chroma_row[src_x0]>>6; cr = src_chroma_row[src_x1]>>6; dst_row[dst_idx++] = (uchar) (y0>>2); dst_row[dst_idx++] = (uchar) (y2>>2); dst_row[dst_idx++] = (uchar) (y1>>2); dst_row[dst_idx++] = (uchar) (y3>>2); dst_row[dst_idx++] = (uchar) (cb>>2); dst_row[dst_idx++] = (uchar) (cr>>2); dst_row[dst_idx++] = (uchar) ((y0 & 0x03) + ((y2 & 0x03) << 2) + ((y1 & 0x03) << 4) + ((y3 & 0x03) << 6)); dst_row[dst_idx++] = (uchar) (((cb & 0x03) + ((cr & 0x03) << 2))); } clEnqueueNDRangeKernelError creating an OpenCL context: %d[ERROR ] %s:%d %s() Error creating the OpenCL Command Queue: %d boxfilter_single_plane[ERROR ] %s:%d %s() ExecuteCopyUBWCTP10ImageToUBWCTP10 failed: error: %d Input format: %d, width: %d, height: %d, inStride: %d, inScanline: %d, planeY: %p, planeUV: %pGPUOpenCL::FlipImage failed to create clBuffer srcUV: error: %dFailed to downscale the first planekailuaCHINODEPROCESSREQUESTINFO is smaller than expected[VERBOSE] %s:%d %s() index:%d, camId:%d, isDiff:%d [VERBOSE] %s:%d %s() Before m_nodeFlags 0x%x [VERBOSE] %s:%d %s() output port id = %d, format = %d sinkbuffer %d m_hBufferManagerDS4Invalid portid %d[VERBOSE] %s:%d %s() bIsPreviewStreamPresent %d, request %llu Flush is in Progress for the session, abortingForce DownscaleImage before flipUnsupported operation (%d), just pass.Request: %li Port to disable: num: %d, portId: %dGpuNodeSetNodeInterfaceubsan: function-type-mismatch ubsan: function-type-mismatchclBuildProgramclCreateSamplerrbRotateImage[VERBOSE] %s:%d %s() Input format: %d, width: %d, height: %d, inStride: %d, inScanline: %d, planeY: %p, planeUV: %p ExecuteDownscaleBy4SinglePlane failed: error: %d[VERBOSE] %s:%d %s() ExecuteBoxFilterSinglePlane CL_TOTAL %f CL_EXECUTE(gpu_hw_time) %f CL_QUEUE %f CL_SUBMIT %f in msecs GPUOpenCL::ConvertP010ImageToUBWCTP10 failed: error: %dGPUOpenCL ReleaseContext error -0 means success-: %dGpuNodeCreate[ERROR ] %s:%d %s() Invalid pProcessRequestInfo argument [VERBOSE] %s:%d %s() camera index:%d, cameraid:%d m_hBufferManagerDS4ForPreviewcameraIdx:%d, bypassCameraId:%d[VERBOSE] %s:%d %s() GPUInstance:%d Request: %lu Master Camera Id: %u, inputIdx:%d(%d)->outputIdx:%d(%d), reso:%dx%d, format:%u [VERBOSE] %s:%d %s() GPU Instance:%d Handle downscaling for Request: %lu [NoBYPASS] ProcessRequest GPU Rotation rotation = %d[MOT_CAMERA_CONTROL]: Failed to retrieve tag id for %s.%s.operator()[ERROR ] %s:%d %s() Could not obtain buffer from: %s ubsan: invalid-objc-castubsan: pointer-overflowFail to load program binary, create Program from source.SaveProgramBinary[ERROR ] %s:%d %s() Error querying for program binary sizes. errNum = %d ExecuteConvertP010ImageToUBWCTP10ExecuteFlipImage failed: error: %dExecuteBoxFilterSinglePlane[VERBOSE] %s:%d %s() Input format: %d, width: %d, height: %d, inStride is %d, output stride is %d GpuNodeGetCapsCHINODECAPSINFO is smaller than expected[ERROR ] %s:%d %s() Invalid argument: pDestroyInfo is NULL Invalid pSetBufferInfo->hNodeSession argumentCHINODERESPONSEINFO is smaller than expected: pInfo->size = %d, sizeof(CHINODEFLUSHREQUESTINFO) = %dInvalid physical camera id:%d [ERROR ] %s:%d %s() Invalid Sensor Dimensions planeId %d input alignment %d %d[VERBOSE] %s:%d %s() One of the o/p ports are invalid so not possible to apply Simo clFlush[ERROR ] %s:%d %s() clCreateProgramWithSource failed: error: %d [ERROR ] %s:%d %s() Unable to allocate memory FWrite failed to write %lu[ERROR ] %s:%d %s() clCreateKernel for the CopyImage Kernel failed: error: %d [VERBOSE] %s:%d %s() mappingEnabledProperty = %d, portID = %d, isOutputport = %d, enableBufferMap = %d Unsupported Chi Format for OpenCL: %d, falling back to NV12[VERBOSE] %s:%d %s() ExecuteBoxFilterSinglePlane crop2 %dx%d ExecuteBoxFilterSinglePlane crop2 %dx%dGPUOpenCL::CopyUBWCTP10ImageToUBWCTP10 failed: error: %dInvalid pQueryBufferInfo->hNodeSessionn argument[ERROR ] %s:%d %s() Invalid pProcessRequestInfo->hNodeSession argument [ERROR ] %s:%d %s() CHINODEPROCESSREQUESTINFO is smaller than expected [ERROR ] %s:%d %s() CHINODERESPONSEINFO is smaller than expected: pInfo->size = %d, sizeof(CHINODEFLUSHREQUESTINFO) = %d Mismatching numbre of m_numOfLinkedCameras:%d <-> Diff Map numCam:%d0Applying SIMO DownscalingmannarBypassCameraIdGPUNodeOnSetupRequestForChiNodeubsan: alignment-assumptionubsan: negate-overflowubsan: load-invalid-value Error Initializing one or more function pointers in Library: %sError getting the requested Device: %dOpenCLObject%s%s%s%X[ERROR ] %s:%d %s() Error querying for devices. errNum = %d [VERBOSE] %s:%d %s() FWrite failed to write %lu InitializeKernelclCreateKernel for the CopyImage Kernel failed: error: %dclCreateKernel for the p010_to_tp10 Kernel failed: error: %dAddEntryinMapTable[ERROR ] %s:%d %s() GPUOpenCL::FlipImage failed to create clBuffer dstUV: error: %d GpuNodeGetFlushResponse[ERROR ] %s:%d %s() ERROR: Buffer Negotiation Failed, W:%d x H:%d! [VERBOSE] %s:%d %s() Change Opt AspectRatio %f to Sensor AR: %f, Updated opt Height %d bIsPreviewStreamPresent %d, request %llu [VERBOSE] %s:%d %s() [BYPASS] ProcessRequest GPU Rotation as rotation = %d UpdateMetaData[ERROR ] %s:%d %s() Unsupported downscale format from Input (Port: %u, Format: %u) to Output (Port: %u, Format: %u) ProcessedFrameNumberubsan: divrem-overflow ubsan: builtin-unreachable clWaitForEventsUnable to allocate memoryclCreateKernel for the DS4SinglePlane Kernel failed: error: %dp010_to_tp10[ERROR ] %s:%d %s() ExecuteRotateImage failed: error: %d ExecuteFlipImageExecuteConvertP010ImageToPD10[VERBOSE] %s:%d %s() Output format: %d, width: %d, height: %d, outStride: %d, OutScanline: %d, planeY: %p, planeUV: %p [ERROR ] %s:%d %s() Failed to downscale the second plane Failed to downscale the second plane[ERROR ] %s:%d %s() ExecuteBoxFilterSinglePlane failed: error: %d Invalid argument: pCapsInfo is NULLGpuNodeSetBufferInfoPhysicalCameraInputConfigSelected aspect ratio is sensor ratio %fplaneId %d align %d %d[VERBOSE] %s:%d %s() This is time to skip request %llu as FD skipped Could not create: %subsan: float-cast-overflow clReleaseProgramclEnqueueCopyBufferFailed to create ds4 sampler with error %dclCreateKernel for the BoxFilterSinglePlane Kernel failed: error: %d[ERROR ] %s:%d %s() clCreateKernel for the p010_to_pd10 Kernel failed: error: %d [VERBOSE] %s:%d %s() Mapped fd %d YUVImage=%p YImage=%p UVImage=%p pAddr=%p GPUOpenCL::FlipImage failed to create clBuffer srcY: error: %d[VERBOSE] %s:%d %s() ExecuteBoxFilterSinglePlane crop %dx%d [ERROR ] %s:%d %s() ExecuteBoxFilter2SinglePlane failed: error: %d Invalid argument: pDestroyInfo is NULLInvalid pProcessRequestInfo->hNodeSession argument[ERROR ] %s:%d %s() CHINODEFLUSHREQUESTINFO is smaller than expected: pInfo->size = %d, sizeof(CHINODEFLUSHREQUESTINFO) = %d Request: %lu Bypass: %u Output Port: %u Fmt: %u - (%u x %u)[VERBOSE] %s:%d %s() GPU Instance:%d Handle 4K EIS for Request: %lu [VERBOSE] %s:%d %s() Couldnt read vendor tag and not skipping frame for FD [VERBOSE] %s:%d %s() ProcessRequest Entry nodeinstance %d m_nodeCaps %u Frame Num %lu Seq number %d numInputs: %u [ERROR ] %s:%d %s() CHINODEPROCREQUESTINFO is smaller than expected CurrentModeubsan: nonnull-arg clReleaseContextInitializeFuncPtrsclCreateKernel for the tp10_to_tp10 Kernel failed: error: %dCreate YUV images[ERROR ] %s:%d %s() Unsupported Chi Format for OpenCL: %d, falling back to NV12 FlipImage[ERROR ] %s:%d %s() GPUOpenCL::FlipImage failed to create clBuffer dstY: error: %d [ERROR ] %s:%d %s() Invalid argument: pCapsInfo is NULL [ERROR ] %s:%d %s() CHINODECAPSINFO is smaller than expected [ERROR ] %s:%d %s() Chi API major version doesn't match (%d:%d) vs (%d:%d) Physical Camera: %u inputPortId:%d[WARN ] %s:%d %s() pCameraConfig: NULL [VERBOSE] %s:%d %s() After m_nodeFlags 0x%x m_bIsSkipProcessing %d [VERBOSE] %s:%d %s() skip request %llu since no preview stream is present [VERBOSE] %s:%d %s() [BYPASS] ProcessRequest GPU NOFLIP %d , rotation = %d ubsan: load-invalid-valueclReleaseSamplerVExecuteCopyImage failed: error: %d[ERROR ] %s:%d %s() ExecuteConvertP010ImageToUBWCTP10 failed: error: %d [ERROR ] %s:%d %s() ExecuteConvertP010ImageToPD10 failed: error: %d GPUOpenCL::DownscaleImage failed: result: %d, error: %dGpuNodeQueryBufferInfoAfter m_nodeFlags 0x%x bIsBufferMappingEnabled %dAspect ratio of Previous %f and present %f [VERBOSE] %s:%d %s() planeId %d align %d %d GpuNodePostPipelineCreateubsan: alignment-assumption clCreateProgramWithSource[ERROR ] %s:%d %s() Error getting the platform ID: %d CL_DEVICE_MAX_WORK_GROUP_SIZE=%dError querying for number of devices. errNum = %dtp10_to_tp10fd %d already maaped but pAddr %p mismatchedDownscaleBy4ImageNDrange %dx%dGPUOpenCL::ConvertP010ImageToPD10 failed to create clBuffer src: error: %dCHINODEDESTROYINFO is smaller than expected[ERROR ] %s:%d %s() CHINODEQUERYBUFFERINFO is smaller than expected InitializeMultiCameraIDMap[ERROR ] %s:%d %s() Mismatching numbre of m_numOfLinkedCameras:%d <-> Diff Map numCam:%d persist.vendor.camera.disableGPUSIMODS[VERBOSE] %s:%d %s() Selected aspect ratio is sensor ratio %f ReadSkipVendorTaglahainaclCreateImageclFinish[VERBOSE] %s:%d %s() CL_DEVICE_MAX_WORK_GROUP_SIZE=%d InitializeResources[WARN ] Create Separate Image input params yuvFormat[0].height = %d yuvFormat[0].width = %d yuvFormat[0].planeStride = %d yuvFormat[0].sliceHeight = %d yuvFormat[1].height = %d yuvFormat[1].width = %d yuvFormat[1].planeStride = %d yuvFormat[1].sliceHeight = %d PortId = %d, isOutputPort = %d format = %dExecuteCopyImage[ERROR ] %s:%d %s() GPUOpenCL::FlipImage failed to create clBuffer srcUV: error: %d [VERBOSE] %s:%d %s() ExecuteBoxFilterSinglePlane Scales %fx%f UninitializeGpuNodeProcRequest[ERROR ] %s:%d %s() Invalid Argument: %p [VERBOSE] %s:%d %s() Input port id = %d, Format = %d m_nodeCaps: %u max %dx%d min %dx%d InitializeDownscaleBufferManagersGPUInstance:%d Request: %lu Master Camera Id: %u, inputIdx:%d(%d)->outputIdx:%d(%d), reso:%dx%d, format:%u[VERBOSE] %s:%d %s() passing DS DSb Skipping frame %llu from DS for FDfillmore[ERROR ] %s:%d %s() Invalid argument: pNodeInterface is NULL GPUIntance:%d ApplyDownscale:inputPort %d->outputPort%d For Request: %lu Eubsan: sub-overflowubsan: negate-overflow ubsan: out-of-boundsubsan: invalid-builtin ubsan: nonnull-returnubsan: nullability-return%s%s%s.%sclGetDeviceInfo[ERROR ] %s:%d %s() clCreateKernel for the BoxFilterSinglePlane Kernel failed: error: %d [VERBOSE] %s:%d %s() Create Separate Image input params yuvFormat[0].height = %d yuvFormat[0].width = %d yuvFormat[0].planeStride = %d yuvFormat[0].sliceHeight = %d yuvFormat[1].height = %d yuvFormat[1].width = %d yuvFormat[1].planeStride = %d yuvFormat[1].sliceHeight = %d PortId = %d, isOutputPort = %d format = %d [ERROR ] %s:%d %s() clCreateImage failed: error: %d ExecuteBoxFilterSinglePlane failed setting arg %u with err %dExecuteBoxFilter2SinglePlane failed: error: %dGPUOpenCL::ConvertP010ImageToPD10 failed: error: %d[VERBOSE] %s:%d %s() GPUOpenCL ReleaseCommandQueue error -0 means success-: %d [VERBOSE] %s:%d %s() GPUOpenCL ReleaseContext error -0 means success-: %d [ERROR ] %s:%d %s() Invalid pQueryBufferInfo argument Invalid pQueryBufferInfo argumentChi API major version doesn't match (%d:%d) vs (%d:%d)Invalid Sensor DimensionsERROR: Buffer Negotiation Failed, W:%d x H:%d! Change Opt AspectRatio %f to Sensor AR: %f, Updated opt Width %d[VERBOSE] %s:%d %s() Applying SIMO Downscaling /sys/devices/soc0/soc_idubsan: shift-out-of-boundsclReleaseMemObjectInitialize[VERBOSE] %s:%d %s() Load program binary %s successful! [VERBOSE] %s:%d %s() fd %d already maaped YUVImage=%p YImage=%p UVImage=%p pAddr=%p fd %d already maaped YUVImage=%p YImage=%p UVImage=%p pAddr=%pConvertChiFormatGPUOpenCL::CopyImage failed: error: %dGPUOpenCL::FlipImage failed: result: %d, err: %dExecuteDownscaleBy4SinglePlane failed setting arg %u with err %dInvalid argument: pTagTypeInfo is NULLPhysicalCameraConfigs not updated for SATnumOutputPorts = %d, numInputPorts = %d[VERBOSE] %s:%d %s() Aspect ratio of Previous %f and present %f passing DS DSb[VERBOSE] %s:%d %s() not skipping FD frame SkipGPUprocessingbasedonFD[VERBOSE] %s:%d %s() GPUIntance:%d ApplyDownscale:inputPort %d->outputPort%d For Request: %lu X ubsan: mul-overflow ubsan: cfi-check-fail ubsan: cfi-check-failclGetDeviceIDsclCreateImage failed: error: %dDeleting fd %d YUVImage=%p YImage=%p UVImage=%p pAddr=%p[VERBOSE] %s:%d %s() Physical Camera: %u inputPortId:%d [ERROR ] %s:%d %s() pPhysicalDiffMap is NULL [VERBOSE] %s:%d %s() bypass set %d, inplace set %d phycam:%d can't bypass, because physical preview dimension not match logical previewGPU Instance:%d Handle downscaling for Request: %lu[VERBOSE] %s:%d %s() FD Frame skip so not possible to apply SIMO sm8450Invalid argument: pNodeInterface is NULL[VERBOSE] %s:%d %s() GPUIntance:%d ApplyDownscale:inputPort %d->outputPort%d For Request: %lu E ubsan: implicit-conversion[VERBOSE] %s:%d %s() Fail to load program binary, create Program from source. Error querying for devices. errNum = %d[ERROR ] %s:%d %s() clCreateKernel for the tp10_to_tp10 Kernel failed: error: %d mappingEnabledProperty = %d, portID = %d, isOutputport = %d, enableBufferMap = %dExecuteCopyUBWCTP10ImageToUBWCTP10 failed: error: %dExecuteBoxFilterSinglePlane failed for WaitForEvents err %dExecuteBoxFilterSinglePlane CL_TOTAL %f CL_EXECUTE(gpu_hw_time) %f CL_QUEUE %f CL_SUBMIT %f in msecs[VERBOSE] %s:%d %s() global WGS %zu x %zu SCales %f and %f Unable to get StreamTypePresent Vendor Tag Id %d[ERROR ] %s:%d %s() Unable to get FD based process skip Tag Id [VERBOSE] %s:%d %s() passing DSb DS passing DSb DS[ERROR ] %s:%d %s() [MOT_CAMERA_CONTROL]: Failed to retrieve tag id for %s.%s. ubsan: type-mismatch clGetPlatformIDs[ERROR ] %s:%d %s() clCreateKernel for the FlipImage Kernel failed: error: %d [VERBOSE] %s:%d %s() Entry %d not found [ERROR ] %s:%d %s() Failed to downscale the first plane Invalid pSetBufferInfo argumentpInfo %p pInfo->hChiSession %pQueryBufferInfo[VERBOSE] %s:%d %s() Sensor width %d height %d aspect ratio %f output port id = %d, format = %d sinkbuffer %dGPU:%d No full inputPort linked to this output port(%d->%d)[VERBOSE] %s:%d %s() [NoBYPASS] ProcessRequest GPU Flip = %d , rotation = %d [VERBOSE] %s:%d %s() Last Request Id flushed %lu Flush request Id %lu from nodeubsan: pointer-overflow clCreateProgramWithBinarywb[ERROR ] %s:%d %s() clCreateKernel for the RotateImage Kernel failed: error: %d [VERBOSE] %s:%d %s() fd %d already maaped but pAddr %p mismatched ExecuteBoxFilterSinglePlane failed: error: %d[ERROR ] %s:%d %s() ExecuteBoxFilterSinglePlane failed for WaitForEvents err %d ExecuteBoxFilterSinglePlaneSIMOInput format: %d, width: %d, height: %d, inStride is %d, output stride is %d [ERROR ] %s:%d %s() GPUOpenCL::ConvertP010ImageToPD10 failed to create clBuffer src: error: %d Soc: %s, kernel: %s, dimensiont: %d, buffer size: {%lu %lu}, localWGSize: {%lu %lu}CHINODECREATEINFO is smaller than expected[ERROR ] %s:%d %s() Invalid pQueryBufferInfo->hNodeSessionn argument Invalid Argument: %pNumber of linked cameras %d, exceeds max supported linked cameras %d StreamTypePresentGPUInstance:%d Handle bypass for Request: %luGPU Instance:%d Handle 4K EIS for Request: %luCouldnt read vendor tag and not skipping frame for FDnetrani/sys/devices/system/soc/soc0/idro.archclSetKernelArg[VERBOSE]ChiNodeGPU.cl.bin.[ERROR ] %s:%d %s() clBuildProgram failed: error: %d, Log: %s [VERBOSE] %s:%d %s() Save program binary successful %s clCreateKernel for the RotateImage Kernel failed: error: %dInvalid argument: pInfo = %ppersist.vendor.camera.disableGPUBufferMap[VERBOSE] %s:%d %s() phycam:%d can't bypass, because physical preview dimension not match logical preview This is time to skip request %llu as FD skipped[VERBOSE] %s:%d %s() Unable to read endorTag IsPreviewStreamPrasent request %llu [NoBYPASS] ProcessRequest GPU Flip = %d , rotation = %d[VERBOSE] %s:%d %s() Request: %li Port to disable: num: %d, portId: %d ubsan: divrem-overflowubsan: vla-bound-not-positive ubsan: invalid-objc-cast clCreateCommandQueueclCreateBufferEclCreateProgramWithBinary failed: error: %d, status: %d[ERROR ] %s:%d %s() clCreateKernel for the DS4SinglePlane Kernel failed: error: %d FindEntryinMapTableExecuteConvertP010ImageToUBWCTP10 failed: error: %dExecuteDownscaleBy4SinglePlane[ERROR ] %s:%d %s() ExecuteBoxFilter2SinglePlane failed setting arg %u with err %d [ERROR ] %s:%d %s() CHINODEDESTROYINFO is smaller than expected Number of linked cameras %d[%p] - InstanceId: %u ReqId: %u pConfig: %p - numConfigurations: %u[VERBOSE] %s:%d %s() After m_nodeFlags 0x%x bIsBufferMappingEnabled %d persist.vendor.camera.enableGPUProfileDataError getting the platform ID: %dEntry %d not foundUnknown rotation angle %d[ERROR ] %s:%d %s() GPUOpenCL::CopyImage failed: error: %d [ERROR ] %s:%d %s() ExecuteDownscaleBy4SinglePlane failed: error: %d ExecuteBoxFilterSinglePlane failed for CLGetEventProfilingInfo err %d[VERBOSE] %s:%d %s() Soc: %s, kernel: %s, dimensiont: %d, buffer size: {%lu %lu}, localWGSize: {%lu %lu} GpuNodeQueryVendorTagGpuNodeFlushRequestcom.qti.chi.cameraconfigurationbypass set %d, inplace set %dCHIGPUNODEm_hBufferManagerFormatConversion[VERBOSE] %s:%d %s() GPUInstance:%d Handle bypass for Request: %lu [ERROR ] %s:%d %s() GPU:%d No full inputPort linked to this output port(%d->%d) skip request %llu since no preview stream is presentGetDisabledPortForRequestsm8550clReleaseKernel[ERROR ] %s:%d %s() Failed to create ds4 weights image with error %d Save program binary successful %scopyImage[ERROR ] %s:%d %s() GPUOpenCL::FlipImage failed: result: %d, err: %d [ERROR ] %s:%d %s() Unable to downscale image with current format ExecuteBoxFilterSinglePlane crop %dx%dGPUOpenCL ReleaseCommandQueue error -0 means success-: %d[ERROR ] %s:%d %s() CHINODECREATEINFO is smaller than expected [ERROR ] %s:%d %s() Invalid argument: pDestroyInfo->hNodeSession is NULL [ERROR ] %s:%d %s() Invalid pSetBufferInfo->hNodeSession argument CHINODESETBUFFERPROPERTIESINFO is smaller than expected[ERROR ] %s:%d %s() PhysicalCameraConfigs not updated for SAT [VERBOSE] %s:%d %s() Change Opt AspectRatio %f to Sensor AR: %f, Updated opt Width %d ubsan: shift-out-of-bounds ubsan: float-cast-overflow[ERROR ] %s:%d %s() Error Initializing one or more function pointers in Library: %s [ERROR ] %s:%d %s() Error getting the requested Device: %d Error querying for program binaries. errNum = %d[ERROR ] %s:%d %s() GPUOpenCL::DownscaleBy4Image failed: result: %d, error: %d DownscaleImageSIMOExecuteBoxFilter2SinglePlane failed setting arg %u with err %dChiNodeEntryPhysical Camera: %u -> Node: %u:%u @ InputPort: %u | Type: %u[ERROR ] %s:%d %s() OpenCL Initialization failed ! [VERBOSE] %s:%d %s() [E] - Preview Dimensions: (%u x %u) [ERROR ] %s:%d %s() ProcessRequest failed for request = %luSeq number %d [VERBOSE] %s:%d %s() Flush request Id %lu from node FlushRequestkona>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>D>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>O>>>>>>>>>>>>>>>>>)>>>>>>>>>>>>>>>>>>>>>>)>>>>0>>>>>>>>)>>>>>>>>>>>>>>>>>>7>>>>>>>>>0>>>>>>>>>>>>>>>>0>>>7>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>>V@@{Gz?ư>H 3[D0+59/vendor/lib64/so@@/data/vendor/camera3A4A5AEAFAGA]A]A]AA]A]A]A9A:A;AOAPAQAKALAMA]A]A]ARASATA   org.quic.camera.streamTypePresentorg.quic.camera.skipgpuprocessingbasedonfdcom.lenovo.moto.controlflip_stillSupportedFeatureChiNodeCapsGpuMemcpycom.qti.node.gpu;(dD\t XL 4lTt@`4Tt<\4T8 t  !X!!!44"Tx"t"#T###($4t$T$t%H%%%$& h&4 &T &t 0' ( ,. . / _D \ad lb \p lr v < ld , \4 t  4 |T < | , < 4bN4 ROBWA_è#_?#O ep#b@R+biAD+b0b?#O `6s"b@RO#bO#_?#Os"a@RbiAT b b?#O `6sa@RO#aO#_?#Osa@RaiApaa?#O f`6sa@RO#aO#_?#Osa@RaiAX?aa?#O B`6s a@RO#ըaO#_?#Os a@RaiAL7aa?#O `6s?~a@RO#ՄaO#_?#Os?oa@RwaiA wa|a?#O `6s0Za@RO#`aO#_?#Os0Ka@RSaiASaXa?#O `6s66a@RO#`O#_?#Osr)`@R1`iA1`6`?#O `6s:2`@RO#`O#_?#Os:2`@R `iA2 ``?#O `6sv_@RO#_O#_?#Osv_@R_iA8+__?#O l`6s_@RO#_O#_?#Os_@R_iAl'__?#O H`6s"_@RO#ծ_O#_?#Os"_@R_iA__?#O $`6s_@RO#Պ_O#_?#Osu_@R}_iA'}__?#O `6s`_@RO#f_O#_?#OsQ_@RY_iAY_^_?#O `6s <_@RO#B_O#_?#Os -_@R5_iA`5_:_?#O `6s!_@RO#_O#_?#Os! _@R_iA\!___$Y9p4A_*@_kaTA5_?#OR _@7*OF(9#_(_qkT!(BRhxH^qh"HqOF(9#__$ՈŔҨULT҉ǿxJ TjJpY* -TMpY- @TDnOGF Ts/7Tag}D?',TMhW#- TMkW# `T7ݾl&l @T-M-4 aT /?T~HҐ0 T$H҄$Te҅կT &8 T  =.ҍN·8m TQҍQ·8Tc nTzdddZa T  ?TF)뚮K+  TSV,H AT Ս? T5ŔҩU  T ʛj|G !T  ? ` T$ґǿxJ TC#N`T  ?@ T}D?/TpIҰmpT ի T\ &? T Ր ?T ?T N ?@T +? T  ? T  ?`T    ?`T Q?T )? aT_ Յ?T_$%_$գ _$wO_$WZ_$N_$w5_$3_$ճJ_$S_$՟6_$70_$c/_$՗L_$ՃZ_$Z_$+M_$ՏK_$&_$գM_$'%?#{O8o9J S;h@CѨ>/>#=~=\i@_? TOQ{P#_9]?#{ R\` @{¨#_?#+m#m{og_WO@ Y;-R%@*R ]B|cD| ƄR[[`6!8[h6@`:!![i6@`>!@( [j6@`B! [k6@`F! [l6@`J!8 [m6@`N!X [n6@`R!-[o6@`V![p6@`Z!"[q6@`^!l[b6@`b!,3|[c6@`f!=v[d6@`j!Hp[e6@`n!j[g6@`r!P d[u6@`v!3^[v6@`z!HX[f6@`~!#R[w6@`!`L[h6@`!-F[i6@`! @[j6@`!, :[k6@`!h3 4[l6@`! .[m6@`!H= ([n6@`!0"[o6@`! [*`g:@sRg:@i>@ kB@ lF@L nJ@ oN@ {V@ jR@J mZ@ h^@ pb@ ff@F vj@ an@ wr@ cv@C bz@ `~@ x@ q@Q d@ e@g@i@Ik@  RZ`B[t:@Ҡ@@ZcC R? 4\A@96Д *`A@?q!TB RR[@!hB\#RRZB RRr[@7!hc'Rn1^A.@9n6`A@qaTB R *RV[У!hBRRZB R *RD[7!hc!RZR@u>@`d#Z@tBR"R?`4^A@9p6bA@_qaTB R *R[@!hBh RRPZB R *R[@7!hc" RCZ5R*j,@]KT*@ OGWF_EgDoC{B#Am+l#_vB@B`y`3R@= U@=;.ZS!R?`` @1 5wJ@lb@צ@UqZaUSb?` @ 5v@ i M Z`@u"RR?Z_AK@ 46ZcAN@qaTB R *RZ@!hB7%RRYB R *RZ@7!hc%R!\A%@96c`Ak@qT*@!hB(D"RRY*@7!hc#D"RY\A@9w7URY`A @qT*@!hBx$$RRYB R *R5Z@Ѕ7!hcP+$RtYUR1(6{cAo@qT*c@!hB.&RRXYB R *R Z@!c<$&RKY*R;ZwACZ@,T\"T;RV@RzS9B| #+TY @\Z`D)T )Tj58@9@B_rNAӨ?T?A T?%^T  @&go_A -ɚ ~7V~bC@|?HT ga_ref!%[% $@vAEuh t @{@ n@-"~ ֕ΓPAʓ"ɓq~Гד8Ɠ.ѓ|ēj H   aG6!\~^fC0UF|n!%_rbN@ӷ B@窌 | eG')%\+ѓʛ'$;|mRՓ, t  yN#Ґ! 6 gA~@F|X&~x 1gA|`~ \X dNҰ]ӷl} w LD `} g "AM|̓B FUʓ) RwWۓ*z> i +~ mA o=]@͓&_˪֓ |A~e1  #f'ՓTƓU̓,@ @J * ʥE}A}xʤ4+ [fAp|| PWʬ: |IV/}O|B\#c|RWqZU!3W/9X BRFXJX'*>XR"R**X##R6{cAo@qaTB R *RTc@!hB8<%cRRTB R *RTc@У7!hc<%4cRl@Lu@T@ _T @ cPR"R**@}T wn@``ScP,Rb}?+4L@96*aA@qaTB R *RTТ!hBFRR*SB R *RqT7!hc$FR*S;R**S**S@.kcP@(j4B(iTzu**ScP**SI@6`A@_qT*}Ђ!hBRT$=RsS*pD/!p*cTx*$=RkS|  @RS|||@@@`@Qi@7oTw!H-CTuv@຋ @mRS`@!?c@`4awz@ % 7 nAS`@?A@96JaAP@qT*"c@Х!hBP<%dfRRS*c@У7!hc<%1dfR!Svv@຋ @mS`@!%?c@`4atz@ % 7 nR`@?D@9D6Є`A@?qT*c@!hB<%DiRRR*c@У7!hc<%D>DiRf!%Suv@຋ @mR`@!(?c@`{4awz@ % 7 nR`@?V@9 6cA@qT*c@Т!hBX?<%lRRR*c@У7!hc<%`lRCcP"R**R }RRcPM4R}R"R*j|*R@i< BcP CTwn@``mRcP,Rb}@? 4N@9.6*aA@qaTB R *RRЂ!hB JRR*9RB R *RR7!hc\JR*,R}!(Svv@຋ @m+R`@!2?c@`H4atz@ % 7 nR`@?I@96еbA@?qT*c@Ђ!hBp-<%nRRQ*c@У7!hc<%h&nR8!2Ruv@຋ @mQ`@!%?c@`j4awz@ % 7 nQ`@?K@9+6ДbA@_qT*c@!hB-<%qRRQ*c@У7!hc<%hqR^cP 4zv T*AT2!%Rvv@຋ @mQ`@!`?c@`4atz@ % 7 nQ`@?M@9M6kaAc@qT`*fc@!hB*<%$tRR[Q`*Xc@`e7!hc<%89$tR*!Q@@}!RjxjxQjxAT*!`/RORRdRo"R' N-R N[A c⃆A@+(6| *1bA*@_qTB RRSOQCI)$RBJ)RA;CGD)N?E)3+e#!h@BP.C ;NB RR5OCI)I+J)GD)Y#E)!l/K;C#c@c.$RC3+; fN@\;CqT@4[qVT @qg@9(6 *g@9(7Y5u9Rg@9/6`A@qTsB Rs *RNЅ@c!hBdRR** .NsB Rs *RNb@!c.dR** Ns^A4 ACF!@k')1JY`j?T,!@k)Tq@9q(6`A@qaT|B R *RN!hB: DzRR*MsB Rs *RNb!c p(DzR*M;Ro>AR@ շ`B@ @ ՅVC'@9R=9@): A)(B)>#0)<')+)N:P7W@j@iM@RrѣѤѥS`?F` ^5W@Ѡj@iG @B'@*6R:_}M@#ѥS?ֹ^5P@ q#TW@j@iM@ѥS?/@`aN_A2P7;@@2@t2Q@? qT/@@1@Z;CqT@4_[q8T  @?q@9(6p*@9 (75/%@+@9B6@`Tm@9(6sbAn@qTsB Rs *RM+@c!hB( DyRR*5MsB Rs *RMb+@!c %DyR*'M\[@`R@+M`?V[@`R@!M?Y[@`R@M?`@ !@k'3YolT#@kT@ @4R@9/6{cAb@_qT|B R *RM@!hBdRR** LsB Rs *RMb@!c.dR** L 4 AC#@kT@  Th@( EF/@|Vt@9T(6`A@qaTxB R *RMM/@;@p@@!h@B xRR*G+@ LzB RZ *R3M;@;EBs@@@!c <xR*kL3@*;@/@@z@@@: RLo $|<<`A@ a@ AJA T V*_A+@[W) @9S'3@,(6ZcAC@qATxB R *RL;@CE@G@ @!hB\'$~RR* +LzB RZ *RL;@GEB@'@,@e!c\'P:$~R*L*Oh A@Tb@h A@aTAATA* T JL \L(L;@@@9w6cA@qaTtB R *RL^eЃ!hB.RRKuB R *RxL^`cХ7!hc.!RK3@ Rm@_TOVWU_TgSoR{Q#_@RLK LL?#{+OU;@**|@`>K`@*R?*5t~@`K`@c!RR?*@5t~@`}K`@SARR?*5t@ ێkoK`@ca@BR*?*4]A@9(6ss *)aA(@q!TB RRKc!hB4䨀RR*6KB RRKb`УХ7!hc4䨀R*+K R@_ TOF+@{D#_QL?#{;WO V;շ@*** q))^A> hT*9 ik8 ֟$R) $ R)$ R$ Rt~@`J`@*R?* 5t~@`J`@!RR?*`5t~@`J`@ARR?* 5t~@`J`@aRR?* 5t~@`J`@sRR?* 5t~@`J`@RcR?*`5t~@`J`@SRR?*5t~@`J`@CRR?*5a@cѣ=~t@ ێkJ`@ca@BR*?*5/@9 7@9-6ss *JaAI@?q!TB RR Kec!hB2d,€RR*EJB RRJb`ce7!hcd,T€R*:J R@_TOIWH;@{F#_֩)aA(@qaTsB Rs *RJbec!hB:d,$RR*JsB Rs *RJb`c7!hcd,($R*J@9 71K?#C{og_WO@!*eJ4a!t1`J@4@)@?T)@I T @i @'@T(@HT@@:Xa!5J` 4!%0J`4a!+J4a!`&J@5i@*1:TRY}Sb@M9T+19Th}Sw@ZXa!J 5`@:RJ|||@@&@@G@/T@G@/4TXa!I5@-==F@RI|||@@%@@G@T@G@/TXa!I4!%I 4a!I4a!`I4]A@9(7h@ 1,T R*}S.@M*T1+T }SB\AB@9(6`A @qATzB RZ *RIBG9@Y@r@?T@ Tl@ z@R|I|||`@"@Ba@(w *bA @_qATB RRIBG9@@r@T@ "Tj@ z@:@G@cbTE@9aB@Z.@!hHRrei*@GBL,1RRHzB RZ *R]IBG9@Y@r@T@  Tm@ v@R I|||`@@a@@F@Tv@6@F@TǢ@9@m>@ӦcF@rg!c1p JRRZHOHWG_FgEoD{CC#_z@RH|||`@!@a@@xF@ TN@9FO@apF@ber!hB 1OR䱁RRHB RRHBG9@@ryD@TK@k Tk@ w@WRH|||`@@a@w@7@-F@"T@9@u"@c?re!c1\IR䱁RG{G?#C{;OU;@**|@`))>G`@*R?*`5t~@`G`@!RR?*5t~@`G`@ARR?*5t~@`G`@aRR?* 5t~@`G`@sRR?*` 5t~@`G`@RcR?* 5t~@`sG`@SRR?*5t~@`eG`@RR?* 5t~@`WG`@RR?*`5t~@`IG`@!RR?*5a@cѣ=2t@ ێk3G`@ca@BR@*?*4]A@9(6ss *)aA(@q!TB RRGabc!hB`3̀RR*FB RRGb`7!hc3̀R*F R@_ TOH;@{FC#_H?#{og_ W O J;I@)0A)$4 sZ_A*s *2Ak!TnA)Cѥ*CR'R>=4A G*4@cѥ*R*A*F5^*]***@@@)@?Yq? qT?q!T@_ qT%R@qT*'dS4*AF@L@9q(6aA@@x9CqTRB RGaecR!hBdR** VFRB R Gbc!cdR.** HFq@T^]@\Z@`R@GF ?XZ@`R@=F?[Z@`R@3F?@:CqaT[qT  @qM@9(7y53*M@9/6aA@qTRB RFaecR!hBdR**0R ERB RFbc!cdR.**1R E5_^@\Z@`R@E ?XZ@`R@E?Z@`R@E?ۊ@ %@͹XE@`?4R;@q 6*.*-O@9o7<R(aA@q!TB RRUFaec!hB$),RR&REB RREFb`cХ7!hc,|R&R<RE @J@__ T*OKWJ_IgHoG{F#_@RFF?#{og_ W O J;I@(0A$4lx )'{_A*yR9 *B RF"cХ!cdR.**0R ?E42A k"T'@CѴrA)*CR'Rq>= 4AE*cѥ*R*A*qE;4*{_AAqE@k@9qK(6)aA:C(@qTzRZ *B REaЂeCR!hBdR** DzRZ *B REBcХ!cdR.** DqT^]@Z@`R@D`?֛{_A\Z@`R@D?ZZ@`R@D?@:CqAT[qT  @?ql@9/6*l@9 /6kaAj@_qTyR9 *B ROEaЂe#R!hBdR**/R D5_^@Z@`R@D`?֛{_AZZ@`R@D?|Z@`R@uD@?)^a]"#@****=5@a@"#@#SDS~SS1 5@ %@͹XSD@`?3R.@qj 6*2*1n@9n74R,ЌaA@qaTsB Rs *RDaТЅc!hB$)3RR&RDuB R *RDТ`Х7!hc3|R&R4R D @J@__ T*OKWJ_IgHoG{F#_@RE*E?#Cs{og_WO?U;ը@+0AH4 !B5`[`_ 2As'@{BDTA)+@[g]A*@@cJ(6RB R`A@`*sDqTHLTz /@7A)Z *@aiХCR!hB3R CRB RVDHTjJ */@?A)B@k!cR' C\A@(7#@}T&`*RB RaA@+DqTHCTt.@FA) *@aЭc@R!hBh4dR WCRB R D?Ti) *.@jA)"@ec@!cdR@ ?C@ ҠB^_@>Cg@W<`:T@*RѤr@?֥]4Є\A@9f6Х`A@qaTtR *B RC]aТR!hBRCtR *B RC]`Уe7!hcRB@ @^@;B@RScѤr?֡]41^A @9`6c`Aj@_qaTtR *B RC]aЂR!hB'RBtR *B RpC]`Уe7!hcRB@ ҠgC^d_@9B@RѤr?֤]4^A@9h6!`A+@qaTtR *B R>C]abR!hB"RxBtR *B R,C]`У7!hcRkB@ Ҡi^e@@7skB@Rr?ִ]t4]A@9j6B`AL@qaTtR *B RB]aТR!hB4R4BtR *B RB]`У7!hcR('BAB4RA{B_@"44RA|B]q_ZZ@`R@B?XZ@`R@B@?YZ@`R@B?\Z@`R@A ?J)*L)?@ `5'@k@l_@_@g KK? qT?q!T@ qT)R@qT*[@L)?@|SW@|S 5@ %@͹XA@?*_@4\A@@97:R@^TI*@^T*s@OTWS_RgQoP{OC#_@*@^T2sbAp@qTvB R *R5B]a!hB6dRR&RnAWB R *R"B]@7!hcdR&R:R_A@^TB@RrBA?#C#m{g_WOY; @ 6@ !***o**v~@*=#M!`(@ _[==<9A`@*R?֚Z_A4*7Rv~@`'A`@!RR?ր4*WRv~@`A`@cARR?ր4*wRtv~@`=-A`@aRR?ր4*RbH@9=-(6aA@qTVB R *RAN-a!h"B$6!",WRR@VB R *RAN-e`""!c,WR@v~@`@`@RR?ր4*R v~@`@`@RsR?ր4*Rv~@`@`@cRR?C4*RI@9i6X *)aA)@?qATB RR6Aab!hB4.,ZRR**o@B RR%A@Х7!hc, ZR**c@L@96kaAk@qTXB R *R Aab!hB4.,\RR**C@XB R *R@@Х7!hc, \R**5@M@9m6aA@qaTXB R *R@aЅ!hB7,_RR*@XB R *R@@Х7!hc,4_R* @6Rmb@q0Tcc!**@?P7w@ q@@C R?4N@96*aA@qaTXB R *R@a!hB,DaRR*?XB R *R@@Х7!hc,0DaR*?w@`Q?+@PR#R?4O@96*aA@qaTXB R *R]@aB!hB/,bRR*?XB R *RK@@Х7!hc,4+bR*?w@`Q?+@!PRR?4P@96*bA@qaTXB R *R"@aB!hB/,bRR*\?XB R *R@@Х7!hc,4+bR*O?w@`QT?+@APRR?4Q@96*1bA1@?qaTTB R *R?aBЃ!hB/,cRR*!?TB R *R?@Х7!hc,4+cR*?w@`Q?+@aPRR?4*@@9`6`A@qaTSB Rs *R?aBc!hB/,DdRR*>XB R *R?@Х7!hc,4+DdR*>?P7B@9 (6`A@?qTTB R *R}?@'@#T@T#@,cT#TcBVcacc!hB6,DeR wR w w# w>SB Rs *RW?@'@bcT@T#@TcT:cB{cccХ!c,@ |1a |DeR | |>"@^_aT*@OTWS_RgQ{O#NmC#_'='K@9 !(g N NG $!.E!.T?P3&.T3;k(6JaAJ@_qaTXB R *R?N)a!hB,$\RRB>VB R *R>N)Х!c,48$\R5>v~@`:>`@RRR?֠4*L@97c&**a@>`P7w@ ێxb@k>`@qa@!TCBR*>>B@9/7wBR*?*>P75P>>>65??#+m#m{ Co g _ W OY;՛(@I@{_A? q5T(0AH:4 &* X_A)ѥѓnA)*#R'R<9R`4 #*#R*'@*#Aa##c# !H(+J#=\5z A AA A *****:4k]Aj@9*6bA @_qATYR9 *B R4>ae#R!hB:'RDo=YR9 *B R#>"@c7!hcD'R)c=9RCy5A<@@ q TqAT@ qT'R*C4Y4'*K@q`T* A AA A S~Se~SfS4s^Ak@9+61bA-@qATWR *B R=aeR!hBH6,RD=[R{ *B R=b@c7!hcD,R47 ='*@ %@͹X=@?*7R's^AAZ=@l@9ql(6aA@@9CqT[R{ *B R=aecR!hBdR** <[R{ *B R=bc!cdR.** <qT\ZZ@`R@< ?\Z@`R@<@?[Z@`R@<?ַ@:CqTZq{_A@T  @qm@9m(6aA@qTYR9 *B R<=ae#R!hBdR**+R r<YR9 *B R&="c!cdR.**,R a<z5*{_A@m@9/7z5]Z@`R@[<`?֛{_A@\Z@`R@N< ?ZZ@`R@D<?'@qO 62A@ kT @*`p@96TД *)aA1@?qTB RR<ae!hB$7D0RR <B RR<@ce7!hcD0R<*Rq@973*1 @<Rq@96JaA@@qTUB R *R<aBe!hBD1RR**;WB R *R<@c7!hcD1R**;)@]? T*ONWM_LgKoJ{I#Hm+Gm#_@R<<?#;m3m+m# m{ g _ W O !*@  @ h@ I@ L!@ #W;@NL# "_A@9@@))((6""V *aA@qTB RR9<A`A`a!hBl34kRRp;B RR&<A`A`Уe!c48kRb;u~@`g;`@*R?ր4*5Rt~@`W;`@!RR?ր4*URzu~@`G;`@ARR?ր4*uRjt~@`7;`@caRR?ր4*RZu~@`?-&;`@RR?ր4*RIt~@`;-;`@RR?ր4*R8u~@`?-;`@RR?ր4*R't~@`;-:`@RRR?`4*u~@`:`@3RR?ր4*5Rt~@`:`@!RsR?ր4*URu~@`:`@ARcR?ր4*uRt~@`:`@aRSR?ր4*Ru~@`:`@CRR?ր4*Rt~@`:`@RR?ր4*Ru~@`:`@RR?ր4*R`!*[@S@TC-B###,!~&!~Ќ1 1U8t8S)** @9(6)aA)@?qTVB R *R:aХ!hB4yRR**8:VB R *R:!c448yR***:t~@`/:`@RR?ր4*RR !(TSB-V#7#,!~!~;] <~ <88S)** @9(6JaAJ@_qTVB R *R:a!hB4{RR**9VB R *R:!c4|{R**9u~@`9`@RR? 4*5R @96kaAk@qTSB Rs *Rp:ac!hB@"4|RR**9VB R *R]:@У7!hc4|R**9 @9l6aA@qaTSB Rs *RB:aХc!hB4~RR*|9VB R *R0:@7!hc4 ~R*o9 R@[? TONWM_LgK{J#Im+Hm3Gm;Fm#_a@#Zt@ ێk\9`@#a@BR*?*@5t:?#C{O9oJ S;h@CR>/>#=~=':i@_? TOP{OC#_G:?#c{ Cog_WOS;j@(0A54!(B*?7S2As1TKjA)tE)OJ]AfE)I@9i(6g`A@qTZRZ *B R9f@AECCR!hBĂR/ 8ZRZ *B Rx9f@BCЅC!c/ĂR 8@#7~3Tڒ@ Ҡ@BA@s?8}`2T@R*Cr@?֫[4m]A@9l6f`A@?qaTYR9 *B R;9[ABE#R!hB!䄁R/u8YR9 *B R)9["@CE7!hc/䄁R h8ْ@ Ҡm@lBA@+[h8O@~?}`)T@R *CCr ?֮[4l]A@96d`A@_qaT[R{ *B R8[AEcR!hB䆁R/-8[R{ *B R8[b@CЅ7!hc/䆁R& 8@Aw8K@48RAx8[qzZ@`R@8`?@Z@`R@8@? O@~@Cxk =< )3{+ 'Z_A* H6K@(6uW *bA@q!TB RRb6@Ab!hB+$-RR5B RRR6@E!c$-R5z@Z_AK@Jy%R #W(6bB`AO@qTY9 *B RR/6@WHAЂ#!hB-RRg5\B R *R6'@'@ЁE@VH!'@c-RV5zЊ@Z_A P7*'@Z@9a2'(7pbA @qAT[B R{ *R5ABec!hB :́RR/5SB Rs *R5b@cЅ7!hc :́R#5R'@_`T.q1bA-@qATZB RZ *R5AЂeC!hB9 :́RR5VB R *R5@cЅ7!hc : ́R45R'@_T!6L@(6wbA@qaTUB R *R5@AbЅ!hB+d.RR4WB R *R5@ЁE!cd.R4zЊ@Z_AL@Qy# (77R'@Z@9a2'(7c``A@qTTB R *R]5@AЂЅ!hB&$/RR'R4[B R{ *RJ5@bЁ!c$/R'R7R4zZ_A'@Z@9a2'(6a!`A"@_qTSB Rs *R)5{BAЂc!hB$g/RRfSa4ZB RZ *R5{BBЁ!c//RSR4/a4CbB їWo<44`4'@b<'B 444'@6R`b+B ѹ4ѽ4`45R+ @@qT@ Տ?aTsJsB@T Rx@Ν. &t?4?o@@BR*24 @qU1^Aq-@?qC T@ 6ssbAf@qT\B R *R4'@AB'@!h|BBh)RRR3UB R *R4B@Ѓ7!hc) RR3v^A@@96y9cA$@qATVB R *Rl4AЂЅ!hB=)RR3YB R9 *R[4"@7!hc)|R3i(6{{cAg@qaTWB R *R@4BAB!hB5)DRRz3TB R *R.4BЁ!c)$DRm3qЉB1^Ai4'@V[ *X R9) RR{* 4@Ё!**K3'@qЉ}B1^AsB "B)T @$@9D(6|МcA@q!T*3@AbR!hBD;**#3*3@Ё!**3'@qi}B1^A|C) TB*@9ZJ(6q1bA)@?q!T*3Z@AbR!hBD;**2*3Z@Ё!**2'@qЩ}B1^A| C)T@pTjE,@9/6g`A@_qT*3@AbR!hBD;**2l'@ p@?TyJ9C@z R8@Z9.t?2?o@@BR*2*3~_A@h(6i)aA;@qT[B R{ *RI3@@ABEc!hB l dRR~2SB Rs *R23@@bЁCЅ!cl %dRl2~_A~B(4'@q+=A!=b=%= P  =A=b===`T'@ q.s / 0 13 s ?/=== ==T'@ qa b cR d e  =A=b===T'@q2r2==b=@== T'@qْR˒6 =="=@=c=`T'@q3s=====T'@qRR =="=@=c=T'@!qdef3gss?====c= T'@%qZTPY\I_@=="==#=` T'@)q*p+,-0.p?@=a==== T'@-qopqP`aD=="==#=T'@1qB1CqDEF1_@=a==== T'@5qP=a=B==#=`T'@9q1qM=!=B=`==T'@=q-Q./01Q?====#=T'@Aq`ab1cqd=!=B=`==ADT@'@N@4X 6T" *#RX Rn R*2g@[@)FkD)a!"R Z1~_A@9?T?>T;O'@P@q@kT:O'@"@C@_kT'@|BJ4KEQA1C<TH#@Lџ TNu~#@  @! AkсE Kkkkk T N@'@ Rl) Bk`TBa k  T1C k))O@9UqT(( D @)@96f`A@qT*1F#@A"R!hB!R0*1F#@ e7!h!RL 0~_A@9*/6pbA @qT*1g@K@)ABF'D)R!hB "R 0P'@}B14'@7\y** *X R! o R#R9?D}BB[ki.T@I?=q'T@ R#B@Z Pa@!**0~_A@rRT (7 (7J (7(7(7 (7/6l*aA@-1q!TGktAR!hBT#**e0*1Gkta!**[0~_A@J/6h*aA@1qATARBT#'@A!h**;0*0Cag@!**00~_A@ /6k*kaAs@0qATEAR@!hBT#**0*0Sag @!**0~_A@/6i*)aA3@0qATIAR' @!hBT#**/*0JaG @!**/~_A@/6m*aA@0qATKARg@!hBT#**/*p0Ma@!**/~_A@J/6e*`A@X0qATLAR@!hBT#**/*E0Na@!**/~_A@/6j*JaAS@-0qTOAR@!hBT#**d/*0j/6 B RR*0@Rw0.6vbA@qAT:B RZ *R0AbEC!hB Rl $R@/6B R *R/A`CED/!p*cl ;$R4/tJҔB@U&t@?4/?'@o@BR*d@a2/d @tB\A?kAT)4'@3VUBs *X R"n R(R6*/a)b!**.d@\A9CCZ_C)BT_CT0C_'@@9C/_Hqb(6d`A@qAT*/~)AbcR!hB+**.*z/a~)b!**.d@\AZ_C)T%_'@@91@H?q/6~cA@qT*S/)AbcR!hB+**.|_A@9U 6wbA@_qAT5B R *R6/AE!hB8$l d)RRq.7B R *R%/ C%7!hcl T>d)Re.1@96``A@?qT9B R9 *R /'@@bE#BP*l &|BA!h(RRC.,/@pPA)^A44B##B#)@9i6maA@qAT9B R9 *RW,Abe#!hBDځRR+6B R *RF, ce7!hclDځRi*@9 7RZ@\_ TOSWR_QgPoO{N#Mm+Lm3Km#_4@96ssbAw@3s *qTB RR,Abec!hB(܁RRU+B RR ,b c%7!hc#܁RK+ RZ@\_T.naA@?qAT;B R{ *R+Aec!hBl ہRR)+$cR<*'RB R( A`A`"`ace!c<$cR*'JJ]A=@#cc< m')ѥ#**(L]A@9g(6LaA@?qaTRB R(w)!"e#R!hBD<hR<*F'RB R'aw)"`ce!c<hR*9'[)* "R[#'JJ]A[@9{/6HaA @_q!TRB R'w)!"e#R!hBD<hR<&R'~'@?@Z@\_qT@RZ@\_pT '1' s 7'[ 0@?@_Ts 7&?#{CO`@j@_qT@ ,?"Tm@@@:@>h@YqT&@l"@.  I)]A(@9 6JJaAH@qAT3B Rs *Rs'!"ec!hBh9߁RR&4B R *Rb'a ce7!hch9;߁R,Kk]Ai@9)6LaA@?qAT3B Rs *RG'!bec!hB;h9dRR&4B R *R6'a ce7!hch9 dRv&ROC@{A#_F\A@96@3s *`A @qTB RR'!"ec!hB=h9RRO&B RR'ab ce7!hch9f"@T@?AT` @*@BR*#@@?jqT?:qT?qaTHR*2o@yx @ 2:)@>f"@T-R2#BbT@ 7@#?ToBsS@s#JR`@Ry @KCLIF@ `A`uB7=1{=o ;B TDB 1 TFS@R 1ؔBC'+/TW@TBsR@\B?CG"C?֘>@oՖ5"@?@_qAT\М_A@99 (6Y9cA8@qAT5B R *RZ#!"E!hBX4-RR"3B Rs *RI#ab`CE!c-4/R"S@7S@4Jc@ TMЭ]A@9(6A!`A;@qT3B Rs *R%#!bEcS@!hB-RR^"9B R9 *R#a"`C%!c-RQ"Cj @o ԗ @sDC RKN"`?S@-Rm^"@"*qT4VoՖҠvC59"@C?@qTTД^A@9)6ZZcAY@?qAT4B R *R"!"%!hBX4|ςRR"6B R *R"a`#E!c|4/ςR! "@R"@D QxqhT e6Pxdx ֟$n"@4Y9cS@h@ywA@?0qT?Hq4T|@{wK@Yq!4T@?iT&@m@yw@O_0q4T_XqA2T@?AiTO3Al4&@*GX3A{kbcTScA)c*#R'R4A"*O@C*R*A*!4*A!C@c\Aqb@9b(6Cc`AO@m@9CqT4R *B R:"!B%ЃR!hBdR** q!R *B R%"A`#e!cdR.** a!q@T@@@YZ@`R@`!?SZ@`R@V! ?XZ@`R@L!`?@S;CqTZqBT . @qY9_A4@94(7W@5:*Y9_A4@94/6HaA@qTR9 *B R!!B%#R!hBdR**9R !R *B R!A`#e!cdR.**-R  W@5@@@YZ@`R@ ?SZ@`R@  ?֘Z@`R@ `?@J@ @@p~@qUs@~ћӝB[c @#*R ?*5~@ @!RR?* 5~@` @ARR`?֯*oϝ5~@ @aRG@R ?*5؂@ ێk @@BR*?*4T^A@96MaA@qaTR *B R!!BEЃR!hBlDR*U R *B R !aCe7!hcDR *H ӊ@ %@͹XL @`?5R@?4T{@&@b{wU*U5S@C @T @?6TC3Ac84&@*sGD3Ak0TYsA)c*#R'R4`Af *O@C*R*iA* W X4*`AY @x@\Aq@9f(6A!`AO@ @8CqTR9 *B R !B%#R!hBdR** R9 *B R A"`#e!cdR.** q@T@@@Y|[@`R@?Sy[@`R@ ?Xs[@`R@`?x@S;CqTZq"Tl  @?qI)]A!@9(759*I)]A!@9A/6E`A@_qTR *B R1 !B%R!hBdR**<R gR *B R A`#e!cdR.**8R V5@@@Y|[@`R@V?Sy[@`R@L ?s[@`R@B`?@k@)@@*nUx@@D}ΛuB՝[s%`@#*R?*5|@`@!RR?*@5y@`@ARR ?*5s@`@aRG@R`?*5x@ ێk`@a@BR*?*4H]A@96I)aA1@?qaTR *B R!B%R!hB%R*R *B Rqa#e7!hcR0*|@ %@͹X`@?5R@q 7@q y6Y*9ccs 7k@zywUs 7S@}ө@@:ikkks 7n@}o@ililOO3A5s 7yS@}u@@jhdjhOC3A5*1$Ձ@ ?zTw@[VH;@q*T@ X?zT^Hz4x @CЄ@c@BR*\@9qa]TKk]Aw@9b(7O$z"@4W[|R6b{Cr *R~*B R2@`A!$ւRl5S@.R*f @)35S@Z3!@_ bT@? TS@*@BR*` @@@:q`TqTjqaT'R*GRgR@2?TI97@\A @9(6Cc`Ag@qT*B R2@!"R!hBd؂R*B R2@`AE!d؂R4@?A Tj@`@S}Ӄ2@Aishs&@M*45s 7s 7I96C2@c\Ao@4(6D`A@qT*B RP2@!"R!hB$ׂR*B RA2@`AE!$ׂR4b"@G@?Tf@@}b@2@h`Bh`&@*4/6Q1bA%@qT*B R2@!BR!hB,$ւRU0s 7as 7[$Ձ@ \?AXT` @aR@Crc@BR*` @q,T4iqT5R2@?T9$Y9_A<@9<6\cA@qT*@!!hBtR|dRB R *R@@ED/!p*c|t0dR-ROq3T9qToR2@?`Ts 72@?TUlRyJv @>C@3b0RBt6?4V^A@9. 6KkaAe@qTB R *R!Bfg!hB6xRRB R{ *RkabEfg7!hcxRAC@c@BR*A2!@ qT q@T q!T2@1RJx_q.*6y"@4 *X R~R qc&@*35S@{3L!@ bT@?aTI9 7Ss^As@9(6ZZcAH@qT*F)!BR!hB?D߂RS* AЇF)`e!D߂RI@?ATS@}ӅF)@@xqajkhkTq@Ts 7@I9J6Q6@1^AS@.@K4(6OaA@q!T*W@!BR!hB?GF)܂R*W@`Ae!F)܂R@}Ӄ@hshhs@ @kW@BTM]A@9h6MaA@qT*!УR!hB'$݂R*aE7!h$݂R/@?! TS@&@I@M@!isis@? TO@P@&@isjsJ#@Q?@?TD@N@F)hsis,yqTqTCR&@*34(W@(6HaA@qT*TF)!BУR!hBdۂR*EAЇF)`!dۂR`&S@0R* @0)s4s 72s 7~s 7xs 7rFR2@? TfA!\A9@9(78RO6ROA!`A7@qATB R *R!BE!hBD T-dRR9B R *RA`Ce!cT-dR-#ROY9cA(@qaTB R{ *R!BEc!hB(;T-佂RR B R *RAЂ`Ce!cT-佂ROE\A@96OaA@qaTB RZ *R!C!hB #1RR&RB R *RaE7!hc1l8R/H]A@9{73RHPbA@qaTB Rs *Rt!"c!hBD$䞁RR&RB R *RbaE7!hcD$䞁R&R3Rs 7u@p"@05 s 7@ts 7,s 7:s 7^H5 RO@ ձ?Tow @C#c0@ Rk@-RS wAVR {=k= _?@ [@ @? qT~B4 *<RR}R*[@+41@4vW@ s *ۄR?7)7 )RB R[@A)b`:CAЦ @E @!*@* [@,1@9? T(@yy:CQPqTjf8k `֟$?*$*$BB\AE@9/6LaA@qTRB R[@A)!:Cc& @R @!hB4** $3*w$/*s$*k$*o*B Rs@`AE!*%~B*{ sBC(TT@HkTH_qTD\A@9(6HaA @qT*B RH@!bR!hB**B R9@`AE!*%y~B*m@A(T@H?kTHqT.]A@9/6-aA@qT*B R@BR!hB*MW@?k )k),!@4#@*+@*pXm:C?q  kT+@*3RO@v"2cq0[@z0#@T(3qAT%\A@@9(6<cA@qTB R *R[@@BУ!h @BD0$RRB Rs *R[@@b@AE' @!c<$R~B'4Bk'T* !!\A(@9&(65 bA@?qTB R *R!hBXRR**B R9 *R{A"@%!c<R***(qTH[@@_qTo@1XTk@1aXT&+\A@9(7@O ) d?aTI96[@72@4*6R(R $"@JZ{3#2@_T!@7xz:C9q"l @zT1T9@@,@a<{s? @!TzB6kT[@O@a @%)<_A@946<cA@W@qTR *B R@:CBЃR!hBT1R*FR *B R@:CAE7!hcR>*6s 7<(RzBj7O@7W@ V!@?!T '@CO@;(Rs 7$<cA@qTB R *R[@@"Ѓ!h @B ?DRRB R *R[@@@AE @!c'DR@O ) ճM?Ts 7I9[7@ .L?T M)** /@@:-Z:D1qA* sN)1@/T?@:D!->D@{@KquqAQT[@OJ_UT\~Ӌ,z!@4O@ @*4ye@:C?ko!@Tz !@ 4O@@*yv @g:C kp!@Tz !@O@4@*yu @s;Ckq !@Tz !@4@*z|@8Ckb  !@ Tz!@W@4@*gzx@8Cke!@Tz!@4@*&{w@8Ckx!@Tz!@4@*"zmP8C@kc!@T *GZ_W4T 1@4y91kz8 *:C@{zA^A)E@!#c䜂R<:3;+ @[@s2 1@ T @Uys:CkT{z1T @W@ @`yt@@?T{BL 6k T[@"Rh!@)W@N,]AW@@9, 6.aA@_qTR{ *B R@:CcR!hB0DR*W@R *B Rk@:CA7!hcDR =*s 7@{B 7W@7@ $?! T&@;@"B\AE@9e/6&`A@q!TR *B R5O@G@J@xz  *[@"A)B=@@{z:C @3;!h+R䜂R# YR9 *B R [@@"O@'@ G@@.yz6s 7U>D!@! 1*I1 4VHqD qaT[@@J[wjK[|++@G @s @@!~t!~!~14@"5 uT6O@ւ*c 3@*4VHq6a4O@ `5'\A@9a(611bA @qT*B!hB'ıRR*@A!cıRZ:Z_AZ@9Z(69*9cA%@qTB R *R[@@B!hB0 g@RRB R{ *Rv[@@b@AE @!cR[@**1@j4$@**+RKh_+k*TeQ@xh9C4QqTiD)K<R\qaTO@^4!!* !!* *S@;R7 @;]"@34))]A7@97<RZP6D*V P7C@@]AT* OFWE_DgCoB{A#l#_6bAS@@?qTB R *R  @B!hB |$RR'RBB R *R @A7!hc|$R'R<R3 W6C@@]T\BkACTb <R;8_A@9(6$Є`A@qT*"!hB RRB Rs *RAb@!c$>R!*> VH?q4O@ 4;{_AO@m@9(6.aA@_qT*ڸB!hBRRB RZ *RAB@!cR!*7**K@ ?ZT****U*&RRGG@?@ G@99R?;q  d@zT[@h+Q#0@qEzb TG@Q T?@w@9:qT*R ? j`T[@0@4*Q?qT:R [@%#@ Ճf)[@K 0@_:k * T[Q@t@9xt$;CkT@"W@B`3@x|@v@?TzBi6k T7W@ a@?aT`&@y;&RRs 7ZR&RzB7s 7P9R [@!@ Ջf)[@)70@9k *IT:Q3@T@zztE;CkT"@+W@kaS@Wx|@@? TzBk6k T7W@ ժ@? aT&@8&RR;s 7R&RzB7s 7W@WI[`RX7R;r-O@ s @9Bq"  @zT[@1@C4*QqT*:R1@CZkTEQ@@ yi9CkTFDK@[@*:R1@CZk TOQ@@zq8CkT:D[@R`T@9BqT"_jT[@1@.4*QqT*:R1@CZ_kTJQ@L@Kxly9C?kTFDK@[@*9R1@#9?kT.Q@@xg:CW@_kT:DO@[@\`RZ3R<rZ-ZVTY@9?Cqg" Ę@z`T[@0@4*8R0@k T Q@.@yn9CkT:DO@[@X@9CqTy"?jO@`T[@0@4*;R0@c{_k TpQ@@&zd8C?kT:D{[@[ @9Cq(Tk"jT[@L0@l4*9RH0@#9kiT8QG@@xtT9CkW@T:DO@Z[@ q@kT2CkT \R[@5+@{@@h[wi[|!~ @$ @^A#!~!~@H#I% (6-ЭaA@?qT*B!hB䯂RRB R *RA@%!c䯂R*J]AO@@@!) T (6/aA@qT*B!hB;dRR*@AE!cdR**_7^Aw@@9/6'`A@_q!T*[@!hB @丂RW@R*[@@AE!F @c|<丂R|O@9*/@* (60bA@?qT*]B!hB(6RRS*P@AE!c6RK**4  q RTC?kT"|RTq QTCkTW@RKqaTW@~W@9DkTW@R?qTW@DkTW@RB5q`TW@D?kTW@R+!q TW@.BEkTW@R!%qTW@EkaTW@<Rbs 7)qTW@EkaTW@\Rs 7-qTW@HF?kaTW@|R"s 7&1qTW@F_kTW@R5qTW@FkTW@R9q`TW@tSGkTW@RB=q TW@GkTW@RW@Aq?T@R?#{ O@j@_aqT@ Ո?T-Э]A@9*(7*OB @{è#_))]A(@97ROB @{è#_+k]Ai@9 7ROB @{è#_/]A@9k60 *bA @qATB RR!f@"E!hBDL.DRRRZB RRf@AC7!hcL.\3DRRN ROB @{è#_*JaAH@qaTB Rs *REc!hBX$L.RR+B R *RACE7!hcL.R*,aA@?qaTB R *Rf@E!hB4 L.dRRB R *Rf@AC7!hcL.2dRROB @{è#_.aA@qaTB R *Rf @BEЃ!hB RRB R *Rf @@!CE!c R*OB @{è#_s 7@?Ts 7?#{ O5^A @@i h(6+ *kaAj@_qATB RRKg@%Ѓ!hB5)RRB RR:g@@!#E!c)$DR @c KqT;7R{_Ak@9+ (7|@8IqTy6@iw?ց@ @?T'@5;*{_Ak@9(7b@?T;Iq{_AaT7Rk@9(7Ws 7= s 77 s 71 @;Iq{_ATy6@7jw? @*k@9(6!!`A*@_qTB RR @g:C:C%!h@@RRJ@"B$ B RR @@:C@!#@g:C!c(R @ w4n@9.6#c`Ak@qTB RR f:Cg@:C"@%!hB-DRR  B RRs f:Cg@:CЗ@A#7!hc<DR  OFWE_DgCoB{A#_?#Ck#m{og_WOT;Չ@*0A+*D4$@ &"s*ӝG @1/BLT@i1KT`#+@*CR'R }Su-}S (*uMc*R*DA* F5@3:'#<[ @*R@? 4*<R'@$\A@9 7x@G @c!RR`? 4*\R'@$\A@9.7e@5 @AR@R?@4'@*|R$\A@97R@" @aR@R? 4*R'@$\A@9n7?@ @сRR`?-4*R'@$\A@960bA @?qTR *B R ER!hB,MR!** R *B R AC%7!hc!MR** '\A@9o7<R-/aA@_qaTR *B Rl BER!hB*OR!* R *B RZ AC%7!hc!OR5* <R@E'5@'@ q TqaT+@k@ qT'R+@L@q`T*A A "A ~SdS~SFS *|5@ %@͹Xy @?GA @\Aq@9(6 aA:C@qTR *B R "R!hBdR** > R *B R ! %!cdR.** . qT?@Y[@`R@- ?S[@`R@#  ?T[@`R@ `?+@@9C?qaTZqBT  @qB\AQ@9(6aA@qTR{ *B R "cR!hBdR**;R R *B R ! %!cdR.**(R sӝ5 *B\AQ@9/7sӝV5Y@V[@`R@?'@V[@`R@`?Y[@`R@@?G@*5+@+1A kT;@ѣC;@ ێk@@BR*?'@*5**@]_ T*k@OTWS_RgQoP{O#NmC#_s^As@97@;R@]_ `T@*@]_ T 1bA@0@qTB R *RB%G@!hB p&RR&R0B R *R!#G@7!hcp&,;R&R;R!@]_ T9@R3 _$$@2?#3 m+m#m{og_WOW;@*bL@ @*a@I@?9YhA T@_IqTa@;Y{? T)@?Iq Tc\Aa@9a`6`A@qATB R *R:C@<;C5@"%!hB-dRR B R *Ry:C@6;C9@!#7!hc<dR @\? YTs 7@l@@@_IqTs 7)@?IqT@ qHT+3AN4 &x&@_*W3{ @VA)#ѥ6kA)t@*CR'R;:9R9]5cѥ*CR'R9RR B4;#A##C#d##!I(-J# $L%y+5;@P3A +@AA ;@bA A $A **** 4!\A-@9-6aA@qATR9 *B R"E#R!hB:;Rd&Rs *B R!bC7!hcd;R)9RXG7Y57@'@@ qTFqT#@-@n@ q qT*XG74y 4_*\F#@@P@qqT*R**A AA bA ~SA ES$A fS~SS~S # 4\A'@@906bA@?qATR *B R|"ER!hBH6ARdR *B Rk!C%7!hcdAR47_*@ %@͹X@@?*&R_'@A@9_Aq1@93@Q(6 aA:C@qTR *B R9"R!hBdR** pR *B R$! %!cdR.** `q T[;@\[@`R@``?S[@`R@V?Z[@`R@L`?@3;CqTZqB,T  @?qB\AB@9B(6aA@_qTR{ *B R"cR!hBdR**:R R *B R! %!cdR.**;R 95*B\AB@9/7Y5[+@\[@`R@ ?7@Y[@`R@@?[[@`R@?#@@t;CqTZqT  @q\A@9C(6 aA@qTR9 *B Ro"#R!hBdR**)R R *B RY! %!cdR.***R Z5*\A@9/7z5Z@|[@`R@`?#@@Z[@`R@`?Y[@`R@x@?_@q 63@,3A kCTu'@Cѥ*R*8Y*;^A@96 *aA@qTB RR"%!hB$7dGRR2B RR!Т#7!hcdGR(*R\A@9`79*7_B*6R\A@96bA@qTB R *R%!hBdGRR**B R9 *R!"#%7!hcdGR**_B* @6@\ 롦TOUWT_SgRoQ{P#Om+Nm3Mm#_@R?#{og_WO @Ja&j`."h6D(m>D @]Ҡ^:@`6Da>D?ֹ@ Ҁ a`6D ?6>  Ҁ ax}C`6D?6i2D)n:D @]Ҡ^:@`2Da:D?ֶ@ Ҁ a`2D?2:  Ҁ a}C`2D ?2jBDt" oFD @]Ҡ^:@s`BDaFD?ַ@ Ҁ ai`BD?֟  Ҁ a}C\@?֟kJDtB" pND @]Ҡ^:@M`JDaND ?ָ@ Ҁ aC`JD?֟  Ҁ a}C6@?֟hV@_AqaTt.A{ TւһvRVB@9@(6`AB@&@@?qTR *B RR!hB8DR0/*R *B R! %!c0/DRp"*_z@@?w@?x@?։@ )@  AQ@?T@ AD@aTe.AATz.p6A #,T`2Al6fR_T`"Ay v`*Ay&"` *f@vv@ %@͹X`@?։@9I(6* *1bA*@_q!TB RR8"%!hB 7$RR*rB RR(! #%!c 78$R*gg@yn@@ f ,i`@ ?֊@9(6*`A@qaTB RZ *R"%C!hB` 7RR*;B R9 *R!" #!c 7L9R*.a@az@ % 7 n0`@@?x@xw@ % 7 n$`@?v@vx@ % 7 n`@?y@yv@ % 7 n `@?z@zy@ % 7 n`@ ?w@wz@ % 7 n`@@?c@cw@ % 7 n`@?b@bx@ % 7 n`@?`A`v@ % 7 n`A?|A|y@`R@`A ?h Ahz@`P` A@? i@iw@`|=`@?`^@  ^4@96*`A@qaTB R *R=%!hB: 7RR*wB R *R+!#7!hc 7$R*j`2AR%`"A7lH9L7OGWF_EgDoC{B#_`AOGWF_EgDoC{B#?#{ w @{¨#_$?#{ @`@ @{¨#__$`?#{ @`@h@9h6`@ @{¨#_?#{#OT;Ո@fA@@5o3s*(q*`qT* I@a!XAI*R1T{*@lq TqTLqT`qT` QqT qT@R 9RpqaT@BR@_? TOE#@{C#__$?(`9T*R ? T(@ a@9 5 @/@@Toa@9-5? a90a9a90@ b@95? a9,a9 b9 @`T< @_@T @D@d @I  @)@? T*%@e @) @M@TA*H K @.RNa9m@a9@fFh @ ` @ @ T m _ m _A( ,@/R/`9@a9@G @ @p@ `Tnn __$)@(@h@I@+RG @@ !T_ T@La@9_TH @*( G @@ TLa@9_ T-@M %@@T'@@GAP -`@9@Ma9 4 4)R @ a@9l@T 4 @-j4@Ec@ a9a9q@qQ+ j @jl @@ T@ a9a9bBKp @ n @@? Tk@c (@ @la@9.4@oa@94 @a9_TOa@9o4H @@ (@b@ h @na@9 4 @ma@94 @a9`@94T@ @_(@(R(a9_֪'@@GA La@9_T 1R`9_@oa@94@!Ra9a9BH @ @@T )R`9_na@94@+Ra9a9D @ @d@Tm#$  @%R@b@9@a9b9a9Cp @ @ @T$ _$ _m @*R%@,b@9@a9*b9a9&F& @/ @@T% _% _?#{@9qT{#_@RA_${vA   zA   ~A   A   A"   AB   Ab   A   A   A   A   A   A"   AB   Ab   A   A   A   A   A   A"   AB   Ab   A   A   A   A   A   A"   AB   Ab   A   A   A   A   B   B"    BB   Bb   B   B   B   B   "B   &B"   *BB   .Bb   2B   6B   :B   >B   BB   FB"   JBB   NBb   RB   VB   ZB   ^B   bB   blD      & . 6o`p`2 oo o؂p8 $ Jo8(oX o0o`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`u`ux'@=ؔ.note.gnu.property.init_array.fini_array.text.got.note.android.ident.got.plt.rela.plt.bss.dynstr.eh_frame_hdr.gnu.version_r.data.rel.ro.relr.dyn.rela.dyn.gnu.version.dynsym.gnu.hash.eh_frame.note.gnu.build-id.dynamic.shstrtab.rodata.gnu_debugdata.data7zXZi"6!|g]?Eh=(aK6ER%_2:J|%$X0u EyD\u ,m qչ.Nvj0Ǽt\\/jL_d#VSQIײaA?yYij|8+*i:Ǜ'/1諈q&}CP$e%O[8L~Dδ iJ]1&l&&VKḒ\nsǎqNB٠u:qSJ?F+z$睔Q(W@4؇ΆIURvF&,<*J?u.,w=_K'mhMr/…Eq:˹LHp#P&䶎GQ;N1(;e࢜l=$WBesWF؆ߎN[=)@3XYP伄2So~FE"z!'VHf}Wpu CYt,<wK*QGmp!RsdZ\"3ڝc&FltvH$~ߐPjB2S{CaZ4tQ]r|{}wqdUy՝Rq)܋{2j:}9IaE"_B˄pFkd|V[? 0– thRxdGԢε"F$ԹSbYJ??E(UsЧWu .xU18ù+;"(<@%7 0&W>0 YZ7   88 oX X yo00`oc$$J`pp2o TB2P$P$²k,@@d, `Y`u`u( ((88@@p2(K؂؂ȔȄ^8