T* pinnedAlloc(const size_t &elements)
{
managerInit();
T* ptr = NULL;
// Allocate the higher megabyte. Overhead of creating pinned memory is
// more so we want more resuable memory.
size_t alloc_bytes = divup(sizeof(T) * elements, 1048576) * 1048576;
if (elements > 0) {
// FIXME: Add better checks for garbage collection
// Perhaps look at total memory available as a metric
if (pinned_maps.size() >= MAX_BUFFERS || pinned_used_bytes >= MAX_BYTES) {
pinnedGarbageCollect();
}
for(mem_iter iter = pinned_maps.begin();
iter != pinned_maps.end(); ++iter) {
mem_info info = iter->second;
if (info.is_free && info.bytes == alloc_bytes) {
iter->second.is_free = false;
pinned_used_bytes += alloc_bytes;
return (T *)iter->first;
}
}
// Perform garbage collection if memory can not be allocated
if (cudaMallocHost((void **)&ptr, alloc_bytes) != cudaSuccess) {
pinnedGarbageCollect();
CUDA_CHECK(cudaMallocHost((void **)(&ptr), alloc_bytes));
}
mem_info info = {false, false, alloc_bytes};
pinned_maps[ptr] = info;
pinned_used_bytes += alloc_bytes;
}
return (T*)ptr;
}
bool op_ring::execute()
{
if( pos != 0 ) switch( pos )
{
// One
case 2:
value=1;
break;
// Zero
case 3:
value=0;
break;
// Load
case 4:
value=*mem_pos;
break;
// Store
case 5:
*mem_pos = value;
break;
// DAdd
case 7:
{
int t = mem_pos - memory.begin();
t += value;
if( t < 0 )
return false; // out of bounds!
if( t < memory.size() )
mem_pos = memory.begin() + t;
else
{
while( memory.size() <= t )
memory.push_back(0);
mem_pos = memory.begin() + t;
}
}
break;
// Logic
case 8:
if( *mem_pos != 0 )
value = value && 1;
else
value = 0;
break;
// If
case 9:
if( *mem_pos == 0 )
break;
// else fall through!
// PAdd
case 6:
{
int t = prog_pos - program.begin();
t += value;
if( t < 0 || t >= program.size() )
return false; // out of bounds!
prog_pos = program.begin() + t;
next_instruction = false;
}
break;
// IntIO
case 10:
if( value == 0 )
{
// read integer
char buf[100];
int c = 0;
while( c < sizeof(buf)-1 )
{
buf[c] = getchar();
c++;
buf[c] = 0;
if( buf[c-1] == '\n' )
break;
}
// swallow, just in case.
if( c == sizeof(buf) )
while( getchar() != '\n' );
(*mem_pos) = atoi( buf );
}
else
printf( "%d", *mem_pos );
break;
// AscIO
case 11:
if( value == 0 )
{
// read character
(*mem_pos) = getchar();
while( getchar() != '\n' );
}
else
printf( "%c", *mem_pos );
break;
// Exit
case 1:
default:
return false;
}
return true;
}