Compare commits
26 Commits
| Author | SHA1 | Date | |
|---|---|---|---|
|
|
6dd7949030 | ||
|
|
a27f8221cd | ||
|
|
8cff776c5e | ||
|
|
bae9e8a0d8 | ||
|
|
bbfad5b9df | ||
|
|
fd24693c6b | ||
|
|
cc9d3595bd | ||
|
|
1b397d8976 | ||
|
|
8c06ad76b6 | ||
|
|
50b735d3a5 | ||
|
|
5660fcf7c5 | ||
|
|
e979c07600 | ||
|
|
cd7d1f9450 | ||
|
|
cdaceef840 | ||
|
|
6ce606977b | ||
|
|
fb72e811d0 | ||
|
|
82197ac0d0 | ||
|
|
959cf0e8d4 | ||
|
|
bcbdc2fe38 | ||
|
|
ca2c7731d4 | ||
|
|
aeb8913a58 | ||
|
|
323a53d808 | ||
|
|
54931fc0e7 | ||
|
|
3024186aa6 | ||
|
|
ceed4ebd83 | ||
|
|
471246df18 |
87
README.md
87
README.md
@@ -1,60 +1,63 @@
|
||||
# MultiPar
|
||||
|
||||
### v1.3.2.9 is public
|
||||
This is the final release of v1.3.2 tree.
|
||||
Because I want to public this as a stable version, I didn't change contents so much.
|
||||
PAR clients are same as previous version.
|
||||
Including long term used applications may be good to avoid false positive at Malware detection.
|
||||
### v1.3.3.1 is public
|
||||
|
||||
I fixed a [compatibility issue in calling 7-Zip](https://github.com/Yutaka-Sawada/MultiPar/issues/92),
|
||||
which I didn't know the change.
|
||||
Thanks Lyoko-Jeremie for bug report.
|
||||
The incident happened, when a user selected many files.
|
||||
This is a testing version to improve speed of PAR2 calculation.
|
||||
Because the new method isn't tested so much, there may be a bug, failure, or mistake.
|
||||
Be careful to use this non-stable version.
|
||||
When you don't want to test by yourself, you should not use this yet.
|
||||
If you see a problem, please report the incident.
|
||||
I will try to solve as possible as I can.
|
||||
|
||||
I made a sample feature to Save & Restore different "base directories".
|
||||
When you put PAR files in another folder from source files, it will set the previous directory automatically.
|
||||
Because this feature was tested little, it's disabled by default at this time.
|
||||
If you want to enable, add section `[Path]` on "MultiPar.ini".
|
||||
Then set `MRUMax` value, which is the maximum number of stored directries.
|
||||
You may set the value upto 26. It's disabled, when the value is 0.
|
||||
These two lines are like below:
|
||||
```
|
||||
[Path]
|
||||
MRUMax=5
|
||||
```
|
||||
CPU's L3 cache optimization depends on hardware environment.
|
||||
It's difficult to guess the best setting for unknown type.
|
||||
It seems to work well on Intel and AMD 's most CPUs.
|
||||
Thanks Anime Tosho and MikeSW17 for long tests.
|
||||
But, I'm not sure the perfomance of rare strange kind CPUs.
|
||||
If you want to compare speed of different settings on your CPU,
|
||||
you may try samples (TestBlock_2023-08-31.zip) in "MultiPar_sample" folder
|
||||
on [OneDrive](https://1drv.ms/u/s!AtGhNMUyvbWOg0cF2UHcs709Icv4).
|
||||
|
||||
While I made MultiPar as an utility tool, I didn't give priority to its speed.
|
||||
If someone wants faster Parchive tool, I suggest to use ParPar tools instead of MultiPar.
|
||||
They are "[High performance PAR2 create client for NodeJS](https://github.com/animetosho/ParPar)" or
|
||||
"[speed focused par2cmdline fork](https://github.com/animetosho/par2cmdline-turbo)".
|
||||
Though the speed depends on hardware environments and user's setting, it would be 50% ~ 100 % faster than my par2j.
|
||||
Only when you have a very fast graphics borad, GPU enabled par2j may be faster.
|
||||
I plan to improve speed of par2j in next v1.3.3 tree.
|
||||
Though it will become 20% ~ 30% faster than old par2j, ParPar would be faster mostly.
|
||||
I improved GPU implementation very much.
|
||||
Thanks [Slava46 and K2M74 for many tests](https://github.com/Yutaka-Sawada/MultiPar/issues/99).
|
||||
While I almost gave up to increase speed, their effort encouraged me to try many ways.
|
||||
Without their aid, I could not implement this GPU function.
|
||||
OpenCL perfomance is varied in every graphics boards.
|
||||
If you have a fast graphics board, enabling "GPU acceleration" would be faster.
|
||||
If it's not so fast (or is slow) on your PC, just un-check the feature.
|
||||
|
||||
I saw a new feature of Inno Setup 6, which changes install mode.
|
||||
It shows a dialog to ask which install mode.
|
||||
Then, a user can install MultiPar in "Program Files" directory by selecting "Install for all users".
|
||||
This method may be easier than starting installer by "Run as administrator".
|
||||
I test the selection dialog at this version.
|
||||
If there is no problem nor complaint from users, I use this style in later versions, too.
|
||||
|
||||
|
||||
[ Changes from 1.3.2.8 to 1.3.2.9 ]
|
||||
[ Changes from 1.3.3.0 to 1.3.3.1 ]
|
||||
|
||||
GUI update
|
||||
- New
|
||||
- Verification may save different base directories in MultiPar.ini file.
|
||||
Installer update
|
||||
- It shows dialog to select "per user" or "per machine" installation.
|
||||
|
||||
- Bug fix
|
||||
- Archiver's option was updated for recent 7-Zip versions.
|
||||
PAR2 client update
|
||||
- Change
|
||||
- Max number of threads to read files on SSD was increased to 6.
|
||||
|
||||
- Improvement
|
||||
- GPU acceleration would become faster.
|
||||
|
||||
|
||||
[ Hash value ]
|
||||
|
||||
MultiPar132.zip
|
||||
MD5: 305D86C8C7A0F5C1A23CEAFFBE4F02BF
|
||||
SHA1: 464BB7AB7D14FD35D2AEF99042EEB8E556DA0417
|
||||
MultiPar1331.zip
|
||||
MD5: ECFC1570C839DD30A2492A7B05C2AD6E
|
||||
SHA1: 5E0E4CC38DAA995294A93ECA10AEB3AE84596170
|
||||
|
||||
MultiPar132_setup.exe
|
||||
MD5: 18F9BE1FF1C6D668E3A3906C691CCB98
|
||||
SHA1: 116C6B2A15FCFD9BB74F0EF9D6C8A4BF78299588
|
||||
MultiPar1331_setup.exe
|
||||
MD5: A55E6FA5A6853CB42E3410F35706BAD9
|
||||
SHA1: 8D46BD6702E82ABA9ACCFA5223B2763B4DCEFE9E
|
||||
To install under "Program Files" or "Program Files (x86)" directory,
|
||||
you must start the installer with administrative privileges by selecting
|
||||
"Run as administrator" on right-click menu.
|
||||
you must select "Install for all users" at the first dialog.
|
||||
|
||||
Old versions and source code packages are available at
|
||||
[GitHub](https://github.com/Yutaka-Sawada/MultiPar/releases) or
|
||||
|
||||
Binary file not shown.
@@ -71,14 +71,13 @@ Don't send current PAR3 files to others, who may not have the same version.
|
||||
|
||||
[ How to install or uninstall with installer package ]
|
||||
|
||||
Double click setup file ( MultiPar131_setup.exe or something like this name ),
|
||||
Double click setup file ( MultiPar133_setup.exe or something like this name ),
|
||||
and follow the installer dialog.
|
||||
At version up, if you want to use previous setting, overwrite install is possible.
|
||||
Before overwrite install, you should un-check "Integrate MultiPar into Shell".
|
||||
You may need to re-start OS after overwrite install or uninstall rarely.
|
||||
To install under "Program Files" or "Program Files (x86)" directory,
|
||||
you must start the installer with administrative privileges by selecting
|
||||
"Run as administrator" on right-click menu.
|
||||
you must select "Install for all users" at the first dialog.
|
||||
|
||||
You can uninstall through the Windows OS's Control Panel,
|
||||
or double click unins000.exe in a folder which MultiPar was installed.
|
||||
@@ -100,7 +99,7 @@ In either case, user made icons and association are available for the user only.
|
||||
|
||||
[ How to install with archive version ]
|
||||
|
||||
Unpack compressed file ( MultiPar131.zip or something like this name ) in a folder.
|
||||
Unpack compressed file ( MultiPar133.zip or something like this name ) in a folder.
|
||||
MultiPar.exe is the interface of MultiPar.
|
||||
|
||||
You can create short-cut icon or send-to link at Option window later.
|
||||
|
||||
@@ -107,12 +107,14 @@ PAR 3.0 仕様のフォーマットは細部が流動的で最終版との互換
|
||||
|
||||
[ インストーラー版のインストールとアンインストール ]
|
||||
|
||||
インストーラー ( MultiPar131_setup.exe みたいな名前 ) をダブル・クリックすると、
|
||||
インストーラー ( MultiPar133_setup.exe みたいな名前 ) をダブル・クリックすると、
|
||||
インストール画面が表示されるので、その指示に従ってください。
|
||||
バージョン・アップ時に、設定項目をそのまま使いたい時は上書きインストールしてもいいです。
|
||||
上書きインストールする前に、「MultiPar をシェルに統合する」のチェックを外してください。
|
||||
上書きインストールやアンインストール後に OS の再起動を求められるかもしれません。
|
||||
「Program Files」や「Program Files (x86)」内にインストールするには、
|
||||
最初のダイアログで「すべてのユーザー用にインストール」を選んでください。
|
||||
|
||||
右クリック・メニューの「管理者として実行」を選んで
|
||||
管理者権限でインストーラーを開始する必要があります。
|
||||
|
||||
@@ -137,7 +139,7 @@ MultiPar をインストールしたフォルダ内の unins000.exe をダブル
|
||||
|
||||
[ アーカイブ版のインストール ]
|
||||
|
||||
配布されてる圧縮ファイル ( MultiPar131.zip みたいな名前 ) を解凍してできたファイルを
|
||||
配布されてる圧縮ファイル ( MultiPar133.zip みたいな名前 ) を解凍してできたファイルを
|
||||
どこか適当なフォルダに全て入れてください。
|
||||
その中の MultiPar.exe というのが MultiPar の実行ファイルです。
|
||||
それをダブル・クリックすると MultiPar が起動します。
|
||||
|
||||
@@ -1,3 +1,37 @@
|
||||
Release note of v1.3.3 tree
|
||||
|
||||
[ Changes from 1.3.3.0 to 1.3.3.1 ] (2023/11/11)
|
||||
|
||||
Installer update
|
||||
It shows dialog to select "per user" or "per machine" installation.
|
||||
|
||||
PAR2 client update
|
||||
Change
|
||||
Max number of threads to read files on SSD was increased to 6.
|
||||
|
||||
Improvement
|
||||
GPU acceleration would become faster.
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
[ Changes from 1.3.2.9 to 1.3.3.0 ] (2023/10/10)
|
||||
|
||||
GUI update
|
||||
Change
|
||||
Option adapted to new "lc" settings.
|
||||
It's possible to add 5th item in "Media size" on Create window.
|
||||
|
||||
PAR2 client update
|
||||
Change
|
||||
Max number of using threads is increased to 32.
|
||||
Threshold to use GPU was decreased.
|
||||
|
||||
Improvement
|
||||
Matrix inversion may use more threads.
|
||||
L3 cache optimization was improved for recent CPUs.
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
Release note of v1.3.2 tree
|
||||
|
||||
I tried to decrease probability of false positive at Malware detection.
|
||||
|
||||
@@ -1,9 +1,21 @@
|
||||
v1.3.2 の更新情報 (2023/08/26)
|
||||
v1.3.3 の更新情報 (2023/11/11)
|
||||
|
||||
[ 1.3.1 から 1.3.2 への変更点 ]
|
||||
まだ動作実験中ですので、不安な人は前のバージョンを使ってください。
|
||||
|
||||
[ 1.3.2 から 1.3.3 への変更点 ]
|
||||
|
||||
・クライアントの変更点
|
||||
CPU Cache の利用方法を改善して速くなりました。
|
||||
GPU による高速化も速くなりました。
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
v1.3.2 の更新情報 (2023/08/26)
|
||||
|
||||
マルウェアとして誤検知されにくいようにしました。
|
||||
|
||||
[ 1.3.1 から 1.3.2 への変更点 ]
|
||||
|
||||
・GUI の変更点
|
||||
動作完了時に Python スクリプトを呼び出せるようにしました。
|
||||
Python スクリプトで作った便利ツールをいくつか追加しました。
|
||||
|
||||
@@ -51,7 +51,7 @@ There are command-line manuals in "<tt>help</tt>" folder.
|
||||
</table>
|
||||
|
||||
<hr>
|
||||
<small>last update 2023/06/13 for version 1.3.2.9</small>
|
||||
<small>last update 2023/11/11 for version 1.3.3.1</small>
|
||||
|
||||
</body>
|
||||
</html>
|
||||
|
||||
@@ -8,14 +8,13 @@
|
||||
|
||||
<h3>Install or uninstall with installer package</h3>
|
||||
<p> 
|
||||
Double click setup file ( <tt>MultiPar131_setup.exe</tt> or something like this name ),
|
||||
Double click setup file ( <tt>MultiPar133_setup.exe</tt> or something like this name ),
|
||||
and follow the installer dialog.
|
||||
At version up, if you want to use previous setting, overwrite install is possible.
|
||||
Before overwrite install, you should un-check "Integrate MultiPar into Shell".
|
||||
You may need to re-start OS after overwrite install or uninstall rarely.
|
||||
To install under "<tt>Program Files</tt>" or "<tt>Program Files (x86)</tt>" directory,
|
||||
you must start the installer with administrative privileges by selecting
|
||||
"Run as administrator" on right-click menu.
|
||||
you must select "Install for all users" at the first dialog.
|
||||
</p>
|
||||
<p> 
|
||||
You can uninstall through the Windows OS's Control Panel,
|
||||
@@ -42,7 +41,7 @@ In either case, user made icons and association are available for the user only.
|
||||
|
||||
<h3>Install with archive version</h3>
|
||||
<p> 
|
||||
Unpack compressed file ( <tt>MultiPar131.zip</tt> or something like this name ) in a folder.
|
||||
Unpack compressed file ( <tt>MultiPar133.zip</tt> or something like this name ) in a folder.
|
||||
<tt>MultiPar.exe</tt> is the interface of MultiPar.
|
||||
</p>
|
||||
<p> 
|
||||
|
||||
@@ -51,7 +51,7 @@
|
||||
</table>
|
||||
|
||||
<hr>
|
||||
<small><EFBFBD>ŏI<EFBFBD>X<EFBFBD>V 2023/02/27 (<28>o<EFBFBD>[<5B>W<EFBFBD><57><EFBFBD><EFBFBD> 1.3.2.8)</small>
|
||||
<small><EFBFBD>ŏI<EFBFBD>X<EFBFBD>V 2023/11/11 (<28>o<EFBFBD>[<5B>W<EFBFBD><57><EFBFBD><EFBFBD> 1.3.3.1)</small>
|
||||
|
||||
</body>
|
||||
</html>
|
||||
|
||||
@@ -8,14 +8,13 @@
|
||||
|
||||
<h3><EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD>[<5B>ł̃C<CC83><43><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD>ƃA<C683><41><EFBFBD>C<EFBFBD><43><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><></h3>
|
||||
<p> 
|
||||
<EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD>[ ( <tt>MultiPar131_setup.exe</tt> <20>݂<EFBFBD><DD82><EFBFBD><EFBFBD>Ȗ<EFBFBD><C896>O ) <20><><EFBFBD>_<EFBFBD>u<EFBFBD><75><EFBFBD>E<EFBFBD>N<EFBFBD><4E><EFBFBD>b<EFBFBD>N<EFBFBD><4E><EFBFBD><EFBFBD><EFBFBD>ƁA
|
||||
<EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD>[ ( <tt>MultiPar133_setup.exe</tt> <20>݂<EFBFBD><DD82><EFBFBD><EFBFBD>Ȗ<EFBFBD><C896>O ) <20><><EFBFBD>_<EFBFBD>u<EFBFBD><75><EFBFBD>E<EFBFBD>N<EFBFBD><4E><EFBFBD>b<EFBFBD>N<EFBFBD><4E><EFBFBD><EFBFBD><EFBFBD>ƁA
|
||||
<EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD><EFBFBD>ʂ<EFBFBD><CA82>\<5C><><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>̂ŁA<C581><41><EFBFBD>̎w<CC8E><77><EFBFBD>ɏ]<5D><><EFBFBD>Ă<EFBFBD><C482><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>B
|
||||
<EFBFBD>o<EFBFBD>[<5B>W<EFBFBD><57><EFBFBD><EFBFBD><EFBFBD>E<EFBFBD>A<EFBFBD>b<EFBFBD>v<EFBFBD><76><EFBFBD>ɁA<C981>ݒ荀<DD92>ڂ<EFBFBD><DA82><EFBFBD><EFBFBD>̂܂g<DC8E><67><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>͏㏑<CD8F><E38F91><EFBFBD>C<EFBFBD><43><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD><EFBFBD>Ă<EFBFBD><C482><EFBFBD><EFBFBD><EFBFBD><EFBFBD>ł<EFBFBD><C582>B
|
||||
<EFBFBD>㏑<EFBFBD><EFBFBD><EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>O<EFBFBD>ɁA<C981>uMultiPar <20><><EFBFBD>V<EFBFBD>F<EFBFBD><46><EFBFBD>ɓ<EFBFBD><C993><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>v<EFBFBD>̃`<60>F<EFBFBD>b<EFBFBD>N<EFBFBD><4E><EFBFBD>O<EFBFBD><4F><EFBFBD>Ă<EFBFBD><C482><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>B
|
||||
<EFBFBD>㏑<EFBFBD><EFBFBD><EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD><EFBFBD>A<EFBFBD><41><EFBFBD>C<EFBFBD><43><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD><EFBFBD><EFBFBD> OS <20>̍ċN<C48B><4E><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>߂<EFBFBD><DF82><EFBFBD><EFBFBD>邩<EFBFBD><E982A9><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>܂<EFBFBD><DC82><EFBFBD><EFBFBD>B
|
||||
<EFBFBD>u<tt>Program Files</tt><EFBFBD>v<EFBFBD><EFBFBD><EFBFBD>u<tt>Program Files (x86)</tt><EFBFBD>v<EFBFBD><EFBFBD><EFBFBD>ɃC<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>ɂ́A
|
||||
<EFBFBD>E<EFBFBD>N<EFBFBD><EFBFBD><EFBFBD>b<EFBFBD>N<EFBFBD>E<EFBFBD><EFBFBD><EFBFBD>j<EFBFBD><EFBFBD><EFBFBD>[<5B>́u<CC81>Ǘ<EFBFBD><C797>҂Ƃ<D282><C682>Ď<EFBFBD><C48E>s<EFBFBD>v<EFBFBD><76><EFBFBD>I<EFBFBD><49><EFBFBD><EFBFBD>
|
||||
<EFBFBD>Ǘ<EFBFBD><EFBFBD>Ҍ<EFBFBD><EFBFBD><EFBFBD><EFBFBD>ŃC<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD>[<5B><><EFBFBD>J<EFBFBD>n<EFBFBD><6E><EFBFBD><EFBFBD><EFBFBD>K<EFBFBD>v<EFBFBD><76><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>܂<EFBFBD><DC82>B
|
||||
<EFBFBD>ŏ<EFBFBD><EFBFBD>̃_<EFBFBD>C<EFBFBD>A<EFBFBD><EFBFBD><EFBFBD>O<EFBFBD>Łu<EFBFBD><EFBFBD><EFBFBD>ׂẴ<EFBFBD><EFBFBD>[<5B>U<EFBFBD>[<5B>p<EFBFBD>ɃC<C983><43><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD>v<EFBFBD><76><EFBFBD>I<EFBFBD><49><EFBFBD>ł<EFBFBD><C582><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>B
|
||||
</p>
|
||||
<p> 
|
||||
<EFBFBD>A<EFBFBD><EFBFBD><EFBFBD>C<EFBFBD><EFBFBD><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><><EFBFBD><EFBFBD> Windows OS <20>̃R<CC83><52><EFBFBD>g<EFBFBD><67><EFBFBD>[<5B><><EFBFBD>E<EFBFBD>p<EFBFBD>l<EFBFBD><6C><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>s<EFBFBD><73><EFBFBD><EFBFBD><EFBFBD>A
|
||||
@@ -43,7 +42,7 @@ MultiPar
|
||||
|
||||
<h3><EFBFBD>A<EFBFBD>[<5B>J<EFBFBD>C<EFBFBD>u<EFBFBD>ł̃C<CC83><43><EFBFBD>X<EFBFBD>g<EFBFBD>[<5B><></h3>
|
||||
<p> 
|
||||
<EFBFBD>z<EFBFBD>z<EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>Ă鈳<EFBFBD>k<EFBFBD>t<EFBFBD>@<40>C<EFBFBD><43> ( <tt>MultiPar131.zip</tt> <20>݂<EFBFBD><DD82><EFBFBD><EFBFBD>Ȗ<EFBFBD><C896>O ) <20><><EFBFBD>𓀂<EFBFBD><F0938082>Ăł<C482><C582><EFBFBD><EFBFBD>t<EFBFBD>@<40>C<EFBFBD><43><EFBFBD><EFBFBD>
|
||||
<EFBFBD>z<EFBFBD>z<EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>Ă鈳<EFBFBD>k<EFBFBD>t<EFBFBD>@<40>C<EFBFBD><43> ( <tt>MultiPar133.zip</tt> <20>݂<EFBFBD><DD82><EFBFBD><EFBFBD>Ȗ<EFBFBD><C896>O ) <20><><EFBFBD>𓀂<EFBFBD><F0938082>Ăł<C482><C582><EFBFBD><EFBFBD>t<EFBFBD>@<40>C<EFBFBD><43><EFBFBD><EFBFBD>
|
||||
<EFBFBD>ǂ<EFBFBD><EFBFBD><EFBFBD><EFBFBD>K<EFBFBD><EFBFBD><EFBFBD>ȃt<EFBFBD>H<EFBFBD><EFBFBD><EFBFBD>_<EFBFBD>ɑS<EFBFBD>ē<EFBFBD><EFBFBD><EFBFBD><EFBFBD>Ă<EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>B
|
||||
<EFBFBD><EFBFBD><EFBFBD>̒<EFBFBD><EFBFBD><EFBFBD> <tt>MultiPar.exe</tt> <20>Ƃ<EFBFBD><C682><EFBFBD><EFBFBD>̂<EFBFBD> MultiPar <20>̎<EFBFBD><CC8E>s<EFBFBD>t<EFBFBD>@<40>C<EFBFBD><43><EFBFBD>ł<EFBFBD><C582>B
|
||||
<EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD>_<EFBFBD>u<EFBFBD><EFBFBD><EFBFBD>E<EFBFBD>N<EFBFBD><EFBFBD><EFBFBD>b<EFBFBD>N<EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD><EFBFBD> MultiPar <20><><EFBFBD>N<EFBFBD><4E><EFBFBD><EFBFBD><EFBFBD>܂<EFBFBD><DC82>B
|
||||
|
||||
@@ -50,7 +50,7 @@
|
||||
</table>
|
||||
|
||||
<hr>
|
||||
<small>最新更新于2023年2月27日,适用于1.3.2.8版本,简体中文化 Deng Shiqing</small>
|
||||
<small>最新更新于2023年11月11日,适用于1.3.3.1版本,简体中文化 Deng Shiqing</small>
|
||||
|
||||
</body>
|
||||
</html>
|
||||
|
||||
@@ -8,7 +8,7 @@
|
||||
|
||||
<h3>使用安装包安装或卸载</h3>
|
||||
<p> 
|
||||
双击安装文件(<tt>MultiPar131_setup.exe</tt>或类似名称文件),然后按照安装程序对话框进行操作。在版本升级时,如果要使用先前的设置,可以进行覆盖安装。在覆盖安装之前, 应取消勾选“将MultiPar整合到右键菜单”。在写入安装或卸载之后,您可能需要重新启动操作系统。“<tt>Program Files</tt>”或“<tt>Program Files (x86)</tt>”目录下,必须在右键菜单上选择“以管理员身份运行”,用管理员权限启动安装程序。
|
||||
双击安装文件(<tt>MultiPar133_setup.exe</tt>或类似名称文件),然后按照安装程序对话框进行操作。在版本升级时,如果要使用先前的设置,可以进行覆盖安装。在覆盖安装之前, 应取消勾选“将MultiPar整合到右键菜单”。在写入安装或卸载之后,您可能需要重新启动操作系统。要在“<tt>Program Files</tt>”或“<tt>Program Files (x86)</tt>”目录下安装,您必须在第一个对话框中选择“为所有用户安装”。
|
||||
</p>
|
||||
<p> 
|
||||
您可以通过Windows操作系统的控制面板卸载程序,或双击MultiPar安装文件夹中的<tt>unins000.exe</tt>。由于卸载程序不会删除设置文件或安装后新添加的文件,因此您可以自行删除它们。
|
||||
@@ -26,7 +26,7 @@
|
||||
|
||||
<h3>使用压缩包安装</h3>
|
||||
<p> 
|
||||
在文件夹中解压压缩文件(<tt>MultiPar131.zip</tt>或类似名称文件)。
|
||||
在文件夹中解压压缩文件(<tt>MultiPar133.zip</tt>或类似名称文件)。
|
||||
<tt>MultiPar.exe</tt>是MultiPar的启动程序。
|
||||
</p>
|
||||
<p> 
|
||||
|
||||
@@ -1,4 +1,4 @@
|
||||
[ MultiPar GUI - version 1.3.2.9 or later ]
|
||||
[ MultiPar GUI - version 1.3.3.0 or later ]
|
||||
|
||||
Usage: MultiPar.exe [command] [/base path] [/list path] [files]
|
||||
|
||||
@@ -106,6 +106,12 @@ Because of alphabet, "MRUMax=26" is the maximum.
|
||||
If you write "MRUMax=0" or remove the line,
|
||||
"Most Resent Used List" is disabled.
|
||||
|
||||
If you want to add 5th item in "Media size" list on Create window,
|
||||
write this line "MediaList4=name:size" under "[Option]" section.
|
||||
Because the name and size are splitted by ":",
|
||||
you cannot include ":" in the name.
|
||||
The max size is 999999999999 bytes. (931 GB)
|
||||
|
||||
|
||||
Example of lines on "MultiPar.ini";
|
||||
|
||||
@@ -114,5 +120,6 @@ FontName=Arial
|
||||
RedundancyMax=10
|
||||
Sort=8
|
||||
RecoveryFileLimit=1
|
||||
MediaList4=7.9GB DVD:8480000000
|
||||
[Path]
|
||||
MRUMax=5
|
||||
|
||||
@@ -1,4 +1,4 @@
|
||||
[ par2j.exe - version 1.3.2.8 or later ]
|
||||
[ par2j.exe - version 1.3.3.0 or later ]
|
||||
|
||||
Type "par2j.exe" to see version, test integrity, and show usage below.
|
||||
|
||||
@@ -359,19 +359,22 @@ the protected archive file is made in the directory.
|
||||
/lc :
|
||||
Set this, if you want to set number of using threads for Multi-Core CPU,
|
||||
or want to disable extra feature. (SSE2 is always used.)
|
||||
The format is "/lc#", # is from 1 to 11 as the number of using threads,
|
||||
12 to use quarter number of physical Cores,
|
||||
13 to use half of physical Cores,
|
||||
14 to use 3/4 number of physical Cores,
|
||||
15 to use the number of physical Cores (disable Hyper Threading),
|
||||
or 15 to use one less number of physical Cores on CPU with 6 or more physical Cores.
|
||||
Without this option (or /lc0),
|
||||
it uses the number of physical Cores on CPU with 6 or more physical Cores,
|
||||
or one more threads on CPU with Hyper Threading and 5 or less physical Cores.
|
||||
The format is "/lc#" (# is from 1 to 32 as the number of using threads).
|
||||
|
||||
You may set additional combinations; +16 to disable SSSE3,
|
||||
+128 to disable CLMUL, +256 to disable JIT, +512 to disable AVX2,
|
||||
+32 or +64 (slower device) to enable GPU acceleration.
|
||||
It's possible to set by rate as following. (It's /lc0 by default.)
|
||||
251: It uses quarter number of physical Cores.
|
||||
252: It uses half of physical Cores.
|
||||
253: It uses 3/4 number of physical Cores.
|
||||
254: It uses one less threads than number of physical Cores.
|
||||
0: It uses the number of physical Cores.
|
||||
255: It uses one more threads than number of physical Cores.
|
||||
|
||||
You may set additional combinations;
|
||||
+1024 to disable CLMUL (and use old SSSE3 code),
|
||||
+2048 to disable JIT (for SSE2),
|
||||
+4096 to disable SSSE3,
|
||||
+8192 to disable AVX2,
|
||||
+256 or +512 (slower device) to enable GPU acceleration.
|
||||
|
||||
for example, /lc1 to use single Core, /lc45 to use half Cores and GPU
|
||||
|
||||
|
||||
BIN
alpha/par2j.exe
BIN
alpha/par2j.exe
Binary file not shown.
Binary file not shown.
@@ -1,4 +1,4 @@
|
||||
[ par2j.exe - version 1.3.2.8 or later ]
|
||||
[ par2j.exe - version 1.3.3.1 or later ]
|
||||
|
||||
Type "par2j.exe" to see version, test integrity, and show usage below.
|
||||
|
||||
@@ -359,19 +359,22 @@ the protected archive file is made in the directory.
|
||||
/lc :
|
||||
Set this, if you want to set number of using threads for Multi-Core CPU,
|
||||
or want to disable extra feature. (SSE2 is always used.)
|
||||
The format is "/lc#", # is from 1 to 11 as the number of using threads,
|
||||
12 to use quarter number of physical Cores,
|
||||
13 to use half of physical Cores,
|
||||
14 to use 3/4 number of physical Cores,
|
||||
15 to use the number of physical Cores (disable Hyper Threading),
|
||||
or 15 to use one less number of physical Cores on CPU with 6 or more physical Cores.
|
||||
Without this option (or /lc0),
|
||||
it uses the number of physical Cores on CPU with 6 or more physical Cores,
|
||||
or one more threads on CPU with Hyper Threading and 5 or less physical Cores.
|
||||
The format is "/lc#" (# is from 1 to 32 as the number of using threads).
|
||||
|
||||
You may set additional combinations; +16 to disable SSSE3,
|
||||
+128 to disable CLMUL, +256 to disable JIT, +512 to disable AVX2,
|
||||
+32 or +64 (slower device) to enable GPU acceleration.
|
||||
It's possible to set by rate as following. (It's /lc0 by default.)
|
||||
251: It uses quarter number of physical Cores.
|
||||
252: It uses half of physical Cores.
|
||||
253: It uses 3/4 number of physical Cores.
|
||||
254: It uses one less threads than number of physical Cores.
|
||||
0: It uses the number of physical Cores.
|
||||
255: It tries to use more threads than number of physical Cores.
|
||||
|
||||
You may set additional combinations;
|
||||
+1024 to disable CLMUL (and use slower SSSE3 code),
|
||||
+2048 to disable JIT (for SSE2),
|
||||
+4096 to disable SSSE3,
|
||||
+8192 to disable AVX2,
|
||||
+256 or +512 (slower device) to enable GPU acceleration.
|
||||
|
||||
for example, /lc1 to use single Core, /lc45 to use half Cores and GPU
|
||||
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// common2.c
|
||||
// Copyright : 2023-03-14 Yutaka Sawada
|
||||
// Copyright : 2023-10-13 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -1848,9 +1848,10 @@ int sqrt32(int num)
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
int cpu_num = 1; // CPU/Core 個数が制限されてる場合は、上位に本来の数を置く
|
||||
// /arch:SSE2, +1=SSSE3, +2=SSE4.1, +4=SSE4.2, +8=CLMUL, +16=AVX2, +128=JIT(SSE2), +256=Old
|
||||
// /arch:SSE2, +1=SSSE3, +2=SSE4.1, +4=SSE4.2, +8=CLMUL, +16=AVX2, +128=JIT(SSE2), +256=ALTMAPなし
|
||||
// 上位 16-bit = L2 cache サイズから計算した制限サイズ
|
||||
unsigned int cpu_flag = 0;
|
||||
unsigned int cpu_cache = 0; // 上位 16-bit = L2 cache * 2, 下位 16-bit = L3 cache
|
||||
unsigned int cpu_cache = 0; // 上位 16-bit = L3 cache の制限サイズ, 下位 16-bit = 同時処理数
|
||||
unsigned int memory_use = 0; // メモリー使用量 0=auto, 1~7 -> 1/8 ~ 7/8
|
||||
|
||||
static int count_bit(DWORD_PTR value)
|
||||
@@ -1869,7 +1870,7 @@ static int count_bit(DWORD_PTR value)
|
||||
void check_cpu(void)
|
||||
{
|
||||
int core_count = 0, use_count;
|
||||
unsigned int CPUInfo[4];
|
||||
unsigned int CPUInfo[4], limit_size = 0;
|
||||
unsigned int returnLength, byteOffset;
|
||||
DWORD_PTR ProcessAffinityMask, SystemAffinityMask; // 32-bit なら 4バイト、64-bit なら 8バイト整数
|
||||
PSYSTEM_LOGICAL_PROCESSOR_INFORMATION buffer = NULL, ptr;
|
||||
@@ -2006,42 +2007,52 @@ void check_cpu(void)
|
||||
//printf("Number of available physical processor cores: %d\n", core_count);
|
||||
if (cache3_size > 0){
|
||||
//printf("L3 cache: %d KB (%d way)\n", cache3_size >> 10 , cache3_way);
|
||||
cache3_size /= cache3_way; // set-associative のサイズにする
|
||||
if (cache3_size < 131072)
|
||||
cache3_size = 128 << 10; // 128 KB 以上にする
|
||||
cpu_cache = cache3_size / cache3_way; // set-associative のサイズにする
|
||||
if (cpu_cache < 131072)
|
||||
cpu_cache = 128 << 10; // 128 KB 以上にする
|
||||
}
|
||||
if (cache2_size > 0){
|
||||
//printf("L2 cache: %d KB (%d way)\n", cache2_size >> 10, cache2_way);
|
||||
cache2_size /= cache2_way; // set-associative のサイズにする
|
||||
if (cache2_size < 32768)
|
||||
cache2_size = 32 << 10; // 32 KB 以上にする
|
||||
//printf("Limit size of Cache Blocking: %d KB\n", cache2_size >> 10);
|
||||
cpu_cache = cache2_size | (cache3_size >> 17);
|
||||
limit_size = cache2_size / cache2_way; // set-associative のサイズにする
|
||||
if (limit_size < 65536)
|
||||
limit_size = 64 << 10; // 64 KB 以上にする
|
||||
// 同時処理数を決める
|
||||
if (cache2_way >= 16){
|
||||
returnLength = cache2_way / 2; // L2 cache の分割数が多い場合は、その半分にする
|
||||
} else {
|
||||
returnLength = 0;
|
||||
}
|
||||
if (cache3_size > 0){ // L2 cache に対する L3 cache のサイズの倍率にする
|
||||
byteOffset = cache3_size / cache2_size;
|
||||
if (returnLength < byteOffset){
|
||||
returnLength = byteOffset;
|
||||
if (cache2_way >= cache3_way) // L2 cache の分割数が L3 cache 以上なら 1.5倍にする
|
||||
returnLength += returnLength / 2;
|
||||
}
|
||||
}
|
||||
cpu_cache |= returnLength & 0x1FFFF;
|
||||
}
|
||||
}
|
||||
|
||||
if (cpu_cache == 0) // キャッシュ・サイズが不明なら、128 KB にする
|
||||
cpu_cache = 128 << 10;
|
||||
if (limit_size == 0) // キャッシュ・サイズが不明なら、128 KB にする
|
||||
limit_size = 128 << 10;
|
||||
//printf("Limit size of Cache Blocking: %d KB\n", limit_size >> 10);
|
||||
// cpu_flag の上位 16-bit にキャッシュの制限サイズを置く
|
||||
cpu_flag |= limit_size & 0xFFFF0000; // 64 KB 未満は無視する
|
||||
|
||||
if (core_count == 0){ // 物理コア数が不明なら、論理コア数と同じにする
|
||||
core_count = cpu_num;
|
||||
use_count = cpu_num;
|
||||
} else if (core_count < cpu_num){ // 物理コア数が共有されてるなら
|
||||
if (core_count >= 6){ // 6 コア以上ならそれ以上増やさない
|
||||
use_count = core_count;
|
||||
} else { // 2~5 コアなら 1個だけ増やす
|
||||
use_count = core_count + 1;
|
||||
}
|
||||
} else if (core_count < cpu_num){ // 物理コアが共有されてるなら
|
||||
use_count = core_count; // 物理コア数と同じにする
|
||||
} else {
|
||||
use_count = core_count;
|
||||
use_count = cpu_num; // 論理コア数と同じにする
|
||||
}
|
||||
if (use_count > MAX_CPU) // 利用するコア数が実装上の制限を越えないようにする
|
||||
use_count = MAX_CPU;
|
||||
//printf("Core count: logical, physical, use = %d, %d, %d\n", cpu_num, core_count, use_count);
|
||||
// 上位に論理コア数と物理コア数、下位に利用するコア数を配置する
|
||||
cpu_num = (cpu_num << 24) | (core_count << 16) | use_count;
|
||||
|
||||
// cpu_flag の上位 17-bit にキャッシュの制限サイズを置く
|
||||
cpu_flag |= cpu_cache & 0xFFFF8000; // 32 KB 未満は無視する
|
||||
}
|
||||
|
||||
// OS が 32-bit か 64-bit かを調べる
|
||||
|
||||
@@ -6,11 +6,11 @@ extern "C" {
|
||||
#endif
|
||||
|
||||
#ifndef _WIN64 // 32-bit 版なら
|
||||
#define MAX_CPU 8 // 32-bit 版は少なくしておく
|
||||
#define MAX_CPU 16 // 32-bit 版は少なくしておく
|
||||
#define MAX_MEM_SIZE 0x7F000000 // 確保するメモリー領域の最大値 2032MB
|
||||
#define MAX_MEM_SIZE32 0x50000000 // 32-bit OS で確保するメモリー領域の最大値 1280MB
|
||||
#else
|
||||
#define MAX_CPU 16 // 最大 CPU/Core 個数 (スレッド本数)
|
||||
#define MAX_CPU 32 // 最大 CPU/Core 個数 (スレッド本数)
|
||||
#endif
|
||||
|
||||
#define MAX_LEN 1024 // ファイル名の最大文字数 (末尾のNULL文字も含む)
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// create.c
|
||||
// Copyright : 2022-02-16 Yutaka Sawada
|
||||
// Copyright : 2023-10-22 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -320,7 +320,7 @@ error_end:
|
||||
return off;
|
||||
}
|
||||
|
||||
#define MAX_MULTI_READ 4 // SSDで同時に読み込む最大ファイル数
|
||||
#define MAX_MULTI_READ 6 // SSDで同時に読み込む最大ファイル数
|
||||
|
||||
// SSD 上で複数ファイルのハッシュ値を同時に求めるバージョン
|
||||
int set_common_packet_multi(
|
||||
@@ -348,11 +348,9 @@ unsigned int time_start = GetTickCount();
|
||||
memset(th, 0, sizeof(FILE_HASH_TH) * MAX_MULTI_READ);
|
||||
// Core数に応じてスレッド数を増やす
|
||||
if ((memory_use & 32) != 0){ // NVMe SSD
|
||||
if (cpu_num >= 8){ // 8 ~ 16 Cores
|
||||
multi_read = 4;
|
||||
} else { // 3 Cores + Hyper-threading, or 4 ~ 7 Cores
|
||||
multi_read = 3;
|
||||
}
|
||||
multi_read = (cpu_num + 2) / 3 + 1; // 3=2, 4~6=3, 7~9=4, 10~12=5, 13~=6
|
||||
if (multi_read > MAX_MULTI_READ)
|
||||
multi_read = MAX_MULTI_READ;
|
||||
} else { // SATA SSD
|
||||
multi_read = 2;
|
||||
}
|
||||
@@ -1282,6 +1280,7 @@ int create_recovery_file_1pass(
|
||||
int footer_size, // 末尾パケットのバッファー・サイズ
|
||||
HANDLE *rcv_hFile, // 各リカバリ・ファイルのハンドル
|
||||
unsigned char *p_buf, // 計算済みのパリティ・ブロック
|
||||
unsigned char *g_buf, // GPU用 (GPUを使わない場合は NULLにすること)
|
||||
unsigned int unit_size)
|
||||
{
|
||||
unsigned char *packet_header, hash[HASH_SIZE];
|
||||
@@ -1438,6 +1437,10 @@ int create_recovery_file_1pass(
|
||||
|
||||
// Recovery Slice packet は後から書き込む
|
||||
for (j = block_start; j < block_start + block_count; j++){
|
||||
if (g_buf != NULL){ // GPUを使った場合
|
||||
// CPUスレッドと GPUスレッドの計算結果を合わせる
|
||||
galois_align_xor(g_buf + (size_t)unit_size * j, p_buf, unit_size);
|
||||
}
|
||||
// パリティ・ブロックのチェックサムを検証する
|
||||
checksum16_return(p_buf, hash, unit_size - HASH_SIZE);
|
||||
if (memcmp(p_buf + unit_size - HASH_SIZE, hash, HASH_SIZE) != 0){
|
||||
|
||||
@@ -82,6 +82,7 @@ int create_recovery_file_1pass(
|
||||
int footer_size, // 末尾パケットのバッファー・サイズ
|
||||
HANDLE *rcv_hFile, // 各リカバリ・ファイルのハンドル
|
||||
unsigned char *p_buf, // 計算済みのパリティ・ブロック
|
||||
unsigned char *g_buf, // GPU用 (GPUを使わない場合は NULLにすること)
|
||||
unsigned int unit_size);
|
||||
|
||||
// 作成中のリカバリ・ファイルを削除する
|
||||
|
||||
File diff suppressed because it is too large
Load Diff
@@ -47,6 +47,15 @@ typedef void (* REGION_MULTIPLY) (
|
||||
int factor); // Number to multiply by
|
||||
REGION_MULTIPLY galois_align_multiply;
|
||||
|
||||
typedef void (* REGION_MULTIPLY2) (
|
||||
unsigned char *src1, // Region to multiply
|
||||
unsigned char *src2,
|
||||
unsigned char *dst, // Products go here
|
||||
unsigned int len, // Byte length
|
||||
int factor1, // Number to multiply by
|
||||
int factor2);
|
||||
REGION_MULTIPLY2 galois_align_multiply2;
|
||||
|
||||
// 領域並び替え用の関数定義
|
||||
typedef void (* REGION_ALTMAP) (unsigned char *data, unsigned int bsize);
|
||||
REGION_ALTMAP galois_altmap_change;
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// lib_opencl.c
|
||||
// Copyright : 2023-06-01 Yutaka Sawada
|
||||
// Copyright : 2023-10-22 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _WIN32_WINNT
|
||||
@@ -72,11 +72,10 @@ typedef cl_int (CL_API_CALL *API_clEnqueueNDRangeKernel)(cl_command_queue, cl_ke
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
// グローバル変数
|
||||
|
||||
extern unsigned int cpu_flag, cpu_cache; // declared in common2.h
|
||||
extern unsigned int cpu_flag; // declared in common2.h
|
||||
extern int cpu_num;
|
||||
|
||||
#define MAX_DEVICE 3
|
||||
#define MAX_GROUP_NUM 64
|
||||
|
||||
HMODULE hLibOpenCL = NULL;
|
||||
|
||||
@@ -104,17 +103,15 @@ API_clEnqueueNDRangeKernel gfn_clEnqueueNDRangeKernel;
|
||||
OpenCL_method : どのデバイスを選ぶか
|
||||
unit_size : ブロックの単位サイズ
|
||||
src_max : ソース・ブロック個数
|
||||
chunk_size = 0: 標準では分割しない
|
||||
|
||||
出力
|
||||
return : エラー番号
|
||||
src_max : 最大で何ブロックまでソースを読み込めるか
|
||||
chunk_size : CPUスレッドの分割サイズ
|
||||
OpenCL_method : 動作フラグいろいろ
|
||||
*/
|
||||
|
||||
// 0=成功, 1~エラー番号
|
||||
int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
int init_OpenCL(int unit_size, int *src_max)
|
||||
{
|
||||
char buf[2048], *p_source;
|
||||
int err = 0, i, j;
|
||||
@@ -141,7 +138,7 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
API_clGetKernelWorkGroupInfo fn_clGetKernelWorkGroupInfo;
|
||||
cl_int ret;
|
||||
cl_uint num_platforms = 0, num_devices = 0, num_groups, param_value;
|
||||
cl_ulong param_value8, cache_size;
|
||||
cl_ulong param_value8;
|
||||
cl_platform_id platform_id[MAX_DEVICE], selected_platform; // Intel, AMD, Nvidia などドライバーの提供元
|
||||
cl_device_id device_id[MAX_DEVICE], selected_device; // CPU や GPU など
|
||||
cl_program program;
|
||||
@@ -285,23 +282,12 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
// printf("Shared Virtual Memory = 0x%I64X\n", param_value8);
|
||||
#endif
|
||||
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_ADDRESS_BITS, sizeof(cl_uint), ¶m_value, NULL);
|
||||
if (ret != CL_SUCCESS)
|
||||
continue;
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof(cl_ulong), ¶m_value8, NULL);
|
||||
if (ret != CL_SUCCESS)
|
||||
continue;
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("ADDRESS_BITS = %d\n", param_value);
|
||||
printf("MAX_MEM_ALLOC_SIZE = %I64d MB\n", param_value8 >> 20);
|
||||
#endif
|
||||
if (param_value == 32){ // CL_DEVICE_ADDRESS_BITS によって確保するメモリー領域の上限を変える
|
||||
if (param_value8 > 0x30000000) // 768MB までにする
|
||||
param_value8 = 0x30000000;
|
||||
} else { // 64-bit OS でも 2GB までにする
|
||||
if (param_value8 > 0x80000000)
|
||||
param_value8 = 0x80000000;
|
||||
}
|
||||
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_MAX_COMPUTE_UNITS, sizeof(cl_uint), &num_groups, NULL);
|
||||
if (ret != CL_SUCCESS)
|
||||
@@ -309,19 +295,14 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(size_t), &data_size, NULL);
|
||||
if (ret != CL_SUCCESS)
|
||||
continue;
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_HOST_UNIFIED_MEMORY, sizeof(cl_uint), ¶m_value, NULL);
|
||||
if (ret != CL_SUCCESS)
|
||||
continue;
|
||||
if (param_value != 0)
|
||||
param_value = 1;
|
||||
// CL_DEVICE_HOST_UNIFIED_MEMORY は OpenCL 2.0 以降で非推奨になったので、参照しない
|
||||
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("MAX_COMPUTE_UNITS = %d\n", num_groups);
|
||||
printf("MAX_WORK_GROUP_SIZE = %zd\n", data_size);
|
||||
printf("HOST_UNIFIED_MEMORY = %d\n", param_value);
|
||||
#endif
|
||||
// MAX_COMPUTE_UNITS * MAX_WORK_GROUP_SIZE で計算力を測る、外付けGPUなら値を倍にする
|
||||
count = (2 - param_value) * (int)data_size * num_groups;
|
||||
// MAX_COMPUTE_UNITS * MAX_WORK_GROUP_SIZE で計算力を測る
|
||||
count = (int)data_size * num_groups;
|
||||
count *= OpenCL_method; // 符号を変える
|
||||
//printf("prev = %d, now = %d\n", gpu_power, count);
|
||||
if ((count > gpu_power) && (data_size >= 256) && // 256以上ないとテーブルを作れない
|
||||
@@ -330,41 +311,19 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
selected_device = device_id[j]; // 使うデバイスの ID
|
||||
selected_platform = platform_id[i];
|
||||
OpenCL_group_num = num_groups; // ワークグループ数は COMPUTE_UNITS 数にする
|
||||
if (OpenCL_group_num > MAX_GROUP_NUM) // 制限を付けてローカルメモリーの消費を抑える
|
||||
OpenCL_group_num = MAX_GROUP_NUM;
|
||||
alloc_max = (size_t)param_value8;
|
||||
|
||||
// AMD Radeon ではメモリー領域が全体の 1/4 とは限らない
|
||||
// AMD や Intel の GPU ではメモリー領域が全体の 1/4 とは限らない
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(cl_ulong), ¶m_value8, NULL);
|
||||
if (ret == CL_SUCCESS){
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("GLOBAL_MEM_SIZE = %I64d MB\n", param_value8 >> 20);
|
||||
#endif
|
||||
// 領域一個あたりのサイズは全体の 1/4 までにする
|
||||
// 領域一個あたりのサイズは全体の 1/4 までにする(VRAMを使いすぎると不安定になる)
|
||||
param_value8 /= 4;
|
||||
if ((cl_ulong)alloc_max > param_value8)
|
||||
alloc_max = (size_t)param_value8;
|
||||
}
|
||||
|
||||
cache_size = 0;
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_GLOBAL_MEM_CACHE_TYPE, sizeof(cl_uint), &num_groups, NULL);
|
||||
if (ret == CL_SUCCESS){
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("GLOBAL_MEM_CACHE_TYPE = %d\n", num_groups);
|
||||
#endif
|
||||
if (num_groups & 3){ // CL_READ_ONLY_CACHE or CL_READ_WRITE_CACHE
|
||||
ret = fn_clGetDeviceInfo(device_id[j], CL_DEVICE_GLOBAL_MEM_CACHE_SIZE, sizeof(cl_ulong), &cache_size, NULL);
|
||||
if (ret == CL_SUCCESS){
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("GLOBAL_MEM_CACHE_SIZE = %I64d KB\n", cache_size >> 10);
|
||||
#endif
|
||||
if (param_value != 0){ // 内蔵 GPU なら CPU との共有キャッシュを活用する
|
||||
if (cache_size >= 1048576) // サイズが小さい場合は分割しない
|
||||
cache_size |= 0x40000000;
|
||||
}
|
||||
}
|
||||
}
|
||||
}
|
||||
}
|
||||
}
|
||||
}
|
||||
@@ -395,58 +354,19 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
return (ret << 8) | 12;
|
||||
|
||||
// 計算方式を選択する
|
||||
gpu_power = unit_size; // unit_size は MEM_UNIT の倍数になってる
|
||||
if ((((cpu_flag & 0x101) == 1) || ((cpu_flag & 16) != 0)) && (sse_unit == 32)){
|
||||
if ((((cpu_flag & 0x101) == 1) || ((cpu_flag & 0x110) == 0x10)) && (sse_unit == 32)){
|
||||
OpenCL_method = 2; // SSSE3 & ALTMAP または AVX2 ならデータの並び替え対応版を使う
|
||||
if (cache_size & 0x40000000){ // 内蔵 GPU でキャッシュを利用できるなら、CPUスレッドと同じにする
|
||||
j = cpu_cache & 0x7FFF8000; // CPUのキャッシュ上限サイズ
|
||||
count = (int)(cache_size & 0x3FFFFFFF) / 4; // ただし、認識できるサイズの 1/4 までにする
|
||||
if ((j == 0) || (j > count))
|
||||
j = count;
|
||||
count = 1;
|
||||
while (gpu_power > j){ // 制限サイズより大きいなら
|
||||
// 分割数を増やして chunk のサイズを試算してみる
|
||||
count++;
|
||||
gpu_power = (unit_size + count - 1) / count;
|
||||
gpu_power = (gpu_power + (MEM_UNIT - 1)) & ~(MEM_UNIT - 1); // MEM_UNITの倍数にする
|
||||
}
|
||||
if (count > 1){
|
||||
*chunk_size = gpu_power;
|
||||
OpenCL_method = 3;
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("gpu cache: limit size = %d, chunk size = %d, split = %d\n", j, gpu_power, count);
|
||||
#endif
|
||||
}
|
||||
/*
|
||||
// 32バイト単位のメモリーアクセスならキャッシュする必要なし?計算速度が半減する・・・
|
||||
} else if ((cache_size & 0x3FFFFFFF) > OpenCL_group_num * 4096){ // 2KB の倍はいるかも?
|
||||
#ifdef DEBUG_OUTPUT
|
||||
printf("gpu: cache size = %d, read size = %d\n", cache_size & 0x3FFFFFFF, OpenCL_group_num * 2048);
|
||||
#endif
|
||||
OpenCL_method = 1;
|
||||
*/
|
||||
}
|
||||
|
||||
} else if (((cpu_flag & 128) != 0) && (sse_unit == 256)){
|
||||
OpenCL_method = 4; // JIT(SSE2) は bit ごとに上位から 16バイトずつ並ぶ
|
||||
// ローカルのテーブルサイズが異なることに注意
|
||||
// XOR 方式以外は 2KB (4バイト * 256項目 * 2個) 使う
|
||||
// XOR (JIT) は 64バイト (4バイト * 16項目) 使う
|
||||
#ifdef DEBUG_OUTPUT
|
||||
// printf("4 KB cache (16-bytes * 256 work items), use if\n");
|
||||
#endif
|
||||
} else {
|
||||
OpenCL_method = 1; // MMX用のコードは遅いので、キャッシュ最適化する必要が無い
|
||||
OpenCL_method = 1; // 並び替えられてないデータ用
|
||||
}
|
||||
|
||||
// work group 数が必要以上に多い場合は減らす
|
||||
/*
|
||||
if (OpenCL_method == 4){
|
||||
// work item 一個が 16バイトずつ計算する、256個なら work group ごとに 4KB 担当する
|
||||
data_size = unit_size / 4096;
|
||||
} else
|
||||
*/
|
||||
if (OpenCL_method & 2){
|
||||
if (OpenCL_method == 2){
|
||||
// work item 一個が 8バイトずつ計算する、256個なら work group ごとに 2KB 担当する
|
||||
data_size = unit_size / 2048;
|
||||
} else {
|
||||
@@ -455,7 +375,7 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
}
|
||||
if (OpenCL_group_num > data_size){
|
||||
OpenCL_group_num = data_size;
|
||||
printf("Number of work groups is reduced to %d\n", (int)OpenCL_group_num);
|
||||
printf("Number of work groups is reduced to %zd\n", OpenCL_group_num);
|
||||
}
|
||||
|
||||
// 最大で何ブロック分のメモリー領域を保持できるのか(ここではまだ確保しない)
|
||||
@@ -574,7 +494,7 @@ int init_OpenCL(int unit_size, int *src_max, int *chunk_size)
|
||||
FreeResource(glob); // not required ?
|
||||
|
||||
// 定数を指定する
|
||||
wsprintfA(buf, "-D BLK_SIZE=%d -D CHK_SIZE=%d", unit_size / 4, gpu_power / 4);
|
||||
wsprintfA(buf, "-cl-fast-relaxed-math -D BLK_SIZE=%d", unit_size / 4);
|
||||
|
||||
// 使用する OpenCL デバイス用にコンパイルする
|
||||
ret = fn_clBuildProgram(program, 1, &selected_device, buf, NULL, NULL);
|
||||
|
||||
@@ -10,14 +10,14 @@ extern "C" {
|
||||
|
||||
extern int OpenCL_method;
|
||||
|
||||
int init_OpenCL(int unit_size, int *src_max, int *chunk_size);
|
||||
int init_OpenCL(int unit_size, int *src_max);
|
||||
int free_OpenCL(void);
|
||||
void info_OpenCL(char *buf, int buf_size);
|
||||
|
||||
int gpu_copy_blocks(
|
||||
unsigned char *data,
|
||||
int unit_size,
|
||||
int src_end);
|
||||
int src_num);
|
||||
|
||||
int gpu_multiply_blocks(
|
||||
int src_num, // Number of multiplying source blocks
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// list.c
|
||||
// Copyright : 2022-10-14 Yutaka Sawada
|
||||
// Copyright : 2023-10-15 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -348,7 +348,7 @@ if (time_start > 0){
|
||||
// SSD 上で複数ファイルを同時に検査する
|
||||
|
||||
// MAX_MULTI_READ の2倍ぐらいにする?
|
||||
#define MAX_READ_NUM 10
|
||||
#define MAX_READ_NUM 12
|
||||
|
||||
int check_file_complete_multi(
|
||||
char *ascii_buf,
|
||||
@@ -370,11 +370,9 @@ unsigned int time_start = GetTickCount();
|
||||
memset(hSub, 0, sizeof(HANDLE) * MAX_READ_NUM);
|
||||
// Core数に応じてスレッド数を増やす
|
||||
if ((memory_use & 32) != 0){ // NVMe SSD
|
||||
if (cpu_num >= 8){ // 8 ~ 16 Cores
|
||||
multi_read = 4;
|
||||
} else { // 3 Cores + Hyper-threading, or 4 ~ 7 Cores
|
||||
multi_read = 3;
|
||||
}
|
||||
multi_read = (cpu_num + 2) / 3 + 1; // 3=2, 4~6=3, 7~9=4, 10~12=5, 13~=6
|
||||
if (multi_read > MAX_READ_NUM / 2)
|
||||
multi_read = MAX_READ_NUM / 2;
|
||||
} else { // SATA SSD
|
||||
multi_read = 2;
|
||||
}
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// md5_crc.c
|
||||
// Copyright : 2022-10-01 Yutaka Sawada
|
||||
// Copyright : 2023-10-29 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -204,6 +204,8 @@ static unsigned int time_start, time1_start;
|
||||
static unsigned int time_total = 0, time2_total = 0, time3_total = 0;
|
||||
#endif
|
||||
|
||||
#define MAX_BUF_SIZE 2097152 // ヒープ領域を使う場合の最大サイズ
|
||||
|
||||
// ファイルのハッシュ値と各スライスのチェックサムを同時に計算する
|
||||
int file_hash_crc(
|
||||
wchar_t *file_name, // ハッシュ値を求めるファイル
|
||||
@@ -671,8 +673,8 @@ time1_start = GetTickCount();
|
||||
}
|
||||
|
||||
// バッファー・サイズが大きいのでヒープ領域を使う
|
||||
for (io_size = IO_SIZE; io_size < 1048576; io_size += IO_SIZE){ // 1 MB までにする
|
||||
if ((io_size + IO_SIZE > (cpu_cache << 17)) || ((__int64)(io_size + IO_SIZE) * 4 > file_left))
|
||||
for (io_size = IO_SIZE; io_size <= MAX_BUF_SIZE; io_size += IO_SIZE){ // IO_SIZE の倍数にする
|
||||
if ((io_size + IO_SIZE > (cpu_cache & 0xFFFE0000)) || ((__int64)(io_size + IO_SIZE) * 4 > file_left))
|
||||
break;
|
||||
}
|
||||
buf1 = _aligned_malloc(io_size * 2, 64);
|
||||
@@ -866,8 +868,8 @@ DWORD WINAPI file_hash_crc2(LPVOID lpParameter)
|
||||
|
||||
// バッファー・サイズが大きいのでヒープ領域を使う
|
||||
prog_tick = 1;
|
||||
for (io_size = IO_SIZE; io_size < 1048576; io_size += IO_SIZE){ // IO_SIZE の倍数で 1 MB までにする
|
||||
if ((io_size + IO_SIZE > (cpu_cache << 17)) || ((__int64)(io_size + IO_SIZE) * 4 > file_left))
|
||||
for (io_size = IO_SIZE; io_size <= MAX_BUF_SIZE; io_size += IO_SIZE){ // IO_SIZE の倍数にする
|
||||
if ((io_size + IO_SIZE > (cpu_cache & 0xFFFE0000)) || ((__int64)(io_size + IO_SIZE) * 4 > file_left))
|
||||
break;
|
||||
prog_tick++;
|
||||
}
|
||||
@@ -1303,8 +1305,8 @@ DWORD WINAPI file_hash_background(LPVOID lpParameter)
|
||||
find_next = files[num].b_off; // 先頭ブロックの番号
|
||||
|
||||
// バッファー・サイズが大きいのでヒープ領域を使う
|
||||
for (io_size = IO_SIZE; io_size < 1048576; io_size += IO_SIZE){ // IO_SIZE の倍数で 1 MB までにする
|
||||
if ((io_size + IO_SIZE > (cpu_cache << 17)) || ((__int64)(io_size + IO_SIZE) * 4 > file_size))
|
||||
for (io_size = IO_SIZE; io_size <= MAX_BUF_SIZE; io_size += IO_SIZE){ // IO_SIZE の倍数にする
|
||||
if ((io_size + IO_SIZE > (cpu_cache & 0xFFFE0000)) || ((__int64)(io_size + IO_SIZE) * 4 > file_size))
|
||||
break;
|
||||
}
|
||||
//printf("\n io_size = %d\n", io_size);
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// par2.c
|
||||
// Copyright : 2023-03-15 Yutaka Sawada
|
||||
// Copyright : 2023-10-15 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -112,12 +112,12 @@ int par2_create(
|
||||
err = -12;
|
||||
} else {
|
||||
// メモリーを確保できるか試す
|
||||
err = read_block_num(parity_num, cpu_num - 1, 0, 256);
|
||||
err = read_block_num(parity_num, 0, 256);
|
||||
if (err == 0)
|
||||
err = -13;
|
||||
}
|
||||
#ifdef TIMER
|
||||
printf("read_block_num = %d\n", read_block_num(parity_num, cpu_num - 1, 0, 256));
|
||||
printf("read_block_num = %d\n", read_block_num(parity_num, 0, 256));
|
||||
#endif
|
||||
if (err > 0){ // 1-pass方式が可能
|
||||
#ifdef TIMER
|
||||
@@ -181,7 +181,7 @@ int par2_create(
|
||||
}
|
||||
} else {
|
||||
// 共通パケットを作成する
|
||||
if ((memory_use & 16) && (cpu_num >= 4) && (entity_num >= 2)){ // SSDなら複数ファイルを同時に処理する
|
||||
if ((memory_use & 16) && (cpu_num >= 3) && (entity_num >= 2)){ // SSDなら複数ファイルを同時に処理する
|
||||
common_size = set_common_packet_multi(common_buf, &packet_num, (switch_p & 2) >> 1, files);
|
||||
} else {
|
||||
common_size = set_common_packet(common_buf, &packet_num, (switch_p & 2) >> 1, files);
|
||||
@@ -529,7 +529,7 @@ int par2_verify(
|
||||
|
||||
// ソース・ファイルが完全かどうかを調べる
|
||||
// ファイルの状態は 完全、消失、追加、破損(完全なブロックの数) の4種類
|
||||
if ((memory_use & 16) && (cpu_num >= 4) && (entity_num >= 2)){ // SSDなら複数ファイルを同時に処理する
|
||||
if ((memory_use & 16) && (cpu_num >= 3) && (entity_num >= 2)){ // SSDなら複数ファイルを同時に処理する
|
||||
err = check_file_complete_multi(ascii_buf, uni_buf, files, s_blk);
|
||||
} else {
|
||||
err = check_file_complete(ascii_buf, uni_buf, files, s_blk);
|
||||
@@ -741,7 +741,7 @@ int par2_repair(
|
||||
|
||||
// ソース・ファイルが完全かどうかを一覧表示する
|
||||
// ファイルの状態は 完全、消失、追加、破損(完全なブロックの数) の4種類
|
||||
if ((memory_use & 16) && (cpu_num >= 4) && (entity_num >= 2)){ // SSDなら複数ファイルを同時に処理する
|
||||
if ((memory_use & 16) && (cpu_num >= 3) && (entity_num >= 2)){ // SSDなら複数ファイルを同時に処理する
|
||||
err = check_file_complete_multi(ascii_buf, uni_buf, files, s_blk);
|
||||
} else {
|
||||
err = check_file_complete(ascii_buf, uni_buf, files, s_blk);
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// par2_cmd.c
|
||||
// Copyright : 2023-03-18 Yutaka Sawada
|
||||
// Copyright : 2023-10-15 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -86,36 +86,38 @@ static void print_environment(void)
|
||||
|
||||
printf("CPU thread\t: %d / %d\n", cpu_num & 0xFFFF, cpu_num >> 24);
|
||||
cpu_num &= 0xFFFF; // 利用するコア数だけにしておく
|
||||
printf("CPU cache limit : %d KB, %d KB\n", (cpu_cache & 0x7FFF8000) >> 10, (cpu_cache & 0x00007FFF) << 7);
|
||||
#ifndef _WIN64 // 32-bit 版は MMX, SSE2, SSSE3 のどれかを表示する
|
||||
printf("CPU cache limit : %d KB, %d KB\n", (cpu_flag & 0xFFFF0000) >> 10, (cpu_cache & 0xFFFE0000) >> 10);
|
||||
#ifndef _WIN64 // 32-bit 版は MMX, SSE2, SSSE3, AVX2 のどれかを表示する
|
||||
printf("CPU extra\t:");
|
||||
if (cpu_flag & 1){
|
||||
if (((cpu_flag & 16) != 0) && ((cpu_flag & 256) == 0)){
|
||||
printf(" AVX2");
|
||||
} else if (cpu_flag & 1){
|
||||
if (cpu_flag & 256){
|
||||
printf(" SSSE3(old)");
|
||||
printf(" SSSE3(slow)");
|
||||
} else {
|
||||
printf(" SSSE3");
|
||||
}
|
||||
} else if (cpu_flag & 128){
|
||||
} else if (((cpu_flag & 128) != 0) && ((cpu_flag & 256) == 0)){
|
||||
printf(" SSE2");
|
||||
} else {
|
||||
printf(" MMX");
|
||||
}
|
||||
#else // 64-bit 版は SSE2, SSSE3 を表示する
|
||||
#else // 64-bit 版は SSE2, SSSE3, AVX2 を表示する
|
||||
printf("CPU extra\t: x64");
|
||||
if (cpu_flag & 1){
|
||||
if (((cpu_flag & 16) != 0) && ((cpu_flag & 256) == 0)){
|
||||
printf(" AVX2");
|
||||
} else if (cpu_flag & 1){
|
||||
if (cpu_flag & 256){
|
||||
printf(" SSSE3(old)");
|
||||
printf(" SSSE3(slow)");
|
||||
} else {
|
||||
printf(" SSSE3");
|
||||
}
|
||||
} else if (cpu_flag & 128){
|
||||
} else if (((cpu_flag & 128) != 0) && ((cpu_flag & 256) == 0)){
|
||||
printf(" SSE2");
|
||||
}
|
||||
#endif
|
||||
if (cpu_flag & 8)
|
||||
printf(" CLMUL");
|
||||
if (cpu_flag & 16)
|
||||
printf(" AVX2");
|
||||
printf("\nMemory usage\t: ");
|
||||
if (memory_use & 7){
|
||||
printf("%d/8", memory_use & 7);
|
||||
@@ -1481,39 +1483,42 @@ ri= switch_set & 0x00040000
|
||||
k = (k * 10) + (tmp_p[j] - '0');
|
||||
j++;
|
||||
}
|
||||
if (k & 32){ // GPU を使う
|
||||
if (k & 256){ // GPU を使う
|
||||
OpenCL_method = 1; // Faster GPU
|
||||
} else if (k & 64){
|
||||
} else if (k & 512){
|
||||
OpenCL_method = -1; // Slower GPU
|
||||
}
|
||||
if (k & 16) // SSSE3 を使わない
|
||||
cpu_flag &= 0xFFFFFFFE;
|
||||
if (k & 128) // CLMUL を使わない、SSSE3 の古いエンコーダーを使う
|
||||
cpu_flag = (cpu_flag & 0xFFFFFFF7) | 0x100;
|
||||
if (k & 256) // JIT(SSE2) を使わない
|
||||
if (k & 1024) // CLMUL と ALTMAP を使わない
|
||||
cpu_flag = (cpu_flag & 0xFFFFFFF7) | 256;
|
||||
if (k & 2048) // JIT(SSE2) を使わない
|
||||
cpu_flag &= 0xFFFFFF7F;
|
||||
if (k & 512) // AVX2 を使わない
|
||||
if (k & 4096) // SSSE3 を使わない
|
||||
cpu_flag &= 0xFFFFFFFE;
|
||||
if (k & 8192) // AVX2 を使わない
|
||||
cpu_flag &= 0xFFFFFFEF;
|
||||
if (k & 15){ // 使用するコア数を変更する
|
||||
k &= 15; // 1~15 の範囲
|
||||
if (k & 255){ // 使用するコア数を変更する
|
||||
k &= 255; // 1~255 の範囲
|
||||
// printf("\n lc# = %d , logical = %d, physical = %d \n", k, cpu_num >> 24, (cpu_num & 0x00FF0000) >> 16);
|
||||
if (k == 12){ // 物理コア数の 1/4 にする
|
||||
if (k == 251){ // 物理コア数の 1/4 にする
|
||||
k = ((cpu_num & 0x00FF0000) >> 16) / 4;
|
||||
} else if (k == 13){ // 物理コア数の半分にする
|
||||
} else if (k == 252){ // 物理コア数の半分にする
|
||||
k = ((cpu_num & 0x00FF0000) >> 16) / 2;
|
||||
} else if (k == 14){ // 物理コア数の 3/4 にする
|
||||
} else if (k == 253){ // 物理コア数の 3/4 にする
|
||||
k = (((cpu_num & 0x00FF0000) >> 16) * 3) / 4;
|
||||
} else if (k == 15){ // 物理コア数にする
|
||||
k = (cpu_num & 0x00FF0000) >> 16;
|
||||
if (k >= 6)
|
||||
k--; // 物理コア数が 6以上なら、1個減らす
|
||||
} else if (k > (cpu_num >> 24)){
|
||||
k = cpu_num >> 24; // 論理コア数を超えないようにする
|
||||
} else if (k == 254){ // 物理コア数より減らす
|
||||
k = ((cpu_num & 0x00FF0000) >> 16) - 1;
|
||||
} else if (k == 255){ // 物理コア数より増やす
|
||||
k = cpu_num >> 16;
|
||||
k = ((k & 0xFF) + (k >> 8)) / 2; // 物理コア数と論理コア数の中間にする?
|
||||
// タスクマネージャーにおける CPU使用率は 100%になるけど、速くはならない・・・
|
||||
// k = (k & 0xFF) + ((k >> 8) - (k & 0xFF)) / 4; // 物理コア数の 5/4 にする?
|
||||
}
|
||||
if (k > MAX_CPU){
|
||||
k = MAX_CPU;
|
||||
} else if (k < 1){
|
||||
k = 1;
|
||||
} else if (k > (cpu_num >> 24)){
|
||||
k = cpu_num >> 24; // 論理コア数を超えないようにする
|
||||
}
|
||||
cpu_num = (cpu_num & 0xFFFF0000) | k; // 指定されたコア数を下位に配置する
|
||||
}
|
||||
|
||||
@@ -1,5 +1,5 @@
|
||||
// reedsolomon.c
|
||||
// Copyright : 2023-05-29 Yutaka Sawada
|
||||
// Copyright : 2023-10-26 Yutaka Sawada
|
||||
// License : GPL
|
||||
|
||||
#ifndef _UNICODE
|
||||
@@ -30,13 +30,13 @@
|
||||
|
||||
// GPU を使う最小データサイズ (MB 単位)
|
||||
// GPU の起動には時間がかかるので、データが小さすぎると逆に遅くなる
|
||||
#define GPU_DATA_LIMIT 512
|
||||
#define GPU_DATA_LIMIT 200
|
||||
|
||||
// GPU を使う最小ブロックサイズとブロック数
|
||||
// CPU と GPU で処理を割り振る為には、ある程度のブロック数を必要とする
|
||||
#define GPU_BLOCK_SIZE_LIMIT 65536
|
||||
#define GPU_SOURCE_COUNT_LIMIT 256
|
||||
#define GPU_PARITY_COUNT_LIMIT 32
|
||||
#define GPU_SOURCE_COUNT_LIMIT 192
|
||||
#define GPU_PARITY_COUNT_LIMIT 8
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
@@ -46,7 +46,7 @@ int try_cache_blocking(int unit_size)
|
||||
int limit_size, chunk_count, chunk_size, cache_line_diff;
|
||||
|
||||
// CPUキャッシュをどのくらいまで使うか
|
||||
limit_size = cpu_flag & 0x7FFF8000; // 最低でも 32KB になる
|
||||
limit_size = cpu_flag & 0x7FFF0000; // 最低でも 64KB になる
|
||||
if (limit_size == 0) // キャッシュ・サイズを取得できなかった場合は最適化しない
|
||||
return unit_size;
|
||||
|
||||
@@ -160,7 +160,6 @@ unsigned int get_io_size(
|
||||
// 何ブロックまとめてファイルから読み込むかを空きメモリー量から計算する
|
||||
int read_block_num(
|
||||
int keep_num, // 保持するパリティ・ブロック数
|
||||
int add_num, // 余裕を見るブロック数
|
||||
size_t trial_alloc, // 確保できるか確認するのか
|
||||
int alloc_unit) // メモリー単位の境界 (sse_unit か MEM_UNIT)
|
||||
{
|
||||
@@ -177,7 +176,7 @@ int read_block_num(
|
||||
|
||||
if (trial_alloc){
|
||||
__int64 possible_size;
|
||||
possible_size = (__int64)unit_size * (source_num + keep_num + add_num);
|
||||
possible_size = (__int64)unit_size * (source_num + keep_num);
|
||||
#ifndef _WIN64 // 32-bit 版なら
|
||||
if (possible_size > MAX_MEM_SIZE) // 確保する最大サイズを 2GB までにする
|
||||
possible_size = MAX_MEM_SIZE;
|
||||
@@ -191,13 +190,13 @@ int read_block_num(
|
||||
}
|
||||
mem_size = get_mem_size(trial_alloc) / unit_size; // 何個分確保できるか
|
||||
|
||||
if (mem_size >= (size_t)(source_num + keep_num + add_num)){ // 最大個数より多い
|
||||
if (mem_size >= (size_t)(source_num + keep_num)){ // 最大個数より多い
|
||||
buf_num = source_num;
|
||||
} else if ((int)mem_size < read_min + keep_num + add_num){ // 少なすぎる
|
||||
} else if ((int)mem_size < read_min + keep_num){ // 少なすぎる
|
||||
buf_num = 0; // メモリー不足の印
|
||||
} else { // ソース・ブロック個数を等分割する
|
||||
int split_num;
|
||||
buf_num = (int)mem_size - (keep_num + add_num);
|
||||
buf_num = (int)mem_size - keep_num;
|
||||
split_num = (source_num + buf_num - 1) / buf_num; // 何回に別けて読み込むか
|
||||
buf_num = (source_num + split_num - 1) / split_num;
|
||||
}
|
||||
@@ -205,6 +204,48 @@ int read_block_num(
|
||||
return buf_num;
|
||||
}
|
||||
|
||||
// 1st encode, decode を何スレッドで実行するか決める
|
||||
int calc_thread_num1(int max_num)
|
||||
{
|
||||
int i, num;
|
||||
|
||||
// 読み込み中はスレッド数を減らす(シングル・スレッドの時は 0にする)
|
||||
num = 0;
|
||||
i = 1;
|
||||
while (i * 2 <= cpu_num){ // 1=0, 2~3=1, 4~7=2, 8~15=3, 16~31=4, 32=5
|
||||
num++;
|
||||
i *= 2;
|
||||
}
|
||||
if (num > max_num)
|
||||
num = max_num;
|
||||
|
||||
return num;
|
||||
}
|
||||
|
||||
// 1st & 2nd encode, decode を何スレッドで実行するか決める
|
||||
int calc_thread_num2(int max_num, int *cpu_num2)
|
||||
{
|
||||
int i, num1, num2;
|
||||
|
||||
// 読み込み中はスレッド数を減らす(シングル・スレッドの時は 0にする)
|
||||
num1 = 0;
|
||||
i = 2;
|
||||
while (i <= cpu_num){ // 1=0, 2~3=1, 4~7=2, 8~15=3, 16~31=4, 32=5
|
||||
num1++;
|
||||
i *= 2;
|
||||
}
|
||||
if (num1 > max_num)
|
||||
num1 = max_num;
|
||||
|
||||
// CPU と GPU で必ず2スレッド使う
|
||||
num2 = cpu_num;
|
||||
if (num2 < 2)
|
||||
num2 = 2;
|
||||
*cpu_num2 = num2;
|
||||
|
||||
return num1;
|
||||
}
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
// 戸川 隼人 の「演習と応用FORTRAN77」の逆行列の計算方法を参考にして
|
||||
// Gaussian Elimination を少し修正して行列の数を一つにしてみた
|
||||
@@ -263,7 +304,7 @@ static int invert_matrix_st(unsigned short *mat,
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
// マルチ・プロセッサー対応
|
||||
|
||||
/*
|
||||
typedef struct { // RS threading control struct
|
||||
unsigned short *mat; // 行列
|
||||
int cols; // 横行の長さ
|
||||
@@ -308,8 +349,57 @@ static DWORD WINAPI thread_func(LPVOID lpParameter)
|
||||
CloseHandle(th->end);
|
||||
return 0;
|
||||
}
|
||||
*/
|
||||
typedef struct { // Maxtrix Inversion threading control struct
|
||||
unsigned short *mat; // 行列
|
||||
int cols; // 横行の長さ
|
||||
volatile int start; // 掛ける行の先頭位置
|
||||
volatile int pivot; // 倍率となる値の位置
|
||||
volatile int skip; // とばす行
|
||||
volatile int now; // 消去する行
|
||||
HANDLE run;
|
||||
HANDLE end;
|
||||
} INV_TH;
|
||||
|
||||
// サブ・スレッド
|
||||
static DWORD WINAPI thread_func(LPVOID lpParameter)
|
||||
{
|
||||
unsigned short *mat;
|
||||
int j, cols, row_start2, factor;
|
||||
HANDLE hRun, hEnd;
|
||||
INV_TH *th;
|
||||
|
||||
th = (INV_TH *)lpParameter;
|
||||
mat = th->mat;
|
||||
cols = th->cols;
|
||||
hRun = th->run;
|
||||
hEnd = th->end;
|
||||
SetEvent(hEnd); // 設定完了を通知する
|
||||
|
||||
WaitForSingleObject(hRun, INFINITE); // 計算開始の合図を待つ
|
||||
while (th->skip >= 0){
|
||||
while ((j = InterlockedDecrement(&(th->now))) >= 0){ // j = --th_now
|
||||
if (j == th->skip)
|
||||
continue;
|
||||
row_start2 = cols * j; // その行の開始位置
|
||||
factor = mat[row_start2 + th->pivot]; // j 行の pivot 列の値
|
||||
mat[row_start2 + th->pivot] = 0; // これが行列を一個で済ます手
|
||||
// 先の計算により、i 行の pivot 列の値は必ず 1なので、この factor が倍率になる
|
||||
galois_region_multiply(mat + th->start, mat + row_start2, cols, factor);
|
||||
}
|
||||
//_mm_sfence(); // メモリーへの書き込みを完了する
|
||||
SetEvent(hEnd); // 計算終了を通知する
|
||||
WaitForSingleObject(hRun, INFINITE); // 計算開始の合図を待つ
|
||||
}
|
||||
|
||||
// 終了処理
|
||||
CloseHandle(hRun);
|
||||
CloseHandle(hEnd);
|
||||
return 0;
|
||||
}
|
||||
|
||||
// マルチ・スレッドで逆行列を計算する (利用するパリティ・ブロックの所だけ)
|
||||
/*
|
||||
static int invert_matrix_mt(unsigned short *mat,
|
||||
int rows, // 横行の数、行列の縦サイズ、失われたソース・ブロックの数 = 利用するパリティ・ブロック数
|
||||
int cols, // 縦列の数、行列の横サイズ、本来のソース・ブロック数
|
||||
@@ -411,6 +501,130 @@ static int invert_matrix_mt(unsigned short *mat,
|
||||
CloseHandle(th->h);
|
||||
return 0;
|
||||
}
|
||||
*/
|
||||
|
||||
static int invert_matrix_mt(unsigned short *mat,
|
||||
int rows, // 横行の数、行列の縦サイズ、失われたソース・ブロックの数 = 利用するパリティ・ブロック数
|
||||
int cols, // 縦列の数、行列の横サイズ、本来のソース・ブロック数
|
||||
source_ctx_r *s_blk) // 各ソース・ブロックの情報
|
||||
{
|
||||
int err = 0, j, row_start2, factor, sub_num;
|
||||
unsigned int time_last = GetTickCount();
|
||||
HANDLE hSub[MAX_CPU / 2], hRun[MAX_CPU / 2], hEnd[MAX_CPU / 2];
|
||||
INV_TH th[1];
|
||||
|
||||
memset(hSub, 0, sizeof(HANDLE) * (MAX_CPU / 2));
|
||||
memset(th, 0, sizeof(INV_TH));
|
||||
|
||||
// サブ・スレッドの数は平方根(切り上げ)にする
|
||||
sub_num = 1;
|
||||
j = 2;
|
||||
while (j < cpu_num){ // 1~2=1, 3~4=2, 5~8=3, 9~16=4, 17~32=5
|
||||
sub_num++;
|
||||
j *= 2;
|
||||
}
|
||||
if (sub_num > rows - 2)
|
||||
sub_num = rows - 2; // 多過ぎても意味ないので制限する
|
||||
#ifdef TIMER
|
||||
// 使うスレッド数は、メイン・スレッドの分も含めるので 1個増える
|
||||
printf("\nMaxtrix Inversion with %d threads\n", sub_num + 1);
|
||||
#endif
|
||||
|
||||
// サブ・スレッドを起動する
|
||||
th->mat = mat;
|
||||
th->cols = cols;
|
||||
for (j = 0; j < sub_num; j++){ // サブ・スレッドごとに
|
||||
// イベントを作成する
|
||||
hRun[j] = CreateEvent(NULL, FALSE, FALSE, NULL); // 両方とも Auto Reset にする
|
||||
if (hRun[j] == NULL){
|
||||
print_win32_err();
|
||||
printf("error, inv-thread\n");
|
||||
err = 1;
|
||||
goto error_end;
|
||||
}
|
||||
hEnd[j] = CreateEvent(NULL, FALSE, FALSE, NULL);
|
||||
if (hEnd[j] == NULL){
|
||||
print_win32_err();
|
||||
CloseHandle(hRun[j]);
|
||||
printf("error, inv-thread\n");
|
||||
err = 1;
|
||||
goto error_end;
|
||||
}
|
||||
// サブ・スレッドを起動する
|
||||
th->run = hRun[j];
|
||||
th->end = hEnd[j];
|
||||
//_mm_sfence(); // メモリーへの書き込みを完了してからスレッドを起動する
|
||||
hSub[j] = (HANDLE)_beginthreadex(NULL, STACK_SIZE, thread_func, (LPVOID)th, 0, NULL);
|
||||
if (hSub[j] == NULL){
|
||||
print_win32_err();
|
||||
CloseHandle(hRun[j]);
|
||||
CloseHandle(hEnd[j]);
|
||||
printf("error, inv-thread\n");
|
||||
err = 1;
|
||||
goto error_end;
|
||||
}
|
||||
WaitForSingleObject(hEnd[j], INFINITE); // 設定終了の合図を待つ (リセットする)
|
||||
}
|
||||
|
||||
// Gaussian Elimination with 1 matrix
|
||||
th->pivot = 0;
|
||||
th->start = 0; // その行の開始位置
|
||||
for (th->skip = 0; th->skip < rows; th->skip++){
|
||||
// 経過表示
|
||||
if (GetTickCount() - time_last >= UPDATE_TIME){
|
||||
if (print_progress((th->skip * 1000) / rows)){
|
||||
err = 2;
|
||||
goto error_end;
|
||||
}
|
||||
time_last = GetTickCount();
|
||||
}
|
||||
|
||||
// その行 (パリティ・ブロック) がどのソース・ブロックの代用か
|
||||
while ((th->pivot < cols) && (s_blk[th->pivot].exist != 0))
|
||||
th->pivot++;
|
||||
|
||||
// Divide the row by element i,pivot
|
||||
factor = mat[th->start + th->pivot];
|
||||
if (factor > 1){
|
||||
mat[th->start + th->pivot] = 1; // これが行列を一個で済ます手
|
||||
galois_region_divide(mat + th->start, cols, factor);
|
||||
} else if (factor == 0){ // factor = 0 だと、その行列の逆行列を計算できない
|
||||
err = (0x00010000 | th->pivot); // どのソース・ブロックで問題が発生したのかを返す
|
||||
goto error_end;
|
||||
}
|
||||
|
||||
// 別の行の同じ pivot 列が 0以外なら、その値を 0にするために、
|
||||
// i 行を何倍かしたものを XOR する
|
||||
th->now = rows; // 初期値 + 1
|
||||
//_mm_sfence(); // メモリーへの書き込みを完了してからスレッドを再開する
|
||||
for (j = 0; j < sub_num; j++)
|
||||
SetEvent(hRun[j]); // サブ・スレッドに計算を開始させる
|
||||
while ((j = InterlockedDecrement(&(th->now))) >= 0){ // j = --th_now
|
||||
if (j == th->skip) // 同じ行はとばす
|
||||
continue;
|
||||
row_start2 = cols * j; // その行の開始位置
|
||||
factor = mat[row_start2 + th->pivot]; // j 行の pivot 列の値
|
||||
mat[row_start2 + th->pivot] = 0; // これが行列を一個で済ます手
|
||||
// 先の計算により、i 行の pivot 列の値は必ず 1なので、この factor が倍率になる
|
||||
galois_region_multiply(mat + th->start, mat + row_start2, cols, factor);
|
||||
}
|
||||
|
||||
WaitForMultipleObjects(sub_num, hEnd, TRUE, INFINITE); // サブ・スレッドの計算終了の合図を待つ
|
||||
th->start += cols;
|
||||
th->pivot++;
|
||||
}
|
||||
|
||||
error_end:
|
||||
InterlockedExchange(&(th->skip), -1); // 終了指示
|
||||
for (j = 0; j < sub_num; j++){
|
||||
if (hSub[j]){ // サブ・スレッドを終了させる
|
||||
SetEvent(hRun[j]);
|
||||
WaitForSingleObject(hSub[j], INFINITE);
|
||||
CloseHandle(hSub[j]);
|
||||
}
|
||||
}
|
||||
return err;
|
||||
}
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
/*
|
||||
@@ -539,11 +753,9 @@ unsigned int time_total = GetTickCount();
|
||||
}
|
||||
|
||||
// パリティ計算用の行列演算の準備をする
|
||||
if (parity_num > source_num){
|
||||
len = sizeof(unsigned short) * (source_num + parity_num);
|
||||
} else {
|
||||
len = sizeof(unsigned short) * source_num * 2;
|
||||
}
|
||||
len = sizeof(unsigned short) * source_num;
|
||||
if (OpenCL_method != 0)
|
||||
len *= 2; // GPU の作業領域も確保しておく
|
||||
constant = malloc(len);
|
||||
if (constant == NULL){
|
||||
printf("malloc, %d\n", len);
|
||||
@@ -551,7 +763,11 @@ unsigned int time_total = GetTickCount();
|
||||
goto error_end;
|
||||
}
|
||||
#ifdef TIMER
|
||||
printf("\nmatrix size = %d.%d KB\n", len >> 10, (len >> 10) % 10);
|
||||
if (len & 0xFFFFF000){
|
||||
printf("\nmatrix size = %u KB\n", len >> 10);
|
||||
} else {
|
||||
printf("\nmatrix size = %u Bytes\n", len);
|
||||
}
|
||||
#endif
|
||||
// パリティ検査行列の基になる定数
|
||||
make_encode_constant(constant);
|
||||
@@ -623,11 +839,9 @@ unsigned int time_total = GetTickCount();
|
||||
}
|
||||
|
||||
// パリティ計算用の行列演算の準備をする
|
||||
if (parity_num > source_num){
|
||||
len = sizeof(unsigned short) * (source_num + parity_num);
|
||||
} else {
|
||||
len = sizeof(unsigned short) * source_num * 2;
|
||||
}
|
||||
len = sizeof(unsigned short) * source_num;
|
||||
if (OpenCL_method != 0)
|
||||
len *= 2; // GPU の作業領域も確保しておく
|
||||
constant = malloc(len);
|
||||
if (constant == NULL){
|
||||
printf("malloc, %d\n", len);
|
||||
@@ -635,7 +849,11 @@ unsigned int time_total = GetTickCount();
|
||||
goto error_end;
|
||||
}
|
||||
#ifdef TIMER
|
||||
printf("\nmatrix size = %d.%d KB\n", len >> 10, (len >> 10) % 10);
|
||||
if (len & 0xFFFFF000){
|
||||
printf("\nmatrix size = %u KB\n", len >> 10);
|
||||
} else {
|
||||
printf("\nmatrix size = %u Bytes\n", len);
|
||||
}
|
||||
#endif
|
||||
// パリティ検査行列の基になる定数
|
||||
make_encode_constant(constant);
|
||||
@@ -719,9 +937,11 @@ unsigned int time_matrix = 0, time_total = GetTickCount();
|
||||
}
|
||||
#ifdef TIMER
|
||||
if (len & 0xFFF00000){
|
||||
printf("\nmatrix size = %d.%d MB\n", len >> 20, (len >> 20) % 10);
|
||||
printf("\nmatrix size = %u MB\n", len >> 20);
|
||||
} else if (len & 0x000FF000){
|
||||
printf("\nmatrix size = %u KB\n", len >> 10);
|
||||
} else {
|
||||
printf("\nmatrix size = %d.%d KB\n", len >> 10, (len >> 10) % 10);
|
||||
printf("\nmatrix size = %u Bytes\n", len);
|
||||
}
|
||||
#endif
|
||||
// 何番目の消失ソース・ブロックがどのパリティで代替されるか
|
||||
@@ -782,8 +1002,7 @@ time_matrix = GetTickCount() - time_matrix;
|
||||
// ブロック数が多いなら、ブロックごとにスレッドを割り当てる (GPU を使う)
|
||||
if (memory_use & 16){
|
||||
err = -4; // SSD なら Read all 方式でブロックが断片化しても速い
|
||||
} else
|
||||
if (read_block_num(block_lost, 2, 0, MEM_UNIT) != 0){
|
||||
} else if (read_block_num(block_lost * 2, 0, MEM_UNIT) != 0){
|
||||
err = -5; // HDD でメモリーが足りてるなら Read some 方式を使う
|
||||
} else {
|
||||
err = -4; // メモリー不足なら Read all 方式でブロックを断片化させる
|
||||
@@ -792,8 +1011,7 @@ time_matrix = GetTickCount() - time_matrix;
|
||||
// ソース・ブロックを全て断片的に読み込むか、いくつかを丸ごと読み込むかを決める
|
||||
if (memory_use & 16){
|
||||
err = -2; // SSD なら Read all 方式でブロックが断片化しても速い
|
||||
} else
|
||||
if (read_block_num(block_lost, cpu_num - 1, 0, sse_unit) != 0){
|
||||
} else if (read_block_num(block_lost, 0, sse_unit) != 0){
|
||||
err = -3; // HDD でメモリーが足りてるなら Read some 方式を使う
|
||||
} else {
|
||||
err = -2; // メモリー不足なら Read all 方式でブロックを断片化させる
|
||||
|
||||
@@ -10,7 +10,6 @@ extern "C" {
|
||||
|
||||
// Read all source & Keep some parity 方式
|
||||
// 部分的なエンコードを行う最低ブロック数
|
||||
#define PART_MAX_RATE 1 // ソース・ブロック数の 1/2 = 50%
|
||||
#define PART_MIN_RATE 5 // ソース・ブロック数の 1/32 = 3.1%
|
||||
|
||||
// Read some source & Keep all parity 方式
|
||||
@@ -18,6 +17,10 @@ extern "C" {
|
||||
#define READ_MIN_RATE 1 // 保持するブロック数の 1/2 = 50%
|
||||
#define READ_MIN_NUM 16
|
||||
|
||||
// CPU cache 最適化のため、同時に処理するブロック数を制限する
|
||||
#define CACHE_MIN_NUM 8
|
||||
#define CACHE_MAX_NUM 128
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
// Cache Blocking を試みる
|
||||
@@ -33,10 +36,15 @@ unsigned int get_io_size(
|
||||
// 何ブロックまとめてファイルから読み込むかを空きメモリー量から計算する
|
||||
int read_block_num(
|
||||
int keep_num, // 保持するパリティ・ブロック数
|
||||
int add_num, // 余裕を見るブロック数
|
||||
size_t trial_alloc, // 確保できるか確認するのか
|
||||
int alloc_unit); // メモリー単位の境界 (sse_unit か MEM_UNIT)
|
||||
|
||||
// 1st encode, decode を何スレッドで実行するか決める
|
||||
int calc_thread_num1(int max_num);
|
||||
|
||||
// 1st & 2nd encode, decode を何スレッドで実行するか決める
|
||||
int calc_thread_num2(int max_num, int *cpu_num2);
|
||||
|
||||
/* * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * */
|
||||
|
||||
// リード・ソロモン符号を使ってエンコードする
|
||||
|
||||
@@ -1,8 +1,8 @@
|
||||
1 RT_STRING ".\\source.cl"
|
||||
|
||||
1 VERSIONINFO
|
||||
FILEVERSION 1,3,2,8
|
||||
PRODUCTVERSION 1,3,2,0
|
||||
FILEVERSION 1,3,3,1
|
||||
PRODUCTVERSION 1,3,3,0
|
||||
FILEOS 0x40004
|
||||
FILETYPE 0x1
|
||||
{
|
||||
@@ -13,8 +13,8 @@ BLOCK "StringFileInfo"
|
||||
VALUE "FileDescription", "PAR2 client"
|
||||
VALUE "LegalCopyright", "Copyright (C) 2023 Yutaka Sawada"
|
||||
VALUE "ProductName", "par2j"
|
||||
VALUE "FileVersion", "1.3.2.8"
|
||||
VALUE "ProductVersion", "1.3.2.0"
|
||||
VALUE "FileVersion", "1.3.3.1"
|
||||
VALUE "ProductVersion", "1.3.3.0"
|
||||
}
|
||||
}
|
||||
|
||||
|
||||
File diff suppressed because it is too large
Load Diff
File diff suppressed because it is too large
Load Diff
@@ -86,59 +86,6 @@ __kernel void method2(
|
||||
}
|
||||
}
|
||||
|
||||
__kernel void method3(
|
||||
__global uint *src,
|
||||
__global uint *dst,
|
||||
__global ushort *factors,
|
||||
int blk_num)
|
||||
{
|
||||
__global uint *blk_src;
|
||||
__local uint mtab[512];
|
||||
int i, blk, chk_size, remain, pos;
|
||||
uint lo, hi, sum1, sum2;
|
||||
const int work_id = get_global_id(0) * 2;
|
||||
const int work_size = get_global_size(0) * 2;
|
||||
const int table_id = get_local_id(0);
|
||||
|
||||
remain = BLK_SIZE;
|
||||
chk_size = CHK_SIZE;
|
||||
while (remain > 0){
|
||||
if (chk_size > remain)
|
||||
chk_size = remain;
|
||||
|
||||
for (i = work_id; i < chk_size; i += work_size){
|
||||
dst[i ] = 0;
|
||||
dst[i + 1] = 0;
|
||||
}
|
||||
|
||||
blk_src = src;
|
||||
for (blk = 0; blk < blk_num; blk++){
|
||||
calc_table(mtab, table_id, factors[blk]);
|
||||
barrier(CLK_LOCAL_MEM_FENCE);
|
||||
|
||||
for (i = work_id; i < chk_size; i += work_size){
|
||||
pos = (i & ~7) + ((i & 7) >> 1);
|
||||
lo = blk_src[pos ];
|
||||
hi = blk_src[pos + 4];
|
||||
sum1 = mtab[(uchar)(lo >> 16)] ^ mtab[256 + (uchar)(hi >> 16)];
|
||||
sum2 = mtab[lo >> 24] ^ mtab[256 + (hi >> 24)];
|
||||
sum1 <<= 16;
|
||||
sum2 <<= 16;
|
||||
sum1 ^= mtab[(uchar)lo] ^ mtab[256 + (uchar)hi];
|
||||
sum2 ^= mtab[(uchar)(lo >> 8)] ^ mtab[256 + (uchar)(hi >> 8)];
|
||||
dst[pos ] ^= (sum1 & 0x00FF00FF) | ((sum2 & 0x00FF00FF) << 8);
|
||||
dst[pos + 4] ^= ((sum1 & 0xFF00FF00) >> 8) | (sum2 & 0xFF00FF00);
|
||||
}
|
||||
blk_src += BLK_SIZE;
|
||||
barrier(CLK_LOCAL_MEM_FENCE);
|
||||
}
|
||||
|
||||
src += CHK_SIZE;
|
||||
dst += CHK_SIZE;
|
||||
remain -= CHK_SIZE;
|
||||
}
|
||||
}
|
||||
|
||||
__kernel void method4(
|
||||
__global uint *src,
|
||||
__global uint *dst,
|
||||
|
||||
@@ -1,2 +1,2 @@
|
||||
#define FILE_VERSION "1.3.2.8" // ファイルのバージョン番号
|
||||
#define PRODUCT_VERSION "1.3.2" // 製品のバージョン番号
|
||||
#define FILE_VERSION "1.3.3.1" // ファイルのバージョン番号
|
||||
#define PRODUCT_VERSION "1.3.3" // 製品のバージョン番号
|
||||
|
||||
Reference in New Issue
Block a user