OpenCL stuff, Fix for eth_call.

exp/CMakeLists.txt

@@ -10,8 +10,26 @@ set(EXECUTABLE exp)
 
 add_executable(${EXECUTABLE} ${SRC_LIST})
 
+target_link_libraries(${EXECUTABLE} ${Boost_REGEX_LIBRARIES})
+
+if (READLINE_FOUND)
+	target_link_libraries(${EXECUTABLE} ${READLINE_LIBRARIES})
+endif()
+
+if (JSONRPC)
+	target_link_libraries(${EXECUTABLE} web3jsonrpc)
+endif()
+
+target_link_libraries(${EXECUTABLE} webthree)
 target_link_libraries(${EXECUTABLE} ethereum)
 target_link_libraries(${EXECUTABLE} p2p)
+target_link_libraries(${EXECUTABLE} ethash-cl)
+target_link_libraries(${EXECUTABLE} ethash)
+target_link_libraries(${EXECUTABLE} OpenCL)
 
 install( TARGETS ${EXECUTABLE} DESTINATION bin) 
 
+
+
+
+

exp/main.cpp

@@ -19,18 +19,25 @@
  * @date 2014
  * Ethereum client.
  */
+#define __CL_ENABLE_EXCEPTIONS
+#define CL_USE_DEPRECATED_OPENCL_2_0_APIS
+#include "libethash-cl/cl.hpp"
+
 #include <functional>
-#include <libethereum/AccountDiff.h>
 #include <libdevcore/RangeMask.h>
 #include <libdevcore/Log.h>
 #include <libdevcore/Common.h>
 #include <libdevcore/CommonData.h>
 #include <libdevcore/RLP.h>
-#include <libdevcrypto/TrieDB.h>
+#include <libdevcore/TransientDirectory.h>
 #include <libdevcore/CommonIO.h>
+#include <libdevcrypto/TrieDB.h>
 #include <libp2p/All.h>
-#include <libethereum/DownloadMan.h>
+#include <libethcore/Ethasher.h>
+#include <libethcore/ProofOfWork.h>
 #include <libethereum/All.h>
+#include <libethereum/AccountDiff.h>
+#include <libethereum/DownloadMan.h>
 #include <liblll/All.h>
 #include <libwhisper/WhisperPeer.h>
 #include <libwhisper/WhisperHost.h>
@@ -101,6 +108,22 @@ int main()
 #else
 int main()
 {
+	std::vector<cl::Platform> platforms;
+	cl::Platform::get(&platforms);
+	if (platforms.empty())
+	{
+		cdebug << "No OpenCL platforms found.";
+		return false;
+	}
+
+	EthashCL ecl;
+	BlockInfo genesis = CanonBlockChain::genesis();
+	TransientDirectory td;
+	std::pair<MineInfo, Ethash::Proof> r;
+	while (!r.first.completed)
+		r = ecl.mine(genesis, 1000);
+	EthashCL::assignResult(r.second, genesis);
+	assert(EthashCPU::verify(genesis));
 	return 0;
 }
 #endif

libdevcore/CommonIO.cpp

@@ -58,7 +58,7 @@ string dev::memDump(bytes const& _bytes, unsigned _width, bool _html)
 }
 
 // Don't forget to delete[] later.
-bytesRef dev::contentsNew(std::string const& _file)
+bytesRef dev::contentsNew(std::string const& _file, bytesRef _dest)
 {
 	std::ifstream is(_file, std::ifstream::binary);
 	if (!is)
@@ -68,8 +68,10 @@ bytesRef dev::contentsNew(std::string const& _file)
 	streamoff length = is.tellg();
 	if (length == 0) // return early, MSVC does not like reading 0 bytes
 		return bytesRef();
+	if (!_dest.empty() && _dest.size() != (unsigned)length)
+		return bytesRef();
 	is.seekg (0, is.beg);
-	bytesRef ret(new byte[length], length);
+	bytesRef ret = _dest.empty() ? bytesRef(new byte[length], length) : _dest;
 	is.read((char*)ret.data(), length);
 	is.close();
 	return ret;

libdevcore/CommonIO.h

@@ -46,7 +46,7 @@ namespace dev
 bytes contents(std::string const& _file);
 std::string contentsString(std::string const& _file);
 /// Retrieve and returns the allocated contents of the given file. If the file doesn't exist or isn't readable, returns nullptr. Don't forget to delete [] when finished.
-bytesRef contentsNew(std::string const& _file);
+bytesRef contentsNew(std::string const& _file, bytesRef _dest = bytesRef());
 
 /// Write the given binary data into the given file, replacing the file if it pre-exists.
 void writeFile(std::string const& _file, bytesConstRef _data);

libdevcore/RangeMask.h

@@ -25,6 +25,7 @@
 #include <utility>
 #include <vector>
 #include <iostream>
+#include <assert.h>
 
 namespace dev
 {

libtestutils/TransientDirectory.cpp → libdevcore/TransientDirectory.cpp

@@ -20,11 +20,11 @@
  */
 
 #include <boost/filesystem.hpp>
-#include <libdevcore/Exceptions.h>
+#include "Exceptions.h"
 #include "TransientDirectory.h"
+#include "CommonIO.h"
 using namespace std;
 using namespace dev;
-using namespace dev::test;
 
 TransientDirectory::TransientDirectory():
 	TransientDirectory((boost::filesystem::temp_directory_path() / "eth_transient" / toString(FixedHash<4>::random())).string())

libtestutils/TransientDirectory.h → libdevcore/TransientDirectory.h

@@ -22,12 +22,9 @@
 #pragma once
 
 #include <string>
-#include "Common.h"
 
 namespace dev
 {
-namespace test
-{
 
 /**
  * @brief temporary directory implementation
@@ -48,4 +45,3 @@ private:
 };
 
 }
-}

libethash/util.h

@@ -29,7 +29,7 @@ extern "C" {
 #ifdef _MSC_VER
 void debugf(const char *str, ...);
 #else
-#define debugf printf
+#define debugf(...) fprintf(stderr, __VA_ARGS__)
 #endif
 
 static inline uint32_t min_u32(uint32_t a, uint32_t b)

libethcore/Ethasher.cpp

@@ -123,6 +123,44 @@ ethash_params Ethasher::params(BlockInfo const& _header)
 	return params((unsigned)_header.number);
 }
 
+void Ethasher::readFull(BlockInfo const& _header, void* _dest)
+{
+	if (!m_fulls.count(_header.seedHash()))
+	{
+		// @memoryleak @bug place it on a pile for deletion - perhaps use shared_ptr.
+/*		if (!m_fulls.empty())
+		{
+			delete [] m_fulls.begin()->second.data();
+			m_fulls.erase(m_fulls.begin());
+		}*/
+
+		try {
+			boost::filesystem::create_directories(getDataDir("ethash"));
+		} catch (...) {}
+
+		auto info = rlpList(c_ethashRevision, _header.seedHash());
+		std::string oldMemoFile = getDataDir("ethash") + "/full";
+		std::string memoFile = getDataDir("ethash") + "/full-R" + toString(c_ethashRevision) + "-" + toHex(_header.seedHash().ref().cropped(0, 8));
+		if (boost::filesystem::exists(oldMemoFile) && contents(oldMemoFile + ".info") == info)
+		{
+			// memofile valid - rename.
+			boost::filesystem::rename(oldMemoFile, memoFile);
+		}
+
+		IGNORE_EXCEPTIONS(boost::filesystem::remove(oldMemoFile));
+		IGNORE_EXCEPTIONS(boost::filesystem::remove(oldMemoFile + ".info"));
+
+		ethash_params p = params((unsigned)_header.number);
+		bytesRef r = contentsNew(memoFile, bytesRef((byte*)_dest, p.full_size));
+		if (!r)
+		{
+			auto c = light(_header);
+			ethash_prep_full(_dest, &p, c);
+			writeFile(memoFile, bytesConstRef((byte*)_dest, p.full_size));
+		}
+	}
+}
+
 ethash_params Ethasher::params(unsigned _n)
 {
 	ethash_params p;

libethcore/Ethasher.h

@@ -56,6 +56,8 @@ public:
 	static ethash_params params(BlockInfo const& _header);
 	static ethash_params params(unsigned _n);
 
+	void readFull(BlockInfo const& _header, void* _dest);
+
 	struct Result
 	{
 		h256 value;

libethcore/ProofOfWork.cpp

@@ -33,32 +33,6 @@
 #include <libdevcore/Common.h>
 #if ETH_ETHASHCL
 #include <libethash-cl/ethash_cl_miner.h>
-#define ETHASH_REVISION REVISION
-#define ETHASH_DATASET_BYTES_INIT DATASET_BYTES_INIT
-#define ETHASH_DATASET_BYTES_GROWTH DATASET_BYTES_GROWTH
-#define ETHASH_CACHE_BYTES_INIT CACHE_BYTES_INIT
-#define ETHASH_CACHE_BYTES_GROWTH CACHE_BYTES_GROWTH
-#define ETHASH_DAGSIZE_BYTES_INIT DAGSIZE_BYTES_INIT
-#define ETHASH_DAG_GROWTH DAG_GROWTH
-#define ETHASH_EPOCH_LENGTH EPOCH_LENGTH
-#define ETHASH_MIX_BYTES MIX_BYTES
-#define ETHASH_HASH_BYTES HASH_BYTES
-#define ETHASH_DATASET_PARENTS DATASET_PARENTS
-#define ETHASH_CACHE_ROUNDS CACHE_ROUNDS
-#define ETHASH_ACCESSES ACCESSES
-#undef REVISION
-#undef DATASET_BYTES_INIT
-#undef DATASET_BYTES_GROWTH
-#undef CACHE_BYTES_INIT
-#undef CACHE_BYTES_GROWTH
-#undef DAGSIZE_BYTES_INIT
-#undef DAG_GROWTH
-#undef EPOCH_LENGTH
-#undef MIX_BYTES
-#undef HASH_BYTES
-#undef DATASET_PARENTS
-#undef CACHE_ROUNDS
-#undef ACCESSES
 #endif
 #include "BlockInfo.h"
 #include "Ethasher.h"
@@ -131,16 +105,16 @@ std::pair<MineInfo, EthashCPU::Proof> EthashCPU::mine(BlockInfo const& _header,
 #if ETH_ETHASHCL
 
 /*
-struct ethash_cl_search_hook
-{
-	// reports progress, return true to abort
-	virtual bool found(uint64_t const* nonces, uint32_t count) = 0;
-	virtual bool searched(uint64_t start_nonce, uint32_t count) = 0;
-};
-
 class ethash_cl_miner
 {
 public:
+	struct search_hook
+	{
+		// reports progress, return true to abort
+		virtual bool found(uint64_t const* nonces, uint32_t count) = 0;
+		virtual bool searched(uint64_t start_nonce, uint32_t count) = 0;
+	};
+
 	ethash_cl_miner();
 
 	bool init(ethash_params const& params, const uint8_t seed[32], unsigned workgroup_size = 64);
@@ -150,58 +124,63 @@ public:
 };
 */
 
-struct EthashCLHook: public ethash_cl_search_hook
+struct EthashCLHook: public ethash_cl_miner::search_hook
 {
-	virtual bool found(uint64_t const* _nonces, uint32_t _count)
+	void abort()
+	{
+		if (m_aborted)
+			return;
+		m_abort = true;
+		for (unsigned timeout = 0; timeout < 100 && !m_aborted; ++timeout)
+			std::this_thread::sleep_for(chrono::milliseconds(30));
+		if (!m_aborted)
+			cwarn << "Couldn't abort. Abandoning OpenCL process.";
+		m_aborted = m_abort = false;
+		m_found.clear();
+	}
+
+	vector<Nonce> fetchFound() { vector<Nonce> ret; Guard l(x_all); std::swap(ret, m_found); return ret; }
+	uint64_t fetchTotal() { Guard l(x_all); auto ret = m_total; m_total = 0; return ret; }
+
+protected:
+	virtual bool found(uint64_t const* _nonces, uint32_t _count) override
 	{
 		Guard l(x_all);
 		for (unsigned i = 0; i < _count; ++i)
-			found.push_back((Nonce)(u64)_nonces[i]);
-		if (abort)
-		{
-			aborted = true;
-			return true;
-		}
-		return false;
+			m_found.push_back((Nonce)(u64)_nonces[i]);
+		m_aborted = true;
+		return true;
 	}
 
-	virtual bool searched(uint64_t _startNonce, uint32_t _count)
+	virtual bool searched(uint64_t _startNonce, uint32_t _count) override
 	{
 		Guard l(x_all);
-		total += _count;
-		last = _startNonce + _count;
-		if (abort)
+		m_total += _count;
+		m_last = _startNonce + _count;
+		if (m_abort)
 		{
-			aborted = true;
+			m_aborted = true;
 			return true;
 		}
 		return false;
 	}
 
-	vector<Nonce> fetchFound() { vector<Nonce> ret; Guard l(x_all); std::swap(ret, found); return ret; }
-	uint64_t fetchTotal() { Guard l(x_all); auto ret = total; total = 0; return ret; }
-
+private:
 	Mutex x_all;
-	vector<Nonce> found;
-	uint64_t total;
-	uint64_t last;
-	bool abort = false;
-	bool aborted = false;
+	vector<Nonce> m_found;
+	uint64_t m_total;
+	uint64_t m_last;
+	bool m_abort = false;
+	bool m_aborted = true;
 };
 
 EthashCL::EthashCL():
-	m_miner(new ethash_cl_miner),
 	m_hook(new EthashCLHook)
 {
 }
 
 EthashCL::~EthashCL()
 {
-	m_hook->abort = true;
-	for (unsigned timeout = 0; timeout < 100 && !m_hook->aborted; ++timeout)
-		std::this_thread::sleep_for(chrono::milliseconds(30));
-	if (!m_hook->aborted)
-		cwarn << "Couldn't abort. Abandoning OpenCL process.";
 }
 
 bool EthashCL::verify(BlockInfo const& _header)
@@ -211,13 +190,16 @@ bool EthashCL::verify(BlockInfo const& _header)
 
 std::pair<MineInfo, Ethash::Proof> EthashCL::mine(BlockInfo const& _header, unsigned _msTimeout, bool, bool)
 {
-	if (m_lastHeader.seedHash() != _header.seedHash())
+	if (!m_lastHeader || m_lastHeader.seedHash() != _header.seedHash())
 	{
-		m_miner->init(Ethasher::params(_header), _header.seedHash().data());
-		// TODO: reinit probably won't work when seed changes.
+		if (m_miner)
+			m_hook->abort();
+		m_miner.reset(new ethash_cl_miner);
+		m_miner->init(Ethasher::params(_header), [&](void* d){ Ethasher::get()->readFull(_header, d); });
 	}
 	if (m_lastHeader != _header)
 	{
+		m_hook->abort();
 		static std::random_device s_eng;
 		uint64_t tryNonce = (uint64_t)(u64)(m_last = Nonce::random(s_eng));
 		m_miner->search(_header.headerHash(WithoutNonce).data(), tryNonce, *m_hook);
@@ -228,10 +210,11 @@ std::pair<MineInfo, Ethash::Proof> EthashCL::mine(BlockInfo const& _header, unsi
 	auto found = m_hook->fetchFound();
 	if (!found.empty())
 	{
-		h256 mixHash; // ?????
-		return std::make_pair(MineInfo{0.0, 1e99, 0, true}, EthashCL::Proof((Nonce)(u64)found[0], mixHash));
+		Nonce n = (Nonce)(u64)found[0];
+		auto result = Ethasher::eval(_header, n);
+		return std::make_pair(MineInfo(true), EthashCL::Proof{n, result.mixHash});
 	}
-	return std::make_pair(MineInfo{0.0, 1e99, 0, false}, EthashCL::Proof());
+	return std::make_pair(MineInfo(false), EthashCL::Proof());
 }
 
 #endif

libethcore/ProofOfWork.h

@@ -42,6 +42,8 @@ namespace eth
 
 struct MineInfo
 {
+	MineInfo() = default;
+	MineInfo(bool _completed): completed(_completed) {}
 	void combine(MineInfo const& _m) { requirement = std::max(requirement, _m.requirement); best = std::min(best, _m.best); hashes += _m.hashes; completed = completed || _m.completed; }
 	double requirement = 0;
 	double best = 1e99;
@@ -67,6 +69,8 @@ protected:
 };
 
 #if ETH_ETHASHCL
+class EthashCLHook;
+
 class EthashCL
 {
 public:
@@ -88,7 +92,7 @@ protected:
 	BlockInfo m_lastHeader;
 	Nonce m_mined;
 	std::unique_ptr<ethash_cl_miner> m_miner;
-	std::unique_ptr<ethash_cl_search_hook> m_hook;
+	std::unique_ptr<EthashCLHook> m_hook;
 };
 
 using Ethash = EthashCL;

libethereum/CMakeLists.txt

@@ -25,19 +25,18 @@ else()
 	add_library(${EXECUTABLE} SHARED ${SRC_LIST} ${HEADERS})
 endif()
 
-target_link_libraries(${EXECUTABLE} ${LEVELDB_LIBRARIES})
-target_link_libraries(${EXECUTABLE} ${Boost_REGEX_LIBRARIES})
-
 target_link_libraries(${EXECUTABLE} evm)
 target_link_libraries(${EXECUTABLE} lll)
 target_link_libraries(${EXECUTABLE} whisper)
 target_link_libraries(${EXECUTABLE} p2p)
 target_link_libraries(${EXECUTABLE} devcrypto)
 target_link_libraries(${EXECUTABLE} ethcore)
+target_link_libraries(${EXECUTABLE} ${LEVELDB_LIBRARIES})
+target_link_libraries(${EXECUTABLE} ${Boost_REGEX_LIBRARIES})
 target_link_libraries(${EXECUTABLE} secp256k1)
-
 if (ETHASHCL)
 	target_link_libraries(${EXECUTABLE} ethash-cl)
+	target_link_libraries(${EXECUTABLE} OpenCL)
 endif ()
 
 if (CMAKE_COMPILER_IS_MINGW)

libethereum/Client.cpp

@@ -334,6 +334,7 @@ void Client::setMiningThreads(unsigned _threads)
 {
 	stopMining();
 #if ETH_ETHASHCL
+	(void)_threads;
 	unsigned t = 1;
 #else
 	auto t = _threads ? _threads : thread::hardware_concurrency();

libethereum/ClientBase.cpp

@@ -75,7 +75,7 @@ ExecutionResult ClientBase::call(Secret _secret, u256 _value, Address _dest, byt
 		u256 n = temp.transactionsFrom(a);
 		Transaction t(_value, _gasPrice, _gas, _dest, _data, n, _secret);
 		if (_ff == FudgeFactor::Lenient)
-			temp.addBalance(a, (u256)(t.gasRequired() * t.gasPrice() + t.value()));
+			temp.addBalance(a, (u256)(t.gas() * t.gasPrice() + t.value()));
 		ret = temp.execute(bc().lastHashes(), t, Permanence::Reverted);
 	}
 	catch (...)

libtestutils/BlockChainLoader.h

@@ -22,9 +22,9 @@
 #pragma once
 #include <string>
 #include <json/json.h>
+#include <libdevcore/TransientDirectory.h>
 #include <libethereum/BlockChain.h>
 #include <libethereum/State.h>
-#include "TransientDirectory.h"
 
 namespace dev
 {

libtestutils/StateLoader.h

@@ -22,8 +22,8 @@
 #pragma once
 
 #include <json/json.h>
+#include <libdevcore/TransientDirectory.h>
 #include <libethereum/State.h>
-#include "TransientDirectory.h"
 
 namespace dev
 {

libweb3jsonrpc/WebThreeStubServerBase.cpp

@@ -449,64 +449,55 @@ static TransactionSkeleton toTransaction(Json::Value const& _json)
 
 string WebThreeStubServerBase::eth_sendTransaction(Json::Value const& _json)
 {
-	TransactionSkeleton t;
-	
 	try
 	{
-		t = toTransaction(_json);
+		string ret;
+		TransactionSkeleton t = toTransaction(_json);
+	
+		if (!t.from)
+			t.from = m_accounts->getDefaultTransactAccount();
+		if (t.creation)
+			ret = toJS(right160(sha3(rlpList(t.from, client()->countAt(t.from)))));;
+		if (!t.gasPrice)
+			t.gasPrice = 10 * dev::eth::szabo;		// TODO: should be determined by user somehow.
+		if (!t.gas)
+			t.gas = min<u256>(client()->gasLimitRemaining(), client()->balanceAt(t.from) / t.gasPrice);
+
+		if (m_accounts->isRealAccount(t.from))
+			authenticate(t, false);
+		else if (m_accounts->isProxyAccount(t.from))
+			authenticate(t, true);
+	
+		return ret;
 	}
 	catch (...)
 	{
 		BOOST_THROW_EXCEPTION(JsonRpcException(Errors::ERROR_RPC_INVALID_PARAMS));
 	}
-	
-	string ret;
-	if (!t.from)
-		t.from = m_accounts->getDefaultTransactAccount();
-	if (t.creation)
-		ret = toJS(right160(sha3(rlpList(t.from, client()->countAt(t.from)))));;
-	if (!t.gasPrice)
-		t.gasPrice = 10 * dev::eth::szabo;		// TODO: should be determined by user somehow.
-	if (!t.gas)
-		t.gas = min<u256>(client()->gasLimitRemaining(), client()->balanceAt(t.from) / t.gasPrice);
-	
-	if (m_accounts->isRealAccount(t.from))
-		authenticate(t, false);
-	else if (m_accounts->isProxyAccount(t.from))
-		authenticate(t, true);
-	
-	return ret;
 }
 
 
 string WebThreeStubServerBase::eth_call(Json::Value const& _json, string const& _blockNumber)
 {
-	TransactionSkeleton t;
-	int number;
-	
 	try
 	{
-		t = toTransaction(_json);
-		number = jsToBlockNumber(_blockNumber);
+		TransactionSkeleton t = toTransaction(_json);
+		if (!t.from)
+			t.from = m_accounts->getDefaultTransactAccount();
+	//	if (!m_accounts->isRealAccount(t.from))
+	//		return ret;
+		if (!t.gasPrice)
+			t.gasPrice = 10 * dev::eth::szabo;
+		if (!t.gas)
+			t.gas = client()->gasLimitRemaining();
+
+		return toJS(client()->call(m_accounts->secretKey(t.from), t.value, t.to, t.data, t.gas, t.gasPrice, jsToBlockNumber(_blockNumber), FudgeFactor::Lenient).output);
 	}
 	catch (...)
 	{
 		BOOST_THROW_EXCEPTION(JsonRpcException(Errors::ERROR_RPC_INVALID_PARAMS));
 	}
 	
-	string ret;
-	if (!t.from)
-		t.from = m_accounts->getDefaultTransactAccount();
-//	if (!m_accounts->isRealAccount(t.from))
-//		return ret;
-	if (!t.gasPrice)
-		t.gasPrice = 10 * dev::eth::szabo;
-	if (!t.gas)
-		t.gas = client()->gasLimitRemaining();
-
-	ret = toJS(client()->call(m_accounts->secretKey(t.from), t.value, t.to, t.data, t.gas, t.gasPrice, number).output);
-	
-	return ret;
 }
 
 bool WebThreeStubServerBase::eth_flush()

test/TestUtils.cpp

@@ -22,6 +22,7 @@
 #include <thread>
 #include <boost/test/unit_test.hpp>
 #include <boost/filesystem.hpp>
+#include <libtestutils/Common.h>
 #include <libtestutils/BlockChainLoader.h>
 #include <libtestutils/FixedClient.h>
 #include "TestUtils.h"

test/blockchain.cpp

@@ -22,7 +22,7 @@
 
 #include <boost/filesystem.hpp>
 #include <libdevcrypto/FileSystem.h>
-#include <libtestutils/TransientDirectory.h>
+#include <libdevcore/TransientDirectory.h>
 #include <libethereum/CanonBlockChain.h>
 #include "TestHelper.h"