From b6362bced4ff3da19a688c1cdcf71ae506a461fd Mon Sep 17 00:00:00 2001 From: kenshirothefist Date: Wed, 15 Jun 2016 23:25:35 +0200 Subject: [PATCH] Add support for EthereumStratum/1.0.0 mode (see -ES option). This adds compatibility with NiceHash.com. --- cmake/scripts/macdeployfix.sh | 0 ethminer/MinerAux.h | 25 +++- extdep/getstuff.bat | 1 + libethash-cl/ethash_cl_miner.cpp | 6 +- libethash-cl/ethash_cl_miner.h | 2 +- libethash-cuda/ethash_cuda_miner.cpp | 73 ++++++++--- libethash-cuda/ethash_cuda_miner.h | 5 +- libethcore/EthashAux.h | 7 +- libethcore/EthashCUDAMiner.cpp | 42 +++++- libethcore/EthashCUDAMiner.h | 3 +- libethcore/EthashGPUMiner.cpp | 9 +- libethcore/EthashGPUMiner.h | 3 +- libethcore/Farm.h | 2 +- libethcore/Miner.cpp | 8 +- libethcore/Miner.h | 5 +- libstratum/EthStratumClient.cpp | 186 ++++++++++++++++++++++----- libstratum/EthStratumClient.h | 10 +- libstratum/EthStratumClientV2.cpp | 179 +++++++++++++++++++++----- libstratum/EthStratumClientV2.h | 10 +- 19 files changed, 469 insertions(+), 107 deletions(-) mode change 100755 => 100644 cmake/scripts/macdeployfix.sh diff --git a/cmake/scripts/macdeployfix.sh b/cmake/scripts/macdeployfix.sh old mode 100755 new mode 100644 diff --git a/ethminer/MinerAux.h b/ethminer/MinerAux.h index 52b70c03d..9c133dc84 100644 --- a/ethminer/MinerAux.h +++ b/ethminer/MinerAux.h @@ -201,6 +201,10 @@ public: BOOST_THROW_EXCEPTION(BadArgument()); } } + else if (arg == "-ES" || arg == "--ethereum-stratum") + { + m_ethereumStratum = true; + } else if ((arg == "-FO" || arg == "--failover-userpass") && i + 1 < argc) { string userpass = string(argv[++i]); @@ -330,6 +334,11 @@ public: string mode = argv[++i]; if (mode == "parallel") m_dagLoadMode = DAG_LOAD_MODE_PARALLEL; else if (mode == "sequential") m_dagLoadMode = DAG_LOAD_MODE_SEQUENTIAL; + else if (mode == "single") + { + m_dagLoadMode = DAG_LOAD_MODE_SINGLE; + m_dagCreateDevice = stol(argv[++i]); + } else { cerr << "Bad " << arg << " option: " << argv[i] << endl; @@ -531,7 +540,8 @@ public: m_clAllowCPU, m_extraGPUMemory, 0, - m_dagLoadMode + m_dagLoadMode, + m_dagCreateDevice )) exit(1); EthashGPUMiner::setNumInstances(m_miningThreads); @@ -557,7 +567,8 @@ public: m_extraGPUMemory, m_cudaSchedule, 0, - m_dagLoadMode + m_dagLoadMode, + m_dagCreateDevice )) exit(1); #else @@ -592,7 +603,8 @@ public: << " -O, --userpass Stratum login credentials" << endl << " -FO, --failover-userpass Failover stratum login credentials (optional, will use normal credentials when omitted)" << endl << " --work-timeout reconnect/failover after n seconds of working on the same (stratum) job. Defaults to 180. Don't set lower than max. avg. block time" << endl - << " -SV, --stratum-version Stratum client version. Defaults to 1 (async client). Use 2 to test new synchronous client." + << " -SV, --stratum-version Stratum client version. Defaults to 1 (async client). Use 2 to test new synchronous client." << endl + << " -ES, --ethereum-stratum Use EthereumStratum/1.0.0 mode." << endl #endif #if ETH_JSONRPC || ETH_STRATUM || !ETH_TRUE << " --farm-recheck Leave n ms between checks for changed work (default: 500). When using stratum, use a high value (i.e. 2000) to get more stable hashrate output" << endl @@ -618,6 +630,7 @@ public: << " -L, --dag-load-mode DAG generation mode." << endl << " parallel - load DAG on all GPUs at the same time (default)" << endl << " sequential - load DAG on GPUs one after another. Use this when the miner crashes during DAG generation" << endl + << " single - generate DAG on device n, then copy to other devices" << endl #if ETH_ETHASHCL || !ETH_TRUE << " --cl-extragpu-mem Set the memory (in MB) you believe your GPU requires for stuff other than mining. default: 0" << endl << " --cl-local-work Set the OpenCL local work size. Default is " << toString(ethash_cl_miner::c_defaultLocalWorkSize) << endl @@ -992,7 +1005,7 @@ private: // this is very ugly, but if Stratum Client V2 tunrs out to be a success, V1 will be completely removed anyway if (m_stratumClientVersion == 1) { - EthStratumClient client(&f, m_minerType, m_farmURL, m_port, m_user, m_pass, m_maxFarmRetries, m_worktimeout); + EthStratumClient client(&f, m_minerType, m_farmURL, m_port, m_user, m_pass, m_maxFarmRetries, m_worktimeout, m_ethereumStratum); if (m_farmFailOverURL != "") { if (m_fuser != "") @@ -1032,7 +1045,7 @@ private: } } else if (m_stratumClientVersion == 2) { - EthStratumClientV2 client(&f, m_minerType, m_farmURL, m_port, m_user, m_pass, m_maxFarmRetries, m_worktimeout); + EthStratumClientV2 client(&f, m_minerType, m_farmURL, m_port, m_user, m_pass, m_maxFarmRetries, m_worktimeout, m_ethereumStratum); if (m_farmFailOverURL != "") { if (m_fuser != "") @@ -1100,6 +1113,7 @@ private: // default value was 350MB of GPU memory for other stuff (windows system rendering, e.t.c.) unsigned m_extraGPUMemory = 0;// 350000000; don't assume miners run desktops... unsigned m_dagLoadMode = 0; // parallel + unsigned m_dagCreateDevice = 0; /// Benchmarking params bool m_phoneHome = false; unsigned m_benchmarkWarmup = 15; @@ -1121,6 +1135,7 @@ private: #if ETH_STRATUM || !ETH_TRUE int m_stratumClientVersion = 1; + bool m_ethereumStratum = false; string m_user; string m_pass; string m_port; diff --git a/extdep/getstuff.bat b/extdep/getstuff.bat index b88260ee4..2bb69792a 100644 --- a/extdep/getstuff.bat +++ b/extdep/getstuff.bat @@ -36,3 +36,4 @@ cd .. goto :EOF + diff --git a/libethash-cl/ethash_cl_miner.cpp b/libethash-cl/ethash_cl_miner.cpp index 1c8d9367a..4cd3a2686 100644 --- a/libethash-cl/ethash_cl_miner.cpp +++ b/libethash-cl/ethash_cl_miner.cpp @@ -504,7 +504,7 @@ typedef struct unsigned buf; } pending_batch; -void ethash_cl_miner::search(uint8_t const* header, uint64_t target, search_hook& hook) +void ethash_cl_miner::search(uint8_t const* header, uint64_t target, search_hook& hook, bool _ethStratum, uint64_t _startN) { try { @@ -531,7 +531,9 @@ void ethash_cl_miner::search(uint8_t const* header, uint64_t target, search_hook unsigned buf = 0; random_device engine; - uint64_t start_nonce = uniform_int_distribution()(engine); + uint64_t start_nonce; + if (_ethStratum) start_nonce = _startN; + else start_nonce = uniform_int_distribution()(engine); for (;; start_nonce += m_globalWorkSize) { // supply output buffer to kernel diff --git a/libethash-cl/ethash_cl_miner.h b/libethash-cl/ethash_cl_miner.h index 794dc2105..a664a15d1 100644 --- a/libethash-cl/ethash_cl_miner.h +++ b/libethash-cl/ethash_cl_miner.h @@ -59,7 +59,7 @@ public: unsigned _deviceId ); void finish(); - void search(uint8_t const* _header, uint64_t _target, search_hook& _hook); + void search(uint8_t const* _header, uint64_t _target, search_hook& _hook, bool _ethStratum, uint64_t _startN); /* -- default values -- */ /// Default value of the local work size. Also known as workgroup size. diff --git a/libethash-cuda/ethash_cuda_miner.cpp b/libethash-cuda/ethash_cuda_miner.cpp index 4d231dcda..033c38cf7 100644 --- a/libethash-cuda/ethash_cuda_miner.cpp +++ b/libethash-cuda/ethash_cuda_miner.cpp @@ -199,7 +199,7 @@ void ethash_cuda_miner::finish() CUDA_SAFE_CALL(cudaDeviceReset()); } -bool ethash_cuda_miner::init(ethash_light_t _light, uint8_t const* _lightData, uint64_t _lightSize, unsigned _deviceId) +bool ethash_cuda_miner::init(ethash_light_t _light, uint8_t const* _lightData, uint64_t _lightSize, unsigned _deviceId, bool _cpyToHost, volatile void** hostDAG) { try { @@ -229,10 +229,14 @@ bool ethash_cuda_miner::init(ethash_light_t _light, uint8_t const* _lightData, u uint32_t lightSize64 = (unsigned)(_lightSize / sizeof(node)); // create buffer for cache - hash64_t * light; - CUDA_SAFE_CALL(cudaMalloc(reinterpret_cast(&light), _lightSize)); - // copy dag cache to CPU. - CUDA_SAFE_CALL(cudaMemcpy(reinterpret_cast(light), _lightData, _lightSize, cudaMemcpyHostToDevice)); + hash64_t * light = NULL; + + if (!*hostDAG) + { + CUDA_SAFE_CALL(cudaMalloc(reinterpret_cast(&light), _lightSize)); + // copy dag cache to CPU. + CUDA_SAFE_CALL(cudaMemcpy(reinterpret_cast(light), _lightData, _lightSize, cudaMemcpyHostToDevice)); + } // create buffer for dag hash128_t * dag; @@ -252,9 +256,28 @@ bool ethash_cuda_miner::init(ethash_light_t _light, uint8_t const* _lightData, u m_sharedBytes = device_props.major * 100 < SHUFFLE_MIN_VER ? (64 * s_blockSize) / 8 : 0 ; + if (!*hostDAG) + { + cout << "Generating DAG for GPU #" << device_num << endl; + ethash_generate_dag(dagSize, s_gridSize, s_blockSize, m_streams[0], device_num); + + if (_cpyToHost) + { + uint8_t* memoryDAG = new uint8_t[dagSize]; + if (!memoryDAG) throw std::runtime_error("Failed to init host memory for DAG, not enough memory?"); + + cout << "Copying DAG from GPU #" << device_num << " to host" << endl; + CUDA_SAFE_CALL(cudaMemcpy(reinterpret_cast(memoryDAG), dag, dagSize, cudaMemcpyDeviceToHost)); - cout << "Generating DAG for GPU #" << device_num << endl; - ethash_generate_dag(dagSize, s_gridSize, s_blockSize, m_streams[0], device_num); + *hostDAG = (void*)memoryDAG; + } + } + else + { + cout << "Copying DAG from host to GPU #" << device_num << endl; + const void* hdag = (const void*)(*hostDAG); + CUDA_SAFE_CALL(cudaMemcpy(reinterpret_cast(dag), hdag, dagSize, cudaMemcpyHostToDevice)); + } return true; } @@ -264,7 +287,7 @@ bool ethash_cuda_miner::init(ethash_light_t _light, uint8_t const* _lightData, u } } -void ethash_cuda_miner::search(uint8_t const* header, uint64_t target, search_hook& hook) +void ethash_cuda_miner::search(uint8_t const* header, uint64_t target, search_hook& hook, bool _ethStratum, uint64_t _startN) { bool initialize = false; bool exit = false; @@ -280,14 +303,34 @@ void ethash_cuda_miner::search(uint8_t const* header, uint64_t target, search_ho set_target(m_current_target); initialize = true; } - if (initialize) + if (_ethStratum) { - random_device engine; - m_current_nonce = uniform_int_distribution()(engine); - m_current_index = 0; - CUDA_SAFE_CALL(cudaDeviceSynchronize()); - for (unsigned int i = 0; i < s_numStreams; i++) - m_search_buf[i][0] = 0; + if (initialize) + { + m_starting_nonce = 0; + m_current_index = 0; + CUDA_SAFE_CALL(cudaDeviceSynchronize()); + for (unsigned int i = 0; i < s_numStreams; i++) + m_search_buf[i][0] = 0; + } + if (m_starting_nonce != _startN) + { + // reset nonce counter + m_starting_nonce = _startN; + m_current_nonce = m_starting_nonce; + } + } + else + { + if (initialize) + { + random_device engine; + m_current_nonce = uniform_int_distribution()(engine); + m_current_index = 0; + CUDA_SAFE_CALL(cudaDeviceSynchronize()); + for (unsigned int i = 0; i < s_numStreams; i++) + m_search_buf[i][0] = 0; + } } uint64_t batch_size = s_gridSize * s_blockSize; for (; !exit; m_current_index++, m_current_nonce += batch_size) diff --git a/libethash-cuda/ethash_cuda_miner.h b/libethash-cuda/ethash_cuda_miner.h index a05ce5d55..99204d71a 100644 --- a/libethash-cuda/ethash_cuda_miner.h +++ b/libethash-cuda/ethash_cuda_miner.h @@ -35,10 +35,10 @@ public: uint64_t _currentBlock ); - bool init(ethash_light_t _light, uint8_t const* _lightData, uint64_t _lightSize, unsigned _deviceId); + bool init(ethash_light_t _light, uint8_t const* _lightData, uint64_t _lightSize, unsigned _deviceId, bool _cpyToHost, volatile void** hostDAG); void finish(); - void search(uint8_t const* header, uint64_t target, search_hook& hook); + void search(uint8_t const* header, uint64_t target, search_hook& hook, bool _ethStratum, uint64_t _startN); /* -- default values -- */ /// Default value of the block size. Also known as workgroup size. @@ -52,6 +52,7 @@ private: hash32_t m_current_header; uint64_t m_current_target; uint64_t m_current_nonce; + uint64_t m_starting_nonce; uint64_t m_current_index; uint32_t m_sharedBytes; diff --git a/libethcore/EthashAux.h b/libethcore/EthashAux.h index 02d2e58df..b729fc638 100644 --- a/libethcore/EthashAux.h +++ b/libethcore/EthashAux.h @@ -53,17 +53,20 @@ struct EthashProofOfWork struct WorkPackage { WorkPackage() = default; - WorkPackage(Ethash::BlockHeader const& _bh): + WorkPackage(Ethash::BlockHeader const& _bh) : boundary(_bh.boundary()), headerHash(_bh.hashWithout()), seedHash(_bh.seedHash()) - {} + { } void reset() { headerHash = h256(); } operator bool() const { return headerHash != h256(); } h256 boundary; h256 headerHash; ///< When h256() means "pause until notified a new work package is available". h256 seedHash; + + uint64_t startNonce = 0; + int exSizeBits = -1; }; static const WorkPackage NullWorkPackage; diff --git a/libethcore/EthashCUDAMiner.cpp b/libethcore/EthashCUDAMiner.cpp index ea75ac2b8..d7a21dd24 100644 --- a/libethcore/EthashCUDAMiner.cpp +++ b/libethcore/EthashCUDAMiner.cpp @@ -144,12 +144,29 @@ void EthashCUDAMiner::workLoop() cnote << "set work; seed: " << "#" + w.seedHash.hex().substr(0, 8) + ", target: " << "#" + w.boundary.hex().substr(0, 16); if (!m_miner || m_minerSeed != w.seedHash) { + unsigned device = s_devices[index()] > -1 ? s_devices[index()] : index(); + if (s_dagLoadMode == DAG_LOAD_MODE_SEQUENTIAL) { while (s_dagLoadIndex < index()) { this_thread::sleep_for(chrono::seconds(1)); } } + else if (s_dagLoadMode == DAG_LOAD_MODE_SINGLE) + { + if (device != s_dagCreateDevice) + { + // wait until DAG is created on selected device + while (s_dagInHostMemory == NULL) { + this_thread::sleep_for(chrono::seconds(1)); + } + } + else + { + // reset load index + s_dagLoadIndex = 0; + } + } cnote << "Initialising miner..."; m_minerSeed = w.seedHash; @@ -157,19 +174,32 @@ void EthashCUDAMiner::workLoop() delete m_miner; m_miner = new ethash_cuda_miner; - unsigned device = s_devices[index()] > -1 ? s_devices[index()] : index(); - EthashAux::LightType light; light = EthashAux::light(w.seedHash); //bytesConstRef dagData = dag->data(); bytesConstRef lightData = light->data(); - m_miner->init(light->light, lightData.data(), lightData.size(), device); + m_miner->init(light->light, lightData.data(), lightData.size(), device, (s_dagLoadMode == DAG_LOAD_MODE_SINGLE), &s_dagInHostMemory); s_dagLoadIndex++; + + if (s_dagLoadMode == DAG_LOAD_MODE_SINGLE) + { + if (s_dagLoadIndex >= s_numInstances && s_dagInHostMemory) + { + // all devices have loaded DAG, we can free now + delete[] s_dagInHostMemory; + s_dagInHostMemory = NULL; + + cout << "Freeing DAG from host" << endl; + } + } } uint64_t upper64OfBoundary = (uint64_t)(u64)((u256)w.boundary >> 192); - m_miner->search(w.headerHash.data(), upper64OfBoundary, *m_hook); + uint64_t startN; + if (w.exSizeBits >= 0) + startN = w.startNonce | ((uint64_t)index() << (64 - 4 - w.exSizeBits)); // this can support up to 16 devices + m_miner->search(w.headerHash.data(), upper64OfBoundary, *m_hook, (w.exSizeBits >= 0), startN); } catch (std::runtime_error const& _e) { @@ -207,10 +237,12 @@ bool EthashCUDAMiner::configureGPU( unsigned _extraGPUMemory, unsigned _scheduleFlag, uint64_t _currentBlock, - unsigned _dagLoadMode + unsigned _dagLoadMode, + unsigned _dagCreateDevice ) { s_dagLoadMode = _dagLoadMode; + s_dagCreateDevice = _dagCreateDevice; _blockSize = ((_blockSize + 7) / 8) * 8; if (!ethash_cuda_miner::configureGPU( diff --git a/libethcore/EthashCUDAMiner.h b/libethcore/EthashCUDAMiner.h index 911b8f460..19031b8d9 100644 --- a/libethcore/EthashCUDAMiner.h +++ b/libethcore/EthashCUDAMiner.h @@ -54,7 +54,8 @@ namespace eth unsigned _extraGPUMemory, unsigned _scheduleFlag, uint64_t _currentBlock, - unsigned _dagLoadMode + unsigned _dagLoadMode, + unsigned _dagCreateDevice ); static void setNumInstances(unsigned _instances) { diff --git a/libethcore/EthashGPUMiner.cpp b/libethcore/EthashGPUMiner.cpp index 2346b718a..30274dfa3 100644 --- a/libethcore/EthashGPUMiner.cpp +++ b/libethcore/EthashGPUMiner.cpp @@ -183,7 +183,10 @@ void EthashGPUMiner::workLoop() } uint64_t upper64OfBoundary = (uint64_t)(u64)((u256)w.boundary >> 192); - m_miner->search(w.headerHash.data(), upper64OfBoundary, *m_hook); + uint64_t startN; + if (w.exSizeBits >= 0) + startN = w.startNonce | ((uint64_t)index() << (64 - 4 - w.exSizeBits)); // this can support up to 16 devices + m_miner->search(w.headerHash.data(), upper64OfBoundary, *m_hook, (w.exSizeBits >= 0), startN); } catch (cl::Error const& _e) { @@ -222,10 +225,12 @@ bool EthashGPUMiner::configureGPU( bool _allowCPU, unsigned _extraGPUMemory, uint64_t _currentBlock, - unsigned _dagLoadMode + unsigned _dagLoadMode, + unsigned _dagCreateDevice ) { s_dagLoadMode = _dagLoadMode; + s_dagCreateDevice = _dagCreateDevice; s_platformId = _platformId; s_deviceId = _deviceId; diff --git a/libethcore/EthashGPUMiner.h b/libethcore/EthashGPUMiner.h index 847799775..c1b998328 100644 --- a/libethcore/EthashGPUMiner.h +++ b/libethcore/EthashGPUMiner.h @@ -53,7 +53,8 @@ public: bool _allowCPU, unsigned _extraGPUMemory, uint64_t _currentBlock, - unsigned _dagLoadMode + unsigned _dagLoadMode, + unsigned _dagCreateDevice ); static void setNumInstances(unsigned _instances) { s_numInstances = std::min(_instances, getNumDevices()); } static void setDevices(unsigned * _devices, unsigned _selectedDeviceCount) diff --git a/libethcore/Farm.h b/libethcore/Farm.h index c236b1e46..084a63422 100644 --- a/libethcore/Farm.h +++ b/libethcore/Farm.h @@ -67,7 +67,7 @@ public: void setWork(WorkPackage const& _wp) { WriteGuard l(x_minerWork); - if (_wp.headerHash == m_work.headerHash) + if (_wp.headerHash == m_work.headerHash && _wp.startNonce == m_work.startNonce) return; m_work = _wp; for (auto const& m: m_miners) diff --git a/libethcore/Miner.cpp b/libethcore/Miner.cpp index 4e9e8c1d6..c182d8456 100644 --- a/libethcore/Miner.cpp +++ b/libethcore/Miner.cpp @@ -8,6 +8,12 @@ template <> unsigned dev::eth::GenericMiner::s_dagLoadMode = 0; template <> -unsigned dev::eth::GenericMiner::s_dagLoadIndex = 0; +volatile unsigned dev::eth::GenericMiner::s_dagLoadIndex = 0; + +template <> +unsigned dev::eth::GenericMiner::s_dagCreateDevice = 0; + +template <> +volatile void* dev::eth::GenericMiner::s_dagInHostMemory = NULL; diff --git a/libethcore/Miner.h b/libethcore/Miner.h index 1c8d00998..5c5e6cc4e 100644 --- a/libethcore/Miner.h +++ b/libethcore/Miner.h @@ -38,6 +38,7 @@ #define DAG_LOAD_MODE_PARALLEL 0 #define DAG_LOAD_MODE_SEQUENTIAL 1 +#define DAG_LOAD_MODE_SINGLE 2 using namespace std; @@ -218,7 +219,9 @@ protected: void accumulateHashes(unsigned _n) { m_hashCount += _n; } static unsigned s_dagLoadMode; - static unsigned s_dagLoadIndex; + static volatile unsigned s_dagLoadIndex; + static unsigned s_dagCreateDevice; + static volatile void* s_dagInHostMemory; private: FarmFace* m_farm = nullptr; unsigned m_index; diff --git a/libstratum/EthStratumClient.cpp b/libstratum/EthStratumClient.cpp index 029c15754..99d8b1c57 100644 --- a/libstratum/EthStratumClient.cpp +++ b/libstratum/EthStratumClient.cpp @@ -1,10 +1,33 @@ #include "EthStratumClient.h" #include +#include using boost::asio::ip::tcp; -EthStratumClient::EthStratumClient(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout) +static void diffToTarget(uint32_t *target, double diff) +{ + uint32_t target2[8]; + uint64_t m; + int k; + + for (k = 6; k > 0 && diff > 1.0; k--) + diff /= 4294967296.0; + m = (uint64_t)(4294901760.0 / diff); + if (m == 0 && k == 6) + memset(target2, 0xff, 32); + else { + memset(target2, 0, 32); + target2[k] = (uint32_t)m; + target2[k + 1] = (uint32_t)(m >> 32); + } + + for (int i = 0; i < 32; i++) + ((uint8_t*)target)[31 - i] = ((uint8_t*)target2)[i]; +} + + +EthStratumClient::EthStratumClient(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout, bool ethstratum) : m_socket(m_io_service) { m_minerType = m; @@ -21,6 +44,8 @@ EthStratumClient::EthStratumClient(GenericFarm * f, MinerType m_maxRetries = retries; m_worktimeout = worktimeout; + m_ethereumStratum = ethstratum; + p_farm = f; p_worktimer = nullptr; connect(); @@ -154,7 +179,10 @@ void EthStratumClient::connect_handler(const boost::system::error_code& ec, tcp: } } std::ostream os(&m_requestBuffer); - os << "{\"id\": 1, \"method\": \"mining.subscribe\", \"params\": []}\n"; + if (m_ethereumStratum) + os << "{\"id\": 1, \"method\": \"mining.subscribe\", \"params\": [\"ethminer/" << ETH_PROJECT_VERSION << "\",\"EthereumStratum/1.0.0\"]}\n"; + else + os << "{\"id\": 1, \"method\": \"mining.subscribe\", \"params\": []}\n"; async_write(m_socket, m_requestBuffer, @@ -236,6 +264,16 @@ void EthStratumClient::readResponse(const boost::system::error_code& ec, std::si } } +void EthStratumClient::processExtranonce(std::string& enonce) +{ + m_extraNonceHexSize = enonce.length(); + + cnote << "Extranonce set to " << enonce; + + for (int i = enonce.length(); i < 16; ++i) enonce += "0"; + m_extraNonce = h64(enonce); +} + void EthStratumClient::processReponse(Json::Value& responseObject) { Json::Value error = responseObject.get("error", new Json::Value); @@ -252,13 +290,29 @@ void EthStratumClient::processReponse(Json::Value& responseObject) case 1: cnote << "Subscribed to stratum server"; - os << "{\"id\": 2, \"method\": \"mining.authorize\", \"params\": [\"" << p_active->user << "\",\"" << p_active->pass << "\"]}\n"; + if (m_ethereumStratum) + { + m_nextWorkDifficulty = 1; + params = responseObject.get("result", Json::Value::null); + if (params.isArray()) + { + std::string enonce = params.get((Json::Value::ArrayIndex)1, "").asString(); + processExtranonce(enonce); + } + + os << "{\"id\": 2, \"method\": \"mining.extranonce.subscribe\", \"params\": []}\n"; + } + + os << "{\"id\": 3, \"method\": \"mining.authorize\", \"params\": [\"" << p_active->user << "\",\"" << p_active->pass << "\"]}\n"; async_write(m_socket, m_requestBuffer, boost::bind(&EthStratumClient::handleResponse, this, boost::asio::placeholders::error)); break; case 2: + // nothing to do... + break; + case 3: m_authorized = responseObject.get("result", Json::Value::null).asBool(); if (!m_authorized) { @@ -286,55 +340,107 @@ void EthStratumClient::processReponse(Json::Value& responseObject) if (params.isArray()) { string job = params.get((Json::Value::ArrayIndex)0, "").asString(); - string sHeaderHash = params.get((Json::Value::ArrayIndex)1, "").asString(); - string sSeedHash = params.get((Json::Value::ArrayIndex)2, "").asString(); - string sShareTarget = params.get((Json::Value::ArrayIndex)3, "").asString(); - //bool cleanJobs = params.get((Json::Value::ArrayIndex)4, "").asBool(); - - // coinmine.pl fix - int l = sShareTarget.length(); - if (l < 66) - sShareTarget = "0x" + string(66 - l, '0') + sShareTarget.substr(2); - - - if (sHeaderHash != "" && sSeedHash != "" && sShareTarget != "") - { - cnote << "Received new job #" + job.substr(0,8); - //cnote << "Header hash: " + sHeaderHash; - //cnote << "Seed hash: " + sSeedHash; - //cnote << "Share target: " + sShareTarget; - h256 seedHash = h256(sSeedHash); - h256 headerHash = h256(sHeaderHash); + if (m_ethereumStratum) + { + string sSeedHash = params.get((Json::Value::ArrayIndex)1, "").asString(); + string sHeaderHash = params.get((Json::Value::ArrayIndex)2, "").asString(); - if (headerHash != m_current.headerHash) + if (sHeaderHash != "" && sSeedHash != "") { - //x_current.lock(); - if (p_worktimer) - p_worktimer->cancel(); + cnote << "Received new job #" + job; + //cnote << "Header hash: " + sHeaderHash; + //cnote << "Seed hash: " + sSeedHash; + //cnote << "Share target: " + sShareTarget; + + h256 seedHash = h256(sSeedHash); + h256 headerHash = h256(sHeaderHash); m_previous.headerHash = m_current.headerHash; m_previous.seedHash = m_current.seedHash; m_previous.boundary = m_current.boundary; + m_previous.startNonce = m_current.startNonce; + m_previous.exSizeBits = m_previous.exSizeBits; m_previousJob = m_job; m_current.headerHash = h256(sHeaderHash); m_current.seedHash = seedHash; - m_current.boundary = h256(sShareTarget);// , h256::AlignRight); + m_current.boundary = h256(); + diffToTarget((uint32_t*)m_current.boundary.data(), m_nextWorkDifficulty); + m_current.startNonce = ethash_swap_u64(*((uint64_t*)m_extraNonce.data())); + m_current.exSizeBits = m_extraNonceHexSize * 4; m_job = job; p_farm->setWork(m_current); - //x_current.unlock(); - p_worktimer = new boost::asio::deadline_timer(m_io_service, boost::posix_time::seconds(m_worktimeout)); - p_worktimer->async_wait(boost::bind(&EthStratumClient::work_timeout_handler, this, boost::asio::placeholders::error)); + } + } + else + { + string sHeaderHash = params.get((Json::Value::ArrayIndex)1, "").asString(); + string sSeedHash = params.get((Json::Value::ArrayIndex)2, "").asString(); + string sShareTarget = params.get((Json::Value::ArrayIndex)3, "").asString(); + //bool cleanJobs = params.get((Json::Value::ArrayIndex)4, "").asBool(); + // coinmine.pl fix + int l = sShareTarget.length(); + if (l < 66) + sShareTarget = "0x" + string(66 - l, '0') + sShareTarget.substr(2); + + + if (sHeaderHash != "" && sSeedHash != "" && sShareTarget != "") + { + cnote << "Received new job #" + job.substr(0, 8); + //cnote << "Header hash: " + sHeaderHash; + //cnote << "Seed hash: " + sSeedHash; + //cnote << "Share target: " + sShareTarget; + + h256 seedHash = h256(sSeedHash); + h256 headerHash = h256(sHeaderHash); + + if (headerHash != m_current.headerHash) + { + //x_current.lock(); + if (p_worktimer) + p_worktimer->cancel(); + + m_previous.headerHash = m_current.headerHash; + m_previous.seedHash = m_current.seedHash; + m_previous.boundary = m_current.boundary; + m_previousJob = m_job; + + m_current.headerHash = h256(sHeaderHash); + m_current.seedHash = seedHash; + m_current.boundary = h256(sShareTarget);// , h256::AlignRight); + m_job = job; + + p_farm->setWork(m_current); + //x_current.unlock(); + p_worktimer = new boost::asio::deadline_timer(m_io_service, boost::posix_time::seconds(m_worktimeout)); + p_worktimer->async_wait(boost::bind(&EthStratumClient::work_timeout_handler, this, boost::asio::placeholders::error)); + + } } } } } - else if (method == "mining.set_difficulty") + else if (method == "mining.set_difficulty" && m_ethereumStratum) { - + params = responseObject.get("params", Json::Value::null); + if (params.isArray()) + { + m_nextWorkDifficulty = params.get((Json::Value::ArrayIndex)0, 1).asDouble(); + if (m_nextWorkDifficulty <= 0.0001) m_nextWorkDifficulty = 0.0001; + cnote << "Difficulty set to " << m_nextWorkDifficulty; + } + } + else if (method == "mining.set_extranonce" && m_ethereumStratum) + { + params = responseObject.get("params", Json::Value::null); + if (params.isArray()) + { + std::string enonce = params.get((Json::Value::ArrayIndex)0, "").asString(); + processExtranonce(enonce); + } } else if (method == "client.get_version") { @@ -364,11 +470,17 @@ bool EthStratumClient::submit(EthashProofOfWork::Solution solution) { x_current.unlock(); cnote << "Solution found; Submitting to" << p_active->host << "..."; - cnote << " Nonce:" << "0x" + solution.nonce.hex(); + string minernonce = solution.nonce.hex().substr(m_extraNonceHexSize, 16 - m_extraNonceHexSize); + if (!m_ethereumStratum) + cnote << " Nonce:" << "0x" + solution.nonce.hex(); if (EthashAux::eval(tempWork.seedHash, tempWork.headerHash, solution.nonce).value < tempWork.boundary) { - string json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; + string json; + if (m_ethereumStratum) + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_job + "\",\"" + minernonce + "\"]}\n"; + else + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; std::ostream os(&m_requestBuffer); os << json; m_stale = false; @@ -379,7 +491,11 @@ bool EthStratumClient::submit(EthashProofOfWork::Solution solution) { } else if (EthashAux::eval(tempPreviousWork.seedHash, tempPreviousWork.headerHash, solution.nonce).value < tempPreviousWork.boundary) { - string json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_previous_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempPreviousWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; + string json; + if (m_ethereumStratum) + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_previous_job + "\",\"" + minernonce + "\"]}\n"; + else + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_previous_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempPreviousWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; std::ostream os(&m_requestBuffer); os << json; m_stale = true; diff --git a/libstratum/EthStratumClient.h b/libstratum/EthStratumClient.h index 3e49b02b8..780b91105 100644 --- a/libstratum/EthStratumClient.h +++ b/libstratum/EthStratumClient.h @@ -21,7 +21,7 @@ using namespace dev::eth; class EthStratumClient { public: - EthStratumClient(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout); + EthStratumClient(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout, bool ethstratum); ~EthStratumClient(); void setFailover(string const & host, string const & port); @@ -87,4 +87,12 @@ private: boost::asio::deadline_timer * p_worktimer; + bool m_ethereumStratum; + + double m_nextWorkDifficulty; + + h64 m_extraNonce; + int m_extraNonceHexSize; + + void processExtranonce(std::string& enonce); }; \ No newline at end of file diff --git a/libstratum/EthStratumClientV2.cpp b/libstratum/EthStratumClientV2.cpp index 2c032cff9..93e906d19 100644 --- a/libstratum/EthStratumClientV2.cpp +++ b/libstratum/EthStratumClientV2.cpp @@ -1,10 +1,33 @@ #include "EthStratumClientV2.h" #include +#include using boost::asio::ip::tcp; -EthStratumClientV2::EthStratumClientV2(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout) +static void diffToTarget(uint32_t *target, double diff) +{ + uint32_t target2[8]; + uint64_t m; + int k; + + for (k = 6; k > 0 && diff > 1.0; k--) + diff /= 4294967296.0; + m = (uint64_t)(4294901760.0 / diff); + if (m == 0 && k == 6) + memset(target2, 0xff, 32); + else { + memset(target2, 0, 32); + target2[k] = (uint32_t)m; + target2[k + 1] = (uint32_t)(m >> 32); + } + + for (int i = 0; i < 32; i++) + ((uint8_t*)target)[31 - i] = ((uint8_t*)target2)[i]; +} + + +EthStratumClientV2::EthStratumClientV2(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout, bool ethstratum) : Worker("stratum"), m_socket(m_io_service) { @@ -21,6 +44,8 @@ EthStratumClientV2::EthStratumClientV2(GenericFarm * f, Miner m_maxRetries = retries; m_worktimeout = worktimeout; + m_ethereumStratum = ethstratum; + p_farm = f; p_worktimer = nullptr; startWorking(); @@ -127,7 +152,10 @@ void EthStratumClientV2::connect() } } std::ostream os(&m_requestBuffer); - os << "{\"id\": 1, \"method\": \"mining.subscribe\", \"params\": []}\n"; + if (m_ethereumStratum) + os << "{\"id\": 1, \"method\": \"mining.subscribe\", \"params\": [\"ethminer/" << ETH_PROJECT_VERSION << "\",\"EthereumStratum/1.0.0\"]}\n"; + else + os << "{\"id\": 1, \"method\": \"mining.subscribe\", \"params\": []}\n"; write(m_socket, m_requestBuffer); } } @@ -184,6 +212,16 @@ void EthStratumClientV2::disconnect() //m_io_service.stop(); } +void EthStratumClientV2::processExtranonce(std::string& enonce) +{ + m_extraNonceHexSize = enonce.length(); + + cnote << "Extranonce set to " << enonce; + + for (int i = enonce.length(); i < 16; ++i) enonce += "0"; + m_extraNonce = h64(enonce); +} + void EthStratumClientV2::processReponse(Json::Value& responseObject) { Json::Value error = responseObject.get("error", new Json::Value); @@ -200,11 +238,27 @@ void EthStratumClientV2::processReponse(Json::Value& responseObject) case 1: cnote << "Subscribed to stratum server"; - os << "{\"id\": 2, \"method\": \"mining.authorize\", \"params\": [\"" << p_active->user << "\",\"" << p_active->pass << "\"]}\n"; + if (m_ethereumStratum) + { + m_nextWorkDifficulty = 1; + params = responseObject.get("result", Json::Value::null); + if (params.isArray()) + { + std::string enonce = params.get((Json::Value::ArrayIndex)1, "").asString(); + processExtranonce(enonce); + } + + os << "{\"id\": 2, \"method\": \"mining.extranonce.subscribe\", \"params\": []}\n"; + } + + os << "{\"id\": 3, \"method\": \"mining.authorize\", \"params\": [\"" << p_active->user << "\",\"" << p_active->pass << "\"]}\n"; write(m_socket, m_requestBuffer); break; case 2: + // nothing to do... + break; + case 3: m_authorized = responseObject.get("result", Json::Value::null).asBool(); if (!m_authorized) { @@ -232,51 +286,104 @@ void EthStratumClientV2::processReponse(Json::Value& responseObject) if (params.isArray()) { string job = params.get((Json::Value::ArrayIndex)0, "").asString(); - string sHeaderHash = params.get((Json::Value::ArrayIndex)1, "").asString(); - string sSeedHash = params.get((Json::Value::ArrayIndex)2, "").asString(); - string sShareTarget = params.get((Json::Value::ArrayIndex)3, "").asString(); - //bool cleanJobs = params.get((Json::Value::ArrayIndex)4, "").asBool(); - - // coinmine.pl fix - int l = sShareTarget.length(); - if (l < 66) - sShareTarget = "0x" + string(66 - l, '0') + sShareTarget.substr(2); - - if (sHeaderHash != "" && sSeedHash != "" && sShareTarget != "") + if (m_ethereumStratum) { - cnote << "Received new job #" + job.substr(0,8); + string job = params.get((Json::Value::ArrayIndex)0, "").asString(); + string sSeedHash = params.get((Json::Value::ArrayIndex)1, "").asString(); + string sHeaderHash = params.get((Json::Value::ArrayIndex)2, "").asString(); - h256 seedHash = h256(sSeedHash); - h256 headerHash = h256(sHeaderHash); - - if (headerHash != m_current.headerHash) + if (sHeaderHash != "" && sSeedHash != "") { - //x_current.lock(); - //if (p_worktimer) - // p_worktimer->cancel(); + cnote << "Received new job #" + job; + //cnote << "Header hash: " + sHeaderHash; + //cnote << "Seed hash: " + sSeedHash; + //cnote << "Share target: " + sShareTarget; + + h256 seedHash = h256(sSeedHash); + h256 headerHash = h256(sHeaderHash); m_previous.headerHash = m_current.headerHash; m_previous.seedHash = m_current.seedHash; m_previous.boundary = m_current.boundary; + m_previous.startNonce = m_current.startNonce; + m_previous.exSizeBits = m_previous.exSizeBits; m_previousJob = m_job; m_current.headerHash = h256(sHeaderHash); m_current.seedHash = seedHash; - m_current.boundary = h256(sShareTarget); + m_current.boundary = h256(); + diffToTarget((uint32_t*)m_current.boundary.data(), m_nextWorkDifficulty); + m_current.startNonce = ethash_swap_u64(*((uint64_t*)m_extraNonce.data())); + m_current.exSizeBits = m_extraNonceHexSize * 4; m_job = job; p_farm->setWork(m_current); - //x_current.unlock(); - //p_worktimer = new boost::asio::deadline_timer(m_io_service, boost::posix_time::seconds(m_worktimeout)); - //p_worktimer->async_wait(boost::bind(&EthStratumClientV2::work_timeout_handler, this, boost::asio::placeholders::error)); + } + } + else + { + string sHeaderHash = params.get((Json::Value::ArrayIndex)1, "").asString(); + string sSeedHash = params.get((Json::Value::ArrayIndex)2, "").asString(); + string sShareTarget = params.get((Json::Value::ArrayIndex)3, "").asString(); + //bool cleanJobs = params.get((Json::Value::ArrayIndex)4, "").asBool(); + + // coinmine.pl fix + int l = sShareTarget.length(); + if (l < 66) + sShareTarget = "0x" + string(66 - l, '0') + sShareTarget.substr(2); + + + if (sHeaderHash != "" && sSeedHash != "" && sShareTarget != "") + { + cnote << "Received new job #" + job.substr(0, 8); + + h256 seedHash = h256(sSeedHash); + h256 headerHash = h256(sHeaderHash); + + if (headerHash != m_current.headerHash) + { + //x_current.lock(); + //if (p_worktimer) + // p_worktimer->cancel(); + + m_previous.headerHash = m_current.headerHash; + m_previous.seedHash = m_current.seedHash; + m_previous.boundary = m_current.boundary; + m_previousJob = m_job; + + m_current.headerHash = h256(sHeaderHash); + m_current.seedHash = seedHash; + m_current.boundary = h256(sShareTarget); + m_job = job; + + p_farm->setWork(m_current); + //x_current.unlock(); + //p_worktimer = new boost::asio::deadline_timer(m_io_service, boost::posix_time::seconds(m_worktimeout)); + //p_worktimer->async_wait(boost::bind(&EthStratumClientV2::work_timeout_handler, this, boost::asio::placeholders::error)); + } } } } } - else if (method == "mining.set_difficulty") + else if (method == "mining.set_difficulty" && m_ethereumStratum) { - + params = responseObject.get("params", Json::Value::null); + if (params.isArray()) + { + m_nextWorkDifficulty = params.get((Json::Value::ArrayIndex)0, 1).asDouble(); + if (m_nextWorkDifficulty <= 0.0001) m_nextWorkDifficulty = 0.0001; + cnote << "Difficulty set to " << m_nextWorkDifficulty; + } + } + else if (method == "mining.set_extranonce" && m_ethereumStratum) + { + params = responseObject.get("params", Json::Value::null); + if (params.isArray()) + { + std::string enonce = params.get((Json::Value::ArrayIndex)0, "").asString(); + processExtranonce(enonce); + } } else if (method == "client.get_version") { @@ -304,11 +411,17 @@ bool EthStratumClientV2::submit(EthashProofOfWork::Solution solution) { x_current.unlock(); cnote << "Solution found; Submitting to" << p_active->host << "..."; - cnote << " Nonce:" << "0x" + solution.nonce.hex(); + string minernonce = solution.nonce.hex().substr(m_extraNonceHexSize, 16 - m_extraNonceHexSize); + if (!m_ethereumStratum) + cnote << " Nonce:" << "0x" + solution.nonce.hex(); if (EthashAux::eval(tempWork.seedHash, tempWork.headerHash, solution.nonce).value < tempWork.boundary) { - string json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; + string json; + if (m_ethereumStratum) + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_job + "\",\"" + minernonce + "\"]}\n"; + else + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; std::ostream os(&m_requestBuffer); os << json; m_stale = false; @@ -317,7 +430,11 @@ bool EthStratumClientV2::submit(EthashProofOfWork::Solution solution) { } else if (EthashAux::eval(tempPreviousWork.seedHash, tempPreviousWork.headerHash, solution.nonce).value < tempPreviousWork.boundary) { - string json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_previous_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempPreviousWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; + string json; + if (m_ethereumStratum) + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_previous_job + "\",\"" + minernonce + "\"]}\n"; + else + json = "{\"id\": 4, \"method\": \"mining.submit\", \"params\": [\"" + p_active->user + "\",\"" + temp_previous_job + "\",\"0x" + solution.nonce.hex() + "\",\"0x" + tempPreviousWork.headerHash.hex() + "\",\"0x" + solution.mixHash.hex() + "\"]}\n"; std::ostream os(&m_requestBuffer); os << json; m_stale = true; diff --git a/libstratum/EthStratumClientV2.h b/libstratum/EthStratumClientV2.h index 38e6741a9..a42cbc8b5 100644 --- a/libstratum/EthStratumClientV2.h +++ b/libstratum/EthStratumClientV2.h @@ -22,7 +22,7 @@ using namespace dev::eth; class EthStratumClientV2 : public Worker { public: - EthStratumClientV2(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout); + EthStratumClientV2(GenericFarm * f, MinerType m, string const & host, string const & port, string const & user, string const & pass, int const & retries, int const & worktimeout, bool ethstratum); ~EthStratumClientV2(); void setFailover(string const & host, string const & port); @@ -81,4 +81,12 @@ private: boost::asio::deadline_timer * p_worktimer; + bool m_ethereumStratum; + + double m_nextWorkDifficulty; + + h64 m_extraNonce; + int m_extraNonceHexSize; + + void processExtranonce(std::string& enonce); }; \ No newline at end of file