Commit graph

278 commits

Author SHA1 Message Date
Arttu Ylä-Outinen 07b5fb9caf Fix out-of-bounds read in encoderstate
When calling encoder_state_encode_leaf with POC 0, index -1 of the GOP
array would be accessed. Fixed by skipping the code for I-frames.
2017-07-20 11:15:30 +03:00
Miika Metsoila e8cc2d8f6a Small fixes 2017-07-07 13:58:19 +03:00
Arttu Ylä-Outinen 67a60a35e3 Fix invalid calls to normalize_lcu_weights
Changes encoder_state_init_new_frame to only call normalize_lcu_weights
when the weights have been written to the array and rate control is
enabled. When rate control is disabled, the weights are not used.
2017-07-07 11:05:31 +03:00
Arttu Ylä-Outinen 2a85f0f5a4 Move hard-coded MV limits to encoder_control_t
Adds field max_inter_ref_lcu to encoder_control_t. It is used to set up
inter-LCU dependencies in encoder_state_encode_leaf and restrict motion
vectors in fracmv_within_tile.
2017-07-05 13:22:53 +03:00
Arttu Ylä-Outinen bb5354f7e2 Relax inter-CTU dependencies when SAO is off
When using WPP and OWF, the first CTU of a row depends on the last CTU
of the row below in the reference frame. This is necessary when SAO is
enabled since we currently do SAO for a whole CTU row at a time. When
SAO is disabled, however, it is unnecessary to wait for the whole row.

Changes CTUs to depend only on the CTU below in the reference frame
instead of the whole row when WPP and OWF are enabled and SAO disabled.
Gives a significant speedup when running on a machine with many CPU
cores.
2017-07-05 13:21:06 +03:00
Arttu Ylä-Outinen 1efa2708b2 Do SAO reconstruction for a single CTU at a time
Moves SAO reconstruction into encoder_state_worker_encode_lcu instead of
doing it in a separate step for the whole CTU row. Reconstruction of the
rightmost 10 pixels and bottommost 10 pixels of a CTU is delayed until
the neighboring CTU has been deblocked.

Doing SAO for the whole CTU row at a time caused unnecessary inter-CTU
dependencies when using WPP and OWF. The first CTU of a row would need
to wait until SAO was done for the row below in the previous frame.
Moving SAO reconstruction to immediately after deblocking each CTU fixes
this problem.
2017-07-04 15:14:31 +03:00
Arttu Ylä-Outinen ec9ff42077 Rewrite SAO recon to handle arbitrary sized blocks
Adds width and height parameters to function kvz_sao_reconstruct and
changes it to take coordinates in units of pixels. This will be useful
for doing SAO for areas smaller than a whole CTU.
2017-06-30 16:09:18 +03:00
Miika Metsoila dcd7acf4fd Fixed crash and incorrect info output 2017-06-27 16:05:15 +03:00
Miika Metsoila f8b6234fdb Changes to refence lists to behave more like L0/L1 lists from the specification 2017-06-27 16:05:15 +03:00
Arttu Ylä-Outinen 00c9f52bd4 Fix setting picture type when using GOP
Changes encoder_state_init_new_frame to set intra frame pictype to
KVZ_NAL_IDR_W_RADL even when using GOP.
2017-06-21 13:21:47 +03:00
Arttu Ylä-Outinen f5eef7f33c Use luma pixel coordinates in encode_coding_tree
Changes functions encode_intra_coding_unit and encode_coding_tree to
take coordinate arguments in units of luma pixels instead of 8 px
blocks. This should make the code easier to understand.
2017-05-24 11:15:31 +03:00
Arttu Ylä-Outinen 2f2c281e8e Fix a memory leak in crypto
A CryptoPP::CFB_Mode<CryptoPP::AES>::Encryption was allocated at the
beginning of encoder_state_encode_leaf and was never freed. This commit
changes encoder_state_worker_encode_lcu to delete the CFB_Mode. Also
moves crypto handle from encoder_state_config_tile_t to encoder_state_t
so that it can be safely deleted without affecting other threads in the
same tile.
2017-05-23 11:51:25 +03:00
Arttu Ylä-Outinen 22155950c1 Rewrite crypto to conform to kvazaar code style 2017-05-23 11:51:25 +03:00
Arttu Ylä-Outinen 1357dd0599 Pass coeffs through encoder state
Changes the way coefficients are passed from kvz_search_lcu to
kvz_encode_coding_tree. Drops fields coeff_y, coeff_u and coeff_v in
videoframe_t and instead passes them through field coeff in
endoder_state_t.
2017-05-12 16:42:41 +03:00
Arttu Ylä-Outinen edfbd6f122 Add field lcu_dqp_enabled to encoder_control_t
Delta QPs for LCUs are enabled when either ROI coding or rate control is
enabled. Having a single field is simpler than always checking whether
ROI or rate control is enabled.
2017-05-11 12:13:47 +03:00
Arttu Ylä-Outinen a9c878b535 Fix crash with WPP when threads are disabled
When WPP is enabled, a reference to SAO reconstruction job is copied
from the wavefront to the main encoder state. However, when threads are
disabled, the job is a null pointer and dereferencing it crashes the
encoder. Fixed by adding a null pointer check.
2017-04-24 12:59:57 +03:00
Arttu Ylä-Outinen 2991962033 Add reference counting to threadequeue_job_t
Both the thread queue and the encoder states hold pointers to the thread
queue jobs. It is possible that a job is removed from the thread queue
and freed while the encoder state is still using it. This commit adds
reference counting to threadqueue_job_t in order to fix the problem.

Fixes #161.
2017-04-12 16:13:52 +03:00
Marko Viitanen 4251607c04 Fix a bug in TMVP reference POC list 2017-02-13 15:19:24 +02:00
Arttu Ylä-Outinen 51786eda67 Drop redundant fields in encoder_control_t
Some of the fields in encoder_control_t were simply copies of the
corresponding fields in kvz_config. This commit drops the copied fields
in favor of using the fields in encoder_control_t.cfg directly.
2017-02-09 14:05:28 +09:00
Arttu Ylä-Outinen e78a8dfcf5 Copy the kvz_config passed to encoder_open
The kvz_config struct is created by the user but kvazaar keeps a pointer
to it. It is easy to break things by modifying the configuration outside
kvazaar. In addition, kvazaar modifies the struct even though it is has
a const modifier.

This commit changes the field cfg in encoder_control_t to be a copy of
the kvz_config struct instead of a pointer, removing modifications to
the const struct and allowing users to do whatever they want with it
after opening the encoder.
2017-02-09 13:23:54 +09:00
Ari Koivula 5513744d24 Merge branch 'slices' 2017-01-31 16:14:30 +02:00
Ari Koivula 52904d3e9f Add --slices=tiles and --slices=wpp
This encapsulates tiles or WPP rows into their own slices, making
it possible to send them as soon as they are done, instead of waiting
for the other substreams to finish and coding the substream offsets
in the slice header.
2017-01-31 15:44:23 +02:00
Ari Koivula f1fc0de2bf Write slice headers to the parent stream
Appending to the child stream doesn't work is the child is a leaf
slice state.

Simplifies flow by removing distinction between tile and slice. Now
that slice headers are written in the parent stream, there is zero
difference between tiles and slices from bitstream point of view.
2017-01-31 13:55:05 +02:00
Ari Koivula 04cd875b2c Move substream finalization to LCU coding job
Having some of the termination bits in the LCU coding and some in the
substream finalization was needlessly confusing. Doing substream
finalization directly after LCU coding makes it easy to verify that the
finalization is done correctly.

Removes one job per WPP row from the job queue.

Removes kvz_cabac_flush, because I don't like bits being put into the
bitstream implicitly. Better to have it all in the open.
2017-01-31 13:01:57 +02:00
Arttu Ylä-Outinen fb10b56b82 Fix checking if a low delay GOP structure is used
Stops assuming that having cfg->gop_lowdelay set means that GOP
structure is used since it is possible that cfg->gop_lowdelay is true
but cfg->gop_len is zero. Adds checks for cfg->gop_len where needed.

Fixes a possible division by zero in kvz_encoder_feed_frame.
2017-01-28 21:56:00 +09:00
Ari Koivula 4a0121ac42 Add --roi parameter
Adds region of interest coding capability.

Works by reading a file of delta QP values which will then be applied
to each frame at LCU level.
2017-01-26 09:14:14 +02:00
Arttu Ylä-Outinen c219d3cd94 Fix deblock when CU QP delta is enabled
Fixes deblock functions so that they use the correct QP for the filtered
edge. Adds field qp to cu_info_t.
2017-01-11 15:53:22 +09:00
Arttu Ylä-Outinen ff5e5ec6d4 Record info about coded LCUs
Adds field lcu_stats to encoder_state_config_frame_t. The following data
is recorded for each LCU:
    - number of bits
    - squared cost
    - used lambda value
    - alpha parameter used for rate control
    - beta parameter used for rate control
2017-01-09 01:24:23 +09:00
Arttu Ylä-Outinen 2a4243acbe Refactor rate control
Moves all code related to setting QP and lambda values to rate_control
module.
2017-01-09 01:24:23 +09:00
Arttu Ylä-Outinen 71633889ce Enable CU QP delta when using rate control
When rate control is enabled, enable cu_qp_delta_enabled_flag in PPS
with diff_cu_qp_delta_depth set to 0. Also adds code for writing the QP
deltas and a new cabac context.
2017-01-09 01:24:23 +09:00
Arttu Ylä-Outinen 640ff94ecd Use separate lambda and QP for each LCU
Adds fields lambda, lambda_sqrt and qp to encoder_state_t. Drops field
cur_lambda_cost_sqrt from encoder_state_config_frame_t and renames
cur_lambda_cost to lambda.
2017-01-09 01:24:23 +09:00
Arttu Ylä-Outinen 6c4f2d196a Move fields from encoder_state_t to frame
Moves fields prepared and frame_done from encoder_state_t to
encoder_state_config_frame_t.
2017-01-09 01:24:23 +09:00
Wassim Hamidouche ea82c38906 correct memory allocation 2016-11-16 12:35:28 +02:00
Wassim Hamidouche da3e2d1d07 resolve parallel encryption 2016-11-16 12:35:28 +02:00
Ari Koivula d18de19d8a Fix DTS and PTS not being passed on through lib API
Fixes "cur_dts is invalid" warning from FFmpeg.
2016-10-28 19:05:47 +03:00
Ari Koivula 4ec039004b Add monochrome encoding
Write bitstream without chroma when encoding with --input-format=P400.
This reduces bitstream size by 0-1 %, compared to coding monochrome in
420 format, and speeds up encoding slightly due to not processing
chroma.
2016-08-25 20:15:26 +03:00
Ari Koivula 032ed30ff4 Add chroma format support to kvz_picture
Add picture_alloc_csp to libkvz api to allocated pictures with chroma
format different from 420.
2016-08-24 19:20:53 +03:00
Marko Viitanen c5f2611a38 Fixes for TMVP to work with the new CU array 2016-08-10 14:09:28 +03:00
Marko Viitanen 39f0165efe Fix a bug in TMVP, the reference cu_array was being overwritten 2016-08-10 14:09:27 +03:00
Arttu Ylä-Outinen 2a946bd88e Rename encoder_state_t.global to frame
"Frame" is more accurate than "global" since when OWF is used, encoder
states for each frame have their own struct.
2016-08-10 13:22:36 +09:00
Arttu Ylä-Outinen aabf6ca3ee Extract encoding code from encoderstate.c
Moves functions kvz_encode_coding_tree and kvz_encode_coeff_nxn from
encoderstate.c to encode_coding_tree.c.
2016-08-09 22:16:50 +09:00
Arttu Ylä-Outinen 803f29be8f Remove reconstructed picture allocation in lossless.
Changes encoder_set_source_picture to set the reconstructed picture to
a copy of the source picture instead of allocating a new picture when
lossless coding is used.
2016-08-03 14:25:08 +09:00
Arttu Ylä-Outinen aaec473a19 Refactor encoder state initialization.
- Moves allocation of the reconstructed picture after the source picture
  is set.
- Extracts main state initialization to a separate function from
  encoder_state_new_frame.
- Changes kvz_encoder_feed_frame to return the frame.
- Renames some functions to better match their purpose.
2016-08-03 14:25:08 +09:00
Arttu Ylä-Outinen 06b82bf888 Disable filters, trskip and signhide in lossless.
When lossless coding is used, deblock and SAO are skipped, transform
skip flag is not written and sign hiding is not used.
2016-08-03 14:25:08 +09:00
Arttu Ylä-Outinen 1dc94663c3 Bypass transform and quantization with --lossless.
When --lossless is given, set cu_transquant_bypass_flag for every CU and
bypass transform and quantization by directly copying reference pixels
to reconstruction and the residual to coefficients.
2016-08-03 14:25:08 +09:00
Arttu Ylä-Outinen 2113b0182d Enable PPS-level tq bypass flag with --lossless.
Sets transquant_bypass_enable_flag to true in PPS when --lossless is
given.
2016-08-03 14:25:08 +09:00
Arttu Ylä-Outinen ae832cda8c Pack cbf flags in cu_info_t to two bytes.
Reduces size of cu_info_t.
2016-06-16 20:24:19 +09:00
Arttu Ylä-Outinen cad2d496b8 Enable 4x8 and 4x16 partition modes
Enables search for 2NxN and Nx2N partition modes for 8x8 CUs and 2NxnU,
2NxnD, nLx2N and nRx2N partition modes for 16x16 CUs.

Changes the loop for copying reconstructed luma pixels in
kvz_inter_recon_lcu to use 4 byte chunks instead of 8 byte chunks since
it is now possible to have 4 pixel wide blocks.
2016-06-16 20:23:16 +09:00
Arttu Ylä-Outinen 2ae260e422 Change width of cells in lcu_t to 4 pixels.
Intra mode info for NxN partition units is now stored in the
corresponding 4x4 cell in lcu_t.cu array.
2016-06-16 18:53:17 +09:00
Arttu Ylä-Outinen 46e8122d27 Add functions for indexing cu_array_t structures.
Replaces macro CU_ARRAY_AT with functions kvz_cu_array_at and
kvz_cu_array_at_const.
2016-06-16 18:52:19 +09:00
Arttu Ylä-Outinen 2c85a00a55 Change kvz_cu_array_alloc to use pixel dimensions.
Changes function kvz_cu_array_alloc to take width and height parameters
in pixels instead of SCUs.
2016-06-15 12:25:11 +09:00
Arttu Ylä-Outinen 8ac1f1986e Move CU array copy to a separate function.
Moves code for copying parts of cu_array_t to a new function
kvz_cu_array_copy in cu module.
2016-06-15 12:25:11 +09:00
Arttu Ylä-Outinen 23fdeeaf10 Move mv_cand and mv_dir into a bitfield in cu_info_t.
Reduces size of cu_info_t.
2016-06-14 12:21:57 +09:00
Arttu Ylä-Outinen b6d793ef33 Drop field inter.mvd from cu_info_t
Instead of storing the mv differences in cu_info_t, they are computed
from the mv candidates and the motion vector. Reduces the size of
cu_info_t.
2016-06-14 12:21:57 +09:00
Arttu Ylä-Outinen ebb10763f1 Drop field inter.mv_ref_coded from cu_info_t.
Storing inter.mv_ref_coded in cu_info_t is unnecessary since it can be
computed from refmap and inter.mv_ref.
2016-06-14 12:21:57 +09:00
Wassim Hamidouche 35634b5596 correct MV sign encryption 2016-06-09 10:49:31 +03:00
Wassim Hamidouche 15abdc6e81 correct sign encryption 2016-06-09 10:49:31 +03:00
Wassim Hamidouche 73c3203a26 encry coef transfs 2016-06-09 10:49:31 +03:00
Wassim Hamidouche 7ad5f8bbe5 encry coef transf sign 2016-06-09 10:49:31 +03:00
Wassim Hamidouche 76cb6dc6c2 add check flags 2016-06-07 10:54:26 +02:00
Wassim Hamidouche 02308d1ba6 add MVs encryption 2016-06-07 10:28:30 +02:00
Wassim Hamidouche 4637c8a828 compile Kvazaar encoder with ITpp library 2016-06-07 08:33:04 +02:00
Ari Koivula 67acead4bc Fix referring over IDR boundary when using --gop
This problem resulted in an illegal bitstream with --gop=lp, because it
uses IDR's. The --gop=8 would not code IDR pictures, even when told to
with -p, which masked this problem.

This fix solves the problem with --gop=lp and also prevents references
across the intra picture in --gop=8. The intra pictures should be set
to IDR in a later fix, or an alternate method of differentiating
between IDR and non-IDR intra should be made.
2016-05-27 13:20:53 +03:00
Ari Koivula a77dc1610e Refactor encoder_state_remove_refs
I needed to debug this, so I rewrote it to make sense. There is an
obvious bug with the IDR handling that I left in place to fix in a
separate commit.
2016-05-27 13:20:45 +03:00
Ari Koivula 61fc3e87ba Run include-what-you-use fix_includes.py fix_includes.py
The includes should make more sense now and not just happen to compile
due to headers included from other headers.

Used a modified version of IWYU. Modifications were to attribute int8_t
and so on to stdint.h instead of sys/types.h and immintrin.h instead of
more specific headers.

include-what-you-use 0.7 (git:b70df35)
based on clang version 3.9.0 (trunk 264728)
2016-04-01 17:46:55 +03:00
Ari Koivula e91ca74733 Refactor kvz_encode_last_significant_xy 2016-03-10 18:47:16 +02:00
Ari Koivula 1fc0e8076c Format kvz_encode_last_significant_xy whitespace 2016-03-10 18:17:45 +02:00
Ari Koivula 4112a4364d Remove g_to_bits table 2016-03-10 15:59:51 +02:00
Ari Koivula 674bfa14ce Comment WPP deblocking and SAO
I was a bit unclear about exactly what happens and when regarding SAO
and deblocking when we do frame-parallel WPP parallelism, so I checked
and commented the bits that were unclear to me.
2016-03-08 19:39:04 +02:00
Ari Koivula fd34dd9bc6 Fix race condition with OWF
There was an off by one error in the dependance setting code, which
resulted in dependencies not being set resulting in checksum errors.
For example if ref_neg=1 and owf=1.
2016-03-07 13:38:23 +02:00
Arttu Ylä-Outinen 626b53ce85 Move sao search from encoderstate to sao.
Moves sao search from function encoder_state_worker_encode_lcu in
encoderstate.c to function kvz_sao_search_lcu in sao.c. Makes functions
kvz_init_sao_info, kvz_sao_search_chroma and kvz_sao_search_luma static
since they are no longer used outside sao.c.
2016-03-01 14:56:16 +02:00
Ari Koivula cfa722e448 Reduce parallelism for tiles
There is still some race-condition with encoding tiles from multiple
frames, so disable this to keep the bitstream deterministic.
2016-02-29 20:20:21 +02:00
Ari Koivula b1adf1576a Add --mv-constraint=frametilemargin
Add an even stricter motion vector constraint to prevent motion vectors
to fractional pixel positions that would need pixels outside the tile.
2016-02-29 19:18:14 +02:00
Ari Koivula f808cbf608 Allow increased parallelism for tiles
When movement vectors are constrained to tiles, only the same tile in
previous frame needs to be depended upon.
2016-02-29 14:33:06 +02:00
Ari Koivula c0dc490dd1 Fix inter non-determinism with tiles
CU data was being copied to the wrong place in the reference frames
cu_array, which led to uninitialized data being used as a starting
point for motion vector search.

Fixes #99.
2016-02-26 17:05:04 +02:00
Ari Koivula c40ede56ad Allow more frame parallelism in LP-gop
Add dependency to the reference frame instead of the previous frame,
in order to allow more frames to be encoded in parallel when temporal
stepping >1 in LP-gop (such as --gop=lp-g8d4r1t2).
2016-02-05 17:08:24 +02:00
Arttu Ylä-Outinen 49677810b5 Rename config module to cfg.
Prevents a conflict with config.h and src/config.h so that the config.h
generated by configure is included in global.h. Fixes problems with
large input files on 32-bit systems.
2016-01-25 12:26:46 +02:00
Ari Koivula 947bae24f9 Update Doxygen documentation
Add module information to all header files.

Update all header file documentations to briefly say what they are, and
to use the javadoc format so the brief actually gets included into the
doxygen documentation.

Remove \file from implementation files, in order to not repeat the info
from the header files.

Add files under strategies and tools to Doxygen and update the Doxygen
settings to be just plain better.

Make README be the main page of Doxygen documentation.
2015-12-17 14:05:50 +02:00
Arttu Ylä-Outinen 7e4f4538a4 Implement encoding AMP part modes.
Also adds parameter --amp for enabling AMP blocks.
2015-12-15 11:21:45 +02:00
Arttu Ylä-Outinen f874c8614e Add part_mode binarization table comment. 2015-12-15 11:21:41 +02:00
Arttu Ylä-Outinen c77074a7ff Implement encoding SMP blocks. 2015-12-15 11:21:41 +02:00
Arttu Ylä-Outinen 98707a1288 Move encoding intra CU to a separate function.
Moves code for encoding a single intra coding unit from function
kvz_encode_coding_tree to a new function encode_intra_coding_unit.
2015-12-15 11:21:40 +02:00
Arttu Ylä-Outinen c336674da3 Move encoding part mode to a separate function.
Moves code for encoding the part mode from function
kvz_encode_coding_tree to a new function encode_part_mode.
2015-12-15 11:21:40 +02:00
Arttu Ylä-Outinen ac952cbb44 Move encoding inter PUs to a separate function.
Moves code for encoding a single inter prediction unit from function
kvz_encode_coding_tree to function encode_inter_prediction_unit.
2015-12-15 11:21:40 +02:00
Ari Koivula 3a80c7de74 Further optimize coefficient coding
Remove the need to count the coefficients by populating the significant
coefficient group map first and finding the last coefficient from the
last group afterward. The speedup is about 2% on ultrafast.

The previous version of this patch was reverted due to a bug, which
has now been fixed.
2015-12-11 16:47:55 +02:00
Ari Koivula b32965925e Revert "Further optimize coefficient coding"
This reverts commit 25462124f8.

That commit broke the bitstream. If it's not good enough to push on Friday
night, it's probably not good enough on Monday morning either.
2015-12-07 15:12:04 +02:00
Ari Koivula 865c86fef2 Remove unused variable 2015-12-07 10:32:18 +02:00
Ari Koivula 25462124f8 Further optimize coefficient coding
Remove the need to count the coefficients by populating the significant
coefficient group map first and finding the last coefficient from the
last group afterward.
2015-12-07 10:23:01 +02:00
Ari Lemmetti 6fe223c4dc Nonzero calculation magic 2015-12-03 18:29:44 +02:00
Ari Koivula cfe834bb53 Merge branch 'lowdelay_GOP'
Conflicts:
	README.md
2015-11-14 00:05:13 +02:00
Ari Koivula a6a713ac02 Use P-slices for lowdelay GOPs 2015-11-13 23:11:11 +02:00
Ari Lemmetti 0816fbea2c Create generic strategy of blit function 2015-11-04 10:07:25 +02:00
Arttu Ylä-Outinen 1c898a2f4a Prefix NAL unit type enum constants with KVZ_. 2015-09-28 10:30:58 +03:00
Arttu Ylä-Outinen 4e5c7fe6e8 Remove function kvz_encoder_compute_stats.
Changes main function to compute frame PSNR by calling
kvz_videoframe_compute_psnr directly with the source and reconstructed
pictures returned from encoder_encode.
2015-09-28 10:30:58 +03:00
Arttu Ylä-Outinen d5dceb45f1 Factor out a function for building ref lists.
The code for building the reference picture lists was duplicated in
functions encoder_state_ref_sort and print_frame_info. This commit moves
it to a new function kvz_encoder_get_ref_lists. Also makes
encoder_ref_insertion_sort static since it is not used outside the
encoderstate module any more.
2015-09-28 10:30:57 +03:00
Arttu Ylä-Outinen 173b70b53f Rename SLICE_* enum constants to KVZ_SLICE_*. 2015-09-28 10:30:56 +03:00
Ari Koivula 1d5cfbdcc2 Remove unused variable. 2015-09-16 18:39:46 +03:00
Ari Koivula 513e80bcca Fix bug causing unnecessary copying of memory
This bug caused a single tiles worth of lcu_info_t structs to be copied
unnecessarily for every LCU in the frame. This obviously caused huge
memory bandwidth issues when coding large frames without tiles. The
effect was minimized somewhat with a large number of tiles, because
only the current tile was copied.

From context it is clear that this piece of code was supposed to copy
a single tile or frame, once the frame was done, but because it was
placed in a function which is called for every LCU, it copied the data
for the LCU, but also lots of extra stuff.

The fix is to copy only the current LCU instead of the whole tile.
2015-09-16 18:23:44 +03:00
Ari Koivula f1ac0e6bc2 Rename _DEBUG to KVZ_DEBUG 2015-09-15 13:04:03 +03:00
Ari Koivula ec2d8d6ad7 Rename _DEBUG_PERF macros to KVZ_PERF
And move them to threadqueue.h, where the things that use them are.
2015-09-15 13:03:32 +03:00