Avisynth Assembly Tutorial?

[MysteryX]

There are 3 types of programmers in this community:

1) Scripters (who know how to use and match plugins)
2) Coders (who know how to write C++ plugins)
3) Assembly guys (who know how to write assembly code)

There are tons of #1, a few #2 like me, and very few #3.

Learning assembly takes a long while, and it's a problem that so few can do it. There are lots of ways to write assembly, and just knowing which road to take requires a lot of research.

Unless... someone could write a tutorial for optimizing Avisynth/VapourSynth filters with assembly? Lining up exactly what to do, what not to do, and listing the required skillset and resources to learn?

If that could shorten the learning curve to optimize filters, everybody would win.

Here are a few basic questions (that I can't answer)

1. Assembly means running commands on 32 or 64 values at once. Can all types of filters be optimized in that way? What can or cannot be optimized?

2. There are various assembly sets (MMX, AVX, etc.) Which ones should or shouldn't be supported?

3. There are various ways to write it in C++, what's the right way to write it and to structure the code for readability?

4. What would be a list of links providing all the knowledge one needs to know to write assembly?

5. Assembly is scary to even get into -- what topics are NOT needed to dig into?

Even if this thread turns into a Q/A database, that would be useful for those who want to optimize their plugins. Would be 2nd best after a well-structured tutorial.

[DJATOM]

I will share my piece of knowledge, I'm not a pro C++ developer or ASM specialist.

If specific filter doing the same computation on entire frame, then it benefits from SIMD optimizations.
For example, if you want to use addition on 2073600 pixels (1920x1080), you will have to compute that 2073600 times! But having SSE2 it's possible to vectorize that calculation with instruction paddb xmm, xmm, which can compute 16 values at once. So only 129600 operations needed. AVX2 version of that instruction (vpaddb ymm, ymm, ymm) can compute 32 values at once - only 64800 operations will be done.

1. SIMD instructions can operate with bytes (8bit), words (16 bit), double-words (32 bit) and quad-words (64 bit). So usage depends on your data.
2. I think most common are SSE2 and AVX(2). Afaik AVX mostly applicable for float while AVX2 extends support for integer types.
3. You can use compiler intrinsics (https://software.intel.com/sites/landingpage/IntrinsicsGuide/) according to your needs and targeting certain CPUs.
4. Well, I mostly use Intel's guide.
5. It's not scary, it's just too complicated if you don't understand how it works. Maybe it's fine to google for some ASM basics and try to understand (as I did).

[johnmeyer]

If Assembly = Assembler, I'm not aware of much, if any, code written in x86 assembler for AVISynth.

I started my career writing assembler code for the HP2116 minicomputer. In my first company, one of my co-founders wrote much of the key code for our desktop publishing program in x86 assembler. Assembler code can be unbelievably fast (our program ran circles around the competition), but it is also not portable at all, meaning that any difference in instruction sets between CPU chips can cause the code to not work.

Because there are so many variations in hardware these days, with CPUs from multiple vendors, and instruction sets that have expanded dramatically over the years, I'm not sure how much assembler programming is done anymore.

You are much better off learning to program in C. You can do some amazing optimizations, both for code space as well as speed, once you know what you're doing. "C" is where the "big boys" play and is what you should spend your time mastering.

[qyot27]

While some (particularly older) AviSynth plugins did use inline assembly, that's not something any of 64-bit plugins could use since it's not portable and MSVC doesn't allow it for 64-bit.

AviSynth+ itself ditched virtually all of its inline assembly in favor of compiler intrinsics many years ago. There are still few areas which use it, but it's only activated in those specific circumstances that it was designed for (32-bit x86). IIRC, there's some in the audio conversion routines, a bit in the plugin manager (which may actually be disabled behind an #if 0 now), and the memcpy_amd source file.

The expected way to do it nowadays, of course, is to use compiler intrinsics, but if you want to dabble in NASM or build out against x264ASM the way other projects have done, have at it. The only thing that upstream AviSynth+ would recommend in that regard is to properly segregate the source paths from each other so that it remains easily portable and you don't have to have huge blocks under arcane #ifdefs (even if you might have to use #ifdefs to control the CPU dispatcher structure). Keep the main C/C++ implementation and CPU dispatch in one file, and put the x86 SIMD in another file or group of files (in the AviSynth+ source, you see this with said x86-specific stuff pushed down under intel/ directories), and then use the proper method to choose the x86 SIMD over the plain C/C++ if the build system or runtime checker detects that it's running on a compatible CPU. Then it doesn't get in the way when you want to add things like AltiVec/VSX (PowerPC/POWER7+), NEON (ARM), VIS (SPARC), etc.

[MysteryX]

These are the kind of things to clarify for people looking into this: compiler intrinsics is what people use nowadays to optimize code. Once you know the keywords, you can Google them. Until then, newbies kind of circle around.

All the questions still stand for compiler intrinsics.

[DTL]

There is no need to write in assembler all of the processing. It is hard and more buggy in compare with compiled C(++). The only what is really benefit is perform the most computing consuming part of the data processing in manual assembler.
The reason to it is still very poor support of SIMD optimizing in todays C-compilers (even the most advanced Intel C for intel chips is very poor in optimizing with SIMD on real 'image processing'. The ordinary program is better to put to optimizing compiler and for most noobs the result most possibly will be better. To outperform it on standard program the asm coder must know more than intel compiler about intel CPUs (and archtecture including memory/cache controller and subsystem) and no need to read such threads.

One reason why is optimizing compilers are too poor on creating outperforming manual SIMD result is also data placement and arrangement. And I think most of C-compilers can not understand even simple image-processing algoriphms and try to auto-put to wide-bits SIMD instrustion sets. Todays optimizing compilers can not rewrite most of pluging code to make program SIMD-optimal at time of compiling and compiler (AI) hints is still poor.

The creating asm part is really not very hard (with small enough asm inserts). Asm itself is very very simple in compare with object-orientrd C++. The hard part is to know about CPU and memory/cache architecture, get (lots of) expirience on optimizing on previous architectures, prepare overall processing pipeline be compatible with current and new architectures (without full redesign). Create design to be well balanced between memory-bound and CPU-bound and also best fit for multithreading. I.e. hard is create design of the required processing to be applicable to current CPU architecture (and fast), then simple is to write asm (intrinsics) text and debug it if needed.
SIMD parallel processing is about good scalable between 64/128/256/512 bit SIMD instructions and it sometime makes easier to copy-paste lots of asm with just renaming operands and instructions. But still no compiler(assembler) exist (I do not know) to auto-scale SIMD asm code to next generation of wider SIMD instructions. That require to create and support lots of equal functions just for 128,256,512 bit-wide instructions.

2. It depends on your target CPUs families. Generally the most advanced SIMD support is in the latest CPUs and it is the least widespread (in general world of poor users of free software). If you plan to buy exact CPU you already know its SIMD instrustions set and question not exist. So currently (to beginning of 202x) most widespread are pre-AVX512(family) instructions sets. The AVX512 are very nice (with complete family) but complete family is possibly not exist even at pro server-class Xeons. So for consumer-grade chips it can be expected AVX512F and may be a bit more. It is also +hardness for asm-designer to support not 1 function for AVX512 but many codepaths for different AVX512 members of familiy. Even if your function have 1 instruction non-supported at target CPU chip you need to make separate version of function. Making branches at high-performance program part is not good idea.

3. For asm it looks like only 2 ways for given build enviroment (and modern x64) - either C/C++ text with intrinsics or just extrernal asm library function with whatever asm syntax authour like and lets linker to build executable. The call/ret to external function may or may not be visibly time-consuming. With intrinsics you allow compiler to be supportive and re-arrange instructions for possibly best performance automatically. But need to check .asm listing and if not agree with compiler - try to re-write intrinsics or switch compiler options. Very typical problem - over-use of SIMD virtual 'registers' so compiler will not warn about it but starts to use cache or even system memory and slo-down alot. With external asm the author 100% responsible with all instructions order and resources using.

4. The most useful links are to intel optimizing guides. It assumes the reader knows the very simple way of explain the program idea with asm instructions to CPU.

5. Most old and almost no to use is the very first SIMD 64 bit sets (MMX). So todays is mostly actual are 128 and 256 bit sets and very perspective (but still rare at most of consumers machines I think) 512 bit sets.
For the very task of plugins processing speed-up you not need 'full-asm' but only need SIMD-asm. And if using C+intrinsincs the 'full-asm' is limited with SIMD-intrinsincs. All other required asm for program to run is as usual produced by C-compiler and it can do it very well.
The all other that speed-up designer need to know: memory, caches, buses, register file, execution units (ports), buffers, uops/cycles (throughoutput), threading, instructions decoding and so on is not directly interconnected with asm. It is pure hardware implementation data of very different but compatible CPU and system architectures.

"All the questions still stand for compiler intrinsics."

It already put to online - just open and use https://software.intel.com/sites/landingpage/IntrinsicsGuide/ . May be it possible to make offline version for the time where online connection is down. No more need to list 1000+ pages instruction set reference (volume 1 and 2) to find the required. Oh - it even have offline version for download - https://software.intel.com/sites/landingpage/IntrinsicsGuide/files/IntelIntrinsicsGuideOffline-3.6.0.zip

Addition: Some possibly useful note - asm design is very easy to write to asm text but much harder to read back. The C/C++ text is easier I think. So it is very recommended to put human-readable description of implemented asm design in software documentation (and extended comments). So this huge amount of different asm implementations for different architectures and different processings will be easier (if at all) to bugfix/support/patch/modify.

[MysteryX]

Easier to write than to read -- I think that's good to take out some of the fear. And lots of asm coders do not like to document code.

Processing 128 pixels at a time for addition or subtraction, that's fine. What about dynamic code processing? Let's take vcm.Median for example, it analyses the noise and expands the grid if necessary. Or, StripeMask (with FrameRateConverter), it processes data on a band-per-band basis in a dynamic way. Are such dynamic logics suitable for asm optimization? I don't see how to structure StripeMask to process many different bands at once.

Is it only suitable for repetitive linear work?

[DTL]

" Are such dynamic logics suitable for asm optimization? "

Much worse. Though you are not restricted to make even self-modified program at runtime.

"expands the grid if necessary. "

Typically if workunit of the same processing is large enough it may be performed with large to small SIMD instructions and last part with simple per-sample processing. Though it may significally decrease the benefit from SIMD because of many 'context switching and branching'. Architecture likes to work with large equal to process datasets with SIMD. Like process most of line with 10..100 SIMD ops 256bit in a loop (unrolled loop for decreasing possible bus direction switching, aligned datasets to read and write from/to cacheline boundary) and last 5 samples in separate last part loop. But processing one 256bit SIMD op and next is last non-aligned and per 1 sample loop like 1..3..7 samples may be much less faster.

[zorr]

Conditional processing is possible with SIMD, usually you use masks to filter calculations so that they only apply to some of the pixels even though you perform the calculations for every one of them.

Let's say we have this kind of loop where c is an array of 8 bit Y values and a is some per pixel offset we'd like to apply to it (also an array):

if (c[i] > x) {
c[i] = c[i] + a[i];
}

You could implement this in SSE2 by first creating a 128-bit mask (processing 16 pixels per instruction) with instruction pcmpgtb. The mask would contain zeros where c[i] <= x and ones where c[i] > x. Then use pand to perform a logical AND operation between the mask and the a vector. This will make a[i] zero wherever c[i] <= x. Finally just add the whole a vector into c with paddb.

In this case we used the fact that zero neutralizes the addition operation. If that doesn't apply, you can still use the mask technique with masked move operation maskmovdqu where you can skip the store where mask is zero.

The code inside the conditional block can be as long as you want, just remember that the calculations are performed for all pixels so it may be wasteful if only a handful actually need to be processed. However avoiding branching is a big performance win and since the loop processes 16 pixels per round it will most likely make the code perform much faster in every situation.

[DTL]

Branches is possible but bad for speed:
•Maintaining steady supply of micro-ops to the execution engine — Mispredicted branches can disrupt
streams of micro-ops, or cause the execution engine to waste execution resources on executing
streams of micro-ops in the non-architected code path. Much of the tuning in this respect focuses on
working with the Branch Prediction Unit. Common techniques are covered in Section 3.4.1, “Branch
Prediction Optimization.”
•Supplying streams of micro-ops to utilize the execution bandwidth and retirement bandwidth as
much as possible — For Intel Core microarchitecture and Intel Core Duo processor family, this aspect
focuses maintaining high decode throughput. In Sandy Bridge microarchitecture, this aspect focuses
on keeping the hot code running from Decoded ICache. Techniques to maximize decode throughput
for Intel Core microarchitecture are covered in Section 3.4.2, “Fetch and Decode Optimization.”
3.4.1 Branch Prediction Optimization
Branch optimizations have a significant impact on performance. By understanding the flow of branches
and improving their predictability, you can increase the speed of code significantly.
Optimizations that help branch prediction are:
•Keep code and data on separate pages. This is very important; see Section 3.6, “Optimizing Memory
Accesses,” for more information.
•Eliminate branches whenever possible.
•Arrange code to be consistent with the static branch prediction algorithm.
•Use the PAUSE instruction in spin-wait loops.
•Unroll as necessary so that repeatedly-executed loops have sixteen or fewer iterations (unless this
causes an excessive code size increase).
•Avoid putting multiple conditional branches in the same 8-byte aligned code block (i.e, have their last
bytes' addresses within the same 8-byte aligned code) if the lower 6 bits of their target IPs are the
same. This restriction has been removed in Ice Lake Client and later microarchitectures.
3.4.1.1 Eliminating Branches
Eliminating branches improves performance because:
•It reduces the possibility of mispredictions.
•It reduces the number of required branch target buffer (BTB) entries. Conditional branches that are
never taken do not consume BTB resources.
There are four principal ways of eliminating branches:
•Arrange code to make basic blocks contiguous.
•Unroll loops,

From intel optimization manual.
The asm is not lowest level control over CPU. Instructions of asm later decoded (compiled to real execution sequence of microops) by instruction decode unit. And it is specific to each chip design and very small covered by documentation.

[MysteryX]

btw if someone is interested, it would be if someone could add optimizations to vcm.Median (https://forum.doom9.org/showthread.php?p=1953877#post1953877); and perhaps make it dual-platform to also work in Avisynth

Its dynamic-radius Median has its uses, but it's very slow right now.

[kedautinh12]

VCMohan had manyplus's Median in avisynth
http://www.avisynth.nl/users/vcmohan/manyPlus/manyPlus.html

[zorr]

The horrors of branching that DTL laid out above are true, but I'd like to point out that they apply to all languages, not just assembly. But it's a good point to bring up because in higher level languages it's the compiler that makes optimization decisions based on the branching rules (and usually does a very good job at it), while in assembly it's the job of the programmer to figure out.

Just to make sure that nobody has the wrong impression, my little example above does NOT have any branching in it except perhaps the outer loop which is correctly predicted pretty much 100% of the time.

[DTL]

I personally interested in possible more optimization of mvtools for 'large' tr about 1 seconds and more when using MDegrainN. Though it already may be good optimized (for old consumer chips about 1..1.5 decades old). The most interesting and promises new offerings of consumer chips manufacturers are in AVX512 - new nice instructions (like replacing many old by new one and possibly great speedup) and new 'large' size register file of about 2 kBytes in size (visible to programmer as 32 zmm 512 bit registers set and available to direct referencing from instructions). That is the fastest memory on chip and directly integrated in each execution core between execution units (ports). The more common used todays chips with AVX256 have only 512 bytes register file for 16 256 bit 'ymm registers'.

[DTL]

Good online tool to inspect asm output of compiler. https://godbolt.org
May shows when it runs out of available 'registers' resources and start to use other memory for store and load data.

[Wilbert]

Some old introduction stuff: http://avisynth.nl/index.php/Filter_SDK/Assembler_optimizing

[DTL]

Some indirect method of estimating software efficiency - measuring chip power consumption or temperature (with fixed cooler efficiency like fixed cooler fan rpm).
The more computational resources used by software at a time - the more switching elements at CMOS logic both math and cache and more energy dissipated. So poorly designed for massive SIMD todays chips software will not heat CPU significally. And best designed may heat to full (and also cause speed-trottling to prevent over-heating). Also the power consumption of system may be controlled by AC Watt meter at power supply inlet. The 'CPU load' data at task manager do not shows actual computation load - it only display CPU busy time but in poorly designed software it may be time waiting for (small pieces of) data from memory. Or time to perform 1bit logic instead of 256/512 bit avx processing.

"horrors of branching that DTL laid out above are true"

I have about 1 day look into mvtools and see the cause of too poor performance today - it was designed with 'logic programmers' at the past when CPU chips were slow and mostly logic-oriented. So MAnalyse try to do lots of logic-search and not massive 'brute-force' computations. In the old days it may be faster in compare with full search field computation.
Now CPUs cores are divided to slow logic unit and much faster SIMD units (with only very basic logic like masks). So with todays chips it seems the execution need to be re-balanced between logic and 'pure brute-force' computation.
As an example it can be compared the 'classic' very logic-loaded and small steps 'ExpandingSearch() + CheckMV()' old days approach and new attempt of 'ExhaustiveSearch8x8_sp2_avx2(WorkingArea& workarea)' at the commit https://github.com/DTL2020/mvtools/commit/4a8ab45f397b642f85b8e9c7cafe14dd72989cc4 .
The more SIMD-oriented approach uses full block vs plane H-search at 12x8 ref-field with register file only and not contain any conditions check and no branching. The V-search may be also fit in register file (without reload ref-data from memory) but require more data move between registers (may be future test which is the best for speed).
Also the write of each sads may be better combine to xmm or ymm register to perform less write operations (that cause additional cache-coherence traffic at multi-core systems and more not very needed operations).

Some sad note from my own practice: Unfortunately the SIMD instructions sets (up to AVX2 and still most of AVX512) still have limited ability of 'horizontal' data shift on the required number of bytes with 'very long vectors'. The typical image-processing task is shift of small enough 'processing frame' like convolution kernel or block of data to SAD across long block of data like part of image line (best is the full line) and with 'sample-granularity' that is usually 8bit or 16 (of 32 for float). And looks like AVX256 (and probably 512 too) do not have direct short instructions of shifting even inside one 256bit 'register' (the used now _mm256_alignr_epi8 with 8bit granularity limited to pair of 128bit lanes only, for AVX512 _mm512_alignr_epi8 it looks divided to 4 lanes of 128bit) and no instructions can shift 'very large vectors' stored in different 'registers'. So current approach is to rotate data as possible (typically down to 32bit units only - not 8 or 16) in one 256bit chuck and use blend of the 'shifted out bits to next 256bit chunk. May be good to write a petition to intel to add new instructions to new chips for shifting of 'very large vectors' (like to several hundreds of bits) between 'registers sets' with granularity down to 8bit. It may make H-stepping at typical image processing easier and faster I think. Though it looks the hardware of register file is also not very suitable to shift data in rows/columns of some 2D memory array at low steps. So typical instructions are about fetching 128/256/512 chunk of data from register file to execution port - perform shift/rotate inside chunk and store back. Also may be 8-bit granularity shift/rotate for 256/512 bit chunks is also not very cheap.

[zorr]

Unfortunately the SIMD instructions sets (up to AVX2 and still most of AVX512) still have limited ability of 'horizontal' data shift on the required number of bytes with 'very long vectors'.

Some related StackOverflow answers, perhaps not faster than what you're already doing.

https://stackoverflow.com/questions/25248766/emulating-shifts-on-32-bytes-with-avx
https://stackoverflow.com/questions/58322652/emulating-shifts-on-64-bytes-with-avx-512

[DTL]

There is also semi-optimal way: Where applicable do not perform completely sequential stepping with 8bit granularity but change stepping to 4*8 bit granularity and use 32bit shifts (permutes). And perform 4 passes with +3 more reloads from memory with shifted to 8bit address. So process first pass 0,3,7,11,.. starting positions; next pass 1,4,8,12; next 2,5,9,13 and so on.

[zorr]

change stepping to 4*8 bit granularity and use 32bit shifts (permutes). And perform 4 passes with +3 more reloads from memory with shifted to 8bit address.

Or do the one byte shift in between passes instead of reading from the memory. I guess it depends on how much data each pass processes, if it's too much the data falls out of cache and reloading will be slower.

[DTL]

Not so bad for AVX512 at desktop CPUs - intel Rocket Lake Q1 2021 chips finally have it. Cnews make some testing of x265 performance of AVX512 - https://www.google.com/amp/s/www.hwcooling.net/en/intel-avx-512-tested-in-x265-how-to-enable-it-and-does-it-help/%3famp=1
The speedup still very poor but it looks development of AVX512 in x265 were slow because there were very few AVX512 chips at endusers available. Hope to make tests soon with i5-11500 of AVX512 versions of some functions in MAnalyse and MDegrainN in mvtools.
The AVX512 have both scatter and gather instructions and AVX2 only gather. And MDegrainN now very poor on memory performance - may scatter/gather of blocks store/load will helps.

[orion44]

I read somewhere that modern optimizing C/C++ compilers are able to beat hand-coded assembly in like 99% of cases.

Is this true in the area of video codec and video filters development?

[DTL]

The speedup is not about compiler vs hand produced assembly executable. It is mostly about compiler vs human "vectorization". As the intel still going into idea of "universal logical core + built-in large vector co-processor" it is required to offload moving pictures data processing to this large vector co-processor. And it reqiure to re-arrange of data structures and variables computations in SIMD-friendly way. But compilers still very poor in producing "auto-vectorized" result. Too much intelligence required for significant program re-write to be compatible with large vector computation on SIMD execution units. Also it may sometime produce different result that is forbidden to optimizing compilers. So in todays work in progress for mvtools the new SIMD processing is separated from old C-version to keep exact compatibility with old versions at output.

Also about non-temporal stores - it require both store writecombining to cacheline size (typically 64bytes) and runtime knowledge if it will help to total execution or not depending on task size and cache size of current CPU. Though possibly profile-guided optimization on intel compiler can somehow approach any help in this problem. But it may help at developer task and CPU and not help at user-side with different task and CPU.

Also as that testing show the AVX512 large vector co-processor is not only 4 times larger in register file size and twice larger in bus width in compare with AVX2 but also very visibly power-hungry. So it is now started two different branches of general purpose CPUs - intel more or less continue to increase size and performance of large vector co-processor near general purpose core and AMD start to increase number of general purpose cores with AVX2 max. So now software for intel need to be optimized for low threads number and AVX512 large vector processing and for AMD - large threads number and AVX2 max. AMD looks like wins for desktop because it is harder to re-write software for AVX512 and much easier to increase threads number.

[DTL]

Heh - the magic of AVX512 time: sad of block 8x8 8bit:


__m512i zmm_src = _mm512_set_epi64(*(pSrc + nSrcPitch * 7), *(pSrc + nSrcPitch * 6), *(pSrc + nSrcPitch * 5), *(pSrc + nSrcPitch * 4), \
* (pSrc + nSrcPitch * 3), *(pSrc + nSrcPitch * 2), *(pSrc + nSrcPitch * 1), *(pSrc + nSrcPitch * 0));

__m512i zmm_ref = _mm512_set_epi64(*(pRef + nRefPitch * 7), *(pRef + nRefPitch * 6), *(pRef + nRefPitch * 5), *(pRef + nRefPitch * 4), \
* (pRef + nRefPitch * 3), *(pRef + nRefPitch * 2), *(pRef + nRefPitch * 1), *(pRef + nRefPitch * 0));

return _mm512_reduce_add_epi64(_mm512_sad_epu8(zmm_src, zmm_ref));


It may be fit even in 1 line: reduce_add(sad(set(src), set(ref))); . 3 'macros' and 1 'real instruction'. Macros expansion is compiler-dependent and possibly can be better at next versions of compiler and target chip.
In poor old times it is 64 subtractions + 64 abs + dozens of additions.

[DTL]

For really high speed motion pictures data processing using SIMD co-processor in host CPU looks like outdated about in the 201x years already. Now best way is understanding Compute Shaders executed by separated HW accelerator in the system (typically with dedicated memory of much higher speed - from GDDR to HBM). It is much more 2D processing-oriented and saves user from dealing with lots of manual dealing with samples scan order and memory-management and applying 2D processing to SIMD co-processor capabilities. Also it is naturally massive-multithreaded inside 'frame' processing for free - no need to deal with MT in OS. So developer of shader only implement its design of real data processing and free from lots of overhead for typical multithreaded program on general-purpose CPU.
In hardware it somehow supported from DirectX-10.x and now more advanced in DX12.
It have API from DirectX and OpenGL. The OpenGL possibly cross-platform compatible with Linux too. For OpenGL some description: https://www.khronos.org/opengl/wiki/Compute_Shader
So for Avisynth interfacing with Compute Shaders processing it currently only required to apply resources upload to HW acc and result download and setup 'compute pipeline' (it is more or less smaller in complexity with 'full-3D rendering pipeline') - https://docs.microsoft.com/en-us/windows/win32/direct3d12/pipelines-and-shaders-with-directx-12 . It may be created some 'universal' CS-Plugin SDK with 'Hello ComputeShader' sample to start with.
The more operations executed in the HW Acc - the less overhead for data uploading-downloading from host memory to accelerator and back. Most benefit is expected for memory-bound operations like massive-multiframe denoisers.