diff options
33 files changed, 465 insertions, 223 deletions
diff --git a/CMakeLists.txt b/CMakeLists.txt index 10f33bd..3b3c7eb 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -1,6 +1,6 @@ project(xmr-stak) -cmake_minimum_required(VERSION 3.1.0) +cmake_minimum_required(VERSION 3.4.0) # enforce C++11 set(CMAKE_CXX_STANDARD_REQUIRED ON) @@ -522,9 +522,11 @@ endif() file(GLOB SRCFILES_CPP "xmrstak/cli/*.cpp") set_source_files_properties(${SRCFILES_CPP} PROPERTIES LANGUAGE CXX) -add_executable(xmr-stak - ${SRCFILES_CPP} -) +if(CMAKE_CXX_COMPILER_ID MATCHES "MSVC") + add_executable(xmr-stak ${SRCFILES_CPP} xmrstak/cli/xmr-stak.manifest) +else() + add_executable(xmr-stak ${SRCFILES_CPP}) +endif() set(EXECUTABLE_OUTPUT_PATH "bin") set(LIBRARY_OUTPUT_PATH "bin") @@ -8,7 +8,7 @@ ENV XMRSTAK_CMAKE_FLAGS -DXMR-STAK_COMPILE=generic -DCUDA_ENABLE=ON -DOpenCL_ENA # Innstall packages RUN apt-get update \ && set -x \ - && apt-get install -qq --no-install-recommends -y ca-certificates cmake cuda-core-9-0 git cuda-cudart-dev-9-0 libhwloc-dev libmicrohttpd-dev libssl-dev \ + && apt-get install -qq --no-install-recommends -y build-essential ca-certificates cmake cuda-core-9-0 git cuda-cudart-dev-9-0 libhwloc-dev libmicrohttpd-dev libssl-dev \ && git clone $GIT_REPOSITORY \ && cd /xmr-stak \ && cmake ${XMRSTAK_CMAKE_FLAGS} . \ @@ -16,7 +16,7 @@ RUN apt-get update \ && cd - \ && mv /xmr-stak/bin/* /usr/local/bin/ \ && rm -rf /xmr-stak \ - && apt-get purge -y -qq cmake cuda-core-9-0 git cuda-cudart-dev-9-0 libhwloc-dev libmicrohttpd-dev libssl-dev \ + && apt-get purge -y -qq build-essential cmake cuda-core-9-0 git cuda-cudart-dev-9-0 libhwloc-dev libmicrohttpd-dev libssl-dev \ && apt-get clean -qq VOLUME /mnt @@ -21,15 +21,15 @@ XMR-Stak is a universal Stratum pool miner. This miner supports CPUs, AMD and NV ## Features - support all common backends (CPU/x86, AMD-GPU and NVIDIA-GPU) -- support all common OS (Linux, Windows and MacOS) +- support all common OS (Linux, Windows and macOS) - supports algorithm cryptonight for Monero (XMR) and cryptonight-light (AEON) - easy to use - guided start (no need to edit a config file for the first start) - auto configuration for each backend - open source software (GPLv3) - TLS support -- HTML statistics -- JSON API for monitoring +- [HTML statistics](doc/usage.md#html-and-json-api-report-configuraton) +- [JSON API for monitoring](doc/usage.md#html-and-json-api-report-configuraton) ## Supported altcoins @@ -62,7 +62,7 @@ fireice-uk: psychocrypt: ``` -43NoJVEXo21hGZ6tDG6Z3g4qimiGdJPE6GRxAmiWwm26gwr62Lqo7zRiCJFSBmbkwTGNuuES9ES5TgaVHceuYc4Y75txCTU +45tcqnJMgd3VqeTznNotiNj4G9PQoK67TGRiHyj6EYSZ31NUbAfs9XdiU5squmZb717iHJLxZv3KfEw8jCYGL5wa19yrVCn ``` ## Release Checksums @@ -1,7 +1,7 @@ # FAQ ## Content Overview -* [SeLockMemoryPrivilege failed](#selockmemoryprivilege-failed) +* ["Obtaining SeLockMemoryPrivilege failed."](#obtaining-selockmemoryprivilege-failed) * [VirtualAlloc failed](#virtualalloc-failed) * [Error msvcp140.dll and vcruntime140.dll not available](#error-msvcp140dll-and-vcruntime140dll-not-available) * [Error: MEMORY ALLOC FAILED: mmap failed](#error-memory-alloc-failed-mmap-failed) @@ -9,23 +9,24 @@ * [Virus Protection Alert](#virus-protection-alert) * [Change Currency to Mine](#change-currency-to-mine) -## SeLockMemoryPrivilege failed +## "Obtaining SeLockMemoryPrivilege failed." -Please see [config.txt](config.txt) under section **LARGE PAGE SUPPORT** +For professional versions of Windows see [this article](https://msdn.microsoft.com/en-gb/library/ms190730.aspx). +Make sure to reboot afterwards! -For Windows 7 pro, or Windows 8 and above see [this article](https://msdn.microsoft.com/en-gb/library/ms190730.aspx) (make sure to reboot afterwards!). +For Windows 7/10 Home: -For Windows 7 Home : +1) Download and install [Windows Server 2003 Resource Kit Tools](https://www.microsoft.com/en-us/download/details.aspx?id=17657). Ignore any incompatibility warning during installation. -1) Download and install [Windows Server 2003 Resource Kit Tools](https://www.microsoft.com/en-us/download/details.aspx?id=17657). Ignore incompatiablity warning during installation. +2) Open cmd or PowerShell as an administrator. -2) In cmd or power shell: `ntrights -u %USERNAME% +r SeLockMemoryPrivilege` (where %USERNAME% is the user that will be running the program. This command needs to be run as admin) +3) Use `ntrights -u %USERNAME% +r SeLockMemoryPrivilege` where %USERNAME% is the user that will be running the program. -3) Reboot. +4) Reboot. Reference: http://rybkaforum.net/cgi-bin/rybkaforum/topic_show.pl?pid=259791#pid259791 -*Warning: do not download ntrights.exe from any other site other then the offical Microsoft download page.* +*Warning: Do not download ntrights.exe from any other site other than the offical Microsoft download page.* ## VirtualAlloc failed diff --git a/doc/Linux_deployment.md b/doc/Linux_deployment.md index 323a97f..3219e8a 100644 --- a/doc/Linux_deployment.md +++ b/doc/Linux_deployment.md @@ -15,6 +15,7 @@ For automatic deployments, please use the steps above to obtain config.txt and u ``` #!/bin/bash +sudo apt install curl curl -O `curl -s https://api.github.com/repos/fireice-uk/xmr-stak/releases/latest | grep -o 'browser_download_url.*xmr-stak-portbin-linux.tar.gz' | sed 's/.*"//'` curl -O http://path.to/your/config.txt tar xzf xmr-stak-portbin-linux.tar.gz diff --git a/doc/compile.md b/doc/compile.md index e97affa..771c9d1 100644 --- a/doc/compile.md +++ b/doc/compile.md @@ -9,7 +9,7 @@ * [Compile on Windows](compile_Windows.md) * [Compile on Linux](compile_Linux.md) * [Compile on FreeBSD](compile_FreeBSD.md) -* [Compile on MacOS](compile_MacOS.md) +* [Compile on macOS](compile_macOS.md) ## Build System @@ -31,7 +31,7 @@ After the configuration you need to compile the miner, follow the guide for your * [Compile in Windows](compile_Windows.md) * [Compile in Linux](compile_Linux.md) * [Compile in FreeBSD](compile_FreeBSD.md) -* [Compile in MacOS](compile_MacOS.md) +* [Compile in macOS](compile_macOS.md) ## Generic Build Options - `CMAKE_INSTALL_PREFIX` install miner to the home folder diff --git a/doc/compile_Linux.md b/doc/compile_Linux.md index b7104ac..8d48e72 100644 --- a/doc/compile_Linux.md +++ b/doc/compile_Linux.md @@ -25,7 +25,7 @@ make install # Arch - sudo pacman -S base-devel hwloc openssl cmake libmicrohttpd + sudo pacman -S --needed base-devel hwloc openssl cmake libmicrohttpd git clone https://github.com/fireice-uk/xmr-stak.git mkdir xmr-stak/build cd xmr-stak/build @@ -64,6 +64,34 @@ cd xmr-stak/build cmake .. make install + + # TinyCore Linux 8.x + # TinyCore is 32-bit only, but there is an x86-64 port, known as "Pure 64," + # hosted on the TinyCore home page, and it works well. + # Beware that huge page support is not enabled in the kernel distributed + # with Pure 64. Consider http://wiki.tinycorelinux.net/wiki:custom_kernel + # Note that as of yet there are no distro packages for microhttpd or hwloc. + # hwloc is easy enough to install manually though, shown below. + # Also note that only CPU mining has been tested on this platform, thus the + # disabling of CUDA and OpenCL shown below. + tce-load -iw openssl-dev.tcz cmake.tcz make.tcz gcc.tcz git.tcz \ + glibc_base-dev.tcz linux-4.8.1_api_headers.tcz \ + glibc_add_lib.tcz + wget https://www.open-mpi.org/software/hwloc/v1.11/downloads/hwloc-1.11.8.tar.gz + tar xzvf hwloc-1.11.8.tar.gz + cd hwloc-1.11.8 + ./configure --prefix=/usr/local + make + sudo make install + cd .. + git clone http://github.com/fireice-uk/xmr-stak + cd xmr-stak + mkdir build + cd build + CC=gcc cmake .. -DCUDA_ENABLE=OFF \ + -DOpenCL_ENABLE=OFF \ + -DMICROHTTPD_ENABLE=OFF + make install ``` - g++ version 5.1 or higher is required for full C++11 support. diff --git a/doc/compile_Windows.md b/doc/compile_Windows.md index 246a2db..d1be866 100644 --- a/doc/compile_Windows.md +++ b/doc/compile_Windows.md @@ -13,6 +13,7 @@ - during the install chose the components - `Desktop development with C++` (left side) - `VC++ 2015.3 v140 toolset for desktop` (right side) + - Since release of VS2017 15.5 (12/04/17), require `VC++ 2017 version 15.4 v14.11 toolset` (under tab `Individual Components`, section `Compilers, build tools, and runtimes`), as CUDA 9.1 is not compatible with compiler 14.12.X ### CMake for Win64 @@ -22,7 +23,7 @@ ### Cuda 8.0+ (only needed to use NVIDIA GPUs) -- donwload and install [https://developer.nvidia.com/cuda-downloads](https://developer.nvidia.com/cuda-downloads) +- download and install [https://developer.nvidia.com/cuda-downloads](https://developer.nvidia.com/cuda-downloads) - for minimal install choose `Custom installation options` during the install and select - CUDA/Development - CUDA/Visual Studio Integration (ignore the warning during the install that VS2017 is not supported) @@ -75,16 +76,21 @@ ## Compile -- download and unzip `xmr-stak` +- download `xmr-stak` [Source Code.zip](https://github.com/fireice-uk/xmr-stak/releases) and save to a location in your home folder (C:\Users\USERNAME\) +- extract the Source Code.zip file - open the command line terminal `cmd` -- `cd` to your unzipped source code directory -- execute the following commands (NOTE: path to VS2017 can be different) +- `cd` to the extracted source code directory +- execute the following commands (NOTE: path to Visual Studio 2017 Community can be different) ``` + # Next line is only if compiling for Cuda 9.1 and using Visual Studio 2017 >= 15.5 (released 12/04/17) + "C:\Program Files (x86)\Microsoft Visual Studio\2017\Community\VC\Auxiliary\Build\vcvarsall.bat" x64 -vcvars_ver=14.11 "C:\Program Files (x86)\Microsoft Visual Studio\2017\Community\Common7\Tools\VsMSBuildCmd.bat" set CMAKE_PREFIX_PATH=C:\xmr-stak-dep\hwloc;C:\xmr-stak-dep\libmicrohttpd;C:\xmr-stak-dep\openssl mkdir build cd build ``` + - Sometimes Windows will `cd` to `C:\Users\USERNAME\source\build\` instead of `C:\Users\USERNAME\xmr-stak-<version>\build`. Ensure you are in the correct `build` directory before proceeding. + - for CUDA 8* ``` cmake -G "Visual Studio 15 2017 Win64" -T v140,host=x64 .. diff --git a/doc/compile_MacOS.md b/doc/compile_macOS.md index 1b0af91..46f1d5b 100644 --- a/doc/compile_MacOS.md +++ b/doc/compile_macOS.md @@ -1,4 +1,4 @@ -# Compile **xmr-stak** for MacOS +# Compile **xmr-stak** for macOS ## Dependencies @@ -18,7 +18,13 @@ make install ### For AMD GPUs -> 🖐 We need help with AMD GPU compilation instructions. Please submit a PR if you managed to install [AMD APP SDK](http://developer.amd.com/amd-accelerated-parallel-processing-app-sdk/) and to compile `xmr-stak` on MacOS. +OpenCL is bundled with Xcode, so no other depedency then the basic ones needed. Just enable OpenCL via the `-DOpenCL_ENABLE=ON` CMake option. + +```shell +brew install hwloc libmicrohttpd gcc openssl cmake +cmake . -DOPENSSL_ROOT_DIR=/usr/local/opt/openssl -DCUDA_ENABLE=OFF -DOpenCL_ENABLE=ON +make install +``` ### For CPU-only mining diff --git a/doc/usage.md b/doc/usage.md index 60cf69b..a810469 100644 --- a/doc/usage.md +++ b/doc/usage.md @@ -23,7 +23,7 @@ The number of files depends on the available backends. `set XMRSTAK_NOWAIT=1` disable the dialog `Press any key to exit.` for non UAC execution. -## Usage on Linux & MacOS +## Usage on Linux & macOS 1) Open a terminal within the folder with the binary 2) Start the miner with `./xmr-stak` diff --git a/xmrstak/backend/amd/amd_gpu/opencl/cryptonight.cl b/xmrstak/backend/amd/amd_gpu/opencl/cryptonight.cl index 255fcbb..ec05712 100644 --- a/xmrstak/backend/amd/amd_gpu/opencl/cryptonight.cl +++ b/xmrstak/backend/amd/amd_gpu/opencl/cryptonight.cl @@ -653,21 +653,11 @@ __kernel void cn2(__global uint4 *Scratchpad, __global ulong *states, __global u for(int i = 0; i < 25; ++i) states[i] = State[i]; - switch(State[0] & 3) - { - case 0: - Branch0[atomic_inc(Branch0 + Threads)] = get_global_id(0) - get_global_offset(0); - break; - case 1: - Branch1[atomic_inc(Branch1 + Threads)] = get_global_id(0) - get_global_offset(0); - break; - case 2: - Branch2[atomic_inc(Branch2 + Threads)] = get_global_id(0) - get_global_offset(0); - break; - case 3: - Branch3[atomic_inc(Branch3 + Threads)] = get_global_id(0) - get_global_offset(0); - break; - } + ulong StateSwitch = State[0] & 3; + __global uint *destinationBranch1 = StateSwitch == 0 ? Branch0 : Branch1; + __global uint *destinationBranch2 = StateSwitch == 2 ? Branch2 : Branch3; + __global uint *destinationBranch = StateSwitch < 2 ? destinationBranch1 : destinationBranch2; + destinationBranch[atomic_inc(destinationBranch + Threads)] = gIdx; } } mem_fence(CLK_GLOBAL_MEM_FENCE); @@ -704,8 +694,7 @@ __kernel void Skein(__global ulong *states, __global uint *BranchBuf, __global u for(uint i = 0; i < 4; ++i) { - if(i < 3) t[0] += 0x40UL; - else t[0] += 0x08UL; + t[0] += i < 3 ? 0x40UL : 0x08UL; t[2] = t[0] ^ t[1]; @@ -715,8 +704,7 @@ __kernel void Skein(__global ulong *states, __global uint *BranchBuf, __global u h = m ^ p; - if(i < 2) t[1] = 0x3000000000000000UL; - else t[1] = 0xB000000000000000UL; + t[1] = i < 2 ? 0x3000000000000000UL : 0xB000000000000000UL; } t[0] = 0x08UL; @@ -744,6 +732,27 @@ __kernel void Skein(__global ulong *states, __global uint *BranchBuf, __global u #define SWAP8(x) as_ulong(as_uchar8(x).s76543210) +#define JHXOR \ + h0h ^= input[0]; \ + h0l ^= input[1]; \ + h1h ^= input[2]; \ + h1l ^= input[3]; \ + h2h ^= input[4]; \ + h2l ^= input[5]; \ + h3h ^= input[6]; \ + h3l ^= input[7]; \ +\ + E8; \ +\ + h4h ^= input[0]; \ + h4l ^= input[1]; \ + h5h ^= input[2]; \ + h5l ^= input[3]; \ + h6h ^= input[4]; \ + h6l ^= input[5]; \ + h7h ^= input[6]; \ + h7l ^= input[7] + __kernel void JH(__global ulong *states, __global uint *BranchBuf, __global uint *output, ulong Target, ulong Threads) { const uint idx = get_global_id(0) - get_global_offset(0); @@ -757,46 +766,27 @@ __kernel void JH(__global ulong *states, __global uint *BranchBuf, __global uint sph_u64 h4h = 0x754D2E7F8996A371UL, h4l = 0x62E27DF70849141DUL, h5h = 0x948F2476F7957627UL, h5l = 0x6C29804757B6D587UL, h6h = 0x6C0D8EAC2D275E5CUL, h6l = 0x0F7A0557C6508451UL, h7h = 0xEA12247067D3E47BUL, h7l = 0x69D71CD313ABE389UL; sph_u64 tmp; - for(int i = 0; i < 5; ++i) + for(int i = 0; i < 3; ++i) { ulong input[8]; - if(i < 3) - { - for(int x = 0; x < 8; ++x) input[x] = (states[(i << 3) + x]); - } - else if(i == 3) - { - input[0] = (states[24]); - input[1] = 0x80UL; - for(int x = 2; x < 8; ++x) input[x] = 0x00UL; - } - else - { - input[7] = 0x4006000000000000UL; - - for(int x = 0; x < 7; ++x) input[x] = 0x00UL; - } - - h0h ^= input[0]; - h0l ^= input[1]; - h1h ^= input[2]; - h1l ^= input[3]; - h2h ^= input[4]; - h2l ^= input[5]; - h3h ^= input[6]; - h3l ^= input[7]; - - E8; - - h4h ^= input[0]; - h4l ^= input[1]; - h5h ^= input[2]; - h5l ^= input[3]; - h6h ^= input[4]; - h6l ^= input[5]; - h7h ^= input[6]; - h7l ^= input[7]; + const int shifted = i << 3; + for(int x = 0; x < 8; ++x) input[x] = (states[shifted + x]); + JHXOR; + } + { + ulong input[8]; + input[0] = (states[24]); + input[1] = 0x80UL; + #pragma unroll 6 + for(int x = 2; x < 8; ++x) input[x] = 0x00UL; + JHXOR; + } + { + ulong input[8]; + for(int x = 0; x < 7; ++x) input[x] = 0x00UL; + input[7] = 0x4006000000000000UL; + JHXOR; } //output[0] = h6h; @@ -832,6 +822,7 @@ __kernel void Blake(__global ulong *states, __global uint *BranchBuf, __global u ((uint8 *)h)[0] = vload8(0U, c_IV256); + #pragma unroll 4 for(uint i = 0, bitlen = 0; i < 4; ++i) { if(i < 3) @@ -907,6 +898,7 @@ __kernel void Groestl(__global ulong *states, __global uint *BranchBuf, __global State[7] = 0x0001000000000000UL; + #pragma unroll 4 for(uint i = 0; i < 4; ++i) { ulong H[8], M[8]; diff --git a/xmrstak/backend/amd/amd_gpu/opencl/wolf-skein.cl b/xmrstak/backend/amd/amd_gpu/opencl/wolf-skein.cl index 868757b..279b652 100644 --- a/xmrstak/backend/amd/amd_gpu/opencl/wolf-skein.cl +++ b/xmrstak/backend/amd/amd_gpu/opencl/wolf-skein.cl @@ -3,6 +3,7 @@ R"===( #define WOLF_SKEIN_CL // Vectorized Skein implementation macros and functions by Wolf +// Updated by taisel #define SKEIN_KS_PARITY 0x1BD11BDAA9FC1A22 @@ -22,11 +23,11 @@ static const __constant ulong SKEIN512_256_IV[8] = 0xC36FBAF9393AD185UL, 0x3EEDBA1833EDFC13UL }; -#define SKEIN_INJECT_KEY(p, s) do { \ +#define SKEIN_INJECT_KEY(p, s, q) do { \ p += h; \ - p.s5 += t[s % 3]; \ - p.s6 += t[(s + 1) % 3]; \ - p.s7 += s; \ + p.s5 += t[s]; \ + p.s6 += t[select(s + 1U, 0U, s == 2U)]; \ + p.s7 += q; \ } while(0) ulong SKEIN_ROT(const uint2 x, const uint y) @@ -35,55 +36,55 @@ ulong SKEIN_ROT(const uint2 x, const uint y) else return(as_ulong(amd_bitalign(x.s10, x, 32 - (y - 32)))); } -void SkeinMix8(ulong4 *pv0, ulong4 *pv1, const uint rc0, const uint rc1, const uint rc2, const uint rc3) +void SkeinMix8(ulong4 *pv0, ulong4 *pv1, const ulong4 rc) { *pv0 += *pv1; - (*pv1).s0 = SKEIN_ROT(as_uint2((*pv1).s0), rc0); - (*pv1).s1 = SKEIN_ROT(as_uint2((*pv1).s1), rc1); - (*pv1).s2 = SKEIN_ROT(as_uint2((*pv1).s2), rc2); - (*pv1).s3 = SKEIN_ROT(as_uint2((*pv1).s3), rc3); + (*pv1).s0 = SKEIN_ROT(as_uint2((*pv1).s0), rc.s0); + (*pv1).s1 = SKEIN_ROT(as_uint2((*pv1).s1), rc.s1); + (*pv1).s2 = SKEIN_ROT(as_uint2((*pv1).s2), rc.s2); + (*pv1).s3 = SKEIN_ROT(as_uint2((*pv1).s3), rc.s3); *pv1 ^= *pv0; } -ulong8 SkeinEvenRound(ulong8 p, const ulong8 h, const ulong *t, const uint s) +ulong8 SkeinEvenRound(ulong8 p, const ulong8 h, const ulong *t, const uint s, const uint q) { - SKEIN_INJECT_KEY(p, s); + SKEIN_INJECT_KEY(p, s, q); ulong4 pv0 = p.even, pv1 = p.odd; - SkeinMix8(&pv0, &pv1, 46, 36, 19, 37); + SkeinMix8(&pv0, &pv1, (ulong4)(46, 36, 19, 37)); pv0 = shuffle(pv0, (ulong4)(1, 2, 3, 0)); pv1 = shuffle(pv1, (ulong4)(0, 3, 2, 1)); - SkeinMix8(&pv0, &pv1, 33, 27, 14, 42); + SkeinMix8(&pv0, &pv1, (ulong4)(33, 27, 14, 42)); pv0 = shuffle(pv0, (ulong4)(1, 2, 3, 0)); pv1 = shuffle(pv1, (ulong4)(0, 3, 2, 1)); - SkeinMix8(&pv0, &pv1, 17, 49, 36, 39); + SkeinMix8(&pv0, &pv1, (ulong4)(17, 49, 36, 39)); pv0 = shuffle(pv0, (ulong4)(1, 2, 3, 0)); pv1 = shuffle(pv1, (ulong4)(0, 3, 2, 1)); - SkeinMix8(&pv0, &pv1, 44, 9, 54, 56); + SkeinMix8(&pv0, &pv1, (ulong4)(44, 9, 54, 56)); return(shuffle2(pv0, pv1, (ulong8)(1, 4, 2, 7, 3, 6, 0, 5))); } -ulong8 SkeinOddRound(ulong8 p, const ulong8 h, const ulong *t, const uint s) +ulong8 SkeinOddRound(ulong8 p, const ulong8 h, const ulong *t, const uint s, const uint q) { - SKEIN_INJECT_KEY(p, s); + SKEIN_INJECT_KEY(p, s, q); ulong4 pv0 = p.even, pv1 = p.odd; - SkeinMix8(&pv0, &pv1, 39, 30, 34, 24); + SkeinMix8(&pv0, &pv1, (ulong4)(39, 30, 34, 24)); pv0 = shuffle(pv0, (ulong4)(1, 2, 3, 0)); pv1 = shuffle(pv1, (ulong4)(0, 3, 2, 1)); - SkeinMix8(&pv0, &pv1, 13, 50, 10, 17); + SkeinMix8(&pv0, &pv1, (ulong4)(13, 50, 10, 17)); pv0 = shuffle(pv0, (ulong4)(1, 2, 3, 0)); pv1 = shuffle(pv1, (ulong4)(0, 3, 2, 1)); - SkeinMix8(&pv0, &pv1, 25, 29, 39, 43); + SkeinMix8(&pv0, &pv1, (ulong4)(25, 29, 39, 43)); pv0 = shuffle(pv0, (ulong4)(1, 2, 3, 0)); pv1 = shuffle(pv1, (ulong4)(0, 3, 2, 1)); - SkeinMix8(&pv0, &pv1, 8, 35, 56, 22); + SkeinMix8(&pv0, &pv1, (ulong4)(8, 35, 56, 22)); return(shuffle2(pv0, pv1, (ulong8)(1, 4, 2, 7, 3, 6, 0, 5))); } @@ -92,20 +93,47 @@ ulong8 Skein512Block(ulong8 p, ulong8 h, ulong h8, const ulong *t) #pragma unroll for(int i = 0; i < 18; ++i) { - p = SkeinEvenRound(p, h, t, i); + p = SkeinEvenRound(p, h, t, 0U, i); ++i; ulong tmp = h.s0; h = shuffle(h, (ulong8)(1, 2, 3, 4, 5, 6, 7, 0)); h.s7 = h8; h8 = tmp; - p = SkeinOddRound(p, h, t, i); + p = SkeinOddRound(p, h, t, 1U, i); + ++i; + tmp = h.s0; + h = shuffle(h, (ulong8)(1, 2, 3, 4, 5, 6, 7, 0)); + h.s7 = h8; + h8 = tmp; + p = SkeinEvenRound(p, h, t, 2U, i); + ++i; + tmp = h.s0; + h = shuffle(h, (ulong8)(1, 2, 3, 4, 5, 6, 7, 0)); + h.s7 = h8; + h8 = tmp; + p = SkeinOddRound(p, h, t, 0U, i); + ++i; + tmp = h.s0; + h = shuffle(h, (ulong8)(1, 2, 3, 4, 5, 6, 7, 0)); + h.s7 = h8; + h8 = tmp; + p = SkeinEvenRound(p, h, t, 1U, i); + ++i; + tmp = h.s0; + h = shuffle(h, (ulong8)(1, 2, 3, 4, 5, 6, 7, 0)); + h.s7 = h8; + h8 = tmp; + p = SkeinOddRound(p, h, t, 2U, i); tmp = h.s0; h = shuffle(h, (ulong8)(1, 2, 3, 4, 5, 6, 7, 0)); h.s7 = h8; h8 = tmp; } - SKEIN_INJECT_KEY(p, 18); + p += h; + p.s5 += t[0]; + p.s6 += t[1]; + p.s7 += 18; return(p); } diff --git a/xmrstak/backend/amd/autoAdjust.hpp b/xmrstak/backend/amd/autoAdjust.hpp index 0bc5239..511a712 100644 --- a/xmrstak/backend/amd/autoAdjust.hpp +++ b/xmrstak/backend/amd/autoAdjust.hpp @@ -101,7 +101,16 @@ private: * sowing down the memory performance because of TLB cache misses */ size_t maxThreads = 1000u; - if(ctx.name.compare("gfx901") == 0) + if( + ctx.name.compare("gfx901") == 0 || + ctx.name.compare("gfx904") == 0 || + // APU + ctx.name.compare("gfx902") == 0 || + // UNKNOWN + ctx.name.compare("gfx900") == 0 || + ctx.name.compare("gfx903") == 0 || + ctx.name.compare("gfx905") == 0 + ) { /* Increase the number of threads for AMD VEGA gpus. * Limit the number of threads based on the issue: https://github.com/fireice-uk/xmr-stak/issues/5#issuecomment-339425089 diff --git a/xmrstak/backend/amd/config.tpl b/xmrstak/backend/amd/config.tpl index af662f8..25b75a1 100644 --- a/xmrstak/backend/amd/config.tpl +++ b/xmrstak/backend/amd/config.tpl @@ -12,6 +12,9 @@ R"===( * [ * { "index" : 0, "intensity" : 1000, "worksize" : 8, "affine_to_cpu" : false, "strided_index" : true }, * ], + * If you do not wish to mine with your AMD GPU(s) then use: + * "gpu_threads_conf" : + * null, */ "gpu_threads_conf" : [ diff --git a/xmrstak/backend/amd/minethd.cpp b/xmrstak/backend/amd/minethd.cpp index e83527c..422c28c 100644 --- a/xmrstak/backend/amd/minethd.cpp +++ b/xmrstak/backend/amd/minethd.cpp @@ -139,7 +139,7 @@ std::vector<iBackend*>* minethd::thread_starter(uint32_t threadOffset, miner_wor if(cfg.cpu_aff >= 0) { #if defined(__APPLE__) - printer::inst()->print_msg(L1, "WARNING on MacOS thread affinity is only advisory."); + printer::inst()->print_msg(L1, "WARNING on macOS thread affinity is only advisory."); #endif printer::inst()->print_msg(L1, "Starting AMD GPU thread %d, affinity: %d.", i, (int)cfg.cpu_aff); @@ -245,7 +245,7 @@ void minethd::work_main() if ( (*((uint64_t*)(bResult + 24))) < oWork.iTarget) executor::inst()->push_event(ex_event(job_result(oWork.sJobID, results[i], bResult, iThreadNo), oWork.iPoolId)); else - executor::inst()->push_event(ex_event("AMD Invalid Result", oWork.iPoolId)); + executor::inst()->push_event(ex_event("AMD Invalid Result", pGpuCtx->deviceIdx, oWork.iPoolId)); } iCount += pGpuCtx->rawIntensity; diff --git a/xmrstak/backend/cpu/autoAdjust.hpp b/xmrstak/backend/cpu/autoAdjust.hpp index 7bdb14e..db805ec 100644 --- a/xmrstak/backend/cpu/autoAdjust.hpp +++ b/xmrstak/backend/cpu/autoAdjust.hpp @@ -33,25 +33,21 @@ class autoAdjust { public: - size_t hashMemSize; - size_t halfHashMemSize; - - autoAdjust() + bool printConfig() { + size_t hashMemSizeKB; + size_t halfHashMemSizeKB; + if(::jconf::inst()->IsCurrencyMonero()) { - hashMemSize = MONERO_MEMORY; - halfHashMemSize = hashMemSize / 2u; + hashMemSizeKB = MONERO_MEMORY / 1024u; + halfHashMemSizeKB = hashMemSizeKB / 2u; } else { - hashMemSize = AEON_MEMORY; - halfHashMemSize = hashMemSize / 2u; + hashMemSizeKB = AEON_MEMORY / 1024u; + halfHashMemSizeKB = hashMemSizeKB / 2u; } - } - - bool printConfig() - { configEditor configTpl{}; @@ -63,9 +59,10 @@ public: std::string conf; - if(!detectL3Size() || L3KB_size < halfHashMemSize || L3KB_size > (halfHashMemSize * 100u)) + + if(!detectL3Size() || L3KB_size < halfHashMemSizeKB || L3KB_size > (halfHashMemSizeKB * 2048u)) { - if(L3KB_size < halfHashMemSize || L3KB_size > (halfHashMemSize * 100)) + if(L3KB_size < halfHashMemSizeKB || L3KB_size > (halfHashMemSizeKB * 2048)) printer::inst()->print_msg(L0, "Autoconf failed: L3 size sanity check failed - %u KB.", L3KB_size); conf += std::string(" { \"low_power_mode\" : false, \"no_prefetch\" : true, \"affine_to_cpu\" : false },\n"); @@ -88,7 +85,7 @@ public: if(L3KB_size <= 0) break; - double_mode = L3KB_size / hashMemSize > (int32_t)(corecnt-i); + double_mode = L3KB_size / hashMemSizeKB > (int32_t)(corecnt-i); conf += std::string(" { \"low_power_mode\" : "); conf += std::string(double_mode ? "true" : "false"); @@ -107,9 +104,9 @@ public: aff_id++; if(double_mode) - L3KB_size -= hashMemSize * 2u; + L3KB_size -= hashMemSizeKB * 2u; else - L3KB_size -= hashMemSize; + L3KB_size -= hashMemSizeKB; } } @@ -142,7 +139,7 @@ private: } L3KB_size = ((get_masked(cpu_info[1], 31, 22) + 1) * (get_masked(cpu_info[1], 21, 12) + 1) * - (get_masked(cpu_info[1], 11, 0) + 1) * (cpu_info[2] + 1)) / halfHashMemSize; + (get_masked(cpu_info[1], 11, 0) + 1) * (cpu_info[2] + 1)) / 1024; return true; } diff --git a/xmrstak/backend/cpu/config.tpl b/xmrstak/backend/cpu/config.tpl index b21a22d..fc4acb9 100644 --- a/xmrstak/backend/cpu/config.tpl +++ b/xmrstak/backend/cpu/config.tpl @@ -24,6 +24,9 @@ R"===( * { "low_power_mode" : false, "no_prefetch" : true, "affine_to_cpu" : 0 }, * { "low_power_mode" : false, "no_prefetch" : true, "affine_to_cpu" : 1 }, * ], + * If you do not wish to mine with your CPU(s) then use: + * "cpu_threads_conf" : + * null, */ "cpu_threads_conf" : diff --git a/xmrstak/backend/cpu/crypto/cryptonight_aesni.h b/xmrstak/backend/cpu/crypto/cryptonight_aesni.h index 9b6e1dc..e4ccbc3 100644 --- a/xmrstak/backend/cpu/crypto/cryptonight_aesni.h +++ b/xmrstak/backend/cpu/crypto/cryptonight_aesni.h @@ -317,10 +317,9 @@ void cryptonight_hash(const void* input, size_t len, void* output, cryptonight_c _mm_store_si128((__m128i *)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx)); idx0 = _mm_cvtsi128_si64(cx); - bx0 = cx; - if(PREFETCH) _mm_prefetch((const char*)&l0[idx0 & MASK], _MM_HINT_T0); + bx0 = cx; uint64_t hi, lo, cl, ch; cl = ((uint64_t*)&l0[idx0 & MASK])[0]; @@ -329,15 +328,15 @@ void cryptonight_hash(const void* input, size_t len, void* output, cryptonight_c lo = _umul128(idx0, cl, &hi); al0 += hi; - ah0 += lo; ((uint64_t*)&l0[idx0 & MASK])[0] = al0; + al0 ^= cl; + if(PREFETCH) + _mm_prefetch((const char*)&l0[al0 & MASK], _MM_HINT_T0); + ah0 += lo; ((uint64_t*)&l0[idx0 & MASK])[1] = ah0; ah0 ^= ch; - al0 ^= cl; - idx0 = al0; - if(PREFETCH) - _mm_prefetch((const char*)&l0[idx0 & MASK], _MM_HINT_T0); + idx0 = al0; } // Optim - 90% time boundary diff --git a/xmrstak/backend/cpu/crypto/cryptonight_common.cpp b/xmrstak/backend/cpu/crypto/cryptonight_common.cpp index 8b2207d..1026b04 100644 --- a/xmrstak/backend/cpu/crypto/cryptonight_common.cpp +++ b/xmrstak/backend/cpu/crypto/cryptonight_common.cpp @@ -31,6 +31,7 @@ extern "C" #include "cryptonight.h" #include "cryptonight_aesni.h" #include "xmrstak/backend/cryptonight.hpp" +#include "xmrstak/misc/console.hpp" #include "xmrstak/jconf.hpp" #include <stdio.h> #include <stdlib.h> @@ -73,6 +74,8 @@ void do_skein_hash(const void* input, size_t len, char* output) { void (* const extra_hashes[4])(const void *, size_t, char *) = {do_blake_hash, do_groestl_hash, do_jh_hash, do_skein_hash}; #ifdef _WIN32 +#include "xmrstak/misc/uac.hpp" + BOOL bRebootDesirable = FALSE; //If VirtualAlloc fails, suggest a reboot BOOL AddPrivilege(TCHAR* pszPrivilege) @@ -176,13 +179,16 @@ size_t cryptonight_init(size_t use_fast_mem, size_t use_mlock, alloc_msg* msg) if(AddPrivilege(TEXT("SeLockMemoryPrivilege")) == 0) { + printer::inst()->print_msg(L0, "Elevating because we need to set up fast memory privileges."); + RequestElevation(); + if(AddLargePageRights()) { msg->warning = "Added SeLockMemoryPrivilege to the current account. You need to reboot for it to work"; bRebootDesirable = TRUE; } else - msg->warning = "Obtaning SeLockMemoryPrivilege failed."; + msg->warning = "Obtaining SeLockMemoryPrivilege failed."; return 0; } @@ -247,6 +253,9 @@ cryptonight_ctx* cryptonight_alloc_ctx(size_t use_fast_mem, size_t use_mlock, al #elif defined(__FreeBSD__) ptr->long_state = (uint8_t*)mmap(0, hashMemSize, PROT_READ | PROT_WRITE, MAP_PRIVATE | MAP_ANONYMOUS | MAP_ALIGNED_SUPER | MAP_PREFAULT_READ, -1, 0); +#elif defined(__OpenBSD__) + ptr->long_state = (uint8_t*)mmap(0, hashMemSize, PROT_READ | PROT_WRITE, + MAP_PRIVATE | MAP_ANON, -1, 0); #else ptr->long_state = (uint8_t*)mmap(0, hashMemSize, PROT_READ | PROT_WRITE, MAP_PRIVATE | MAP_ANONYMOUS | MAP_HUGETLB | MAP_POPULATE, 0, 0); diff --git a/xmrstak/backend/cpu/minethd.cpp b/xmrstak/backend/cpu/minethd.cpp index 143b66f..1d9165e 100644 --- a/xmrstak/backend/cpu/minethd.cpp +++ b/xmrstak/backend/cpu/minethd.cpp @@ -73,7 +73,16 @@ namespace cpu bool minethd::thd_setaffinity(std::thread::native_handle_type h, uint64_t cpu_id) { #if defined(_WIN32) - return SetThreadAffinityMask(h, 1ULL << cpu_id) != 0; + // we can only pin up to 64 threads + if(cpu_id < 64) + { + return SetThreadAffinityMask(h, 1ULL << cpu_id) != 0; + } + else + { + printer::inst()->print_msg(L0, "WARNING: Windows supports only affinity up to 63."); + return false; + } #elif defined(__APPLE__) thread_port_t mach_thread; thread_affinity_policy_data_t policy = { static_cast<integer_t>(cpu_id) }; @@ -84,6 +93,8 @@ bool minethd::thd_setaffinity(std::thread::native_handle_type h, uint64_t cpu_id CPU_ZERO(&mn); CPU_SET(cpu_id, &mn); return pthread_setaffinity_np(h, sizeof(cpuset_t), &mn) == 0; +#elif defined(__OpenBSD__) + printer::inst()->print_msg(L0,"WARNING: thread pinning is not supported under OPENBSD."); #else cpu_set_t mn; CPU_ZERO(&mn); @@ -307,7 +318,7 @@ std::vector<iBackend*> minethd::thread_starter(uint32_t threadOffset, miner_work if(cfg.iCpuAff >= 0) { #if defined(__APPLE__) - printer::inst()->print_msg(L1, "WARNING on MacOS thread affinity is only advisory."); + printer::inst()->print_msg(L1, "WARNING on macOS thread affinity is only advisory."); #endif printer::inst()->print_msg(L1, "Starting %dx thread, affinity: %d.", cfg.iMultiway, (int)cfg.iCpuAff); diff --git a/xmrstak/backend/nvidia/config.tpl b/xmrstak/backend/nvidia/config.tpl index 5479172..f489956 100644 --- a/xmrstak/backend/nvidia/config.tpl +++ b/xmrstak/backend/nvidia/config.tpl @@ -26,6 +26,9 @@ R"===( * "affine_to_cpu" : false, "sync_mode" : 3, * }, * ], + * If you do not wish to mine with your nVidia GPU(s) then use: + * "gpu_threads_conf" : + * null, */ "gpu_threads_conf" : diff --git a/xmrstak/backend/nvidia/minethd.cpp b/xmrstak/backend/nvidia/minethd.cpp index 5564596..9fd08fb 100644 --- a/xmrstak/backend/nvidia/minethd.cpp +++ b/xmrstak/backend/nvidia/minethd.cpp @@ -166,7 +166,7 @@ std::vector<iBackend*>* minethd::thread_starter(uint32_t threadOffset, miner_wor if(cfg.cpu_aff >= 0) { #if defined(__APPLE__) - printer::inst()->print_msg(L1, "WARNING on MacOS thread affinity is only advisory."); + printer::inst()->print_msg(L1, "WARNING on macOS thread affinity is only advisory."); #endif printer::inst()->print_msg(L1, "Starting NVIDIA GPU thread %d, affinity: %d.", i, (int)cfg.cpu_aff); @@ -287,7 +287,7 @@ void minethd::work_main() if ( (*((uint64_t*)(bResult + 24))) < oWork.iTarget) executor::inst()->push_event(ex_event(job_result(oWork.sJobID, foundNonce[i], bResult, iThreadNo), oWork.iPoolId)); else - executor::inst()->push_event(ex_event("NVIDIA Invalid Result", oWork.iPoolId)); + executor::inst()->push_event(ex_event("NVIDIA Invalid Result", ctx.device_id, oWork.iPoolId)); } iCount += h_per_round; diff --git a/xmrstak/cli/cli-miner.cpp b/xmrstak/cli/cli-miner.cpp index b84b783..7dc0f2f 100644 --- a/xmrstak/cli/cli-miner.cpp +++ b/xmrstak/cli/cli-miner.cpp @@ -86,6 +86,9 @@ void help() cout<<" --noNVIDIA disable the NVIDIA miner backend"<<endl; cout<<" --nvidia FILE NVIDIA backend miner config file"<<endl; #endif +#ifndef CONF_NO_HTTPD + cout<<" -i --httpd HTTP_PORT HTTP interface port"<<endl; +#endif cout<<" "<<endl; cout<<"The following options can be used for automatic start without a guided config,"<<endl; cout<<"If config exists then this pool will be top priority."<<endl; @@ -209,6 +212,28 @@ void do_guided_config() currency = tmp; } + auto& http_port = params::inst().httpd_port; + if(http_port == params::httpd_port_unset) + { +#if defined(CONF_NO_HTTPD) + http_port = params::httpd_port_disabled; +#else + std::cout<<"- Do you want to use the HTTP interface?" <<std::endl; + std::cout<<"Unlike the screen display, browser interface is not affected by the GPU lag." <<std::endl; + std::cout<<"If you don't want to use it, please enter 0, otherwise enter port number that the miner should listen on" <<std::endl; + + int32_t port; + while(!(std::cin >> port) || port < 0 || port > 65535) + { + std::cin.clear(); + std::cin.ignore(INT_MAX, '\n'); + std::cout << "Invalid port number. Please enter a number between 0 and 65535." << std::endl; + } + + http_port = port; +#endif + } + auto& pool = params::inst().poolURL; bool userSetPool = true; if(pool.empty()) @@ -306,6 +331,7 @@ void do_guided_config() configTpl.replace("POOLCONF", pool_table); configTpl.replace("CURRENCY", currency); + configTpl.replace("HTTP_PORT", std::to_string(http_port)); configTpl.write(params::inst().configFile); std::cout<<"Configuration stored in file '"<<params::inst().configFile<<"'"<<std::endl; } @@ -342,7 +368,14 @@ int main(int argc, char *argv[]) params::inst().executablePrefix += seperator; } - bool uacDialog = true; + params::inst().minerArg0 = argv[0]; + params::inst().minerArgs.reserve(argc * 16); + for(int i = 1; i < argc; i++) + { + params::inst().minerArgs += " "; + params::inst().minerArgs += argv[i]; + } + bool pool_url_set = false; for(size_t i = 1; i < argc-1; i++) { @@ -504,9 +537,31 @@ int main(int argc, char *argv[]) } params::inst().configFile = argv[i]; } + else if(opName.compare("-i") == 0 || opName.compare("--httpd") == 0) + { + ++i; + if( i >=argc ) + { + printer::inst()->print_msg(L0, "No argument for parameter '-i/--httpd' given"); + win_exit(); + return 1; + } + + char* endp = nullptr; + long int ret = strtol(argv[i], &endp, 10); + + if(endp == nullptr || ret < 0 || ret > 65535) + { + printer::inst()->print_msg(L0, "Argument for parameter '-i/--httpd' must be a number between 0 and 65535"); + win_exit(); + return 1; + } + + params::inst().httpd_port = ret; + } else if(opName.compare("--noUAC") == 0) { - uacDialog = false; + params::inst().allowUAC = false; } else { @@ -516,20 +571,6 @@ int main(int argc, char *argv[]) } } -#ifdef _WIN32 - if(uacDialog && !IsElevated()) - { - std::string minerArgs; - for(int i = 1; i < argc; i++) - { - minerArgs += " "; - minerArgs += argv[i]; - } - - SelfElevate(argv[0], minerArgs); - } -#endif - // check if we need a guided start if(!configEditor::file_exist(params::inst().configFile)) do_guided_config(); @@ -540,22 +581,35 @@ int main(int argc, char *argv[]) return 1; } +#ifdef _WIN32 + /* For Windows 7 and 8 request elevation at all times unless we are using slow memory */ + if(jconf::inst()->GetSlowMemSetting() != jconf::slow_mem_cfg::always_use && !IsWindows10OrNewer()) + { + printer::inst()->print_msg(L0, "Elevating due to Windows 7 or 8. You need Windows 10 to use fast memory without UAC elevation."); + RequestElevation(); + } +#endif + if (!BackendConnector::self_test()) { win_exit(); return 1; } -#ifndef CONF_NO_HTTPD - if(jconf::inst()->GetHttpdPort() != 0) + if(jconf::inst()->GetHttpdPort() != uint16_t(params::httpd_port_disabled)) { +#ifdef CONF_NO_HTTPD + printer::inst()->print_msg(L0, "HTTPD port is enabled but this binary was compiled without HTTP support!"); + win_exit(); + return 1; +#else if (!httpd::inst()->start_daemon()) { win_exit(); return 1; } - } #endif + } printer::inst()->print_str("-------------------------------------------------------------------\n"); printer::inst()->print_str(get_version_str_short().c_str()); diff --git a/xmrstak/cli/xmr-stak.manifest b/xmrstak/cli/xmr-stak.manifest new file mode 100644 index 0000000..ed65c97 --- /dev/null +++ b/xmrstak/cli/xmr-stak.manifest @@ -0,0 +1,34 @@ +<?xml version="1.0" encoding="UTF-8" standalone="yes"?> +<assembly xmlns="urn:schemas-microsoft-com:asm.v1" manifestVersion="1.0" xmlns:asmv3="urn:schemas-microsoft-com:asm.v3"> +<assemblyIdentity + version="1.0.0.0" + processorArchitecture="amd64" + name="xmr-stak" + type="win32" +/> + <description>XMR-Stak Monero Miner</description> + <trustInfo xmlns="urn:schemas-microsoft-com:asm.v3"> + <security> + <requestedPrivileges> + <requestedExecutionLevel + level="asInvoker" + uiAccess="false" + /> + </requestedPrivileges> + </security> + </trustInfo> + <compatibility xmlns="urn:schemas-microsoft-com:compatibility.v1"> + <application> + <!-- Windows 10 --> + <supportedOS Id="{8e0f7a12-bfb3-4fe8-b9a5-48fd50a15a9a}"/> + <!-- Windows 8.1 --> + <supportedOS Id="{1f676c76-80e1-4239-95bb-83d0f6d0da78}"/> + <!-- Windows Vista --> + <supportedOS Id="{e2011457-1546-43c5-a5fe-008deee3d3f0}"/> + <!-- Windows 7 --> + <supportedOS Id="{35138b9a-5d96-4fbd-8e2d-a2440225f93a}"/> + <!-- Windows 8 --> + <supportedOS Id="{4a2f28e3-53b9-4441-ba9c-d69d4a4a6e38}"/> + </application> + </compatibility> +</assembly> diff --git a/xmrstak/config.tpl b/xmrstak/config.tpl index ae97190..2c7bd41 100644 --- a/xmrstak/config.tpl +++ b/xmrstak/config.tpl @@ -159,7 +159,7 @@ POOLCONF], * * httpd_port - Port we should listen on. Default, 0, will switch off the server. */ -"httpd_port" : 0, +"httpd_port" : HTTP_PORT, /* * HTTP Authentication diff --git a/xmrstak/jconf.cpp b/xmrstak/jconf.cpp index f279f52..a1db451 100644 --- a/xmrstak/jconf.cpp +++ b/xmrstak/jconf.cpp @@ -242,7 +242,10 @@ uint64_t jconf::GetAutohashTime() uint16_t jconf::GetHttpdPort() { - return prv->configValues[iHttpdPort]->GetUint(); + if(xmrstak::params::inst().httpd_port == xmrstak::params::httpd_port_unset) + return prv->configValues[iHttpdPort]->GetUint(); + else + return uint16_t(xmrstak::params::inst().httpd_port); } const char* jconf::GetHttpUsername() diff --git a/xmrstak/misc/executor.cpp b/xmrstak/misc/executor.cpp index a3088a5..055739b 100644 --- a/xmrstak/misc/executor.cpp +++ b/xmrstak/misc/executor.cpp @@ -594,7 +594,7 @@ void executor::ex_main() break; case EV_GPU_RES_ERROR: - log_result_error(std::string(ev.oGpuError.error_str)); + log_result_error(std::string(ev.oGpuError.error_str + std::string(" GPU ID ") + std::to_string(ev.oGpuError.idx))); break; case EV_PERF_TICK: @@ -761,6 +761,7 @@ void executor::hashrate_report(std::string& out) else out.append(1, '\n'); + double fTotalCur[3] = { 0.0, 0.0, 0.0}; for (i = 0; i < nthd; i++) { double fHps[3]; @@ -775,10 +776,14 @@ void executor::hashrate_report(std::string& out) out.append(hps_format(fHps[0], num, sizeof(num))).append(" |"); out.append(hps_format(fHps[1], num, sizeof(num))).append(" |"); out.append(hps_format(fHps[2], num, sizeof(num))).append(1, ' '); - - fTotal[0] += fHps[0]; - fTotal[1] += fHps[1]; - fTotal[2] += fHps[2]; + + fTotal[0] += (std::isnormal(fHps[0])) ? fHps[0] : 0.0; + fTotal[1] += (std::isnormal(fHps[1])) ? fHps[1] : 0.0; + fTotal[2] += (std::isnormal(fHps[2])) ? fHps[2] : 0.0; + + fTotalCur[0] += (std::isnormal(fHps[0])) ? fHps[0] : 0.0; + fTotalCur[1] += (std::isnormal(fHps[1])) ? fHps[1] : 0.0; + fTotalCur[2] += (std::isnormal(fHps[2])) ? fHps[2] : 0.0; if((i & 0x1) == 1) //Odd i's out.append("|\n"); @@ -786,21 +791,25 @@ void executor::hashrate_report(std::string& out) if((i & 0x1) == 1) //We had odd number of threads out.append("|\n"); - - if(nthd != 1) - out.append("-----------------------------------------------------\n"); - else - out.append("---------------------------\n"); + + out.append("Totals (").append(name).append("): "); + out.append(hps_format(fTotalCur[0], num, sizeof(num))); + out.append(hps_format(fTotalCur[1], num, sizeof(num))); + out.append(hps_format(fTotalCur[2], num, sizeof(num))); + out.append(" H/s\n"); + + out.append("-----------------------------------------------------------------\n"); } } - out.append("Totals: "); + out.append("Totals (ALL): "); out.append(hps_format(fTotal[0], num, sizeof(num))); out.append(hps_format(fTotal[1], num, sizeof(num))); out.append(hps_format(fTotal[2], num, sizeof(num))); out.append(" H/s\nHighest: "); out.append(hps_format(fHighestHps, num, sizeof(num))); out.append(" H/s\n"); + out.append("-----------------------------------------------------------------\n"); } char* time_format(char* buf, size_t len, std::chrono::system_clock::time_point time) diff --git a/xmrstak/misc/executor.hpp b/xmrstak/misc/executor.hpp index c2caa39..fbaa265 100644 --- a/xmrstak/misc/executor.hpp +++ b/xmrstak/misc/executor.hpp @@ -177,7 +177,6 @@ private: iPoolCallTimes.clear(); tPoolConnTime = std::chrono::system_clock::now(); iPoolHashes = 0; - iPoolDiff = 0; } double fHighestHps = 0.0; diff --git a/xmrstak/misc/uac.cpp b/xmrstak/misc/uac.cpp new file mode 100644 index 0000000..ad9d394 --- /dev/null +++ b/xmrstak/misc/uac.cpp @@ -0,0 +1,77 @@ +#ifdef _WIN32 +#include "xmrstak/misc/console.hpp" +#include "xmrstak/params.hpp" + +#include <string> +#include <windows.h> + +BOOL IsElevated() +{ + BOOL fRet = FALSE; + HANDLE hToken = NULL; + if (OpenProcessToken(GetCurrentProcess(), TOKEN_QUERY, &hToken)) + { + TOKEN_ELEVATION Elevation; + DWORD cbSize = sizeof(TOKEN_ELEVATION); + if (GetTokenInformation(hToken, TokenElevation, &Elevation, sizeof(Elevation), &cbSize)) + fRet = Elevation.TokenIsElevated; + } + if (hToken) + CloseHandle(hToken); + return fRet; +} + +BOOL SelfElevate(const std::string& my_path, const std::string& params) +{ + if (IsElevated()) + return FALSE; + + SHELLEXECUTEINFO shExecInfo = { 0 }; + shExecInfo.cbSize = sizeof(SHELLEXECUTEINFO); + shExecInfo.fMask = SEE_MASK_NOCLOSEPROCESS; + shExecInfo.hwnd = NULL; + shExecInfo.lpVerb = "runas"; + shExecInfo.lpFile = my_path.c_str(); + shExecInfo.lpParameters = params.c_str(); + shExecInfo.lpDirectory = NULL; + shExecInfo.nShow = SW_SHOW; + shExecInfo.hInstApp = NULL; + + if (!ShellExecuteEx(&shExecInfo)) + return FALSE; + + // Loiter in the background to make scripting easier + printer::inst()->print_msg(L0, "This window has been opened because xmr-stak needed to run as administrator. It can be safely closed now."); + WaitForSingleObject(shExecInfo.hProcess, INFINITE); + std::exit(0); + + return TRUE; +} + +VOID RequestElevation() +{ + if(IsElevated()) + return; + + if(!xmrstak::params::inst().allowUAC) + { + printer::inst()->print_msg(L0, "The miner needs to run as administrator, but you passed --noUAC option. Please remove it or set use_slow_memory to always."); + win_exit(); + return; + } + + SelfElevate(xmrstak::params::inst().minerArg0, xmrstak::params::inst().minerArgs); +} + +BOOL IsWindows10OrNewer() +{ + OSVERSIONINFOEX osvi = { 0 }; + osvi.dwOSVersionInfoSize = sizeof(OSVERSIONINFOEX); + osvi.dwMajorVersion = 10; + osvi.dwMinorVersion = 0; + DWORDLONG dwlConditionMask = 0; + VER_SET_CONDITION(dwlConditionMask, VER_MAJORVERSION, VER_GREATER_EQUAL); + VER_SET_CONDITION(dwlConditionMask, VER_MINORVERSION, VER_GREATER_EQUAL); + return ::VerifyVersionInfo(&osvi, VER_MAJORVERSION | VER_MINORVERSION, dwlConditionMask); +} +#endif diff --git a/xmrstak/misc/uac.hpp b/xmrstak/misc/uac.hpp index 55c5f1a..33c79ae 100644 --- a/xmrstak/misc/uac.hpp +++ b/xmrstak/misc/uac.hpp @@ -1,51 +1,10 @@ #pragma once #ifdef _WIN32 -#include "xmrstak/misc/console.hpp" - #include <string> -#include <windows.h> - -BOOL IsElevated() -{ - BOOL fRet = FALSE; - HANDLE hToken = NULL; - if (OpenProcessToken(GetCurrentProcess(), TOKEN_QUERY, &hToken)) - { - TOKEN_ELEVATION Elevation; - DWORD cbSize = sizeof(TOKEN_ELEVATION); - if (GetTokenInformation(hToken, TokenElevation, &Elevation, sizeof(Elevation), &cbSize)) - fRet = Elevation.TokenIsElevated; - } - if (hToken) - CloseHandle(hToken); - return fRet; -} - -BOOL SelfElevate(const char* my_path, const std::string& params) -{ - if (IsElevated()) - return FALSE; - - SHELLEXECUTEINFO shExecInfo = { 0 }; - shExecInfo.cbSize = sizeof(SHELLEXECUTEINFO); - shExecInfo.fMask = SEE_MASK_NOCLOSEPROCESS; - shExecInfo.hwnd = NULL; - shExecInfo.lpVerb = "runas"; - shExecInfo.lpFile = my_path; - shExecInfo.lpParameters = params.c_str(); - shExecInfo.lpDirectory = NULL; - shExecInfo.nShow = SW_SHOW; - shExecInfo.hInstApp = NULL; - - if (!ShellExecuteEx(&shExecInfo)) - return FALSE; - - // Loiter in the background to make scripting easier - printer::inst()->print_msg(L0, "This window has been opened because xmr-stak needed to run as administrator. It can be safely closed now."); - WaitForSingleObject(shExecInfo.hProcess, INFINITE); - std::exit(0); - return TRUE; -} +BOOL IsElevated(); +BOOL SelfElevate(const std::string& my_path, const std::string& params); +VOID RequestElevation(); +BOOL IsWindows10OrNewer(); #endif diff --git a/xmrstak/net/msgstruct.hpp b/xmrstak/net/msgstruct.hpp index a5affc8..8c4bdbe 100644 --- a/xmrstak/net/msgstruct.hpp +++ b/xmrstak/net/msgstruct.hpp @@ -66,8 +66,9 @@ struct sock_err // Unlike socket errors, GPU errors are read-only strings struct gpu_res_err { + size_t idx; // GPU index const char* error_str; - gpu_res_err(const char* error_str) : error_str(error_str) {} + gpu_res_err(const char* error_str, size_t idx) : error_str(error_str), idx(idx) {} }; enum ex_event_name { EV_INVALID_VAL, EV_SOCK_READY, EV_SOCK_ERROR, EV_GPU_RES_ERROR, @@ -99,7 +100,7 @@ struct ex_event }; ex_event() { iName = EV_INVALID_VAL; iPoolId = 0;} - ex_event(const char* gpu_err, size_t id) : iName(EV_GPU_RES_ERROR), iPoolId(id), oGpuError(gpu_err) {} + ex_event(const char* gpu_err, size_t gpu_idx, size_t id) : iName(EV_GPU_RES_ERROR), iPoolId(id), oGpuError(gpu_err, gpu_idx) {} ex_event(std::string&& err, bool silent, size_t id) : iName(EV_SOCK_ERROR), iPoolId(id), oSocketError(std::move(err), silent) { } ex_event(job_result dat, size_t id) : iName(EV_MINER_HAVE_RESULT), iPoolId(id), oJobResult(dat) {} ex_event(pool_job dat, size_t id) : iName(EV_POOL_HAVE_JOB), iPoolId(id), oPoolJob(dat) {} diff --git a/xmrstak/net/socket.cpp b/xmrstak/net/socket.cpp index e19d1d4..89e9902 100644 --- a/xmrstak/net/socket.cpp +++ b/xmrstak/net/socket.cpp @@ -190,7 +190,7 @@ void tls_socket::print_error() if(jconf::inst()->TlsSecureAlgos()) pCallback->set_socket_error("Unknown TLS error. Secure TLS maybe unsupported, try setting tls_secure_algo to false."); else - pCallback->set_socket_error("Unknown TLS error."); + pCallback->set_socket_error("Unknown TLS error. You might be trying to connect to a non-TLS port."); } else pCallback->set_socket_error(buf, len); diff --git a/xmrstak/params.hpp b/xmrstak/params.hpp index bc32612..d0d6685 100644 --- a/xmrstak/params.hpp +++ b/xmrstak/params.hpp @@ -31,6 +31,10 @@ struct params std::string poolUsername; bool nicehashMode = false; + static constexpr int32_t httpd_port_unset = -1; + static constexpr int32_t httpd_port_disabled = 0; + int32_t httpd_port = httpd_port_unset; + std::string currency; std::string configFile; @@ -38,6 +42,10 @@ struct params std::string configFileNVIDIA; std::string configFileCPU; + bool allowUAC = true; + std::string minerArg0; + std::string minerArgs; + params() : binaryName("xmr-stak"), executablePrefix(""), |
