From ec2d235c6927d5e322ad7077a02845f626e00287 Mon Sep 17 00:00:00 2001 From: ZwX1616 Date: Fri, 8 Mar 2024 16:43:18 -0800 Subject: [PATCH] camerad: OS04C10 operational (#31674) * it's something * backup * 16:10 * cleanup * this is fine * close * remove some junk * no heck * disos * real 10 * for some reason this is flipped * 20hz * no return * ae * tear * need curve laster * correct real gains * fix time * cleanup * why the scam * disable for now * 0.7 * hdr * that doesnt work * what * hugeoof * clean up * cleanup * fix regs * welp cant * is this corrent * it is sq * remove * back * add base comment * clean up * make orders clear * not abcd --------- Co-authored-by: Comma Device old-commit-hash: 7eb1e958848c6fdb0865da95eaba991252ca9fd6 --- system/camerad/cameras/camera_common.cc | 8 +- system/camerad/cameras/real_debayer.cl | 175 ++++++++++++++------- system/camerad/sensors/ar0231.cc | 8 +- system/camerad/sensors/os04c10.cc | 53 +++---- system/camerad/sensors/os04c10_registers.h | 155 +++++++++--------- system/camerad/sensors/ox03c10.cc | 6 +- system/camerad/sensors/sensor.h | 4 - 7 files changed, 235 insertions(+), 174 deletions(-) diff --git a/system/camerad/cameras/camera_common.cc b/system/camerad/cameras/camera_common.cc index fff4a62d31..aa815bfadf 100644 --- a/system/camerad/cameras/camera_common.cc +++ b/system/camerad/cameras/camera_common.cc @@ -27,14 +27,18 @@ public: "-cl-fast-relaxed-math -cl-denorms-are-zero " "-DFRAME_WIDTH=%d -DFRAME_HEIGHT=%d -DFRAME_STRIDE=%d -DFRAME_OFFSET=%d " "-DRGB_WIDTH=%d -DRGB_HEIGHT=%d -DYUV_STRIDE=%d -DUV_OFFSET=%d " - "-DIS_OX=%d -DCAM_NUM=%d%s", + "-DIS_OX=%d -DIS_OS=%d -DIS_BGGR=%d -DCAM_NUM=%d%s", ci->frame_width, ci->frame_height, ci->frame_stride, ci->frame_offset, b->rgb_width, b->rgb_height, buf_width, uv_offset, - ci->image_sensor == cereal::FrameData::ImageSensor::OX03C10, s->camera_num, s->camera_num==1 ? " -DVIGNETTING" : ""); + ci->image_sensor == cereal::FrameData::ImageSensor::OX03C10, + ci->image_sensor == cereal::FrameData::ImageSensor::OS04C10, + ci->image_sensor == cereal::FrameData::ImageSensor::OS04C10, + s->camera_num, s->camera_num==1 ? " -DVIGNETTING" : ""); const char *cl_file = "cameras/real_debayer.cl"; cl_program prg_debayer = cl_program_from_file(context, device_id, cl_file, args); krnl_ = CL_CHECK_ERR(clCreateKernel(prg_debayer, "debayer10", &err)); CL_CHECK(clReleaseProgram(prg_debayer)); + } void queue(cl_command_queue q, cl_mem cam_buf_cl, cl_mem buf_cl, int width, int height, cl_event *debayer_event) { diff --git a/system/camerad/cameras/real_debayer.cl b/system/camerad/cameras/real_debayer.cl index e15a873d6d..5f8d046cb5 100644 --- a/system/camerad/cameras/real_debayer.cl +++ b/system/camerad/cameras/real_debayer.cl @@ -8,7 +8,7 @@ float3 color_correct(float3 rgb) { // color correction - #if IS_OX + #if IS_OX | IS_OS float3 x = rgb.x * (float3)(1.5664815 , -0.29808738, -0.03973474); x += rgb.y * (float3)(-0.48672447, 1.41914433, -0.40295248); x += rgb.z * (float3)(-0.07975703, -0.12105695, 1.44268722); @@ -20,6 +20,8 @@ float3 color_correct(float3 rgb) { #if IS_OX return -0.507089*exp(-12.54124638*x)+0.9655*powr(x,0.5)-0.472597*x+0.507089; + #elif IS_OS + return powr(x,0.7); #else // tone mapping params const float gamma_k = 0.75; @@ -35,6 +37,9 @@ float3 color_correct(float3 rgb) { } float get_vignetting_s(float r) { + #if IS_OS + r = r / 2.2545f; + #endif if (r < 62500) { return (1.0f + 0.0000008f*r); } else if (r < 490000) { @@ -85,6 +90,24 @@ float4 val4_from_12(uchar8 pvs, float gain) { } +float4 val4_from_10(uchar8 pvs, uchar ext, bool aligned, float gain) { + uint4 parsed; + if (aligned) { + parsed = (uint4)(((uint)pvs.s0 << 2) + (pvs.s1 & 0b00000011), + ((uint)pvs.s2 << 2) + ((pvs.s6 & 0b11000000) / 64), + ((uint)pvs.s3 << 2) + ((pvs.s6 & 0b00110000) / 16), + ((uint)pvs.s4 << 2) + ((pvs.s6 & 0b00001100) / 4)); + } else { + parsed = (uint4)(((uint)pvs.s0 << 2) + ((pvs.s3 & 0b00110000) / 16), + ((uint)pvs.s1 << 2) + ((pvs.s3 & 0b00001100) / 4), + ((uint)pvs.s2 << 2) + ((pvs.s3 & 0b00000011)), + ((uint)pvs.s4 << 2) + ((ext & 0b11000000) / 64)); + } + + float4 pv = convert_float4(parsed) / 1024.0; + return clamp(pv*gain, 0.0, 1.0); +} + float get_k(float a, float b, float c, float d) { return 2.0 - (fabs(a - b) + fabs(c - d)); } @@ -94,20 +117,51 @@ __kernel void debayer10(const __global uchar * in, __global uchar * out) const int gid_x = get_global_id(0); const int gid_y = get_global_id(1); - const int y_top_mod = (gid_y == 0) ? 2: 0; - const int y_bot_mod = (gid_y == (RGB_HEIGHT/2 - 1)) ? 1: 3; + const int row_before_offset = (gid_y == 0) ? 2 : 0; + const int row_after_offset = (gid_y == (RGB_HEIGHT/2 - 1)) ? 1 : 3; float3 rgb; uchar3 rgb_out[4]; - int start = (2 * gid_y - 1) * FRAME_STRIDE + (3 * gid_x - 2) + (FRAME_STRIDE * FRAME_OFFSET); + #if IS_BGGR + constant int row_read_order[] = {3, 2, 1, 0}; + constant int rgb_write_order[] = {2, 3, 0, 1}; + #else + constant int row_read_order[] = {0, 1, 2, 3}; + constant int rgb_write_order[] = {0, 1, 2, 3}; + #endif + + int start_idx; + #if IS_10BIT + bool aligned10; + if (gid_x % 2 == 0) { + aligned10 = true; + start_idx = (2 * gid_y - 1) * FRAME_STRIDE + (5 * gid_x / 2 - 2) + (FRAME_STRIDE * FRAME_OFFSET); + } else { + aligned10 = false; + start_idx = (2 * gid_y - 1) * FRAME_STRIDE + (5 * (gid_x - 1) / 2 + 1) + (FRAME_STRIDE * FRAME_OFFSET); + } + #else + start_idx = (2 * gid_y - 1) * FRAME_STRIDE + (3 * gid_x - 2) + (FRAME_STRIDE * FRAME_OFFSET); + #endif // read in 8x4 chars uchar8 dat[4]; - dat[0] = vload8(0, in + start + FRAME_STRIDE*y_top_mod); - dat[1] = vload8(0, in + start + FRAME_STRIDE*1); - dat[2] = vload8(0, in + start + FRAME_STRIDE*2); - dat[3] = vload8(0, in + start + FRAME_STRIDE*y_bot_mod); + dat[0] = vload8(0, in + start_idx + FRAME_STRIDE*row_before_offset); + dat[1] = vload8(0, in + start_idx + FRAME_STRIDE*1); + dat[2] = vload8(0, in + start_idx + FRAME_STRIDE*2); + dat[3] = vload8(0, in + start_idx + FRAME_STRIDE*row_after_offset); + + // need extra bit for 10-bit + #if IS_10BIT + uchar extra[4]; + if (!aligned10) { + extra[0] = in[start_idx + FRAME_STRIDE*row_before_offset + 8]; + extra[1] = in[start_idx + FRAME_STRIDE*1 + 8]; + extra[2] = in[start_idx + FRAME_STRIDE*2 + 8]; + extra[3] = in[start_idx + FRAME_STRIDE*row_after_offset + 8]; + } + #endif // correct vignetting #if VIGNETTING @@ -118,60 +172,69 @@ __kernel void debayer10(const __global uchar * in, __global uchar * out) const float gain = 1.0; #endif - // process them to floats - float4 va = val4_from_12(dat[0], gain); - float4 vb = val4_from_12(dat[1], gain); - float4 vc = val4_from_12(dat[2], gain); - float4 vd = val4_from_12(dat[3], gain); + float4 v_rows[4]; + // parse into floats + #if IS_10BIT + v_rows[row_read_order[0]] = val4_from_10(dat[0], extra[0], aligned10, 1.0); + v_rows[row_read_order[1]] = val4_from_10(dat[1], extra[1], aligned10, 1.0); + v_rows[row_read_order[2]] = val4_from_10(dat[2], extra[2], aligned10, 1.0); + v_rows[row_read_order[3]] = val4_from_10(dat[3], extra[3], aligned10, 1.0); + #else + v_rows[row_read_order[0]] = val4_from_12(dat[0], gain); + v_rows[row_read_order[1]] = val4_from_12(dat[1], gain); + v_rows[row_read_order[2]] = val4_from_12(dat[2], gain); + v_rows[row_read_order[3]] = val4_from_12(dat[3], gain); + #endif + // mirror padding if (gid_x == 0) { - va.s0 = va.s2; - vb.s0 = vb.s2; - vc.s0 = vc.s2; - vd.s0 = vd.s2; + v_rows[0].s0 = v_rows[0].s2; + v_rows[1].s0 = v_rows[1].s2; + v_rows[2].s0 = v_rows[2].s2; + v_rows[3].s0 = v_rows[3].s2; } else if (gid_x == RGB_WIDTH/2 - 1) { - va.s3 = va.s1; - vb.s3 = vb.s1; - vc.s3 = vc.s1; - vd.s3 = vd.s1; + v_rows[0].s3 = v_rows[0].s1; + v_rows[1].s3 = v_rows[1].s1; + v_rows[2].s3 = v_rows[2].s1; + v_rows[3].s3 = v_rows[3].s1; } // a simplified version of https://opensignalprocessingjournal.com/contents/volumes/V6/TOSIGPJ-6-1/TOSIGPJ-6-1.pdf - const float k01 = get_k(va.s0, vb.s1, va.s2, vb.s1); - const float k02 = get_k(va.s2, vb.s1, vc.s2, vb.s1); - const float k03 = get_k(vc.s0, vb.s1, vc.s2, vb.s1); - const float k04 = get_k(va.s0, vb.s1, vc.s0, vb.s1); - rgb.x = (k02*vb.s2+k04*vb.s0)/(k02+k04); // R_G1 - rgb.y = vb.s1; // G1(R) - rgb.z = (k01*va.s1+k03*vc.s1)/(k01+k03); // B_G1 - rgb_out[0] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); - - const float k11 = get_k(va.s1, vc.s1, va.s3, vc.s3); - const float k12 = get_k(va.s2, vb.s1, vb.s3, vc.s2); - const float k13 = get_k(va.s1, va.s3, vc.s1, vc.s3); - const float k14 = get_k(va.s2, vb.s3, vc.s2, vb.s1); - rgb.x = vb.s2; // R - rgb.y = (k11*(va.s2+vc.s2)*0.5+k13*(vb.s3+vb.s1)*0.5)/(k11+k13); // G_R - rgb.z = (k12*(va.s3+vc.s1)*0.5+k14*(va.s1+vc.s3)*0.5)/(k12+k14); // B_R - rgb_out[1] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); - - const float k21 = get_k(vb.s0, vd.s0, vb.s2, vd.s2); - const float k22 = get_k(vb.s1, vc.s0, vc.s2, vd.s1); - const float k23 = get_k(vb.s0, vb.s2, vd.s0, vd.s2); - const float k24 = get_k(vb.s1, vc.s2, vd.s1, vc.s0); - rgb.x = (k22*(vb.s2+vd.s0)*0.5+k24*(vb.s0+vd.s2)*0.5)/(k22+k24); // R_B - rgb.y = (k21*(vb.s1+vd.s1)*0.5+k23*(vc.s2+vc.s0)*0.5)/(k21+k23); // G_B - rgb.z = vc.s1; // B - rgb_out[2] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); - - const float k31 = get_k(vb.s1, vc.s2, vb.s3, vc.s2); - const float k32 = get_k(vb.s3, vc.s2, vd.s3, vc.s2); - const float k33 = get_k(vd.s1, vc.s2, vd.s3, vc.s2); - const float k34 = get_k(vb.s1, vc.s2, vd.s1, vc.s2); - rgb.x = (k31*vb.s2+k33*vd.s2)/(k31+k33); // R_G2 - rgb.y = vc.s2; // G2(B) - rgb.z = (k32*vc.s3+k34*vc.s1)/(k32+k34); // B_G2 - rgb_out[3] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); + const float k01 = get_k(v_rows[0].s0, v_rows[1].s1, v_rows[0].s2, v_rows[1].s1); + const float k02 = get_k(v_rows[0].s2, v_rows[1].s1, v_rows[2].s2, v_rows[1].s1); + const float k03 = get_k(v_rows[2].s0, v_rows[1].s1, v_rows[2].s2, v_rows[1].s1); + const float k04 = get_k(v_rows[0].s0, v_rows[1].s1, v_rows[2].s0, v_rows[1].s1); + rgb.x = (k02*v_rows[1].s2+k04*v_rows[1].s0)/(k02+k04); // R_G1 + rgb.y = v_rows[1].s1; // G1(R) + rgb.z = (k01*v_rows[0].s1+k03*v_rows[2].s1)/(k01+k03); // B_G1 + rgb_out[rgb_write_order[0]] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); + + const float k11 = get_k(v_rows[0].s1, v_rows[2].s1, v_rows[0].s3, v_rows[2].s3); + const float k12 = get_k(v_rows[0].s2, v_rows[1].s1, v_rows[1].s3, v_rows[2].s2); + const float k13 = get_k(v_rows[0].s1, v_rows[0].s3, v_rows[2].s1, v_rows[2].s3); + const float k14 = get_k(v_rows[0].s2, v_rows[1].s3, v_rows[2].s2, v_rows[1].s1); + rgb.x = v_rows[1].s2; // R + rgb.y = (k11*(v_rows[0].s2+v_rows[2].s2)*0.5+k13*(v_rows[1].s3+v_rows[1].s1)*0.5)/(k11+k13); // G_R + rgb.z = (k12*(v_rows[0].s3+v_rows[2].s1)*0.5+k14*(v_rows[0].s1+v_rows[2].s3)*0.5)/(k12+k14); // B_R + rgb_out[rgb_write_order[1]] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); + + const float k21 = get_k(v_rows[1].s0, v_rows[3].s0, v_rows[1].s2, v_rows[3].s2); + const float k22 = get_k(v_rows[1].s1, v_rows[2].s0, v_rows[2].s2, v_rows[3].s1); + const float k23 = get_k(v_rows[1].s0, v_rows[1].s2, v_rows[3].s0, v_rows[3].s2); + const float k24 = get_k(v_rows[1].s1, v_rows[2].s2, v_rows[3].s1, v_rows[2].s0); + rgb.x = (k22*(v_rows[1].s2+v_rows[3].s0)*0.5+k24*(v_rows[1].s0+v_rows[3].s2)*0.5)/(k22+k24); // R_B + rgb.y = (k21*(v_rows[1].s1+v_rows[3].s1)*0.5+k23*(v_rows[2].s2+v_rows[2].s0)*0.5)/(k21+k23); // G_B + rgb.z = v_rows[2].s1; // B + rgb_out[rgb_write_order[2]] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); + + const float k31 = get_k(v_rows[1].s1, v_rows[2].s2, v_rows[1].s3, v_rows[2].s2); + const float k32 = get_k(v_rows[1].s3, v_rows[2].s2, v_rows[3].s3, v_rows[2].s2); + const float k33 = get_k(v_rows[3].s1, v_rows[2].s2, v_rows[3].s3, v_rows[2].s2); + const float k34 = get_k(v_rows[1].s1, v_rows[2].s2, v_rows[3].s1, v_rows[2].s2); + rgb.x = (k31*v_rows[1].s2+k33*v_rows[3].s2)/(k31+k33); // R_G2 + rgb.y = v_rows[2].s2; // G2(B) + rgb.z = (k32*v_rows[2].s3+k34*v_rows[2].s1)/(k32+k34); // B_G2 + rgb_out[rgb_write_order[3]] = convert_uchar3_sat(color_correct(clamp(rgb, 0.0, 1.0)) * 255.0); // write ys uchar2 yy = (uchar2)( diff --git a/system/camerad/sensors/ar0231.cc b/system/camerad/sensors/ar0231.cc index 1ca4b3f1ad..5c4934fb61 100644 --- a/system/camerad/sensors/ar0231.cc +++ b/system/camerad/sensors/ar0231.cc @@ -80,14 +80,14 @@ float ar0231_parse_temp_sensor(uint16_t calib1, uint16_t calib2, uint16_t data_r AR0231::AR0231() { image_sensor = cereal::FrameData::ImageSensor::AR0231; data_word = true; - frame_width = FRAME_WIDTH; - frame_height = FRAME_HEIGHT; - frame_stride = FRAME_STRIDE; + frame_width = 1928; + frame_height = 1208; + frame_stride = (frame_width * 12 / 8) + 4; extra_height = AR0231_REGISTERS_HEIGHT + AR0231_STATS_HEIGHT; registers_offset = 0; frame_offset = AR0231_REGISTERS_HEIGHT; - stats_offset = AR0231_REGISTERS_HEIGHT + FRAME_HEIGHT; + stats_offset = AR0231_REGISTERS_HEIGHT + frame_height; start_reg_array.assign(std::begin(start_reg_array_ar0231), std::end(start_reg_array_ar0231)); init_reg_array.assign(std::begin(init_array_ar0231), std::end(init_array_ar0231)); diff --git a/system/camerad/sensors/os04c10.cc b/system/camerad/sensors/os04c10.cc index 449e06be83..aaef9986b5 100644 --- a/system/camerad/sensors/os04c10.cc +++ b/system/camerad/sensors/os04c10.cc @@ -10,14 +10,11 @@ const float sensor_analog_gains_OS04C10[] = { 10.5, 11.0, 11.5, 12.0, 12.5, 13.0, 13.5, 14.0, 14.5, 15.0, 15.5}; const uint32_t os04c10_analog_gains_reg[] = { - 0x100, 0x110, 0x120, 0x130, 0x140, 0x150, 0x160, 0x170, 0x180, 0x190, 0x1B0, - 0x1D0, 0x1F0, 0x200, 0x220, 0x240, 0x260, 0x280, 0x2A0, 0x2C0, 0x2E0, 0x300, - 0x320, 0x360, 0x3A0, 0x3E0, 0x400, 0x440, 0x480, 0x4C0, 0x500, 0x540, 0x580, - 0x5C0, 0x600, 0x640, 0x680, 0x700, 0x780, 0x800, 0x880, 0x900, 0x980, 0xA00, - 0xA80, 0xB00, 0xB80, 0xC00, 0xC80, 0xD00, 0xD80, 0xE00, 0xE80, 0xF00, 0xF80}; - -const uint32_t VS_TIME_MIN_OS04C10 = 1; -//const uint32_t VS_TIME_MAX_OS04C10 = 34; // vs < 35 + 0x080, 0x088, 0x090, 0x098, 0x0A0, 0x0A8, 0x0B0, 0x0B8, 0x0C0, 0x0C8, 0x0D8, + 0x0E8, 0x0F8, 0x100, 0x110, 0x120, 0x130, 0x140, 0x150, 0x160, 0x170, 0x180, + 0x190, 0x1B0, 0x1D0, 0x1F0, 0x200, 0x220, 0x240, 0x260, 0x280, 0x2A0, 0x2C0, + 0x2E0, 0x300, 0x320, 0x340, 0x380, 0x3C0, 0x400, 0x440, 0x480, 0x4C0, 0x500, + 0x540, 0x580, 0x5C0, 0x600, 0x640, 0x680, 0x6C0, 0x700, 0x740, 0x780, 0x7C0}; } // namespace @@ -25,15 +22,9 @@ OS04C10::OS04C10() { image_sensor = cereal::FrameData::ImageSensor::OS04C10; data_word = false; - frame_width = 1920; - frame_height = 1080; - frame_stride = (1920*10/8); - - /* - frame_width = 0xa80; - frame_height = 0x5f0; - frame_stride = 0xd20; - */ + frame_width = 2688; + frame_height = 1520; + frame_stride = (frame_width * 12 / 8); // no alignment extra_height = 0; frame_offset = 0; @@ -42,17 +33,17 @@ OS04C10::OS04C10() { init_reg_array.assign(std::begin(init_array_os04c10), std::end(init_array_os04c10)); probe_reg_addr = 0x300a; probe_expected_data = 0x5304; - mipi_format = CAM_FORMAT_MIPI_RAW_10; - frame_data_type = 0x2b; + mipi_format = CAM_FORMAT_MIPI_RAW_12; + frame_data_type = 0x2c; mclk_frequency = 24000000; // Hz - dc_gain_factor = 7.32; + dc_gain_factor = 1; dc_gain_min_weight = 1; // always on is fine dc_gain_max_weight = 1; dc_gain_on_grey = 0.9; dc_gain_off_grey = 1.0; exposure_time_min = 2; // 1x - exposure_time_max = 2016; + exposure_time_max = 2200; analog_gain_min_idx = 0x0; analog_gain_rec_idx = 0x0; // 1x analog_gain_max_idx = 0x36; @@ -62,30 +53,22 @@ OS04C10::OS04C10() { for (int i = 0; i <= analog_gain_max_idx; i++) { sensor_analog_gains[i] = sensor_analog_gains_OS04C10[i]; } - min_ev = (exposure_time_min + VS_TIME_MIN_OS04C10) * sensor_analog_gains[analog_gain_min_idx]; + min_ev = (exposure_time_min) * sensor_analog_gains[analog_gain_min_idx]; max_ev = exposure_time_max * dc_gain_factor * sensor_analog_gains[analog_gain_max_idx]; target_grey_factor = 0.01; } std::vector OS04C10::getExposureRegisters(int exposure_time, int new_exp_g, bool dc_gain_enabled) const { - // t_HCG&t_LCG + t_VS on LPD, t_SPD on SPD - uint32_t hcg_time = exposure_time; - //uint32_t lcg_time = hcg_time; - //uint32_t spd_time = std::min(std::max((uint32_t)exposure_time, (exposure_time_max + VS_TIME_MAX_OS04C10) / 3), exposure_time_max + VS_TIME_MAX_OS04C10); - //uint32_t vs_time = std::min(std::max((uint32_t)exposure_time / 40, VS_TIME_MIN_OS04C10), VS_TIME_MAX_OS04C10); - + uint32_t long_time = exposure_time; uint32_t real_gain = os04c10_analog_gains_reg[new_exp_g]; - hcg_time = 100; - real_gain = 0x320; + // uint32_t short_time = long_time > exposure_time_min*8 ? long_time / 8 : exposure_time_min; return { - {0x3501, hcg_time>>8}, {0x3502, hcg_time&0xFF}, - //{0x3581, lcg_time>>8}, {0x3582, lcg_time&0xFF}, - //{0x3541, spd_time>>8}, {0x3542, spd_time&0xFF}, - //{0x35c2, vs_time&0xFF}, - + {0x3501, long_time>>8}, {0x3502, long_time&0xFF}, + // {0x3511, short_time>>8}, {0x3512, short_time&0xFF}, {0x3508, real_gain>>8}, {0x3509, real_gain&0xFF}, + // {0x350c, real_gain>>8}, {0x350d, real_gain&0xFF}, }; } diff --git a/system/camerad/sensors/os04c10_registers.h b/system/camerad/sensors/os04c10_registers.h index ad91a02950..f2388d91b8 100644 --- a/system/camerad/sensors/os04c10_registers.h +++ b/system/camerad/sensors/os04c10_registers.h @@ -4,43 +4,33 @@ const struct i2c_random_wr_payload start_reg_array_os04c10[] = {{0x100, 1}}; const struct i2c_random_wr_payload stop_reg_array_os04c10[] = {{0x100, 0}}; const struct i2c_random_wr_payload init_array_os04c10[] = { - // OS04C10_AA_00_02_17_wAO_1920x1080_MIPI728Mbps_Linear12bit_20FPS_4Lane_MCLK24MHz + // OS04C10_AA_00_02_17_wAO_2688x1524_MIPI728Mbps_Linear12bit_20FPS_4Lane_MCLK24MHz {0x0103, 0x01}, - {0x0301, 0x84}, + + // PLL + {0x0301, 0xe4}, {0x0303, 0x01}, - {0x0305, 0x5b}, + {0x0305, 0xb6}, {0x0306, 0x01}, {0x0307, 0x17}, {0x0323, 0x04}, {0x0324, 0x01}, {0x0325, 0x62}, + {0x3012, 0x06}, {0x3013, 0x02}, {0x3016, 0x72}, {0x3021, 0x03}, {0x3106, 0x21}, {0x3107, 0xa1}, - {0x3500, 0x00}, - {0x3501, 0x00}, - {0x3502, 0x40}, - {0x3503, 0x88}, - {0x3508, 0x07}, - {0x3509, 0xc0}, - {0x350a, 0x04}, - {0x350b, 0x00}, - {0x350c, 0x07}, - {0x350d, 0xc0}, - {0x350e, 0x04}, - {0x350f, 0x00}, - {0x3510, 0x00}, - {0x3511, 0x00}, - {0x3512, 0x20}, + + // ? {0x3624, 0x00}, {0x3625, 0x4c}, - {0x3660, 0x00}, + {0x3660, 0x04}, {0x3666, 0xa5}, {0x3667, 0xa5}, - {0x366a, 0x64}, + {0x366a, 0x50}, {0x3673, 0x0d}, {0x3672, 0x0d}, {0x3671, 0x0d}, @@ -63,22 +53,22 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x36a0, 0x12}, {0x36a1, 0x5d}, {0x36a2, 0x66}, - {0x370a, 0x00}, + {0x370a, 0x02}, {0x370e, 0x0c}, {0x3710, 0x00}, {0x3713, 0x00}, {0x3725, 0x02}, {0x372a, 0x03}, {0x3738, 0xce}, - {0x3748, 0x00}, - {0x374a, 0x00}, - {0x374c, 0x00}, - {0x374e, 0x00}, + {0x3748, 0x02}, + {0x374a, 0x02}, + {0x374c, 0x02}, + {0x374e, 0x02}, {0x3756, 0x00}, - {0x3757, 0x0e}, + {0x3757, 0x00}, {0x3767, 0x00}, {0x3771, 0x00}, - {0x377b, 0x20}, + {0x377b, 0x28}, {0x377c, 0x00}, {0x377d, 0x0c}, {0x3781, 0x03}, @@ -111,6 +101,8 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x3d8d, 0xe2}, {0x3f00, 0x0b}, {0x3f06, 0x04}, + + // BLC {0x400a, 0x01}, {0x400b, 0x50}, {0x400e, 0x08}, @@ -118,7 +110,7 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x4045, 0x7e}, {0x4047, 0x7e}, {0x4049, 0x7e}, - {0x4090, 0x14}, + {0x4090, 0x04}, {0x40b0, 0x00}, {0x40b1, 0x00}, {0x40b2, 0x00}, @@ -128,24 +120,25 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x40b7, 0x00}, {0x40b8, 0x00}, {0x40b9, 0x00}, - {0x40ba, 0x00}, + {0x40ba, 0x01}, + {0x4301, 0x00}, {0x4303, 0x00}, {0x4502, 0x04}, {0x4503, 0x00}, {0x4504, 0x06}, {0x4506, 0x00}, - {0x4507, 0x64}, + {0x4507, 0x47}, {0x4803, 0x00}, {0x480c, 0x32}, - {0x480e, 0x00}, - {0x4813, 0x00}, + {0x480e, 0x04}, + {0x4813, 0xe4}, {0x4819, 0x70}, {0x481f, 0x30}, {0x4823, 0x3f}, {0x4825, 0x30}, {0x4833, 0x10}, - {0x484b, 0x07}, + {0x484b, 0x27}, {0x488b, 0x00}, {0x4d00, 0x04}, {0x4d01, 0xad}, @@ -156,31 +149,37 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x4d0b, 0x01}, {0x4e00, 0x2a}, {0x4e0d, 0x00}, + + // ISP {0x5001, 0x09}, {0x5004, 0x00}, {0x5080, 0x04}, - {0x5036, 0x00}, + {0x5036, 0x80}, {0x5180, 0x70}, {0x5181, 0x10}, + + // DPC {0x520a, 0x03}, {0x520b, 0x06}, {0x520c, 0x0c}, + {0x580b, 0x0f}, {0x580d, 0x00}, {0x580f, 0x00}, {0x5820, 0x00}, {0x5821, 0x00}, + {0x301c, 0xf8}, {0x301e, 0xb4}, - {0x301f, 0xd0}, - {0x3022, 0x01}, + {0x301f, 0xf0}, + {0x3022, 0x61}, {0x3109, 0xe7}, {0x3600, 0x00}, {0x3610, 0x65}, {0x3611, 0x85}, {0x3613, 0x3a}, {0x3615, 0x60}, - {0x3621, 0x90}, + {0x3621, 0xb0}, {0x3620, 0x0c}, {0x3629, 0x00}, {0x3661, 0x04}, @@ -194,9 +193,9 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x3701, 0x12}, {0x3703, 0x28}, {0x3704, 0x0e}, - {0x3706, 0x4a}, + {0x3706, 0x9d}, {0x3709, 0x4a}, - {0x370b, 0xa2}, + {0x370b, 0x48}, {0x370c, 0x01}, {0x370f, 0x04}, {0x3714, 0x24}, @@ -206,19 +205,19 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x3720, 0x00}, {0x3724, 0x13}, {0x373f, 0xb0}, - {0x3741, 0x4a}, - {0x3743, 0x4a}, - {0x3745, 0x4a}, - {0x3747, 0x4a}, - {0x3749, 0xa2}, - {0x374b, 0xa2}, - {0x374d, 0xa2}, - {0x374f, 0xa2}, + {0x3741, 0x9d}, + {0x3743, 0x9d}, + {0x3745, 0x9d}, + {0x3747, 0x9d}, + {0x3749, 0x48}, + {0x374b, 0x48}, + {0x374d, 0x48}, + {0x374f, 0x48}, {0x3755, 0x10}, {0x376c, 0x00}, - {0x378d, 0x30}, - {0x3790, 0x4a}, - {0x3791, 0xa2}, + {0x378d, 0x3c}, + {0x3790, 0x01}, + {0x3791, 0x01}, {0x3798, 0x40}, {0x379e, 0x00}, {0x379f, 0x04}, @@ -249,29 +248,25 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x4041, 0x07}, {0x4008, 0x02}, {0x4009, 0x0d}, - {0x3800, 0x01}, - {0x3801, 0x80}, - {0x3802, 0x00}, - {0x3803, 0xdc}, - {0x3804, 0x09}, - {0x3805, 0x0f}, - {0x3806, 0x05}, - {0x3807, 0x23}, - {0x3808, 0x07}, - {0x3809, 0x80}, - {0x380a, 0x04}, - {0x380b, 0x38}, - {0x380c, 0x04}, - {0x380d, 0x2e}, - {0x380e, 0x12}, - {0x380f, 0x70}, + + // 2704x1536 -> 2688x1520 out + {0x3800, 0x00}, {0x3801, 0x00}, + {0x3802, 0x00}, {0x3803, 0x00}, + {0x3804, 0x0a}, {0x3805, 0x8f}, + {0x3806, 0x05}, {0x3807, 0xff}, + {0x3808, 0x0a}, {0x3809, 0x80}, + {0x380a, 0x05}, {0x380b, 0xf0}, {0x3811, 0x08}, {0x3813, 0x08}, {0x3814, 0x01}, {0x3815, 0x01}, {0x3816, 0x01}, {0x3817, 0x01}, - {0x3820, 0xB0}, + + {0x380c, 0x08}, {0x380d, 0x5c}, // HTS + {0x380e, 0x09}, {0x380f, 0x38}, // VTS + + {0x3820, 0xb0}, {0x3821, 0x00}, {0x3880, 0x25}, {0x3882, 0x20}, @@ -281,12 +276,12 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x3cae, 0x00}, {0x4000, 0xf3}, {0x4001, 0x60}, - {0x4003, 0x40}, + {0x4003, 0x80}, {0x4300, 0xff}, {0x4302, 0x0f}, {0x4305, 0x83}, {0x4505, 0x84}, - {0x4809, 0x1e}, + {0x4809, 0x0e}, {0x480a, 0x04}, {0x4837, 0x15}, {0x4c00, 0x08}, @@ -294,5 +289,25 @@ const struct i2c_random_wr_payload init_array_os04c10[] = { {0x4c04, 0x00}, {0x4c05, 0x00}, {0x5000, 0xf9}, - {0x3c8c, 0x10}, + {0x3822, 0x14}, + + // initialize exposure + {0x3503, 0x88}, + + // long + {0x3500, 0x00}, {0x3501, 0x00}, {0x3502, 0x80}, + {0x3508, 0x00}, {0x3509, 0x80}, + {0x350a, 0x04}, {0x350b, 0x00}, + + // short + // {0x3510, 0x00}, {0x3511, 0x00}, {0x3512, 0x10}, + // {0x350c, 0x00}, {0x350d, 0x80}, + // {0x350e, 0x04}, {0x350f, 0x00}, + + // wb + {0x5100, 0x06}, {0x5101, 0xcb}, + {0x5102, 0x04}, {0x5103, 0x00}, + {0x5104, 0x08}, {0x5105, 0xde}, + + {0x5106, 0x02}, {0x5107, 0x00}, }; diff --git a/system/camerad/sensors/ox03c10.cc b/system/camerad/sensors/ox03c10.cc index 1f0609820b..c74274872f 100644 --- a/system/camerad/sensors/ox03c10.cc +++ b/system/camerad/sensors/ox03c10.cc @@ -24,9 +24,9 @@ const uint32_t VS_TIME_MAX_OX03C10 = 34; // vs < 35 OX03C10::OX03C10() { image_sensor = cereal::FrameData::ImageSensor::OX03C10; data_word = false; - frame_width = FRAME_WIDTH; - frame_height = FRAME_HEIGHT; - frame_stride = FRAME_STRIDE; // (0xa80*12//8) + frame_width = 1928; + frame_height = 1208; + frame_stride = (frame_width * 12 / 8) + 4; extra_height = 16; // top 2 + bot 14 frame_offset = 2; diff --git a/system/camerad/sensors/sensor.h b/system/camerad/sensors/sensor.h index 4e2194d914..d97fd32a9c 100644 --- a/system/camerad/sensors/sensor.h +++ b/system/camerad/sensors/sensor.h @@ -12,10 +12,6 @@ #include "system/camerad/sensors/os04c10_registers.h" #define ANALOG_GAIN_MAX_CNT 55 -const size_t FRAME_WIDTH = 1928; -const size_t FRAME_HEIGHT = 1208; -const size_t FRAME_STRIDE = 2896; // for 12 bit output. 1928 * 12 / 8 + 4 (alignment) - class SensorInfo { public: