github/AMD-SHARK-Studio
1
1
Fork 0
mirror of https://github.com/nod-ai/AMD-SHARK-Studio.git synced 2026-04-03 03:00:17 -04:00
Code Issues Packages Projects Releases Wiki Activity

Remove old dshark

Browse Source
This commit is contained in:
anush elangovan
2022-04-13 13:15:25 -07:00
parent c586564356
commit e3befc0e32
25 changed files with 0 additions and 1597 deletions
Download Patch File Download Diff File Expand all files Collapse all files

17
CMakeLists.txt
View File

@@ -1,17 +0,0 @@
cmake_minimum_required(VERSION 3.13.4)
project(iree-wrapper VERSION 1.0 LANGUAGES CXX C)
set(CMAKE_C_STANDARD 11)
set(CMAKE_CXX_STANDARD 17)
set_property(GLOBAL PROPERTY USE_FOLDERS ON)
list(APPEND CMAKE_MODULE_PATH "${PROJECT_BINARY_DIR}/lib/cmake/mlir")
add_library(asio INTERFACE)
target_compile_options(asio INTERFACE ASIO_STANDALONE)
target_include_directories(asio INTERFACE thirdparty/asio/asio/include)
target_link_libraries(asio INTERFACE pthread)
add_subdirectory(thirdparty/iree EXCLUDE_FROM_ALL)
add_subdirectory(dSHARK)

7
dSHARK/CMakeLists.txt
View File

@@ -1,7 +0,0 @@
project(dSHARK CXX)
cmake_minimum_required(VERSION 3.13.4)
add_subdirectory(dshark_network_lib)
add_subdirectory(dshark_run_module)
add_subdirectory(dshark_client)
add_subdirectory(dshark_server)

9
dSHARK/dshark_client/CMakeLists.txt
View File

@@ -1,9 +0,0 @@
cmake_minimum_required(VERSION 3.13.4)
project(dshark_client)
add_executable(dshark_client dshark_client.cpp)
target_include_directories(dshark_client PRIVATE ../dshark_network_lib)
target_link_directories(dshark_client PRIVATE ../dshark_network_lib)
target_link_libraries(dshark_client dshark_network_lib pthread)

123
dSHARK/dshark_client/dshark_client.cpp
View File

@@ -1,123 +0,0 @@
#include <iostream>
#include "dshark_common.h"
#include "dshark_message.h"
#include "dshark_client.h"
#include "dshark_server.h"
#include <asio.hpp>
#include <fstream>
#include <iterator>
enum class dSHARKMessageType : uint32_t
{
EvaluateBinary,
ServerAccept,
ServerDeny,
ServerPing,
MessageAll,
ServerMessage,
};
class CustomClient : public dshark::client_interface<dSHARKMessageType>
{
public:
void PingServer()
{
dshark::message<dSHARKMessageType> msg;
msg.header.id = dSHARKMessageType::ServerPing;
// Caution with this...
std::chrono::system_clock::time_point timeNow = std::chrono::system_clock::now();
msg << timeNow;
Send(msg);
}
void MessageAll()
{
dshark::message<dSHARKMessageType> msg;
msg.header.id = dSHARKMessageType::MessageAll;
Send(msg);
}
void EvaluateBinary(std::string filepath)
{
std::cout << "Loading Binary" << std::endl;
std::ifstream input(filepath, std::ios::binary);
std::vector<unsigned char> buffer(std::istreambuf_iterator<char>(input), {});
dshark::message<dSHARKMessageType> msg;
msg.header.id = dSHARKMessageType::EvaluateBinary;
for (int i = 0; i < buffer.size(); i++)
{
msg << buffer[i];
}
Send(msg);
}
};
int main(int argc, char* argv[])
{
CustomClient c;
c.Connect("127.0.0.1", 60000);
bool bQuit = false;
std::cout << "Sending File " << argv[1] << std::endl;
c.EvaluateBinary((std::string)argv[1]);
while (!bQuit)
{
if (c.IsConnected())
{
if (!c.Incoming().empty())
{
auto msg = c.Incoming().pop_front().msg;
switch (msg.header.id)
{
case dSHARKMessageType::ServerAccept:
{
// Server has responded to a ping request
std::cout << "Server Accepted Connection\n";
}
break;
case dSHARKMessageType::EvaluateBinary:
{
std::cout << "Binary File Sent" << std::endl;
}
break;
case dSHARKMessageType::ServerPing:
{
// Server has responded to a ping request
std::chrono::system_clock::time_point timeNow = std::chrono::system_clock::now();
std::chrono::system_clock::time_point timeThen;
msg >> timeThen;
std::cout << "Ping: " << std::chrono::duration<double>(timeNow - timeThen).count() << "\n";
}
break;
case dSHARKMessageType::ServerMessage:
{
// Server has responded to a ping request
uint32_t clientID;
msg >> clientID;
std::cout << "Hello from [" << clientID << "]\n";
}
break;
}
}
}
else
{
std::cout << "Server Down\n";
bQuit = true;
}
}
return 0;
}

5
dSHARK/dshark_network_lib/CMakeLists.txt
View File

@@ -1,5 +0,0 @@
cmake_minimum_required(VERSION 3.13.4)
add_library(dshark_network_lib dshark_common.h dshark_queue.h dshark_message.h dshark_connection.h dshark_client.h dshark_server.h)
set_target_properties(dshark_network_lib PROPERTIES LINKER_LANGUAGE CXX)

96
dSHARK/dshark_network_lib/dshark_client.h
View File

@@ -1,96 +0,0 @@
#pragma once
#include "dshark_common.h"
#include "dshark_message.h"
#include "dshark_queue.h"
#include "dshark_connection.h"
namespace dshark {
template <typename T>
class client_interface
{
public:
client_interface() : m_socket(m_context)
{
}
virtual ~client_interface()
{
Disconnect();
}
public:
bool Connect(const std::string& host, const uint16_t port)
{
try
{
// Resolve hostname/ip-address into tangiable physical address
asio::ip::tcp::resolver resolver(m_context);
asio::ip::tcp::resolver::results_type endpoints = resolver.resolve(host, std::to_string(port));
// Create connection
m_connection = std::make_unique<connection<T>>(connection<T>::owner::client, m_context, asio::ip::tcp::socket(m_context), m_qMessagesIn);
// Tell the connection object to connect to server
m_connection->ConnectToServer(endpoints);
// Start Context Thread
thrContext = std::thread([this]() { m_context.run(); });
}
catch (std::exception& e)
{
std::cerr << "Client Exception: " << e.what() << "\n";
return false;
}
return true;
}
void Disconnect()
{
// If connection exists, and it's connected then...
if (IsConnected())
{
// ...disconnect from server gracefully
m_connection->Disconnect();
}
// Either way, we're also done with the asio context...
m_context.stop();
// ...and its thread
if (thrContext.joinable())
thrContext.join();
// Destroy the connection object
m_connection.release();
}
bool IsConnected()
{
if (m_connection) {
return m_connection->IsConnected();
}
else {
return false;
}
}
void Send(const message<T>& msg)
{
if (IsConnected())
m_connection->Send(msg);
}
// Retrieve queue of messages from server
dsqueue<owned_message<T>>& Incoming()
{
return m_qMessagesIn;
}
protected:
asio::io_context m_context;
std::thread thrContext;
asio::ip::tcp::socket m_socket;
std::unique_ptr<connection<T>> m_connection;
private:
dsqueue<owned_message<T>> m_qMessagesIn;
};
}

18
dSHARK/dshark_network_lib/dshark_common.h
View File

@@ -1,18 +0,0 @@
#pragma once
#include <memory>
#include <thread>
#include <mutex>
#include <deque>
#include <optional>
#include <vector>
#include <iostream>
#include <algorithm>
#include <cstdint>
#include <chrono>
#include <stdio.h>
#define ASIO_STANDALONE
#include <asio.hpp>
#include <asio/ts/buffer.hpp>
#include <asio/ts/internet.hpp>

291
dSHARK/dshark_network_lib/dshark_connection.h
View File

@@ -1,291 +0,0 @@
#pragma once
#include "dshark_common.h"
#include "dshark_message.h"
#include "dshark_queue.h"
namespace dshark {
template <typename T>
class connection : public std::enable_shared_from_this<connection<T>>
{
public:
// A connection is "owned" by either a server or a client, and its
// behaviour is slightly different bewteen the two.
enum class owner
{
server,
client
};
public:
// Constructor: Specify Owner, connect to context, transfer the socket
// Provide reference to incoming message queue
connection(owner parent, asio::io_context& asioContext, asio::ip::tcp::socket socket, dsqueue<owned_message<T>>& qIn)
: m_asioContext(asioContext), m_socket(std::move(socket)), m_qMessagesIn(qIn)
{
m_nOwnerType = parent;
}
virtual ~connection()
{}
// This ID is used system wide - its how clients will understand other clients
// exist across the whole system.
uint32_t GetID() const
{
return id;
}
public:
void ConnectToClient(uint32_t uid = 0)
{
if (m_nOwnerType == owner::server)
{
if (m_socket.is_open())
{
id = uid;
ReadHeader();
}
}
}
void ConnectToServer(const asio::ip::tcp::resolver::results_type& endpoints)
{
// Only clients can connect to servers
if (m_nOwnerType == owner::client)
{
// Request asio attempts to connect to an endpoint
asio::async_connect(m_socket, endpoints,
[this](std::error_code ec, asio::ip::tcp::endpoint endpoint)
{
if (!ec)
{
ReadHeader();
}
});
}
}
void Disconnect()
{
if (IsConnected())
asio::post(m_asioContext, [this]() { m_socket.close(); });
}
bool IsConnected() const
{
return m_socket.is_open();
}
// Prime the connection to wait for incoming messages
void StartListening()
{
}
public:
// ASYNC - Send a message, connections are one-to-one so no need to specifiy
// the target, for a client, the target is the server and vice versa
void Send(const message<T>& msg)
{
asio::post(m_asioContext,
[this, msg]()
{
// If the queue has a message in it, then we must
// assume that it is in the process of asynchronously being written.
// Either way add the message to the queue to be output. If no messages
// were available to be written, then start the process of writing the
// message at the front of the queue.
bool bWritingMessage = !m_qMessagesOut.empty();
m_qMessagesOut.push_back(msg);
if (!bWritingMessage)
{
WriteHeader();
}
});
}
private:
// ASYNC - Prime context to write a message header
void WriteHeader()
{
// If this function is called, we know the outgoing message queue must have
// at least one message to send. So allocate a transmission buffer to hold
// the message, and issue the work - asio, send these bytes
asio::async_write(m_socket, asio::buffer(&m_qMessagesOut.front().header, sizeof(message_header<T>)),
[this](std::error_code ec, std::size_t length)
{
// asio has now sent the bytes - if there was a problem
// an error would be available...
if (!ec)
{
// ... no error, so check if the message header just sent also
// has a message body...
if (m_qMessagesOut.front().body.size() > 0)
{
// ...it does, so issue the task to write the body bytes
WriteBody();
}
else
{
// ...it didnt, so we are done with this message. Remove it from
// the outgoing message queue
m_qMessagesOut.pop_front();
// If the queue is not empty, there are more messages to send, so
// make this happen by issuing the task to send the next header.
if (!m_qMessagesOut.empty())
{
WriteHeader();
}
}
}
else
{
// ...asio failed to write the message, we could analyse why but
// for now simply assume the connection has died by closing the
// socket. When a future attempt to write to this client fails due
// to the closed socket, it will be tidied up.
std::cout << "[" << id << "] Write Header Fail.\n";
m_socket.close();
}
});
}
// ASYNC - Prime context to write a message body
void WriteBody()
{
// If this function is called, a header has just been sent, and that header
// indicated a body existed for this message. Fill a transmission buffer
// with the body data, and send it!
asio::async_write(m_socket, asio::buffer(m_qMessagesOut.front().body.data(), m_qMessagesOut.front().body.size()),
[this](std::error_code ec, std::size_t length)
{
if (!ec)
{
// Sending was successful, so we are done with the message
// and remove it from the queue
m_qMessagesOut.pop_front();
// If the queue still has messages in it, then issue the task to
// send the next messages' header.
if (!m_qMessagesOut.empty())
{
WriteHeader();
}
}
else
{
// Sending failed, see WriteHeader() equivalent for description :P
std::cout << "[" << id << "] Write Body Fail.\n";
m_socket.close();
}
});
}
// ASYNC - Prime context ready to read a message header
void ReadHeader()
{
// If this function is called, we are expecting asio to wait until it receives
// enough bytes to form a header of a message. We know the headers are a fixed
// size, so allocate a transmission buffer large enough to store it. In fact,
// we will construct the message in a "temporary" message object as it's
// convenient to work with.
asio::async_read(m_socket, asio::buffer(&m_msgTemporaryIn.header, sizeof(message_header<T>)),
[this](std::error_code ec, std::size_t length)
{
if (!ec)
{
// A complete message header has been read, check if this message
// has a body to follow...
if (m_msgTemporaryIn.header.size > 0)
{
// ...it does, so allocate enough space in the messages' body
// vector, and issue asio with the task to read the body.
m_msgTemporaryIn.body.resize(m_msgTemporaryIn.header.size);
ReadBody();
}
else
{
// it doesn't, so add this bodyless message to the connections
// incoming message queue
AddToIncomingMessageQueue();
}
}
else
{
// Reading form the client went wrong, most likely a disconnect
// has occurred. Close the socket and let the system tidy it up later.
std::cout << "[" << id << "] Read Header Fail.\n";
m_socket.close();
}
});
}
// ASYNC - Prime context ready to read a message body
void ReadBody()
{
// If this function is called, a header has already been read, and that header
// request we read a body, The space for that body has already been allocated
// in the temporary message object, so just wait for the bytes to arrive...
asio::async_read(m_socket, asio::buffer(m_msgTemporaryIn.body.data(), m_msgTemporaryIn.body.size()),
[this](std::error_code ec, std::size_t length)
{
if (!ec)
{
// ...and they have! The message is now complete, so add
// the whole message to incoming queue
AddToIncomingMessageQueue();
}
else
{
// As above!
std::cout << "[" << id << "] Read Body Fail.\n";
m_socket.close();
}
});
}
// Once a full message is received, add it to the incoming queue
void AddToIncomingMessageQueue()
{
// Shove it in queue, converting it to an "owned message", by initialising
// with the a shared pointer from this connection object
if (m_nOwnerType == owner::server)
m_qMessagesIn.push_back({ this->shared_from_this(), m_msgTemporaryIn });
else
m_qMessagesIn.push_back({ nullptr, m_msgTemporaryIn });
// We must now prime the asio context to receive the next message. It
// wil just sit and wait for bytes to arrive, and the message construction
// process repeats itself. Clever huh?
ReadHeader();
}
protected:
// Each connection has a unique socket to a remote
asio::ip::tcp::socket m_socket;
// This context is shared with the whole asio instance
asio::io_context& m_asioContext;
// This queue holds all messages to be sent to the remote side
// of this connection
dsqueue<message<T>> m_qMessagesOut;
// This references the incoming queue of the parent object
dsqueue<owned_message<T>>& m_qMessagesIn;
// Incoming messages are constructed asynchronously, so we will
// store the part assembled message here, until it is ready
message<T> m_msgTemporaryIn;
// The "owner" decides how some of the connection behaves
owner m_nOwnerType = owner::server;
uint32_t id = 0;
};
}

90
dSHARK/dshark_network_lib/dshark_message.h
View File

@@ -1,90 +0,0 @@
#pragma once
#include "dshark_common.h"
namespace dshark
{
template <typename T>
struct message_header
{
T id{};
uint32_t size = 0;
};
template <typename T>
struct message
{
message_header<T> header{};
std::vector<uint8_t> body;
size_t size() const
{
return body.size();
}
friend std::ostream& operator << (std::ostream& os, const message<T>& msg)
{
os << "ID:" << int(msg.header.id) << " Size:" << msg.header.size;
return os;
}
template<typename DataType>
friend message<T>& operator << (message<T>& msg, const DataType& data)
{
// Check that the type of the data being pushed is trivially copyable
static_assert(std::is_standard_layout<DataType>::value, "Data is too complex to be pushed into vector");
// Cache current size of vector, as this will be the point we insert the data
size_t i = msg.body.size();
// Resize the vector by the size of the data being pushed
msg.body.resize(msg.body.size() + sizeof(DataType));
// Physically copy the data into the newly allocated vector space
std::memcpy(msg.body.data() + i, &data, sizeof(DataType));
// Recalculate the message size
msg.header.size = msg.size();
// Return the target message so it can be "chained"
return msg;
}
template<typename DataType>
friend message<T>& operator >> (message<T>& msg, DataType& data)
{
// Check that the type of the data being pushed is trivially copyable
static_assert(std::is_standard_layout<DataType>::value, "Data is too complex to be pulled from vector");
// Cache the location towards the end of the vector where the pulled data starts
size_t i = msg.body.size() - sizeof(DataType);
// Physically copy the data from the vector into the user variable
std::memcpy(&data, msg.body.data() + i, sizeof(DataType));
// Shrink the vector to remove read bytes, and reset end position
msg.body.resize(i);
// Recalculate the message size
msg.header.size = msg.size();
// Return the target message so it can be "chained"
return msg;
}
};
template <typename T>
class connection;
template <typename T>
struct owned_message
{
std::shared_ptr<connection<T>> remote = nullptr;
message<T> msg;
// Again, a friendly string maker
friend std::ostream& operator<<(std::ostream& os, const owned_message<T>& msg)
{
os << msg.msg;
return os;
}
};
}

99
dSHARK/dshark_network_lib/dshark_queue.h
View File

@@ -1,99 +0,0 @@
#pragma once
#include "dshark_common.h"
namespace dshark {
template<typename T>
class dsqueue
{
public:
dsqueue() = default;
dsqueue(const dsqueue<T>&) = delete;
virtual ~dsqueue() { clear(); }
public:
const T& front()
{
//std::scoped_lock lock(muxQueue);
return deqQueue.front();
}
const T& back()
{
//std::scoped_lock lock(muxQueue);
return deqQueue.back();
}
T pop_front()
{
//std::scoped_lock lock(muxQueue);
auto t = std::move(deqQueue.front());
deqQueue.pop_front();
return t;
}
// Removes and returns item from back of Queue
T pop_back()
{
//std::scoped_lock lock(muxQueue);
auto t = std::move(deqQueue.back());
deqQueue.pop_back();
return t;
}
// Adds an item to back of Queue
void push_back(const T& item)
{
//std::scoped_lock lock(muxQueue);
deqQueue.emplace_back(std::move(item));
//std::unique_lock<std::mutex> ul(muxBlocking);
cvBlocking.notify_one();
}
// Adds an item to front of Queue
void push_front(const T& item)
{
//std::scoped_lock lock(muxQueue);
deqQueue.emplace_front(std::move(item));
//std::unique_lock<std::mutex> ul(muxBlocking);
cvBlocking.notify_one();
}
// Returns true if Queue has no items
bool empty()
{
//std::scoped_lock lock(muxQueue);
return deqQueue.empty();
}
// Returns number of items in Queue
size_t count()
{
//std::scoped_lock lock(muxQueue);
return deqQueue.size();
}
//Clears Queue
void clear()
{
//std::scoped_lock lock(muxQueue);
return deqQueue.clear();
}
void wait()
{
while (empty())
{
std::unique_lock<std::mutex> ul(muxBlocking);
cvBlocking.wait(ul);
}
}
protected:
//std::mutex muxQueue;
std::deque<T> deqQueue;
std::condition_variable cvBlocking;
std::mutex muxBlocking;
};
}

237
dSHARK/dshark_network_lib/dshark_server.h
View File

@@ -1,237 +0,0 @@
#pragma once
#include "dshark_common.h"
#include "dshark_queue.h"
#include "dshark_message.h"
#include "dshark_connection.h"
namespace dshark {
template<typename T>
class server_interface
{
public:
// Create a server, ready to listen on specified port
server_interface(uint16_t port)
: m_asioAcceptor(m_asioContext, asio::ip::tcp::endpoint(asio::ip::tcp::v4(), port))
{
}
virtual ~server_interface()
{
// May as well try and tidy up
Stop();
}
// Starts the server!
bool Start()
{
try
{
// Issue a task to the asio context - This is important
// as it will prime the context with "work", and stop it
// from exiting immediately. Since this is a server, we
// want it primed ready to handle clients trying to
// connect.
WaitForClientConnection();
// Launch the asio context in its own thread
m_threadContext = std::thread([this]() { m_asioContext.run(); });
}
catch (std::exception& e)
{
// Something prohibited the server from listening
std::cerr << "[SERVER] Exception: " << e.what() << "\n";
return false;
}
std::cout << "[SERVER] Started!\n";
return true;
}
// Stops the server!
void Stop()
{
// Request the context to close
m_asioContext.stop();
// Tidy up the context thread
if (m_threadContext.joinable()) m_threadContext.join();
// Inform someone, anybody, if they care...
std::cout << "[SERVER] Stopped!\n";
}
// ASYNC - Instruct asio to wait for connection
void WaitForClientConnection()
{
// Prime context with an instruction to wait until a socket connects. This
// is the purpose of an "acceptor" object. It will provide a unique socket
// for each incoming connection attempt
m_asioAcceptor.async_accept(
[this](std::error_code ec, asio::ip::tcp::socket socket)
{
// Triggered by incoming connection request
if (!ec)
{
// Display some useful(?) information
std::cout << "[SERVER] New Connection: " << socket.remote_endpoint() << "\n";
// Create a new connection to handle this client
std::shared_ptr<connection<T>> newconn =
std::make_shared<connection<T>>(connection<T>::owner::server,
m_asioContext, std::move(socket), m_qMessagesIn);
// Give the user server a chance to deny connection
if (OnClientConnect(newconn))
{
// Connection allowed, so add to container of new connections
m_deqConnections.push_back(std::move(newconn));
// And very important! Issue a task to the connection's
// asio context to sit and wait for bytes to arrive!
m_deqConnections.back()->ConnectToClient(nIDCounter++);
std::cout << "[" << m_deqConnections.back()->GetID() << "] Connection Approved\n";
}
else
{
std::cout << "[-----] Connection Denied\n";
// Connection will go out of scope with no pending tasks, so will
// get destroyed automagically due to the wonder of smart pointers
}
}
else
{
// Error has occurred during acceptance
std::cout << "[SERVER] New Connection Error: " << ec.message() << "\n";
}
// Prime the asio context with more work - again simply wait for
// another connection...
WaitForClientConnection();
});
}
// Send a message to a specific client
void MessageClient(std::shared_ptr<connection<T>> client, const message<T>& msg)
{
// Check client is legitimate...
if (client && client->IsConnected())
{
// ...and post the message via the connection
client->Send(msg);
}
else
{
// If we cant communicate with client then we may as
// well remove the client - let the server know, it may
// be tracking it somehow
OnClientDisconnect(client);
// Off you go now, bye bye!
client.reset();
// Then physically remove it from the container
m_deqConnections.erase(
std::remove(m_deqConnections.begin(), m_deqConnections.end(), client), m_deqConnections.end());
}
}
// Send message to all clients
void MessageAllClients(const message<T>& msg, std::shared_ptr<connection<T>> pIgnoreClient = nullptr)
{
bool bInvalidClientExists = false;
// Iterate through all clients in container
for (auto& client : m_deqConnections)
{
// Check client is connected...
if (client && client->IsConnected())
{
// ..it is!
if (client != pIgnoreClient)
client->Send(msg);
}
else
{
// The client couldnt be contacted, so assume it has
// disconnected.
OnClientDisconnect(client);
client.reset();
// Set this flag to then remove dead clients from container
bInvalidClientExists = true;
}
}
// Remove dead clients, all in one go - this way, we dont invalidate the
// container as we iterated through it.
if (bInvalidClientExists)
m_deqConnections.erase(
std::remove(m_deqConnections.begin(), m_deqConnections.end(), nullptr), m_deqConnections.end());
}
// Force server to respond to incoming messages
void Update(size_t nMaxMessages = -1, bool bWait = false)
{
if (bWait) m_qMessagesIn.wait();
// Process as many messages as you can up to the value
// specified
size_t nMessageCount = 0;
while (nMessageCount < nMaxMessages && !m_qMessagesIn.empty())
{
// Grab the front message
auto msg = m_qMessagesIn.pop_front();
// Pass to message handler
OnMessage(msg.remote, msg.msg);
nMessageCount++;
}
}
protected:
// This server class should override thse functions to implement
// customised functionality
// Called when a client connects, you can veto the connection by returning false
virtual bool OnClientConnect(std::shared_ptr<connection<T>> client)
{
return false;
}
// Called when a client appears to have disconnected
virtual void OnClientDisconnect(std::shared_ptr<connection<T>> client)
{
}
// Called when a message arrives
virtual void OnMessage(std::shared_ptr<connection<T>> client, message<T>& msg)
{
}
protected:
// Thread Safe Queue for incoming message packets
dsqueue<owned_message<T>> m_qMessagesIn;
// Container of active validated connections
std::deque<std::shared_ptr<connection<T>>> m_deqConnections;
// Order of declaration is important - it is also the order of initialisation
asio::io_context m_asioContext;
std::thread m_threadContext;
// These things need an asio context
asio::ip::tcp::acceptor m_asioAcceptor; // Handles new incoming connection attempts...
// Clients will be identified in the "wider system" via an ID
uint32_t nIDCounter = 10000;
};
}

16
dSHARK/dshark_run_module/CMakeLists.txt
View File

@@ -1,16 +0,0 @@
cmake_minimum_required(VERSION 3.13.4)
add_library(run_module run_module.h run_module.c dshark_driver_module.c)
target_link_libraries(run_module iree_base_base
iree_hal_hal
iree_hal_cuda_registration_registration
iree_modules_hal_hal
iree_vm_vm
iree_vm_bytecode_module
#iree_runtime_runtime
iree_hal_local_loaders_system_library_loader
iree_hal_local_sync_driver
)
set_target_properties(run_module PROPERTIES LINKER_LANGUAGE C)

123
dSHARK/dshark_run_module/dshark_driver_module.c
View File

@@ -1,123 +0,0 @@
// Copyright 2021 The IREE Authors
//
// Licensed under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
#include <inttypes.h>
#include <stddef.h>
#include "iree/base/api.h"
#include "iree/base/internal/flags.h"
#include "iree/base/tracing.h"
#include "iree/hal/cuda/api.h"
#define IREE_HAL_CUDA_DRIVER_ID 0x43554441u // CUDA
// Force using CUDA streams until we support command buffer caching to avoid the
// overhead of graph creation.
IREE_FLAG(
bool, cuda_use_streams, true,
"Use CUDA streams for executing command buffers (instead of graphs).");
IREE_FLAG(bool, cuda_allow_inline_execution, false,
"Allow command buffers to execute inline against CUDA streams when "
"possible.");
IREE_FLAG(int32_t, cuda_default_index, 0, "Index of the default CUDA device.");
static iree_status_t iree_hal_cuda_driver_factory_enumerate(
void* self, const iree_hal_driver_info_t** out_driver_infos,
iree_host_size_t* out_driver_info_count) {
// NOTE: we could query supported cuda versions or featuresets here.
static const iree_hal_driver_info_t driver_infos[1] = {{
.driver_id = IREE_HAL_CUDA_DRIVER_ID,
.driver_name = iree_string_view_literal("cuda"),
.full_name = iree_string_view_literal("CUDA (dynamic)"),
}};
*out_driver_info_count = IREE_ARRAYSIZE(driver_infos);
*out_driver_infos = driver_infos;
return iree_ok_status();
}
static iree_status_t iree_hal_cuda_driver_factory_try_create0(
void* self, iree_hal_driver_id_t driver_id, iree_allocator_t allocator,
iree_hal_driver_t** out_driver) {
IREE_ASSERT_ARGUMENT(out_driver);
*out_driver = NULL;
if (driver_id != IREE_HAL_CUDA_DRIVER_ID) {
return iree_make_status(IREE_STATUS_UNAVAILABLE,
"no driver with ID %016" PRIu64
" is provided by this factory",
driver_id);
}
IREE_TRACE_ZONE_BEGIN(z0);
iree_hal_cuda_device_params_t default_params;
iree_hal_cuda_device_params_initialize(&default_params);
if (FLAG_cuda_use_streams) {
default_params.command_buffer_mode =
IREE_HAL_CUDA_COMMAND_BUFFER_MODE_STREAM;
}
default_params.allow_inline_execution = FLAG_cuda_allow_inline_execution;
iree_hal_cuda_driver_options_t driver_options;
iree_hal_cuda_driver_options_initialize(&driver_options);
driver_options.default_device_index = 0;
iree_string_view_t identifier = iree_make_cstring_view("cuda");
iree_status_t status = iree_hal_cuda_driver_create(
identifier, &default_params, &driver_options, allocator, out_driver);
IREE_TRACE_ZONE_END(z0);
return status;
}
static iree_status_t iree_hal_cuda_driver_factory_try_create1(
void* self, iree_hal_driver_id_t driver_id, iree_allocator_t allocator,
iree_hal_driver_t** out_driver) {
IREE_ASSERT_ARGUMENT(out_driver);
*out_driver = NULL;
if (driver_id != IREE_HAL_CUDA_DRIVER_ID) {
return iree_make_status(IREE_STATUS_UNAVAILABLE,
"no driver with ID %016" PRIu64
" is provided by this factory",
driver_id);
}
IREE_TRACE_ZONE_BEGIN(z0);
iree_hal_cuda_device_params_t default_params;
iree_hal_cuda_device_params_initialize(&default_params);
if (FLAG_cuda_use_streams) {
default_params.command_buffer_mode =
IREE_HAL_CUDA_COMMAND_BUFFER_MODE_STREAM;
}
default_params.allow_inline_execution = FLAG_cuda_allow_inline_execution;
iree_hal_cuda_driver_options_t driver_options;
iree_hal_cuda_driver_options_initialize(&driver_options);
driver_options.default_device_index = 1;
iree_string_view_t identifier = iree_make_cstring_view("cuda");
iree_status_t status = iree_hal_cuda_driver_create(
identifier, &default_params, &driver_options, allocator, out_driver);
IREE_TRACE_ZONE_END(z0);
return status;
}
IREE_API_EXPORT iree_status_t
iree_hal_cuda_driver_module_register(iree_hal_driver_registry_t* registry, int index) {
if(index == 1){
static const iree_hal_driver_factory_t factory = {
.self = NULL,
.enumerate = iree_hal_cuda_driver_factory_enumerate,
.try_create = iree_hal_cuda_driver_factory_try_create1,
};
return iree_hal_driver_registry_register_factory(registry, &factory);
} else {
static const iree_hal_driver_factory_t factory = {
.self = NULL,
.enumerate = iree_hal_cuda_driver_factory_enumerate,
.try_create = iree_hal_cuda_driver_factory_try_create0,
};
return iree_hal_driver_registry_register_factory(registry, &factory);
}
}

228
dSHARK/dshark_run_module/run_module.c
View File

@@ -1,228 +0,0 @@
// Copyright 2021 The IREE Authors
//
// Licensed under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
// A example of setting up the HAL module to run simple pointwise array
// multiplication with the device implemented by different backends via
//_sample_driver().
//
// NOTE: this file does not properly handle error cases and will leak on
// failure. Applications that are just going to exit()/abort() on failure can
// probably get away with the same thing but really should prefer not to.
#include <stdio.h>
#include <string.h>
#include "iree/base/api.h"
#include "iree/hal/api.h"
#include "iree/modules/hal/module.h"
#include "iree/vm/api.h"
#include "iree/vm/bytecode_module.h"
#include "iree/base/internal/flags.h"
#include "dshark_driver_module.c"
#include "iree/base/tracing.h"
#include "iree/hal/cuda/api.h"
//#include "iree/hal/cuda/cuda_driver.c"
#include "iree/hal/cuda/cuda_device.h"
//#include "iree/hal/cuda/cuda_device.c"
#include "iree/base/internal/file_io.h"
//iree_status_t create_sample_device(iree_allocator_t host_allocator,
/// iree_hal_device_t** out_device, int index) {
// Only register the CUDA HAL driver.
// IREE_RETURN_IF_ERROR(
// iree_hal_cuda_driver_module_register(iree_hal_driver_registry_default(), index));
// Create the HAL driver from the name.
// iree_hal_driver_t* driver = NULL;
// iree_string_view_t identifier = iree_make_cstring_view("cuda");
// iree_status_t status = iree_hal_driver_registry_try_create_by_name(
// iree_hal_driver_registry_default(), identifier, host_allocator, &driver);
// Create the default device (primary GPU).
/// if (iree_status_is_ok(status)) {
//status = iree_hal_driver_create_default_device(driver, host_allocator,
// out_device);
/// CreateDevice("cuda", &out_device);
// }
// iree_hal_driver_release(driver);
// return iree_ok_status();
//}
iree_status_t CreateDevice(const char* driver_name, iree_hal_device_t** out_device) {
iree_hal_driver_t* driver = NULL;
IREE_RETURN_IF_ERROR(iree_hal_driver_registry_try_create_by_name(
iree_hal_driver_registry_default(),
iree_make_cstring_view(driver_name),
iree_allocator_system(), &driver),
"creating driver '%s'", driver_name);
IREE_RETURN_IF_ERROR(iree_hal_driver_create_default_device(
driver, iree_allocator_system(), out_device),
"creating default device for driver '%s'", driver_name);
iree_hal_driver_release(driver);
return iree_ok_status();
}
//const iree_const_byte_span_t load_bytecode_module_data() {
// const struct iree_file_toc_t* module_file_toc =
// simple_embedding_test_bytecode_module_cuda_c_create();
// return iree_make_const_byte_span(module_file_toc->data,
// module_file_toc->size);
//}
iree_status_t Run(char* module_file, int index) {
// TODO(benvanik): move to instance-based registration.
IREE_RETURN_IF_ERROR(iree_hal_module_register_types());
iree_vm_instance_t* instance = NULL;
IREE_RETURN_IF_ERROR(
iree_vm_instance_create(iree_allocator_system(), &instance));
//create module here
iree_file_contents_t* module_data = NULL;
IREE_RETURN_IF_ERROR(iree_file_read_contents(module_file, iree_allocator_system(), &module_data));
iree_vm_module_t* input_module = NULL;
IREE_RETURN_IF_ERROR(iree_vm_bytecode_module_create(
module_data->const_buffer, iree_file_contents_deallocator(module_data),
iree_allocator_system(), &input_module));
IREE_RETURN_IF_ERROR(
iree_hal_cuda_driver_module_register(iree_hal_driver_registry_default(), index));
iree_hal_device_t* device = NULL;
IREE_RETURN_IF_ERROR(CreateDevice("cuda", &device),
"create device");
//device = (iree_hal_device_t)device;
iree_vm_module_t* hal_module = NULL;
IREE_RETURN_IF_ERROR(
iree_hal_module_create(device, iree_allocator_system(), &hal_module));
// Load bytecode module from the embedded data.
//const iree_const_byte_span_t module_data = load_bytecode_module_data();
//iree_vm_module_t* bytecode_module = NULL;
//IREE_RETURN_IF_ERROR(iree_vm_bytecode_module_create(
// module_data, iree_allocator_null(), iree_allocator_system(),
// &bytecode_module));
// Allocate a context that will hold the module state across invocations.
iree_vm_context_t* context = NULL;
iree_vm_module_t* modules[] = {hal_module, input_module};
IREE_RETURN_IF_ERROR(iree_vm_context_create_with_modules(
instance, IREE_VM_CONTEXT_FLAG_NONE, &modules[0], IREE_ARRAYSIZE(modules),
iree_allocator_system(), &context));
iree_vm_module_release(hal_module);
iree_vm_module_release(input_module);
// Lookup the entry point function.
// Note that we use the synchronous variant which operates on pure type/shape
// erased buffers.
const char kMainFunctionName[] = "module.simple_mul";
iree_vm_function_t main_function;
IREE_RETURN_IF_ERROR(iree_vm_context_resolve_function(
context, iree_make_cstring_view(kMainFunctionName), &main_function));
// Initial buffer contents for 4 * 2 = 8.
const float kFloat4[] = {4.0f, 4.0f, 4.0f, 4.0f};
const float kFloat2[] = {2.0f, 2.0f, 2.0f, 2.0f};
// Allocate buffers in device-local memory so that if the device has an
// independent address space they live on the fast side of the fence.
iree_hal_dim_t shape[1] = {IREE_ARRAYSIZE(kFloat4)};
iree_hal_buffer_view_t* arg0_buffer_view = NULL;
iree_hal_buffer_view_t* arg1_buffer_view = NULL;
IREE_RETURN_IF_ERROR(iree_hal_buffer_view_allocate_buffer(
iree_hal_device_allocator(device), shape, IREE_ARRAYSIZE(shape),
IREE_HAL_ELEMENT_TYPE_FLOAT_32, IREE_HAL_ENCODING_TYPE_DENSE_ROW_MAJOR,
(iree_hal_buffer_params_t){
.type = IREE_HAL_MEMORY_TYPE_DEVICE_LOCAL |
IREE_HAL_MEMORY_TYPE_HOST_VISIBLE,
.usage = IREE_HAL_BUFFER_USAGE_DISPATCH |
IREE_HAL_BUFFER_USAGE_TRANSFER |
IREE_HAL_BUFFER_USAGE_MAPPING,
},
iree_make_const_byte_span(kFloat4, sizeof(kFloat4)), &arg0_buffer_view));
IREE_RETURN_IF_ERROR(iree_hal_buffer_view_allocate_buffer(
iree_hal_device_allocator(device), shape, IREE_ARRAYSIZE(shape),
IREE_HAL_ELEMENT_TYPE_FLOAT_32, IREE_HAL_ENCODING_TYPE_DENSE_ROW_MAJOR,
(iree_hal_buffer_params_t){
.type = IREE_HAL_MEMORY_TYPE_DEVICE_LOCAL |
IREE_HAL_MEMORY_TYPE_HOST_VISIBLE,
.usage = IREE_HAL_BUFFER_USAGE_DISPATCH |
IREE_HAL_BUFFER_USAGE_TRANSFER |
IREE_HAL_BUFFER_USAGE_MAPPING,
},
iree_make_const_byte_span(kFloat2, sizeof(kFloat2)), &arg1_buffer_view));
// Setup call inputs with our buffers.
iree_vm_list_t* inputs = NULL;
IREE_RETURN_IF_ERROR(iree_vm_list_create(
/*element_type=*/NULL,
/*capacity=*/2, iree_allocator_system(), &inputs),
"can't allocate input vm list");
iree_vm_ref_t arg0_buffer_view_ref =
iree_hal_buffer_view_move_ref(arg0_buffer_view);
iree_vm_ref_t arg1_buffer_view_ref =
iree_hal_buffer_view_move_ref(arg1_buffer_view);
IREE_RETURN_IF_ERROR(
iree_vm_list_push_ref_move(inputs, &arg0_buffer_view_ref));
IREE_RETURN_IF_ERROR(
iree_vm_list_push_ref_move(inputs, &arg1_buffer_view_ref));
// Prepare outputs list to accept the results from the invocation.
// The output vm list is allocated statically.
iree_vm_list_t* outputs = NULL;
IREE_RETURN_IF_ERROR(iree_vm_list_create(
/*element_type=*/NULL,
/*capacity=*/1, iree_allocator_system(), &outputs),
"can't allocate output vm list");
// Synchronously invoke the function.
IREE_RETURN_IF_ERROR(iree_vm_invoke(
context, main_function, IREE_VM_INVOCATION_FLAG_NONE,
/*policy=*/NULL, inputs, outputs, iree_allocator_system()));
// Get the result buffers from the invocation.
iree_hal_buffer_view_t* ret_buffer_view =
(iree_hal_buffer_view_t*)iree_vm_list_get_ref_deref(
outputs, 0, iree_hal_buffer_view_get_descriptor());
if (ret_buffer_view == NULL) {
return iree_make_status(IREE_STATUS_NOT_FOUND,
"can't find return buffer view");
}
// Read back the results and ensure we got the right values.
float results[] = {0.0f, 0.0f, 0.0f, 0.0f};
IREE_RETURN_IF_ERROR(iree_hal_device_transfer_d2h(
device, iree_hal_buffer_view_buffer(ret_buffer_view), 0, results,
sizeof(results), IREE_HAL_TRANSFER_BUFFER_FLAG_DEFAULT,
iree_infinite_timeout()));
for (iree_host_size_t i = 0; i < IREE_ARRAYSIZE(results); ++i) {
if (results[i] != 8.0f) {
return iree_make_status(IREE_STATUS_UNKNOWN, "result mismatches");
}
}
iree_vm_list_release(inputs);
iree_vm_list_release(outputs);
iree_hal_device_release(device);
iree_vm_context_release(context);
iree_vm_instance_release(instance);
return iree_ok_status();
}
int run_module(char* filename, int index) {
// fread command
//iree_sample_state_t* sample_state = setup_sample();
//iree_program_state_t* program_state = load_program(sample_state, vmfb_data, vmfb_data_length);
//call_function(program_state, function_name, inputs)
const iree_status_t result = Run(filename, index);
int ret = (int)iree_status_code(result);
if (!iree_status_is_ok(result)) {
iree_status_fprint(stderr, result);
iree_status_free(result);
}
fprintf(stdout, "simple_embedding done\n");
return ret;
}

4
dSHARK/dshark_run_module/run_module.h
View File

@@ -1,4 +0,0 @@
#pragma once
#include "run_module.c"
int run_module(char* filename[], int index);

11
dSHARK/dshark_server/CMakeLists.txt
View File

@@ -1,11 +0,0 @@
cmake_minimum_required(VERSION 3.13.4)
project(dshark_server CXX C)
add_executable(dshark_server dshark_server.cpp)
target_include_directories(dshark_server PRIVATE ../dshark_network_lib ../dshark_run_module)
target_link_directories(dshark_server PRIVATE ../dshark_network_lib ../dshark_run_module)
target_link_libraries(dshark_server PRIVATE dshark_network_lib pthread run_module)
set_target_properties(dshark_server PROPERTIES LINKER_LANGUAGE CXX)

103
dSHARK/dshark_server/dshark_server.cpp
View File

@@ -1,103 +0,0 @@
#include "dshark_common.h"
#include "dshark_message.h"
#include "dshark_client.h"
#include "dshark_server.h"
#include "run_module.h"
#include <fstream>
enum class dSHARKMessageType : uint32_t
{
EvaluateBinary,
ServerAccept,
ServerDeny,
ServerPing,
MessageAll,
ServerMessage,
};
class CustomServer : public dshark::server_interface<dSHARKMessageType>
{
public:
CustomServer(uint16_t nPort) : dshark::server_interface<dSHARKMessageType>(nPort)
{
}
protected:
virtual bool OnClientConnect(std::shared_ptr<dshark::connection<dSHARKMessageType>> client)
{
dshark::message<dSHARKMessageType> msg;
msg.header.id = dSHARKMessageType::ServerAccept;
client->Send(msg);
return true;
}
// Called when a client appears to have disconnected
virtual void OnClientDisconnect(std::shared_ptr<dshark::connection<dSHARKMessageType>> client)
{
std::cout << "Removing client [" << client->GetID() << "]\n";
}
// Called when a message arrives
virtual void OnMessage(std::shared_ptr<dshark::connection<dSHARKMessageType>> client, dshark::message<dSHARKMessageType>& msg)
{
switch (msg.header.id)
{
case dSHARKMessageType::EvaluateBinary:
{
std::cout << "[" << client->GetID() << "]: evaluate binary\n";
auto myfile = std::fstream("output/file.vmfb", std::ios::out | std::ios::binary);
int n = msg.body.size();
std::cout << n << "\n";
myfile.write((char*)&msg.body[0], n);
myfile.close();
char* f_ = "output/file.vmfb";
run_module(f_, 0);
client->Send(msg);
}
break;
case dSHARKMessageType::ServerPing:
{
std::cout << "[" << client->GetID() << "]: Server Ping\n";
// Simply bounce message back to client
client->Send(msg);
}
break;
case dSHARKMessageType::MessageAll:
{
std::cout << "[" << client->GetID() << "]: Message All\n";
// Construct a new message and send it to all clients
dshark::message<dSHARKMessageType> msg;
msg.header.id = dSHARKMessageType::ServerMessage;
msg << client->GetID();
MessageAllClients(msg, client);
}
break;
}
}
};
int main()
{
CustomServer server(60000);
server.Start();
while (1)
{
server.Update(-1, true);
}
return 0;
}

BIN
dSHARK/mount-origin/.DS_Store vendored
View File

Binary file not shown.

9
dSHARK/mount-origin/404.html
View File

@@ -1,9 +0,0 @@
<meta charset="UTF-8">
<html>
<body>
<img src="libwebsockets.org-logo.svg"><br>
<h1>404</h1>
Sorry, that file doesn't exist.
</body>
</html>

14
dSHARK/mount-origin/after-form1.html
View File

@@ -1,14 +0,0 @@
<html>
<head>
<meta charset=utf-8 http-equiv="Content-Language" content="en"/>
</head>
<body>
<img src="libwebsockets.org-logo.svg">
<img src="strict-csp.svg"><br>
Thanks for posting the form.<br>
<br>
The file you uploaded should have been saved in the current directory.
</body>
</html>

15
dSHARK/mount-origin/index.html
View File

@@ -1,15 +0,0 @@
<html>
<head>
<meta charset=utf-8 http-equiv="Content-Language" content="en"/>
</head>
<body>
<form name=multipart action="/form1" method="post" enctype="multipart/form-data">
Select a file to upload:
<input type="file" name="file" id="file" size="20">&nbsp;
<span id=file_info style="font-size:12pt;"></span><br>
<br>
<input type="submit" name="send" value="Submit">
</form>
</body>
</html>

5
dSHARK/simple_embedding_test.mlir
View File

@@ -1,5 +0,0 @@
func @simple_mul(%arg0: tensor<4xf32>, %arg1: tensor<4xf32>) -> tensor<4xf32>
{
%0 = "mhlo.multiply"(%arg0, %arg1) {name = "mul.1"} : (tensor<4xf32>, tensor<4xf32>) -> tensor<4xf32>
return %0 : tensor<4xf32>
}

47
dSHARK/simple_embedding_test_bytecode_module_cuda_c.c
View File

@@ -1,47 +0,0 @@
#include <stddef.h>
#include <stdint.h>
#if !defined(IREE_DATA_ALIGNAS_PTR)
#if defined(_MSC_VER)
#define IREE_DATA_ALIGNAS_PTR __declspec(align(8))
#else
#include <stdalign.h>
#define IREE_DATA_ALIGNAS_PTR alignas(alignof(void*))
#endif // _MSC_VER
#endif // !IREE_DATA_ALIGNAS_PTR
#ifndef IREE_FILE_TOC
#define IREE_FILE_TOC
#if __cplusplus
extern "C" {
#endif // __cplusplus
typedef struct iree_file_toc_t {
const char* name; // the file's original name
const char* data; // beginning of the file
size_t size; // length of the file
} iree_file_toc_t;
#if __cplusplus
}
#endif // __cplusplus
#endif // IREE_FILE_TOC
IREE_DATA_ALIGNAS_PTR static uint8_t const file_0[] = {
8,0,0,0,73,82,69,69,186,235,255,255,36,0,0,0,204,2,0,0,244,2,0,0,60,3,0,0,180,2,0,0,64,3,0,0,128,11,0,0,152,11,0,0,16,0,0,0,6,0,0,0,109,111,100,117,108,101,0,0,0,236,255,255,12,0,0,0,16,0,0,0,32,0,0,0,4,0,0,0,2,2,1,0,4,0,0,0,44,2,0,0,24,2,0,0,12,2,0,0,0,0,0,0,2,0,0,0,204,0,0,0,4,0,0,0,90,236,255,255,4,0,0,0,23,0,0,0,0,0,0,0,2,0,0,0,7,0,0,0,2,0,0,0,10,0,0,0,2,0,0,0,17,0,0,0,2,0,0,0,24,0,0,0,2,0,0,0,27,0,0,0,2,0,0,0,38,0,0,0,2,0,0,0,45,0,0,0,2,0,0,0,52,0,0,0,2,0,0,0,72,0,0,0,2,0,0,0,79,0,0,0,2,0,0,0,114,0,0,0,2,0,0,0,123,0,0,0,2,0,0,0,130,0,0,0,2,0,0,0,156,0,0,0,2,0,0,0,167,0,0,0,2,0,0,0,182,0,0,0,2,0,0,0,193,0,0,0,2,0,0,0,200,0,0,0,2,0,0,0,230,0,0,0,2,0,0,0,238,0,0,0,2,0,0,0,246,0,0,0,2,0,0,0,1,1,0,0,3,0,0,0,30,237,255,255,4,0,0,0,37,0,0,0,0,0,0,0,2,0,0,0,7,0,0,0,1,0,0,0,14,0,0,0,2,0,0,0,21,0,0,0,1,0,0,0,28,0,0,0,1,0,0,0,35,0,0,0,2,0,0,0,42,0,0,0,2,0,0,0,49,0,0,0,2,0,0,0,56,0,0,0,2,0,0,0,63,0,0,0,2,0,0,0,70,0,0,0,2,0,0,0,73,0,0,0,2,0,0,0,80,0,0,0,2,0,0,0,87,0,0,0,2,0,0,0,94,0,0,0,2,0,0,0,105,0,0,0,2,0,0,0,116,0,0,0,1,0,0,0,123,0,0,0,1,0,0,0,154,0,0,0,1,0,0,0,166,0,0,0,1,0,0,0,180,0,0,0,1,0,0,0,194,0,0,0,1,0,0,0,216,0,0,0,1,0,0,0,248,0,0,0,1,0,0,0,6,1,0,0,1,0,0,0,28,1,0,0,2,0,0,0,48,1,0,0,2,0,0,0,66,1,0,0,2,0,0,0,78,1,0,0,2,0,0,0,94,1,0,0,2,0,0,0,144,1,0,0,2,0,0,0,166,1,0,0,2,0,0,0,184,1,0,0,2,0,0,0,196,1,0,0,2,0,0,0,210,1,0,0,1,0,0,0,240,1,0,0,1,0,0,0,246,1,0,0,2,0,0,0,38,238,255,255,1,0,0,0,56,238,255,255,28,0,0,0,1,0,0,0,1,0,0,0,72,238,255,255,12,0,0,0,3,0,0,0,8,0,0,0,52,0,0,0,47,104,111,109,101,47,101,108,105,97,115,47,100,83,72,65,82,75,47,100,83,72,65,82,75,47,115,105,109,112,108,101,95,101,109,98,101,100,100,105,110,103,95,116,101,115,116,46,109,108,105,114,0,0,0,0,166,238,255,255,4,0,0,0,2,0,0,0,10,0,0,0,148,8,0,0,116,8,0,0,88,8,0,0,52,8,0,0,16,8,0,0,228,7,0,0,200,7,0,0,168,7,0,0,128,7,0,0,100,7,0,0,18,0,0,0,20,7,0,0,196,6,0,0,96,6,0,0,0,6,0,0,152,5,0,0,56,5,0,0,224,4,0,0,124,4,0,0,44,4,0,0,224,3,0,0,120,3,0,0,8,3,0,0,160,2,0,0,48,2,0,0,216,1,0,0,116,1,0,0,12,1,0,0,160,0,0,0,2,0,0,0,76,0,0,0,24,0,0,0,4,0,0,0,56,8,0,0,44,8,0,0,32,8,0,0,20,8,0,0,56,239,255,255,32,0,0,0,8,0,0,0,1,0,0,0,116,239,255,255,4,0,0,0,4,0,0,0,48,118,95,118,0,0,0,0,6,0,0,0,95,95,105,110,105,116,0,0,122,239,255,255,56,0,0,0,4,0,0,0,116,239,255,255,24,0,0,0,32,0,0,0,4,0,0,0,5,0,0,0,48,114,114,95,114,0,0,0,2,0,0,0,3,0,0,0,3,0,0,0,1,0,0,0,3,0,0,0,10,0,0,0,115,105,109,112,108,101,95,109,117,108,0,0,198,239,255,255,64,0,0,0,4,0,0,0,192,239,255,255,
28,0,0,0,40,0,0,0,4,0,0,0,8,0,0,0,48,114,105,67,114,68,95,114,0,0,0,0,3,0,0,0,6,0,0,0,0,0,0,0,5,0,0,0,1,0,0,0,8,0,0,0,28,0,0,0,104,97,108,46,101,120,101,99,117,116,97,98,108,101,95,108,97,121,111,117,116,46,99,114,101,97,116,101,0,0,0,0,46,240,255,255,68,0,0,0,4,0,0,0,40,240,255,255,28,0,0,0,44,0,0,0,4,0,0,0,9,0,0,0,48,114,114,114,67,114,68,95,114,0,0,0,4,0,0,0,6,0,0,0,9,0,0,0,9,0,0,0,8,0,0,0,1,0,0,0,7,0,0,0,21,0,0,0,104,97,108,46,101,120,101,99,117,116,97,98,108,101,46,99,114,101,97,116,101,0,0,0,146,240,255,255,64,0,0,0,4,0,0,0,140,240,255,255,24,0,0,0,36,0,0,0,4,0,0,0,7,0,0,0,48,114,114,114,95,105,105,0,3,0,0,0,6,0,0,0,9,0,0,0,9,0,0,0,2,0,0,0,0,0,0,0,0,0,0,0,20,0,0,0,104,97,108,46,100,101,118,105,99,101,46,113,117,101,114,121,46,105,51,50,0,0,0,0,242,240,255,255,52,0,0,0,4,0,0,0,236,240,255,255,24,0,0,0,28,0,0,0,4,0,0,0,4,0,0,0,48,114,95,114,0,0,0,0,1,0,0,0,6,0,0,0,1,0,0,0,1,0,0,0,20,0,0,0,104,97,108,46,100,101,118,105,99,101,46,97,108,108,111,99,97,116,111,114,0,0,0,0,70,241,255,255,64,0,0,0,4,0,0,0,64,241,255,255,28,0,0,0,40,0,0,0,4,0,0,0,9,0,0,0,48,114,105,67,105,105,68,95,114,0,0,0,3,0,0,0,6,0,0,0,0,0,0,0,0,0,0,0,1,0,0,0,5,0,0,0,32,0,0,0,104,97,108,46,100,101,115,99,114,105,112,116,111,114,95,115,101,116,95,108,97,121,111,117,116,46,99,114,101,97,116,101,0,0,0,0,178,241,255,255,64,0,0,0,4,0,0,0,182,241,255,255,24,0,0,0,4,0,0,0,9,0,0,0,48,114,114,105,105,105,105,95,118,0,0,0,6,0,0,0,4,0,0,0,7,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,27,0,0,0,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,46,100,105,115,112,97,116,99,104,0,22,242,255,255,60,0,0,0,4,0,0,0,26,242,255,255,28,0,0,0,4,0,0,0,12,0,0,0,48,114,114,105,67,105,114,105,105,68,95,118,0,0,0,0,4,0,0,0,4,0,0,0,8,0,0,0,0,0,0,0,0,0,0,0,38,0,0,0,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,46,112,117,115,104,95,100,101,115,99,114,105,112,116,111,114,95,115,101,116,0,0,130,242,255,255,52,0,0,0,4,0,0,0,134,242,255,255,20,0,0,0,4,0,0,0,7,0,0,0,48,114,105,105,105,95,118,0,4,0,0,0,4,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,36,0,0,0,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,46,101,120,101,99,117,116,105,111,110,95,98,97,114,114,105,101,114,0,0,0,0,230,242,255,255,40,0,0,0,4,0,0,0,234,242,255,255,20,0,0,0,4,0,0,0,4,0,0,0,48,114,95,118,0,0,0,0,1,0,0,0,4,0,0,0,22,0,0,0,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,46,101,110,100,0,0,46,243,255,255,40,0,0,0,4,0,0,0,50,243,255,255,20,0,0,0,4,0,0,0,4,0,0,0,48,114,95,118,0,0,0,0,1,0,0,0,4,0,0,0,24,0,0,0,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,46,98,101,103,105,110,0,0,0,0,122,243,255,255,60,0,0,0,4,0,0,0,116,243,255,255,24,0,0,0,36,0,0,0,4,0,0,0,6,0,0,0,48,114,105,105,95,114,0,0,3,0,0,0,6,0,0,0,0,0,0,0,0,0,0,0,1,0,0,0,4,0,0,0,25,0,0,0,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,46,99,114,101,97,116,
101,0,0,0,218,243,255,255,52,0,0,0,4,0,0,0,212,243,255,255,24,0,0,0,28,0,0,0,4,0,0,0,4,0,0,0,48,114,95,114,0,0,0,0,1,0,0,0,3,0,0,0,1,0,0,0,2,0,0,0,22,0,0,0,104,97,108,46,98,117,102,102,101,114,95,118,105,101,119,46,98,117,102,102,101,114,0,0,46,244,255,255,60,0,0,0,4,0,0,0,50,244,255,255,24,0,0,0,4,0,0,0,10,0,0,0,48,114,114,105,105,67,105,68,95,118,0,0,5,0,0,0,3,0,0,0,9,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,22,0,0,0,104,97,108,46,98,117,102,102,101,114,95,118,105,101,119,46,97,115,115,101,114,116,0,0,138,244,255,255,68,0,0,0,4,0,0,0,132,244,255,255,28,0,0,0,44,0,0,0,4,0,0,0,9,0,0,0,48,114,105,105,67,105,68,95,114,0,0,0,4,0,0,0,2,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,1,0,0,0,3,0,0,0,22,0,0,0,104,97,108,46,98,117,102,102,101,114,95,118,105,101,119,46,99,114,101,97,116,101,0,0,238,244,255,255,64,0,0,0,4,0,0,0,242,244,255,255,24,0,0,0,4,0,0,0,9,0,0,0,48,114,114,114,105,105,105,95,118,0,0,0,6,0,0,0,2,0,0,0,9,0,0,0,1,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,17,0,0,0,104,97,108,46,98,117,102,102,101,114,46,97,115,115,101,114,116,0,0,0,74,245,255,255,64,0,0,0,4,0,0,0,68,245,255,255,24,0,0,0,40,0,0,0,4,0,0,0,7,0,0,0,48,114,105,105,105,95,114,0,4,0,0,0,1,0,0,0,0,0,0,0,0,0,0,0,0,0,0,0,1,0,0,0,2,0,0,0,22,0,0,0,104,97,108,46,97,108,108,111,99,97,116,111,114,46,97,108,108,111,99,97,116,101,0,0,170,245,255,255,44,0,0,0,4,0,0,0,174,245,255,255,20,0,0,0,4,0,0,0,5,0,0,0,48,114,114,95,118,0,0,0,2,0,0,0,6,0,0,0,4,0,0,0,22,0,0,0,104,97,108,46,101,120,46,115,117,98,109,105,116,95,97,110,100,95,119,97,105,116,0,0,246,245,255,255,40,0,0,0,4,0,0,0,12,246,255,255,20,0,0,0,4,0,0,0,4,0,0,0,48,118,95,114,0,0,0,0,1,0,0,0,6,0,0,0,20,0,0,0,104,97,108,46,101,120,46,115,104,97,114,101,100,95,100,101,118,105,99,101,0,0,0,0,78,246,255,255,4,0,0,0,10,0,0,0,33,118,109,46,98,117,102,102,101,114,0,0,102,246,255,255,4,0,0,0,22,0,0,0,33,104,97,108,46,101,120,101,99,117,116,97,98,108,101,95,108,97,121,111,117,116,0,0,138,246,255,255,4,0,0,0,15,0,0,0,33,104,97,108,46,101,120,101,99,117,116,97,98,108,101,0,166,246,255,255,4,0,0,0,11,0,0,0,33,104,97,108,46,100,101,118,105,99,101,0,190,246,255,255,4,0,0,0,26,0,0,0,33,104,97,108,46,100,101,115,99,114,105,112,116,111,114,95,115,101,116,95,108,97,121,111,117,116,0,0,230,246,255,255,4,0,0,0,19,0,0,0,33,104,97,108,46,99,111,109,109,97,110,100,95,98,117,102,102,101,114,0,6,247,255,255,4,0,0,0,16,0,0,0,33,104,97,108,46,98,117,102,102,101,114,95,118,105,101,119,0,0,0,0,38,247,255,255,4,0,0,0,11,0,0,0,33,104,97,108,46,98,117,102,102,101,114,0,62,247,255,255,4,0,0,0,14,0,0,0,33,104,97,108,46,97,108,108,111,99,97,116,111,114,0,0,90,247,255,255,4,0,0,0,3,0,0,0,105,51,50,0,116,247,255,255,124,8,0,0,124,247,255,255,96,8,0,0,132,247,255,255,60,8,0,0,140,247,255,255,92,3,0,0,2,0,0,0,0,0,0,0,48,2,0,0,14,0,7,0,48,2,0,0,8,1,0,0,6,0,4,0,56,3,0,0,9,4,0,0,0,0,0,9,32,0,0,33,1,0,9,1,0,0,0,2,0,9,16,0,0,0,3,0,9,10,0,0,0,4,0,9,50,0,0,0,5,0,9,14,0,0,0,6,0,9,17,0,0,0,7,0,9,3,0,0,
0,8,0,9,2,0,0,0,9,0,8,10,0,9,28,0,0,0,11,0,9,13,0,0,0,12,0,0,0,0,0,0,13,0,4,0,0,0,0,8,0,0,0,2,128,4,1,0,0,0,7,0,0,0,3,128,11,3,0,0,0,4,128,83,5,0,0,128,5,0,255,255,255,255,255,255,255,255,1,0,5,0,0,128,4,128,1,0,2,0,0,0,0,0,82,0,0,0,128,0,0,0,1,0,5,128,82,6,0,0,128,0,1,0,0,192,1,0,0,128,82,14,0,0,128,0,1,0,5,128,1,0,6,128,82,3,0,0,128,0,6,0,0,128,4,128,6,128,3,0,3,0,4,0,0,0,83,5,0,0,128,0,5,0,255,255,255,255,255,255,255,255,1,0,5,0,1,128,4,128,1,0,2,0,0,0,0,0,82,6,0,0,128,0,1,0,1,192,1,0,1,128,82,3,0,0,128,0,6,0,1,128,4,192,6,128,3,0,3,0,4,0,0,0,82,2,0,0,128,0,4,0,6,192,5,0,6,0,3,0,1,0,4,128,82,7,0,0,128,0,3,0,5,128,7,0,8,0,1,0,6,128,82,8,0,0,128,0,1,0,6,128,0,0,81,13,0,94,1,0,0,0,0,0,246,1,0,0,0,0,83,11,0,0,128,0,4,0,255,255,255,255,255,255,3,0,15,0,6,128,2,192,10,0,10,0,0,192,10,0,3,0,2,0,1,192,10,0,3,0,9,0,4,128,10,0,3,0,0,0,82,12,0,0,128,0,6,0,6,128,3,192,10,0,2,0,2,0,2,0,0,0,82,10,0,0,128,0,4,0,6,128,11,0,12,0,10,0,0,0,82,9,0,0,128,0,1,0,6,128,0,0,82,1,0,0,128,0,2,0,5,192,6,192,0,0,83,4,0,0,128,0,4,0,255,255,255,255,255,255,1,0,4,0,4,192,1,0,2,0,0,0,1,0,0,128,84,0,1,0,0,192,85,9,0,50,0,100,101,118,105,99,101,32,110,111,116,32,115,117,112,112,111,114,116,101,100,32,105,110,32,116,104,101,32,99,111,109,112,105,108,101,100,32,99,111,110,102,105,103,117,114,97,116,105,111,110,0,0,0,9,1,0,0,0,0,0,8,1,0,9,2,0,0,0,2,0,9,7,0,0,0,3,0,10,0,128,82,0,0,0,128,0,0,1,0,1,128,11,1,0,0,0,2,128,11,2,0,0,0,3,128,82,15,0,0,128,0,3,0,1,128,2,192,3,128,2,0,4,0,5,0,49,5,0,0,0,5,0,83,13,0,0,128,3,0,255,255,255,255,3,0,8,0,1,128,0,0,1,0,3,0,0,0,3,0,2,0,3,0,1,0,2,128,30,4,0,5,0,1,0,0,0,1,0,0,0,0,0,0,83,17,0,0,128,0,3,0,255,255,255,255,1,0,3,0,1,128,1,0,2,192,1,0,2,128,5,0,0,0,0,8,0,0,0,2,192,81,0,0,182,0,0,0,0,0,238,0,0,0,0,0,4,0,0,0,0,8,0,0,0,0,128,11,0,0,0,0,2,128,83,16,0,0,128,0,4,0,255,255,255,255,255,255,1,0,4,0,1,192,3,192,2,192,0,192,1,0,0,128,80,246,0,0,0,0,0,0,80,246,0,0,0,0,0,0,5,1,0,0,0,7,0,0,0,0,192,84,0,0,0,0,0,0,204,4,0,0,8,0,0,0,67,85,68,65,72,251,255,255,40,0,0,0,28,0,0,0,8,0,0,0,64,0,0,0,1,0,0,0,64,0,0,0,1,0,0,0,1,0,0,0,1,0,0,0,0,0,0,0,1,0,0,0,4,0,0,0,21,0,0,0,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,0,0,0,98,4,0,0,47,47,10,47,47,32,71,101,110,101,114,97,116,101,100,32,98,121,32,76,76,86,77,32,78,86,80,84,88,32,66,97,99,107,45,69,110,100,10,47,47,10,10,46,118,101,114,115,105,111,110,32,54,46,48,10,46,116,97,114,103,101,116,32,115,109,95,51,53,10,46,97,100,100,114,101,115,115,95,115,105,122,101,32,54,52,10,10,9,47,47,32,46,103,108,111,98,108,9,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,10,10,46,118,105,115,105,98,108,101,32,46,101,110,116,114,121,32,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,40,10,9,46,112,
97,114,97,109,32,46,117,54,52,32,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,95,112,97,114,97,109,95,48,44,10,9,46,112,97,114,97,109,32,46,117,54,52,32,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,95,112,97,114,97,109,95,49,44,10,9,46,112,97,114,97,109,32,46,117,54,52,32,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,95,112,97,114,97,109,95,50,10,41,10,123,10,9,46,114,101,103,32,46,112,114,101,100,32,9,37,112,60,50,62,59,10,9,46,114,101,103,32,46,98,51,50,32,9,37,114,60,51,62,59,10,9,46,114,101,103,32,46,102,51,50,32,9,37,102,60,52,62,59,10,9,46,114,101,103,32,46,98,54,52,32,9,37,114,100,60,49,54,62,59,10,10,9,109,111,118,46,117,51,50,32,9,37,114,49,44,32,37,99,116,97,105,100,46,120,59,10,9,109,117,108,46,119,105,100,101,46,117,51,50,32,9,37,114,100,49,48,44,32,37,114,49,44,32,54,52,59,10,9,109,111,118,46,117,54,52,32,9,37,114,100,49,49,44,32,52,59,10,9,115,117,98,46,115,54,52,32,9,37,114,100,49,50,44,32,37,114,100,49,49,44,32,37,114,100,49,48,59,10,9,109,111,118,46,117,51,50,32,9,37,114,50,44,32,37,116,105,100,46,120,59,10,9,99,118,116,46,117,54,52,46,117,51,50,32,9,37,114,100,49,51,44,32,37,114,50,59,10,9,115,101,116,112,46,108,101,46,115,54,52,32,9,37,112,49,44,32,37,114,100,49,50,44,32,37,114,100,49,51,59,10,9,64,37,112,49,32,98,114,97,32,9,76,66,66,48,95,50,59,10,9,108,100,46,112,97,114,97,109,46,117,54,52,32,9,37,114,100,52,44,32,91,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,95,112,97,114,97,109,95,48,93,59,10,9,108,100,46,112,97,114,97,109,46,117,54,52,32,9,37,114,100,53,44,32,91,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,95,112,97,114,97,109,95,50,93,59,10,9,99,118,116,97,46,116,111,46,103,108,111,98,97,108,46,117,54,52,32,9,37,114,100,54,44,32,37,114,100,53,59,10,9,108,100,46,112,97,114,97,109,46,117,54,52,32,9,37,114,100,55,44,32,91,115,105,109,112,108,101,95,109,117,108,95,100,105,115,112,97,116,99,104,95,48,95,112,97,114,97,109,95,49,93,59,10,9,99,118,116,97,46,116,111,46,103,108,111,98,97,108,46,117,54,52,32,9,37,114,100,56,44,32,37,114,100,55,59,10,9,99,118,116,97,46,116,111,46,103,108,111,98,97,108,46,117,54,52,32,9,37,114,100,57,44,32,37,114,100,52,59,10,9,97,100,100,46,115,54,52,32,9,37,114,100,49,52,44,32,37,114,100,49,48,44,32,37,114,100,49,51,59,10,9,115,104,108,46,98,54,52,32,9,37,114,100,49,53,44,32,37,114,100,49,52,44,32,50,59,10,9,97,100,100,46,115,54,52,32,9,37,114,100,49,44,32,37,114,100,57,44,32,37,114,100,49,53,59,10,9,97,100,100,46,115,54,52,32,9,37,114,100,50,44,32,37,114,100,56,44,32,37,114,100,49,53,59,10,9,97,100,100,46,115,54,52,32,9,37,114,100,51,44,32,37,114,100,54,44,32,37,114,100,49,53,59,10,9,108,100,46,103,108,111,98,97,108,46,102,51,50,32,9,37,102,49,44,32,91,37,114,100,49,93,59,10,9,108,100,46,103,108,111,98,97,108,46,102,51,50,32,9,37,102,50,44,32,91,37,114,100,50,93,59,10,9,109,117,108,46,114,110,46,102,51,50,32,9,37,102,51,44,32,37,102,49,44,32,37,102,50,59,10,9,115,116,46,103,108,111,98,97,108,46,102,51,50,32,9,91,37,114,100,51,93,44,32,37,102,51,59,10,76,66,66,48,95,50,58,10,9,114,101,116,59,10,10,125,10,0,0,12,0,20,0,4,0,12,0,8,0,16,0,0,0,0,0,0,0,0,0,21,0,0,0,104,97,108,46,101,120,101,99,117,116,97,98,108,101,46,102,111,114,109,97,116,0,0,0,13,0,0,0,99,117,100,97,45,110,118,112,116,120,45,102,
98,0,0,0,6,0,0,0,116,101,110,115,111,114,0,0,10,0,8,0,0,0,0,0,4,0,6,0,8,0,4,0,10,0,12,0,0,0,4,0,8,0,8,0,12,0,4,0,8,0,10,0,12,0,4,0,0,0,8,0,10,0,16,0,4,0,8,0,12,0,8,0,8,0,0,0,4,0,24,0,40,0,4,0,8,0,12,0,16,0,24,0,0,0,20,0,28,0,32,0,36,0,
0,
};
static const struct iree_file_toc_t toc[] = {
{
"simple_embedding_test_bytecode_module_cuda_c.vmfb",
(const char*)file_0,
sizeof(file_0) - 1
},
{NULL, NULL, 0},
};
const struct iree_file_toc_t* simple_embedding_test_bytecode_module_cuda_c_create() {
return &toc[0];
}

30
dSHARK/simple_embedding_test_bytecode_module_cuda_c.h
View File

@@ -1,30 +0,0 @@
#pragma once
#include <stddef.h>
#ifndef IREE_FILE_TOC
#define IREE_FILE_TOC
#if __cplusplus
extern "C" {
#endif // __cplusplus
typedef struct iree_file_toc_t {
const char* name; // the file's original name
const char* data; // beginning of the file
size_t size; // length of the file
} iree_file_toc_t;
#if __cplusplus
}
#endif // __cplusplus
#endif // IREE_FILE_TOC
#if __cplusplus
extern "C" {
#endif // __cplusplus
const iree_file_toc_t* simple_embedding_test_bytecode_module_cuda_c_create();
static inline size_t simple_embedding_test_bytecode_module_cuda_c_size() {
return 1;
}
#if __cplusplus
}
#endif // __cplusplus

BIN
dSHARK/simple_embedding_test_bytecode_module_cuda_c.vmfb
View File

Binary file not shown.