Browse Source

relaxed local work size requirements

added mutex to stratum pending calls. untested
cl-refactor
Genoil 9 years ago
parent
commit
4e2f2f37a4
  1. 2
      libethash-cuda/ethash_cuda_miner_kernel.cu
  2. 5
      libethcore/EthashCUDAMiner.cpp
  3. 4
      libethcore/EthashGPUMiner.cpp
  4. 9
      libstratum/EthStratumClient.cpp
  5. 1
      libstratum/EthStratumClient.h

2
libethash-cuda/ethash_cuda_miner_kernel.cu

@ -11,7 +11,7 @@
#include "dagger_shuffled.cuh" #include "dagger_shuffled.cuh"
__global__ void __global__ void
__launch_bounds__(128, 7) __launch_bounds__(896, 1)
ethash_search( ethash_search(
volatile uint32_t* g_output, volatile uint32_t* g_output,
uint64_t start_nonce uint64_t start_nonce

5
libethcore/EthashCUDAMiner.cpp

@ -219,13 +219,14 @@ bool EthashCUDAMiner::configureGPU(
uint64_t _currentBlock uint64_t _currentBlock
) )
{ {
_blockSize = ((_blockSize + 7) / 8) * 8;
/*
if (_blockSize != 32 && _blockSize != 64 && _blockSize != 128) if (_blockSize != 32 && _blockSize != 64 && _blockSize != 128)
{ {
cout << "Given localWorkSize of " << toString(_blockSize) << "is invalid. Must be either 32,64 or 128" << endl; cout << "Given localWorkSize of " << toString(_blockSize) << "is invalid. Must be either 32,64 or 128" << endl;
return false; return false;
} }
*/
if (!ethash_cuda_miner::configureGPU( if (!ethash_cuda_miner::configureGPU(
s_devices, s_devices,
_blockSize, _blockSize,

4
libethcore/EthashGPUMiner.cpp

@ -213,12 +213,14 @@ bool EthashGPUMiner::configureGPU(
s_platformId = _platformId; s_platformId = _platformId;
s_deviceId = _deviceId; s_deviceId = _deviceId;
_localWorkSize = ((_localWorkSize + 7) / 8) * 8;
/*
if (_localWorkSize != 32 && _localWorkSize != 64 && _localWorkSize != 128 && _localWorkSize != 256) if (_localWorkSize != 32 && _localWorkSize != 64 && _localWorkSize != 128 && _localWorkSize != 256)
{ {
cout << "Given localWorkSize of " << toString(_localWorkSize) << "is invalid. Must be either 32,64,128 or 256" << endl; cout << "Given localWorkSize of " << toString(_localWorkSize) << "is invalid. Must be either 32,64,128 or 256" << endl;
return false; return false;
} }
*/
if (!ethash_cl_miner::configureGPU( if (!ethash_cl_miner::configureGPU(
_platformId, _platformId,
_localWorkSize, _localWorkSize,

9
libstratum/EthStratumClient.cpp

@ -16,8 +16,6 @@ EthStratumClient::EthStratumClient(GenericFarm<EthashProofOfWork> * f, MinerType
m_connected = false; m_connected = false;
m_precompute = true; m_precompute = true;
m_pending = 0; m_pending = 0;
p_farm = f; p_farm = f;
connect(); connect();
} }
@ -119,13 +117,16 @@ void EthStratumClient::connect_handler(const boost::system::error_code& ec, tcp:
} }
void EthStratumClient::readline() { void EthStratumClient::readline() {
m_mtx.lock();
if (m_pending == 0) { if (m_pending == 0) {
async_read_until(m_socket, m_responseBuffer, "\n", async_read_until(m_socket, m_responseBuffer, "\n",
boost::bind(&EthStratumClient::readResponse, this, boost::bind(&EthStratumClient::readResponse, this,
boost::asio::placeholders::error, boost::asio::placeholders::bytes_transferred)); boost::asio::placeholders::error, boost::asio::placeholders::bytes_transferred));
m_pending++; m_pending++;
} }
m_mtx.unlock();
} }
void EthStratumClient::handleResponse(const boost::system::error_code& ec) { void EthStratumClient::handleResponse(const boost::system::error_code& ec) {
@ -141,7 +142,9 @@ void EthStratumClient::handleResponse(const boost::system::error_code& ec) {
void EthStratumClient::readResponse(const boost::system::error_code& ec, std::size_t bytes_transferred) void EthStratumClient::readResponse(const boost::system::error_code& ec, std::size_t bytes_transferred)
{ {
m_mtx.lock();
m_pending = m_pending > 0 ? m_pending - 1 : 0; m_pending = m_pending > 0 ? m_pending - 1 : 0;
m_mtx.unlock();
if (!ec) if (!ec)
{ {

1
libstratum/EthStratumClient.h

@ -48,6 +48,7 @@ private:
bool m_connected; bool m_connected;
bool m_precompute; bool m_precompute;
boost::mutex m_mtx;
int m_pending; int m_pending;
string m_response; string m_response;

Loading…
Cancel
Save