Merge pull request #2155 from LefterisJP/opencl_smarter_gpu_memorycheck

OpenCL smarter GPU memorycheck

ethminer/MinerAux.h

@@ -30,6 +30,7 @@
 
 #include <boost/algorithm/string.hpp>
 #include <boost/algorithm/string/trim_all.hpp>
+#include <boost/optional.hpp>
 
 #include <libdevcore/FileSystem.h>
 #include <libevmcore/Instruction.h>
@@ -131,6 +132,8 @@ public:
 			m_shouldListDevices = true;
 		else if (arg == "--allow-opencl-cpu")
 			m_clAllowCPU = true;
+		else if (arg == "--cl-extragpu-mem" && i + 1 < argc)
+			m_extraGPUMemory = 1000000 * stol(argv[++i]);
 		else if (arg == "--phone-home" && i + 1 < argc)
 		{
 			string m = argv[++i];
@@ -175,6 +178,8 @@ public:
 			m_minerType = MinerType::CPU;
 		else if (arg == "-G" || arg == "--opencl")
 			m_minerType = MinerType::GPU;
+		else if (arg == "--current-block" && i + 1 < argc)
+			m_currentBlock = stol(argv[++i]);
 		else if (arg == "--no-precompute")
 		{
 			m_precompute = false;
@@ -261,7 +266,13 @@ public:
 		else if (m_minerType == MinerType::GPU)
 		{
 			ProofOfWork::GPUMiner::setNumInstances(m_miningThreads);
-			if (!ProofOfWork::GPUMiner::configureGPU(m_openclPlatform, m_openclDevice, m_clAllowCPU))
+			if (!ProofOfWork::GPUMiner::configureGPU(
+					m_openclPlatform,
+					m_openclDevice,
+					m_clAllowCPU,
+					m_extraGPUMemory,
+					m_currentBlock
+				))
 			{
 				cout << "No GPU device with sufficient memory was found. Can't GPU mine. Remove the -G argument" << endl;
 				exit(1);
@@ -303,6 +314,10 @@ public:
 			<< "    --opencl-platform <n>  When mining using -G/--opencl use OpenCL platform n (default: 0)." << endl
 			<< "    --opencl-device <n>  When mining using -G/--opencl use OpenCL device n (default: 0)." << endl
 			<< "    -t, --mining-threads <n> Limit number of CPU/GPU miners to n (default: use everything available on selected platform)" << endl
+			<< "    --allow-opencl-cpu Allows CPU to be considered as an OpenCL device if the OpenCL platform supports it." << endl
+			<< "    --list-devices List the detected OpenCL devices and exit." <<endl
+			<< "    --current-block Let the miner know the current block number at configuration time. Will help determine DAG size and required GPU memory." <<endl
+			<< "    --cl-extragpu-mem Set the memory (in MB) you believe your GPU requires for stuff other than mining. Windows rendering e.t.c.." <<endl
 			;
 	}
 
@@ -492,6 +507,9 @@ private:
 	unsigned m_miningThreads = UINT_MAX;
 	bool m_shouldListDevices = false;
 	bool m_clAllowCPU = false;
+	boost::optional<uint64_t> m_currentBlock;
+	// default value is 350MB of GPU memory for other stuff (windows system rendering, e.t.c.)
+	unsigned m_extraGPUMemory = 350000000;
 
 	/// DAG initialisation param.
 	unsigned m_initDAG = 0;

libethash-cl/ethash_cl_miner.cpp

@@ -32,11 +32,11 @@
 #include <vector>
 #include <libethash/util.h>
 #include <libethash/ethash.h>
+#include <libethash/internal.h>
 #include "ethash_cl_miner.h"
 #include "ethash_cl_miner_kernel.h"
 
 #define ETHASH_BYTES 32
-#define ETHASH_CL_MINIMUM_MEMORY 2000000000
 
 // workaround lame platforms
 #if !CL_VERSION_1_2
@@ -137,14 +137,18 @@ unsigned ethash_cl_miner::getNumDevices(unsigned _platformId)
 	return devices.size();
 }
 
-bool ethash_cl_miner::configureGPU(bool _allowCPU)
+bool ethash_cl_miner::configureGPU(bool _allowCPU, unsigned _extraGPUMemory, boost::optional<uint64_t> _currentBlock)
 {
 	s_allowCPU = _allowCPU;
-	return searchForAllDevices([](cl::Device const _device) -> bool
+	s_extraRequiredGPUMem = _extraGPUMemory;
+	// by default let's only consider the DAG of the first epoch
+	uint64_t dagSize = _currentBlock ? ethash_get_datasize(*_currentBlock) : 1073739904U;
+	uint64_t requiredSize =  dagSize + _extraGPUMemory;
+	return searchForAllDevices([&requiredSize](cl::Device const _device) -> bool
 		{
 			cl_ulong result;
 			_device.getInfo(CL_DEVICE_GLOBAL_MEM_SIZE, &result);
-			if (result >= ETHASH_CL_MINIMUM_MEMORY)
+			if (result >= requiredSize)
 			{
 				ETHCL_LOG(
 					"Found suitable OpenCL device [" << _device.getInfo<CL_DEVICE_NAME>()
@@ -156,7 +160,7 @@ bool ethash_cl_miner::configureGPU(bool _allowCPU)
 			ETHCL_LOG(
 				"OpenCL device " << _device.getInfo<CL_DEVICE_NAME>()
 				<< " has insufficient GPU memory." << result <<
-				" bytes of memory found < " << ETHASH_CL_MINIMUM_MEMORY << " bytes of memory required"
+				" bytes of memory found < " << requiredSize << " bytes of memory required"
 			);
 			return false;
 		}
@@ -164,6 +168,7 @@ bool ethash_cl_miner::configureGPU(bool _allowCPU)
 }
 
 bool ethash_cl_miner::s_allowCPU = false;
+unsigned ethash_cl_miner::s_extraRequiredGPUMem;
 
 bool ethash_cl_miner::searchForAllDevices(function<bool(cl::Device const&)> _callback)
 {
@@ -279,7 +284,7 @@ bool ethash_cl_miner::init(
 		// configure chunk number depending on max allocateable memory
 		cl_ulong result;
 		device.getInfo(CL_DEVICE_MAX_MEM_ALLOC_SIZE, &result);
-		if (result >= ETHASH_CL_MINIMUM_MEMORY)
+		if (result >= _dagSize)
 		{
 			m_dagChunksNum = 1;
 			ETHCL_LOG("Using 1 big chunk. Max OpenCL allocateable memory is" << result);

libethash-cl/ethash_cl_miner.h

@@ -12,6 +12,7 @@
 #include "cl.hpp"
 #endif
 
+#include <boost/optional.hpp>
 #include <time.h>
 #include <functional>
 #include <libethash/ethash.h>
@@ -40,7 +41,7 @@ public:
 	static unsigned getNumDevices(unsigned _platformId = 0);
 	static std::string platform_info(unsigned _platformId = 0, unsigned _deviceId = 0);
 	static void listDevices();
-	static bool configureGPU(bool _allowCPU);
+	static bool configureGPU(bool _allowCPU, unsigned _extraGPUMemory, boost::optional<uint64_t> _currentBlock);
 
 	bool init(
 		uint8_t const* _dag,
@@ -73,5 +74,9 @@ private:
 	unsigned m_workgroup_size;
 	bool m_opencl_1_1;
 
+	/// Allow CPU to appear as an OpenCL device or not. Default is false
 	static bool s_allowCPU;
+	/// GPU memory required for other things, like window rendering e.t.c.
+	/// User can set it via the --cl-extragpu-mem argument.
+	static unsigned s_extraRequiredGPUMem;
 };

libethcore/Ethash.cpp

@@ -381,11 +381,17 @@ void Ethash::GPUMiner::listDevices()
 	return ethash_cl_miner::listDevices();
 }
 
-bool Ethash::GPUMiner::configureGPU(unsigned _platformId, unsigned _deviceId, bool _allowCPU)
+bool Ethash::GPUMiner::configureGPU(
+	unsigned _platformId,
+	unsigned _deviceId,
+	bool _allowCPU,
+	unsigned _extraGPUMemory,
+	boost::optional<uint64_t> _currentBlock
+)
 {
 	s_platformId = _platformId;
 	s_deviceId = _deviceId;
-	return ethash_cl_miner::configureGPU(_allowCPU);
+	return ethash_cl_miner::configureGPU(_allowCPU, _extraGPUMemory, _currentBlock);
 }
 
 #endif

libethcore/Ethash.h

@@ -88,7 +88,7 @@ public:
 		static unsigned instances() { return s_numInstances > 0 ? s_numInstances : std::thread::hardware_concurrency(); }
 		static std::string platformInfo();
 		static void listDevices() {}
-		static bool configureGPU(unsigned, unsigned, bool) { return false; }
+		static bool configureGPU(unsigned, unsigned, bool, unsigned, boost::optional<uint64_t>) { return false; }
 		static void setNumInstances(unsigned _instances) { s_numInstances = std::min<unsigned>(_instances, std::thread::hardware_concurrency()); }
 	protected:
 		void kickOff() override
@@ -117,7 +117,13 @@ public:
 		static std::string platformInfo();
 		static unsigned getNumDevices();
 		static void listDevices();
-		static bool configureGPU(unsigned _platformId, unsigned _deviceId, bool _allowCPU);
+		static bool configureGPU(
+			unsigned _platformId,
+			unsigned _deviceId,
+			bool _allowCPU,
+			unsigned _extraGPUMemory,
+			boost::optional<uint64_t> _currentBlock
+		);
 		static void setNumInstances(unsigned _instances) { s_numInstances = std::min<unsigned>(_instances, getNumDevices()); }
 
 	protected:

libethcore/EthashAux.cpp

@@ -54,6 +54,11 @@ uint64_t EthashAux::cacheSize(BlockInfo const& _header)
 	return ethash_get_cachesize((uint64_t)_header.number);
 }
 
+uint64_t EthashAux::dataSize(uint64_t _blockNumber)
+{
+	return ethash_get_datasize(_blockNumber);
+}
+
 h256 EthashAux::seedHash(unsigned _number)
 {
 	unsigned epoch = _number / ETHASH_EPOCH_LENGTH;

libethcore/EthashAux.h

@@ -66,6 +66,7 @@ public:
 	static h256 seedHash(unsigned _number);
 	static uint64_t number(h256 const& _seedHash);
 	static uint64_t cacheSize(BlockInfo const& _header);
+	static uint64_t dataSize(uint64_t _blockNumber);
 
 	static LightType light(h256 const& _seedHash);