             }
             
             // Encode 8x8 block
-            if ( gpujpeg_huffman_gpu_decoder_decode_block(get_buff, get_bits, component_dc, data_compressed, data_compressed_size, block, d_table_dc, d_table_ac) != 0 )
+            if ( gpujpeg_huffman_gpu_decoder_decode_block(component_dc, block, d_table_dc, d_table_ac, r_bit, r_bit_count, s_byte, s_byte_idx, d_byte, d_byte_chunk_count) != 0 )
                 break;
         } 
     }

/** Natural order in constant memory */
__constant__ int gpujpeg_huffman_gpu_decoder_order_natural[GPUJPEG_ORDER_NATURAL_SIZE];

/**
 * Fill more bit to current get buffer
 * 
 * @param get_bits
 * @param get_buff
 * @param data
 * @param data_size
 * @return void
 */
__device__ inline void
gpujpeg_huffman_gpu_decoder_decode_fill_bit_buffer(int & get_bits, int & get_buff, uint8_t* & data, int & data_size)
{
    while ( get_bits < 25 ) {
        //Are there some data?
        if( data_size > 0 ) { 
            // Attempt to read a byte
            //printf("read byte %X 0x%X\n", (int)data, (unsigned char)*data);
            unsigned char uc = *data++;
            data_size--;            

            // If it's 0xFF, check and discard stuffed zero byte
            if ( uc == 0xFF ) {
                while ( uc == 0xFF ) {
                    //printf("read byte %X 0x%X\n", (int)data, (unsigned char)*data);
                    uc = *data++;
                    data_size--;
                }

                if ( uc == 0 ) {
                    // Found FF/00, which represents an FF data byte
                    uc = 0xFF;
                } else {                
                    // There should be enough bits still left in the data segment;
                    // if so, just break out of the outer while loop.
                    //if (m_nGetBits >= nbits)
                    if ( get_bits >= 0 )
                        break;
                }
            }

            get_buff = (get_buff << 8) | ((int) uc);
            get_bits += 8;            
        }
        else
            break;
    }
}

/**
 * Get bits

 * @param data_size
 * @return bits
 */
__device__ inline unsigned int
gpujpeg_huffman_gpu_decoder_get_bits(
                const unsigned int nbits, unsigned int & r_bit, unsigned int & r_bit_count, uint4 * const s_byte, 
                unsigned int & s_byte_idx, const uint4 * & d_byte, unsigned int & d_byte_chunk_count)
{
    // load bits into the register if haven't got enough
    gpujpeg_huffman_gpu_decoder_load_bits(nbits, r_bit, r_bit_count, s_byte, s_byte_idx, d_byte, d_byte_chunk_count);
    
    // update remaining bit count
    r_bit_count -= nbits;
    
    // return bits 
    return (r_bit >> r_bit_count) & ((1 << nbits) - 1);
}


